第一章:CUDA入门——GPU架构演进、CUDA编程模型、开发环境搭建
各位同学,欢迎来到《CUDA编程与显存管理实战》的第一课。
说实话,我刚开始接触GPU编程那会儿,也是一头雾水。那时候CUDA刚出来没多久,文档少得可怜,踩坑全靠自己摔。现在回头看,很多弯路其实可以避免。今天这一章,咱们就把地基打牢。
1.1 GPU架构演进:从图形卡到通用计算利器
先聊聊GPU的“发家史”。你想想看,最早的GPU就是专门画图的。显卡嘛,打游戏、做渲染,跟科学计算八竿子打不着。
转折点出现在2006年。NVIDIA发布了CUDA,GPU开始能跑通用计算了。我印象特别深,当时实验室里有人用GPU做矩阵乘法,速度比CPU快了两个数量级——所有人都惊呆了。
咱们快速过一下关键架构:
| 架构代号 | 发布年份 | 核心变化 | 我的评价 |
|---|---|---|---|
| Fermi | 2010 | 首个完整计算架构,支持ECC | 经典之作,但功耗感人 |
| Kepler | 2012 | SMX单元,动态并行 | 能效比大幅提升 |
| Maxwell | 2014 | 第三代流处理器,低功耗 | 我项目里用得最多的架构 |
| Pascal | 2016 | NVLink,统一内存 | 终于不用手动拷贝数据了 |
| Volta | 2017 | Tensor Core,独立线程调度 | AI训练的神器 |
| Turing | 2018 | RT Core,混合渲染 | 光线追踪走进现实 |
| Ampere | 2020 | 第三代Tensor Core,MIG | 云上多租户的福音 |
嗯,这里要注意一个核心概念:SM(Streaming Multiprocessor)。它是GPU的基本计算单元。每个SM里有一堆CUDA Core,还有共享内存、寄存器、调度器。说白了,SM就是GPU里的“小CPU”,只不过它擅长的是并行干活。
核心要点:GPU架构演进的本质,就是让越来越多的计算单元能高效地并行工作。从Fermi的16个SM到Ampere的128个SM,数量在涨,每个SM的能力也在涨。
1.2 CUDA编程模型:你只需要记住三个层次
很多新手一上来就被Grid、Block、Thread搞晕了。其实没那么复杂。我习惯用一个比喻:
- Grid = 整个工厂的车间
- Block = 车间里的工作小组
- Thread = 小组里的工人
每个工人(Thread)干自己的活,小组(Block)内可以互相帮忙(共享内存),车间之间(Grid)基本独立。
来看一个最简单的CUDA程序:
// 核函数:每个线程加1
__global__ void add_one(int *data, int n) {
int idx = threadIdx.x + blockIdx.x * blockDim.x;
if (idx < n) {
data[idx] += 1;
}
}
int main() {
int n = 1024;
int *d_data;
cudaMalloc(&d_data, n * sizeof(int));
// 启动1024个线程,分成4个block,每个block 256个线程
add_one<<<4, 256>>>(d_data, n);
cudaDeviceSynchronize();
cudaFree(d_data);
return 0;
}
你看,代码结构其实很清晰:
- 用
__global__声明核函数 - 用
<<<grid, block>>>指定线程布局 - 在核函数里用
threadIdx、blockIdx定位当前线程
个人经验:我刚开始写CUDA时,总喜欢把block设得很大,比如1024个线程一个block。后来发现这样反而慢。为什么呢?因为每个SM的资源有限,block太大反而限制了并行度。一般256或512比较稳妥。
1.3 开发环境搭建:别让工具成为绊脚石
环境搭建这事儿,说简单也简单,说坑也坑。我曾经在一台Ubuntu 18.04上折腾了整整两天,就为了把CUDA 10.2和驱动对上号。
咱们一步步来:
1.3.1 安装NVIDIA驱动
先确认你的显卡型号:
lspci | grep -i nvidia
然后去NVIDIA官网下载对应驱动。我个人建议用nvidia-driver-xxx这种包管理器安装,省心。
避坑指南:我曾经直接在官网下载.run文件安装,结果和系统自带的nouveau驱动冲突,黑屏了。后来学乖了,先禁用nouveau再装。命令如下:
sudo bash -c "echo blacklist nouveau > /etc/modprobe.d/blacklist-nvidia-nouveau.conf"
sudo update-initramfs -u
reboot
1.3.2 安装CUDA Toolkit
CUDA Toolkit包含了编译器(nvcc)、库文件、头文件等。我建议用runfile方式安装,可以自定义路径:
wget https://developer.download.nvidia.com/compute/cuda/12.1.0/local_installers/cuda_12.1.0_530.30.02_linux.run
sudo sh cuda_12.1.0_530.30.02_linux.run
安装完后,记得配环境变量:
export PATH=/usr/local/cuda-12.1/bin:$PATH
export LD_LIBRARY_PATH=/usr/local/cuda-12.1/lib64:$LD_LIBRARY_PATH
1.3.3 安装Nsight工具
Nsight是NVIDIA的性能分析神器。我强烈建议装上Nsight Systems和Nsight Compute:
- Nsight Systems:看整体时间线,找瓶颈
- Nsight Compute:深入分析每个核函数的性能
安装很简单,CUDA Toolkit里已经包含了。或者从NVIDIA官网单独下载。
验证安装:跑一下nvidia-smi看驱动是否正常,再跑nvcc --version看CUDA版本。最后编译一个示例:
cd /usr/local/cuda-12.1/samples/1_Utilities/deviceQuery
make
./deviceQuery
如果看到你的显卡型号和CUDA版本,恭喜你,环境搭好了!
1.4 本章知识体系总览
为了让你对本章内容有个整体印象,我画了一张图:
这张图把本章的三个核心模块串起来了。你从左往右看:先了解GPU硬件怎么演变的,再学CUDA软件模型怎么用,最后把环境搭好。三件事做完,你就可以开始写真正的CUDA程序了。
我的建议:别急着跑大程序。先拿deviceQuery和bandwidthTest这两个示例跑一跑,看看你的显卡参数。了解你的硬件,才能写出高效的代码。
好了,第一章就到这里。记住:GPU编程的核心思维是“并行”。把一个大任务拆成成千上万个独立的小任务,让每个线程处理一小块。这个思维转变过来,后面就顺了。
想要更多资料,联系可可