【详细】CUDA开发学习教程清单

以下是一份系统化的CUDA开发学习教程清单，涵盖基础到进阶内容，并附带关键代码示例：

---

📚 一、基础环境搭建（Ubuntu示例）

1. ​​安装CUDA Toolkit​​

# 下载.run文件后执行安装
sudo sh cuda_12.3.2_524.91.01_linux.run
# 配置环境变量
echo 'export PATH=/usr/local/cuda/bin:$PATH' >> ~/.bashrc
echo 'export LD_LIBRARY_PATH=/usr/local/cuda/lib64:$LD_LIBRARY_PATH' >> ~/.bashrc
source ~/.bashrc
# 验证安装
nvcc --version  # 输出版本信息即成功

2. ​​验证GPU支持​​

#include <cuda_runtime.h>
#include <stdio.h>
int main() {
    int deviceCount;
    cudaGetDeviceCount(&deviceCount);
    printf("可用GPU数量: %d\n", deviceCount);
    return 0;
}
// 编译: nvcc device_query.cu -o device_query

---

⚙️ 二、核心编程模型

1. ​​Hello GPU（首个核函数）​​

#include <stdio.h>
__global__ void helloGPU() {
    printf("Hello from GPU thread %d!\n", threadIdx.x);
}
int main() {
    helloGPU<<<1, 5>>>();  // 1个块，5个线程
    cudaDeviceSynchronize();  // 等待GPU执行
    return 0;
}
// 输出5行问候语

2. ​​向量加法（线程组织）​​

__global__ void vectorAdd(int *a, int *b, int *c, int n) {
    int idx = blockIdx.x * blockDim.x + threadIdx.x; // 全局索引
    if (idx < n) c[idx] = a[idx] + b[idx];
}
int main() {
    int n = 1000;
    int *d_a, *d_b, *d_c;
    cudaMalloc(&d_a, n*sizeof(int));
    cudaMalloc(&d_b, n*sizeof(int));
    cudaMalloc(&d_c, n*sizeof(int));
    // 数据初始化及传输省略...
    dim3 blockSize(256);  // 每块256线程
    dim3 gridSize((n + 255)/256); // 计算所需块数
    vectorAdd<<<gridSize, blockSize>>>(d_a, d_b, d_c, n);
    cudaMemcpy(h_c, d_c, n*sizeof(int), cudaMemcpyDeviceToHost);
    // 释放资源省略...
}

---

🧠 三、内存管理进阶

1. ​​共享内存加速矩阵乘法​​

__global__ void matMul(float *A, float *B, float *C, int N) {
    __shared__ float sA[16][16];  // 块内共享内存
    __shared__ float sB[16][16];
    int tx = threadIdx.x, ty = threadIdx.y;
    int row = blockIdx.y * blockDim.y + ty;
    int col = blockIdx.x * blockDim.x + tx;
    float sum = 0;
    for (int i = 0; i < N; i += blockDim.x) {
        sA[ty][tx] = A[row*N + i + tx];  // 协作加载数据块
        sB[ty][tx] = B[(i+ty)*N + col];
        __syncthreads();  // 同步线程确保数据就绪
        for (int k = 0; k < blockDim.x; k++) 
            sum += sA[ty][k] * sB[k][tx];
        __syncthreads();
    }
    C[row*N+col] = sum;
}
// 调用示例: dim3 block(16,16); matMul<<<grid, block>>>(A, B, C, N);

---

🚀 四、实战应用场景

1. ​​OpenCV图像处理（CUDA加速）​​

#include <opencv2/cudaimgproc.hpp>
int main() {
    cv::Mat src_host = cv::imread("image.jpg");
    cv::cuda::GpuMat src, gray;
    src.upload(src_host);  // 上传到GPU
    cv::cuda::cvtColor(src, gray, cv::COLOR_BGR2GRAY); // GPU灰度转换
    cv::Mat gray_host;
    gray.download(gray_host);  // 下载结果
}

2. ​​并行归约求最大值​​

__global__ void reduceMax(float *d_in, float *d_out) {
    extern __shared__ float sdata[];
    int tid = threadIdx.x;
    int i = blockIdx.x * blockDim.x + tid;
    sdata[tid] = d_in[i];
    __syncthreads();
    for (int s = blockDim.x/2; s>0; s>>=1) {
        if (tid < s) sdata[tid] = fmaxf(sdata[tid], sdata[tid+s]);
        __syncthreads();
    }
    if (tid == 0) d_out[blockIdx.x] = sdata[0];
}
// 调用: reduceMax<<<grid, block, smem_size>>>(d_input, d_output);

---

📖 五、推荐学习资源

1. ​​官方文档​​

   • CUDA C++ Programming Guide - 架构详解
   • CUDA Best Practices Guide - 性能优化手册

2. ​​中文教程​​

   • 谭升博客《CUDA入门教程》：逐行注释的GitHub代码
   • 知乎专栏《CUDA编程入门》：从Hello World到内存优化
   • 《》中文版：系统化概念+工业级案例

3. ​​开源项目实践​​

   • LightSeq : Transformer推理加速库
   • CUDA Samples : NVIDIA官方示例集

---

💡 六、调试与优化工具

工具名称	用途	命令示例
cuda-gdb	GPU线程级调试	cuda-gdb ./my_program
nsys	性能分析（时间线）	nsys profile ./my_program
ncu	内核指令级分析	ncu --set full ./my_program

---

学习路径建议：从向量加法 → 矩阵运算 → 共享内存优化 → 结合深度学习框架。所有代码需用nvcc编译，注意GPU架构匹配（如RTX 4090需-arch=sm_89）。遇到内存错误时，使用cuda-memcheck ./program检测越界访问。