第一章 CUDA入门:GPU架构概述、CUDA编程模型、开发环境搭建
各位同学好,我是你们这门课的老朋友。今天咱们正式开篇,聊聊CUDA入门那些事儿。说实话,我当年刚接触GPU编程时,也是一头雾水——CPU玩得挺溜,一上GPU就懵了。后来踩了不少坑,才慢慢摸清门道。这一章,我就把最基础、最核心的东西掰开揉碎讲给你听。
1.1 GPU架构概述:为什么GPU能“并行”到飞起?
先问个问题:CPU和GPU到底有啥区别?你想想看,CPU就像个全能博士,啥都能干,但一次只能处理一两件事。GPU呢?它更像一支建筑队——每个工人只会搬砖,但几千人一起干,效率惊人。
GPU的核心设计理念就是“大量简单计算单元并行干活”。以NVIDIA的架构为例,从Fermi到Ampere,再到Hopper,核心逻辑没变:
- SM(流多处理器):GPU的基本调度单元。每个SM里有一堆CUDA核心、共享内存、寄存器等。
- Warp(线程束):32个线程为一组,一起执行指令。这是硬件调度的最小单位。
- Global Memory(全局内存):显存,所有线程都能访问,但延迟高(几百个周期)。
- Shared Memory(共享内存):每个SM内部的高速缓存,延迟低(几个周期),但容量小。
我在项目中遇到过一个问题:有个同事写了个核函数,数据全从全局内存读,结果性能惨不忍睹。后来我把热点数据搬到共享内存里,速度直接翻了5倍。嗯,这里要注意——访存优化是CUDA调优的重中之重。
核心概念速览:
- GPU适合数据并行任务(图像处理、矩阵运算、深度学习)
- CPU适合控制密集型任务(分支多、逻辑复杂)
- Warp divergence(线程束分歧)会严重拖慢性能——同一个warp里别搞太多if-else
下面这张图是我自己画的,帮你理清GPU的层级结构:
1.2 CUDA编程模型:你写的代码到底怎么跑?
说白了,CUDA编程就是“CPU管控制,GPU管计算”。CPU叫Host(主机),GPU叫Device(设备)。你写的核函数(Kernel)在GPU上跑,剩下的逻辑在CPU上跑。
我习惯把CUDA程序分成三步走:
- 数据准备:在CPU上分配内存,把数据从CPU拷贝到GPU
- 执行计算:启动核函数,GPU上的成千上万个线程并行干活
- 结果回传:把计算结果从GPU拷回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() {
int n = 1024;
float *d_a, *d_b, *d_c;
// 1. 在GPU上分配显存
cudaMalloc(&d_a, n * sizeof(float));
cudaMalloc(&d_b, n * sizeof(float));
cudaMalloc(&d_c, n * sizeof(float));
// 2. 把数据从CPU拷到GPU(省略了CPU端数据初始化)
cudaMemcpy(d_a, h_a, n * sizeof(float), cudaMemcpyHostToDevice);
cudaMemcpy(d_b, h_b, n * sizeof(float), cudaMemcpyHostToDevice);
// 3. 启动核函数:256个线程一个块,一共4个块
vecAdd<<<4, 256>>>(d_a, d_b, d_c, n);
// 4. 结果拷回CPU
cudaMemcpy(h_c, d_c, n * sizeof(float), cudaMemcpyDeviceToHost);
// 5. 释放显存
cudaFree(d_a); cudaFree(d_b); cudaFree(d_c);
return 0;
}
小技巧:我刚开始写CUDA时,经常忘了检查cudaMalloc和cudaMemcpy的返回值。后来养成习惯,每次调用CUDA API后都加一句cudaGetLastError()检查,省了好多调试时间。
这里有个关键概念——线程层次结构:
- Thread(线程):最小的执行单元
- Block(线程块):一组线程,共享shared memory,可以同步
- Grid(网格):一组block,构成整个核函数
你想想看,<<<gridDim, blockDim>>>这个语法,gridDim是网格维度,blockDim是块维度。我一般用一维索引处理数组,二维索引处理图像,三维索引处理体数据。
1.3 开发环境搭建:别让环境卡住你
说实话,环境搭建是很多新手的第一道坎。我当年装CUDA Toolkit时,因为驱动版本不对,折腾了一整天。下面我把步骤拆清楚,你照着来就行。
1.3.1 NVIDIA驱动安装
先确认你的显卡型号。在终端里跑:
lspci | grep -i nvidia # Linux
wmic path win32_videocontroller get name # Windows
然后去NVIDIA官网下载对应驱动。我个人建议装Game Ready驱动或者Studio驱动都行,关键是版本别太老。
避坑指南:我曾经在Ubuntu上装了驱动后黑屏,原因是nouveau开源驱动没禁用。记得在/etc/modprobe.d/blacklist-nouveau.conf里加上blacklist nouveau,然后重启。
1.3.2 CUDA Toolkit安装
CUDA Toolkit包含了编译器nvcc、CUDA库、调试工具等。推荐用runfile(本地安装包)方式安装,可控性高:
wget https://developer.download.nvidia.com/compute/cuda/12.3.0/local_installers/cuda_12.3.0_linux.run
sudo sh cuda_12.3.0_linux.run
安装时注意:
- 不要勾选Driver(如果你已经装好了驱动)
- 记住安装路径,默认是
/usr/local/cuda-12.3 - 配置环境变量:
export PATH=/usr/local/cuda-12.3/bin:$PATH - 配置库路径:
export LD_LIBRARY_PATH=/usr/local/cuda-12.3/lib64:$LD_LIBRARY_PATH
验证安装:
nvcc --version
# 应该输出类似:Cuda compilation tools, release 12.3, V12.3.107
1.3.3 cuDNN安装
cuDNN是深度神经网络的加速库。如果你做AI相关开发,这个必装。步骤:
- 去NVIDIA官网下载cuDNN(需要注册账号)
- 解压后把文件拷贝到CUDA目录:
tar -xzvf cudnn-linux-x86_64-8.9.7.29_cuda12-archive.tar.xz
sudo cp cudnn-*-archive/include/cudnn*.h /usr/local/cuda/include
sudo cp -P cudnn-*-archive/lib/libcudnn* /usr/local/cuda/lib64
sudo chmod a+r /usr/local/cuda/include/cudnn*.h /usr/local/cuda/lib64/libcudnn*
验证cuDNN版本:
cat /usr/local/cuda/include/cudnn_version.h | grep CUDNN_MAJOR -A 2
版本兼容性速查表(以CUDA 12.x为例):
| 组件 | 推荐版本 | 备注 |
|---|---|---|
| NVIDIA驱动 | ≥ 545.23 | 向下兼容 |
| CUDA Toolkit | 12.3 | LTS版本,稳定 |
| cuDNN | 8.9.7 | 匹配CUDA 12.x |
嗯,到这里,你的CUDA开发环境就搭好了。我建议你写个简单的hello.cu跑一下,确认一切正常:
// hello.cu
#include <stdio.h>
__global__ void hello() {
printf("Hello from thread %d in block %d\n",
threadIdx.x, blockIdx.x);
}
int main() {
hello<<<2, 4>>>();
cudaDeviceSynchronize();
return 0;
}
编译运行:
nvcc hello.cu -o hello
./hello
看到每个线程打印自己的ID,说明环境没问题了。记住,环境搭建只是起点,真正的挑战在后面。下一章我们会深入线程模型和内存模型,那才是CUDA性能调优的核心。
想要更多资料,联系可可