1、CUDA入门：GPU计算的历史与CUDA架构概述、GPU与CPU的区别、CUDA编程模型与Hello World

1.1 GPU计算的历史：从图形卡到通用计算

说起GPU计算的历史，我其实挺感慨的。2000年初我刚入行时，GPU还只是个「打游戏」的硬件。那时候大家管它叫图形加速卡，主要任务就是把三角形画得又快又漂亮。

转折点出现在2003年左右。斯坦福大学的研究人员发现，GPU里那些成百上千的小核心，其实特别适合做科学计算。矩阵乘法、流体模拟、分子动力学……这些在CPU上跑得慢吞吞的任务，放到GPU上竟然快了几十倍。

我记得当时有个叫Brook的项目，是第一个把GPU当通用计算用的编程框架。说实话，那玩意儿用起来真痛苦——你得把计算问题伪装成图形渲染问题，用像素着色器去算数据。嗯，那会儿的工程师都是狠人。

2006年，NVIDIA推出了CUDA。这玩意儿彻底改变了游戏规则。它让开发者可以直接用C语言写GPU程序，再也不用把计算问题伪装成画三角形了。我个人觉得，这是GPU计算历史上最重要的里程碑。

到了2010年代，深度学习爆发了。GPU成了AI训练的标配硬件。你想想看，没有GPU，今天的大模型、自动驾驶、AlphaGo，可能都还停留在实验室里。

1.2 GPU与CPU的区别：为什么GPU这么快？

很多人问我：「GPU和CPU到底有啥区别？不都是处理器吗？」

这个问题问得好。我打个比方你就明白了：CPU就像是一个全能型选手，什么活都能干，但一次只能干一两件事。GPU呢，就像是一支千人合唱团，每个人只会唱一个音符，但一千个人一起唱，气势就出来了。

具体来说，它们的区别主要体现在这几个方面：

对比维度	CPU	GPU
核心数量	4-16个（高性能）	数千个
核心设计	大而强，复杂控制逻辑	小而简，专注计算
缓存大小	大（MB级别）	小（KB级别）
内存带宽	50-100 GB/s	500-2000 GB/s
适用场景	串行任务、复杂逻辑	并行任务、大规模数据
功耗	几十瓦	几百瓦

我在项目中遇到过不少新手，总想把所有逻辑都扔到GPU上跑。结果呢？性能反而更差了。为什么会这样？因为GPU擅长的是「数据并行」——同样的操作，对大量数据同时执行。如果你的任务逻辑分支多、依赖性强，那还是老老实实用CPU吧。

1.3 CUDA架构概述：软硬件的桥梁

CUDA的全称是Compute Unified Device Architecture，统一计算设备架构。说白了，它就是NVIDIA给GPU编程搭的一座桥。

从硬件角度看，CUDA架构有几个关键概念：

• Streaming Multiprocessor (SM)：GPU的基本计算单元，每个SM包含多个CUDA核心
• CUDA Core：最基础的计算单元，执行算术运算
• Global Memory：全局显存，所有线程都能访问，但延迟高
• Shared Memory：共享内存，同一个Block内的线程共享，延迟低
• Register：寄存器，每个线程私有，速度最快

从软件角度看，CUDA把程序组织成三层结构：

1. Grid（网格）：一个完整的计算任务
2. Block（线程块）：Grid中的子任务，Block内的线程可以协作
3. Thread（线程）：最小的执行单元

我建议你把这个三层结构记在心里。写CUDA程序时，你就是在决定：开多少个Grid？每个Grid里放多少个Block？每个Block里放多少个Thread？

1.4 CUDA编程模型：Host与Device

CUDA编程模型的核心思想，就是「CPU负责控制，GPU负责计算」。CPU这边叫Host，GPU那边叫Device。

典型的CUDA程序流程是这样的：

1. 在Host上准备数据
2. 把数据从Host拷贝到Device
3. 在Device上启动Kernel（GPU函数）
4. 把结果从Device拷贝回Host
5. 在Host上处理结果

嗯，这里要注意：数据拷贝是开销很大的操作。我曾经犯过一个错误，在循环里反复拷贝小数据，结果性能比纯CPU还慢。后来我改成一次拷贝大块数据，速度直接翻了10倍。

Kernel函数的写法也有讲究。它前面要加一个 __global__ 修饰符，表示这个函数是在GPU上运行的。调用时用尖括号指定Grid和Block的维度：

// Kernel定义
__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];
    }
}

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

你看，每个线程通过 blockIdx.x、blockDim.x 和 threadIdx.x 计算出自己负责的数据索引。这就是CUDA的「单指令多线程」模型——所有线程执行同一段代码，但处理不同的数据。

1.5 Hello World：你的第一个CUDA程序

好了，理论说够了，我们来写个真正的CUDA程序。虽然「Hello World」有点老套，但用它来理解CUDA的基本流程，再合适不过了。

#include <stdio.h>
#include <cuda_runtime.h>

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

int main() {
    // 打印CPU信息
    printf("Hello World from CPU!\n");

    // 启动Kernel：2个Block，每个Block4个线程
    helloFromGPU<<<2, 4>>>();

    // 等待GPU执行完毕
    cudaDeviceSynchronize();

    // 检查错误
    cudaError_t err = cudaGetLastError();
    if (err != cudaSuccess) {
        printf("CUDA Error: %s\n", cudaGetErrorString(err));
    }

    return 0;
}

编译和运行：

nvcc hello.cu -o hello
./hello

输出结果：

Hello World from CPU!
Hello World from GPU! Block 0, Thread 0
Hello World from GPU! Block 0, Thread 1
Hello World from GPU! Block 0, Thread 2
Hello World from GPU! Block 0, Thread 3
Hello World from GPU! Block 1, Thread 0
Hello World from GPU! Block 1, Thread 1
Hello World from GPU! Block 1, Thread 2
Hello World from GPU! Block 1, Thread 3

你注意到没有？GPU线程的执行顺序是不确定的。每次运行，打印的顺序可能都不一样。这就是并行计算的特性——线程之间是并发执行的。

我个人习惯在写每个CUDA程序时，都加上 cudaDeviceSynchronize() 和错误检查。别嫌麻烦，这能帮你省下大量调试时间。

记住一句话：CUDA不难，难的是写出高效的CUDA。慢慢来，我们一起把基础打扎实。