第一章:CUDA基础与环境搭建

各位同学,咱们今天聊聊CUDA的入门。说实话,很多新手一上来就急着写核函数,结果环境都没配好,编译报错一堆。我当年也踩过这个坑——记得第一次装CUDA Toolkit,折腾了整整一个下午,最后发现是环境变量没配对。

所以这一章,咱们把基础打扎实。从GPU架构讲起,再到工具安装、编译器使用,最后跑通第一个程序。你跟着我的节奏走,保证不迷路。

1.1 GPU架构概述:为什么它这么快?

先问个问题:CPU和GPU到底有啥区别?

CPU像是一个全能型选手——什么活都能干,但一次只能干几件事。GPU呢?它更像是一群流水线上的工人——每个人只干一件事,但人多力量大。

具体来说,NVIDIA的GPU架构有几个关键概念:

  • SM(Streaming Multiprocessor):流式多处理器,这是GPU的核心计算单元。每个SM里包含多个CUDA核心。
  • Warp:线程束,32个线程为一组。这是GPU调度和执行的最小单位。记住这个数字——32,后面你会经常见到它。
  • Global Memory:全局显存,所有线程都能访问,但延迟高。我习惯把它比作「仓库」——容量大,但取东西慢。
  • Shared Memory:共享内存,同一个Block内的线程共享,延迟低。这就像「工位上的工具架」——随手就能拿到。

我个人习惯把GPU架构理解成一个「三层结构」:

核心记忆点:

  • Thread(线程)→ 最细粒度的工作单元
  • Block(线程块)→ 一组线程,共享Shared Memory
  • Grid(网格)→ 所有Block的集合,对应一个核函数调用

下面这张图,是我自己画的一个简化版架构图,帮你理清关系:

Grid(网格) Block 0 Block 1 Block 2 Thread 0 Thread 1 Thread 2 Thread 3 ... ... Shared Memory Thread Block Grid

小提示:刚开始接触这些概念时,别死记硬背。你只需要记住:Thread是干活的人,Block是一个小组,Grid是整个项目组。后面写代码时,你会慢慢理解它们的关系。

1.2 CUDA Toolkit安装:别走弯路

安装CUDA Toolkit,说白了就是三步:下载、安装、配环境。但每一步都有坑,我一个个说。

第一步:确认你的硬件

先看看你的显卡支不支持CUDA。打开终端(Windows下是命令提示符),输入:

nvidia-smi

如果能看到显卡型号和驱动版本,恭喜你,硬件没问题。如果报错「命令未找到」,那你得先去NVIDIA官网装驱动。

注意:我曾经遇到过一台机器,nvidia-smi能跑,但CUDA程序死活编译不过。查了半天,发现是驱动版本太老,不支持新版的CUDA Toolkit。所以,驱动版本和Toolkit版本要匹配——这个后面会讲。

第二步:下载并安装

去NVIDIA官网下载CUDA Toolkit。我建议选LTS(长期支持)版本,稳定。比如CUDA 11.8或12.x,看你的项目需求。

安装时,Windows用户选「自定义安装」,把「Visual Studio Integration」勾上——如果你用VS的话。Linux用户更简单,用runfile或deb包都行。我个人习惯用runfile,因为可以指定安装路径。

第三步:配置环境变量

这是最容易出问题的一步。安装完后,需要把以下路径加到系统环境变量里:

操作系统 需要添加的路径
Windows C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v12.0\bin
Linux /usr/local/cuda-12.0/bin

验证安装是否成功,打开终端输入:

nvcc --version

如果能看到版本号,说明装好了。如果报错,八成是环境变量没配对。

避坑指南:我曾经帮一个同事排查问题,他nvcc --version能跑,但编译时总说找不到cuda_runtime.h。后来发现,他忘了把include路径加到环境变量里。所以,除了bin目录,还要检查一下CUDA_PATH这个变量有没有设置。

1.3 nvcc编译器使用:不只是个编译器

nvcc是NVIDIA的CUDA编译器。它能把你的.cu文件编译成可在GPU上运行的代码。说白了,它是个「二合一」工具——既处理CPU代码(用主机编译器),也处理GPU代码(用设备编译器)。

最基本的用法:

nvcc -o my_program my_program.cu

这条命令会把my_program.cu编译成可执行文件my_program。

常用的编译选项,我列个表:

选项 作用 示例
-arch=sm_xx 指定GPU架构,比如sm_75对应Turing架构 -arch=sm_86(RTX 30系列)
-O2 优化等级,一般用-O2就够了 nvcc -O2 test.cu
-G 生成调试信息,配合cuda-gdb使用 nvcc -G test.cu
-lineinfo 生成行号信息,用于性能分析 nvcc -lineinfo test.cu

重点: -arch参数一定要和你显卡的架构匹配。用错了,程序可能跑不起来,或者性能很差。怎么查?nvidia-smi能看到显卡型号,然后去NVIDIA官网查对应的Compute Capability(计算能力)。比如RTX 3090是sm_86。

1.4 第一个CUDA程序:Hello, GPU!

终于到了写代码的环节。咱们写一个最简单的程序:在GPU上打印一句话。

#include <stdio.h>

// 核函数:在GPU上执行
__global__ void helloFromGPU() {
    printf("Hello from GPU! Thread ID: %d\n", threadIdx.x);
}

int main() {
    // 打印CPU信息
    printf("Hello from CPU!\n");
    
    // 启动核函数:1个Block,每个Block里5个线程
    helloFromGPU<<<1, 5>>>();
    
    // 等待GPU执行完毕
    cudaDeviceSynchronize();
    
    return 0;
}

保存为hello.cu,然后编译:

nvcc -o hello hello.cu

运行:

./hello

你会看到类似这样的输出:

Hello from CPU!
Hello from GPU! Thread ID: 0
Hello from GPU! Thread ID: 1
Hello from GPU! Thread ID: 2
Hello from GPU! Thread ID: 3
Hello from GPU! Thread ID: 4

注意看,线程ID是从0到4,一共5个。这就是我们指定的线程数量。

解释一下:

  • __global__:告诉编译器,这个函数是在GPU上跑的核函数。
  • <<<1, 5>>>:第一个参数是Block数量,第二个是每个Block里的线程数。这里我们用了1个Block,5个线程。
  • cudaDeviceSynchronize():等GPU干完活再继续。不加这行,CPU可能先跑完,程序就退出了,GPU上的打印可能看不到。

常见错误:我第一次写CUDA程序时,忘了加cudaDeviceSynchronize(),结果GPU的打印信息时有时无。后来才明白,CPU和GPU是异步执行的——你想想看,CPU发完指令就跑了,GPU还在后面慢慢执行。不加同步,程序可能已经结束了,GPU的输出还没到。

好了,到这里你已经跑通了第一个CUDA程序。虽然简单,但核心流程都在了:写核函数、配置线程、编译运行。后面的章节,我们会在这个基础上不断加东西——内存管理、线程同步、性能优化……

嗯,今天就到这儿。记住,环境搭建是基础中的基础,别嫌麻烦。配好了环境,后面才能跑得顺。

本章要点回顾:

  • GPU架构:Thread → Block → Grid 三层结构
  • 安装CUDA Toolkit:确认硬件 → 下载安装 → 配环境变量
  • nvcc编译:-arch指定架构,-O2优化,-G调试
  • 第一个程序:核函数用__global__声明,用<<<grid, block>>>启动

可可精品资料库