CUDA 编程简介

CUDA（Compute Unified Device Architecture）是由 NVIDIA 提供的一种并行计算平台和编程模型。它允许开发者利用 NVIDIA GPU 的并行计算能力，编写可以在 GPU 上高效运行的代码，从而加速计算密集型任务。

CUDA 通过扩展标准的 C/C++ 语言，提供了编译工具链和库，使程序员可以轻松编写并行代码。

---

1. CUDA 的核心概念

GPU 和 CPU 的协同计算

• CPU：负责运行主程序（主机代码，Host Code），将任务分配给 GPU。
• GPU：负责运行数据并行的任务（设备代码，Device Code），执行大量并行计算。

---

CUDA 核函数 (Kernel)

• Kernel 是 CUDA 程序中的核心函数，它运行在 GPU 上的所有线程中。
• 通过 __global__ 关键字定义：

  __global__ void myKernel(int *a) {
      int i = threadIdx.x; // 当前线程索引
      a[i] = a[i] * 2;     // 并行计算
  }
  

---

线程、线程块和网格 (Grid)

CUDA 提供了一种分层的 线程模型，用于在 GPU 上组织并行计算：

• 线程 (Thread)：执行计算的基本单元。
• 线程块 (Thread Block)：由一组线程组成，可以共享数据。
• 网格 (Grid)：由多个线程块组成。

线程的组织结构可以用一维、二维或三维方式定义：

dim3 blockSize(256);       // 每个线程块 256 个线程
dim3 gridSize((N + 255)/256); // 网格大小 (任务总数 / 线程数)
myKernel<<<gridSize, blockSize>>>(d_data);

---

内存模型

CUDA 提供多种内存空间，分为 主机内存（CPU） 和 设备内存（GPU）：

1. 全局内存（Global Memory）：GPU 所有线程都可以访问，访问速度慢但容量大。
2. 共享内存（Shared Memory）：线程块内的线程共享的数据，速度快但容量小。
3. 寄存器（Registers）：线程私有的内存，速度最快。
4. 常量内存（Constant Memory）：只读内存，所有线程共享。
5. 纹理内存 / 表面内存：用于特殊优化的数据访问场景。

---

2. CUDA 程序结构

一个典型的 CUDA 程序分为以下几个步骤：

1. 主机代码初始化（CPU）：

   • 分配和初始化内存。

2. 从CPU内存中拷贝数据到GPU内存

   • 将数据传输到 GPU（cudaMemcpy）。

3. 内核函数执行（GPU）：

   • 调用核函数（<<<gridSize, blockSize>>>），GPU 执行并行任务。

4. 结果传回主机：

   • 将 GPU 上的数据传回 CPU。

5. 释放资源：

   • 释放主机和设备内存。

---

3. 简单 CUDA 示例

3.1 helloworld示例

#include<stdio.h>
__global__ void helloWorldFromGPU(){
    printf("hello world from GPU!\n\n");
}

int main(int argc, char const *argv[])
{
    helloWorldFromGPU<<<1,10>>>();
    return 0;
}

• 执行：

nvcc helloworld.cu -o main
main

cudaDeviceReset函数

cudaDeviceReset 是一个用于清理 CUDA 设备状态的函数，它会释放与当前 CUDA 上下文（Context）相关联的所有资源，并将设备状态重置为初始状态。如果程序不调用它，CUDA 运行时可能会保留一些状态或资源，直到程序完全结束。

1. 调用 cudaDeviceReset 的情况：

• 释放 GPU 上的资源：cudaDeviceReset 会确保设备上的内存和上下文被完全释放，设备状态恢复为初始状态。
• 干净退出：CUDA 上下文释放后，程序会正常退出，不会留有残余资源。

2. 移除 cudaDeviceReset 的情况：

• 资源未完全释放：在程序结束时，GPU 上的资源会保留，直到操作系统自动回收。
• 影响多次执行的程序：如果后续还有其他 CUDA 程序，设备资源状态可能会受到前一个程序的影响。
• 调试环境行为：在调试工具（如 cuda-memcheck）中，未调用 cudaDeviceReset 可能会产生警告，指出设备资源未被显式释放。

对于 小型程序，两种情况下的运行结果几乎没有差异，程序可以正常执行。但在以下情况下，差异会更加明显：

• 长时间运行的程序：不调用 cudaDeviceReset 可能导致设备资源持续占用。
• 反复调用 CUDA 程序：如果 GPU 设备状态未重置，后续程序可能会遇到资源分配失败或初始化冲突。
• 工具检测：使用 CUDA 调试工具（如 cuda-memcheck）时，不调用 cudaDeviceReset 可能会报告内存泄漏。

#pragma execution_character_set("utf-8")
#include <cuda_runtime.h>
#include <iostream>

__global__ void vectorAdd(int *a, int *b, int *c, int n) {
    int idx = threadIdx.x + blockIdx.x * blockDim.x;
    if (idx < n) {
        c[idx] = a[idx] + b[idx];
    }
}

int main() {
    const int N = 1000000; // 数据规模
    size_t size = N * sizeof(int);

    // 分配主机内存
    int *h_a = (int *)malloc(size);
    int *h_b = (int *)malloc(size);
    int *h_c = (int *)malloc(size);

    // 初始化数据
    for (int i = 0; i < N; i++) {
        h_a[i] = i;
        h_b[i] = 2 * i;
    }

    // 分配设备内存
    int *d_a, *d_b, *d_c;
    cudaMalloc(&d_a, size);
    cudaMalloc(&d_b, size);
    cudaMalloc(&d_c, size);

    // 将数据从主机传输到设备
    cudaMemcpy(d_a, h_a, size, cudaMemcpyHostToDevice);
    cudaMemcpy(d_b, h_b, size, cudaMemcpyHostToDevice);

    // 启动核函数
    int blockSize = 256;
    int gridSize = (N + blockSize - 1) / blockSize;
    vectorAdd<<<gridSize, blockSize>>>(d_a, d_b, d_c, N);

    // 将结果传回主机
    cudaMemcpy(h_c, d_c, size, cudaMemcpyDeviceToHost);

    // 验证结果
    for (int i = 0; i < 10; i++) {
        std::cout << h_c[i] << " ";
    }
    std::cout << std::endl;

    // 释放资源
    cudaFree(d_a);
    cudaFree(d_b);
    cudaFree(d_c);
    free(h_a);
    free(h_b);
    free(h_c);

    // cudaDeviceReset 调用 (可以移除)
    cudaDeviceReset();

    return 0;
}

---

cudaDeviceSynchronize函数

cudaDeviceSynchronize 与 cudaDeviceReset 的区别

1. cudaDeviceSynchronize：

   • 作用：确保所有之前提交到 GPU 的任务（如核函数、数据传输）都执行完成，起到 同步 的作用。
   • 执行后：程序继续执行后续的代码，但不会清理 GPU 上的资源。

2. cudaDeviceReset：

   • 作用：除了同步任务外，还会将 GPU 设备状态重置并释放所有资源（如设备内存、上下文等）。
   • 执行后：GPU 状态重置为初始状态，所有分配的资源会被释放。

---

为什么不能直接用 cudaDeviceSynchronize 替换 cudaDeviceReset？

• cudaDeviceSynchronize 仅起到同步作用，它不会释放 GPU 资源。
• 如果程序在多次运行 CUDA 任务时，没有调用 cudaDeviceReset，GPU 上的资源可能无法及时释放，导致 资源泄漏 或 初始化冲突。

---

替换示例

以下是代码中用 cudaDeviceSynchronize 替换 cudaDeviceReset 的示例：

#pragma execution_character_set("utf-8")
#include <cuda_runtime.h>
#include <iostream>

__global__ void vectorAdd(int *a, int *b, int *c, int n) {
    int idx = threadIdx.x + blockIdx.x * blockDim.x;
    if (idx < n) {
        c[idx] = a[idx] + b[idx];
    }
}

int main() {
    const int N = 1000000; // 数据规模
    size_t size = N * sizeof(int);

    // 分配主机内存
    int *h_a = (int *)malloc(size);
    int *h_b = (int *)malloc(size);
    int *h_c = (int *)malloc(size);

    // 初始化数据
    for (int i = 0; i < N; i++) {
        h_a[i] = i;
        h_b[i] = 2 * i;
    }

    // 分配设备内存
    int *d_a, *d_b, *d_c;
    cudaMalloc(&d_a, size);
    cudaMalloc(&d_b, size);
    cudaMalloc(&d_c, size);

    // 将数据从主机传输到设备
    cudaMemcpy(d_a, h_a, size, cudaMemcpyHostToDevice);
    cudaMemcpy(d_b, h_b, size, cudaMemcpyHostToDevice);

    // 启动核函数
    int blockSize = 256;
    int gridSize = (N + blockSize - 1) / blockSize;
    vectorAdd<<<gridSize, blockSize>>>(d_a, d_b, d_c, N);

    // 将结果传回主机
    cudaMemcpy(h_c, d_c, size, cudaMemcpyDeviceToHost);

    // 验证结果
    for (int i = 0; i < 10; i++) {
        std::cout << h_c[i] << " ";
    }
    std::cout << std::endl;

    // 释放资源
    cudaFree(d_a);
    cudaFree(d_b);
    cudaFree(d_c);
    free(h_a);
    free(h_b);
    free(h_c);

    // cudaDeviceReset 调用 (可以移除)
    // cudaDeviceReset();
    cudaDeviceSynchronize();

    return 0;
}

打印线程索引

#include<stdio.h>
__global__ void helloWorldFromGPU(){
    //  需要添加线程索引
    int threadIndex = threadIdx.x;
    printf("hello world from GPU! threadIndex: %d\n\n", threadIndex);
}

int main(int argc, char const *argv[])
{
    helloWorldFromGPU<<<1,10>>>();  // 在GPU上执行helloWorldFromGPU函数
    return 0;
}

---

1. 调用 cudaDeviceReset：

   • GPU 资源会被完全释放。
   • 在多次运行程序时，GPU 设备状态重新初始化。

2. 调用 cudaDeviceSynchronize：

   • 程序会等待所有任务完成，但不会释放 GPU 资源。
   • 如果不手动释放内存（如 cudaFree），资源会在程序结束时才被操作系统回收。

---

• 如果程序结束时需要 释放 GPU 资源，请使用 cudaDeviceReset。
• 如果只需要确保 任务完成 而不重置设备状态，使用 cudaDeviceSynchronize。

---

• cudaDeviceSynchronize：同步 GPU 任务，确保执行完成，但不释放资源。
• cudaDeviceReset：同步任务并释放 GPU 资源，重置设备状态。

两者不能完全互换，具体使用取决于程序对资源管理的需求。

4. CUDA 的应用场景

• 科学计算：例如矩阵乘法、数值模拟等。
• 深度学习：使用 CUDA 加速框架如 TensorFlow 和 PyTorch。
• 图像处理：高性能图像滤波、边缘检测等。
• 金融计算：蒙特卡洛模拟、风险分析等。
• 物理仿真：粒子模拟、流体动力学等。

---

5. CUDA 生态系统

CUDA 提供了丰富的库和工具，方便开发者进行高性能编程：

• cuBLAS：GPU 加速的线性代数库。
• cuDNN：用于深度学习的加速库。
• Thrust：类似于 C++ STL 的并行模板库。
• Nsight：性能分析和调试工具。
• NCCL：用于多 GPU 通信的库。

---

总结

CUDA 是一种强大的并行编程工具，通过利用 NVIDIA GPU 的计算能力，使开发者能够高效地加速各种计算任务。它通过扩展 C/C++ 语言，提供了易于上手的编程模型，广泛应用于科学计算、深度学习和高性能计算领域。