第一章：CUDA入门——GPU架构概览、CUDA编程模型、CUDA开发环境搭建

各位同学，欢迎来到《CUDA高性能AI训练从零开始》的第一章。

说实话，每次带新人入门CUDA，我最怕的就是一上来就堆概念。什么线程束、共享内存、SM调度……听得人头皮发麻。所以我换个路子——咱们先搞清楚GPU到底是个什么玩意儿，再动手把环境搭起来，最后写个程序跑通。嗯，就这么干。

1.1 GPU架构概览：为什么GPU能“并行”到飞起？

先问个问题：CPU和GPU到底差在哪？

CPU像是一个全能博士——什么都会，但一次只能处理一两件事。GPU呢？它像是一千个小学生——每个只会简单的加减法，但一千个一起算，速度就上去了。

我当年第一次接触GPU编程时，最大的震撼就是：这玩意儿天生就是为矩阵乘法生的。AI训练里，动不动就是几百万个神经元同时计算，CPU一个个算得累死，GPU一挥手就搞定了。

核心概念：GPU拥有成百上千个计算核心（CUDA Core），每个核心虽然频率不高，但胜在数量多。AI训练中的矩阵运算、卷积操作，天然适合这种“大量简单计算并行执行”的模式。

具体到NVIDIA的GPU架构，我挑几个关键点说说：

  • SM（Streaming Multiprocessor，流式多处理器）——GPU的基本计算单元。每个SM里包含多个CUDA Core、共享内存、寄存器等。你可以把SM想象成一个“小分队”，里面几十个士兵（CUDA Core）一起干活。
  • 全局内存（Global Memory）——显存，容量大（8GB、16GB甚至80GB），但速度相对慢。所有线程都能访问，但要注意访问模式，不然性能会崩。
  • 共享内存（Shared Memory）——每个SM内部的小块高速缓存，容量小（几十KB到上百KB），但速度极快。我经常跟学员说：用好共享内存，你的CUDA程序就成功了一半。
  • 线程束（Warp）——GPU执行指令的最小单位，32个线程一组。如果32个线程走不同的分支（if-else），性能会大幅下降。这玩意儿叫“线程束发散”，我在项目中踩过这个坑，后面会细讲。

个人经验：我刚学CUDA时，总以为线程越多越好。后来发现，如果线程访问全局内存的模式是“乱跳”的，性能反而比CPU还差。记住：合并访问（Coalesced Access）是GPU内存优化的第一课。

1.2 CUDA编程模型：你只需要记住三个层次

CUDA编程模型，说白了就是三个层次：Grid（网格）→ Block（线程块）→ Thread（线程）。

我画个图帮你理解：

Grid（网格） Block (0,0) Block (1,0) Block (2,0) T0 T1 T2 T3 T4 T5 ... ... ... 一个Grid包含多个Block，一个Block包含多个Thread

你看，一个Grid就是一个完整的计算任务。Grid里划分成多个Block，每个Block里又有多个Thread。实际编程时，你只需要告诉CUDA：我要开多少个Block，每个Block里有多少个Thread。剩下的调度，CUDA帮你搞定。

举个例子，你要计算两个长度为10000的向量相加：

// 核函数定义
__global__ void vecAdd(float* A, float* B, float* C, int N) {
    int i = blockIdx.x * blockDim.x + threadIdx.x;
    if (i < N) {
        C[i] = A[i] + B[i];
    }
}

// 调用核函数
int N = 10000;
int threadsPerBlock = 256;
int blocksPerGrid = (N + threadsPerBlock - 1) / threadsPerBlock;
vecAdd<<<blocksPerGrid, threadsPerBlock>>>(d_A, d_B, d_C, N);

这里有个细节：blockIdx.x是当前Block的索引，threadIdx.x是当前Thread在Block内的索引，blockDim.x是每个Block的线程数。三者一组合，就得到了全局唯一的线程ID。我刚开始写的时候老搞混，后来记住一句话：全局ID = Block索引 × Block大小 + 线程内索引。

避坑指南：我曾经犯过一个低级错误——核函数里忘了加if (i < N)这个边界检查。当数据量不是线程数的整数倍时，多余的线程会访问越界内存，轻则计算结果错误，重则程序崩溃。记住：永远要做边界检查。

1.3 CUDA开发环境搭建：别怕，三步搞定

环境搭建这块，我踩过的坑比写过的代码还多。咱们一步步来：

第一步：安装NVIDIA驱动

这是基础中的基础。去NVIDIA官网下载对应你显卡型号的驱动。装完后，在命令行输入nvidia-smi，如果能看到显卡信息，就说明驱动装好了。

个人习惯：我建议装驱动时选择“自定义安装”，然后勾选“清洁安装”。这样可以避免旧驱动残留导致的各种奇怪问题。

第二步：安装CUDA Toolkit

CUDA Toolkit包含了编译器（nvcc）、CUDA运行时库、开发工具等。下载地址在NVIDIA官网，选择对应你操作系统的版本。

装完后，验证一下：

nvcc --version

如果能看到版本号，恭喜你，CUDA环境已经就绪。

第三步：安装cuDNN

cuDNN是NVIDIA专门为深度学习优化的库，封装了卷积、池化、激活函数等操作。AI训练基本离不开它。

安装步骤：

  1. 去NVIDIA官网下载cuDNN（需要注册账号）
  2. 解压后，把文件复制到CUDA Toolkit的安装目录下
  3. 配置环境变量，把cuDNN的lib路径加到LD_LIBRARY_PATH（Linux）或PATH（Windows）

验证cuDNN是否装好：

# 查看cuDNN版本
cat /usr/local/cuda/include/cudnn_version.h | grep CUDNN_MAJOR -A 2

重要提醒：CUDA Toolkit、驱动、cuDNN三者版本必须匹配。我见过太多人因为版本不兼容，编译报错一整天。建议去NVIDIA官网查一下“CUDA Compatibility Matrix”，确认你的组合没问题。

1.4 第一个CUDA程序：Hello World from GPU

环境搭好了，咱们写个程序庆祝一下。这个程序会在GPU上打印一句话：

#include <stdio.h>

__global__ void helloFromGPU() {
    printf("Hello World from GPU! Block %d, Thread %d\n", 
           blockIdx.x, threadIdx.x);
}

int main() {
    // 启动10个Block，每个Block有10个线程
    helloFromGPU<<<10, 10>>>();
    
    // 等待GPU执行完毕
    cudaDeviceSynchronize();
    
    printf("Hello World from CPU!\n");
    return 0;
}

编译运行：

nvcc hello.cu -o hello
./hello

你会看到GPU打印了100行“Hello World”，每个线程都喊了一嗓子。嗯，虽然简单，但这是你CUDA生涯的第一步。

小技巧：如果程序运行后什么都没输出，别慌。检查一下cudaDeviceSynchronize()有没有加。GPU是异步执行的，不加这行，CPU可能直接跑完了，GPU的输出还没来得及打印。

本章总结

咱们这一章干了三件事：

  • 搞懂了GPU为什么能并行——靠的是成百上千个CUDA Core和SM架构
  • 学会了CUDA编程模型——Grid、Block、Thread三层结构
  • 搭好了开发环境——驱动、CUDA Toolkit、cuDNN一步到位

说实话，第一章的内容不算多，但都是地基。地基打不牢，后面学什么都飘。下一章咱们会深入聊聊内存管理——这可是CUDA性能优化的核心。到时候见。

想要更多资料，联系可可

← 上一章 📖 返回目录 下一章 →