一、CUDA编程模型概述：GPU架构演进、CUDA编程模型、主机与设备、核函数概念

1.1 GPU架构演进：从图形卡到通用计算利器

说实话，我刚入行那会儿，GPU还是个纯粹的“画图卡”。大家用它打游戏、做渲染，跟科学计算八竿子打不着。但后来事情变了——2006年NVIDIA发布了CUDA，GPU正式从图形专用处理器，变成了通用计算加速器。

GPU架构的演进，我把它分成几个关键阶段：

• 早期架构（Tesla/ Fermi）：CUDA刚诞生，硬件上还没有独立的L1/L2缓存，程序员得手动管理共享内存。我记得第一次写Fermi的代码，被bank conflict折磨得够呛。
• Kepler/Maxwell：引入了动态并行、Hyper-Q等技术。说白了，就是让GPU自己也能调度自己，不用啥事都靠CPU来指挥。
• Pascal/Volta：统一内存、独立线程调度。Volta的独立线程调度是个大杀器——以前一个warp里所有线程必须执行同一条指令，现在可以各自飞了。
• Turing/Ampere/Hopper：Tensor Core、MIG、DPX指令集。这些架构开始面向AI和高性能计算做深度定制。

你想想看，从最初的128个CUDA核心，到现在H100的上万个核心，GPU的并行能力提升了两个数量级。但核心思想没变——用大量轻量级线程，掩盖访存延迟。

1.2 CUDA编程模型：主机与设备的分工协作

CUDA编程模型，说白了就是一套“CPU+GPU”的协作框架。CPU叫主机（Host），GPU叫设备（Device）。主机负责控制流、I/O、逻辑判断；设备负责大规模并行计算。

我在项目中遇到过不少新手，上来就想把整个程序扔到GPU上跑。结果呢？数据传输开销比计算本身还大。记住一个原则：CPU做CPU擅长的事，GPU做GPU擅长的事。

主机与设备的关系，可以用这张图来理解：

嗯，这里要注意：主机和设备有各自独立的内存空间。你不能直接在GPU上访问CPU的变量，反之亦然。必须通过cudaMemcpy来搬运数据。

1.3 核函数概念：GPU上的并行执行单元

核函数（Kernel），就是跑在GPU上的函数。它用__global__修饰符声明，由CPU调用，但在GPU上执行。

核函数有几个关键特征：

• 由CPU调用，GPU执行：调用时用三括号语法 <<<grid, block>>> 指定线程组织方式
• 所有线程执行同一段代码：但每个线程有自己的线程ID，可以处理不同的数据
• 异步执行：CPU发起核函数后，默认不会等待它完成，而是继续执行后面的代码

来看一个最简单的例子：

// 核函数声明：向量加法
__global__ void vecAdd(float *A, float *B, float *C, int N) {
    int idx = threadIdx.x + blockIdx.x * blockDim.x;
    if (idx < N) {
        C[idx] = A[idx] + B[idx];
    }
}

// 主机端调用
int main() {
    // ... 分配内存、拷贝数据 ...
    
    // 启动核函数：1个grid，256个线程/block
    vecAdd<<<1, 256>>>(d_A, d_B, d_C, N);
    
    // ... 拷贝结果回主机 ...
    return 0;
}

这段代码里，threadIdx.x 是线程在block内的索引，blockIdx.x 是block在grid内的索引，blockDim.x 是每个block的线程数。三者组合起来，就能唯一确定一个全局线程ID。

1.4 线程层次结构：Grid、Block、Thread

CUDA的线程组织方式，是一个三层结构：

层级	说明	最大数量（典型）
Grid	一个核函数启动的所有线程	2³¹-1（x维度）
Block	一组可以协作的线程，共享共享内存	1024个线程/block
Thread	最小的执行单元	每个block最多1024个

为什么要有这个层次结构？说白了，是为了适配GPU的硬件架构。一个block内的线程跑在同一个SM（流多处理器）上，可以共享数据、同步执行。不同block的线程，则可能跑在不同的SM上，彼此独立。

我个人的习惯是：先确定block大小，再算grid大小。block大小通常取32的倍数（warp大小），比如128、256。然后根据总数据量，算出需要多少个block。

// 确定block大小
int blockSize = 256;
// 计算grid大小（向上取整）
int gridSize = (N + blockSize - 1) / blockSize;

// 启动核函数
myKernel<<<gridSize, blockSize>>>(d_data, N);

1.5 总结：CUDA编程模型的核心思想

回顾一下，CUDA编程模型的核心就三件事：

1. 主机与设备分工：CPU管控制，GPU管计算
2. 核函数是并行入口：用 __global__ 声明，用三括号启动
3. 线程层次组织：Grid → Block → Thread，适配硬件架构

你想想看，这套模型其实很优雅。它把并行计算的复杂性，封装成了程序员熟悉的函数调用形式。你不需要关心GPU内部有多少个核心、怎么调度线程——你只需要写好核函数，CUDA runtime会帮你搞定剩下的事。

当然，想写出高性能的CUDA代码，光知道这些还不够。后面我们会深入讨论内存层次、warp调度、原子操作、流并发等话题。但今天这第一课，先把地基打牢。