第一个CUDA程序：Hello World from GPU

说实话，学任何一门新技术，我最期待的就是跑通第一个程序。那种「亮了」的感觉，特别爽。CUDA也不例外。今天我们就来写一个最经典的Hello World，不过这次——是从GPU里打印出来的。

核函数（Kernel）到底是什么？

先别急着写代码。我们得搞清楚一个核心概念：核函数。

核函数，说白了就是在GPU上执行的函数。CPU上的普通函数，你调用它，它在CPU上跑。核函数呢，你调用它，它在GPU的成百上千个线程上同时跑。

我个人习惯把核函数想象成一个「分身术」——你写一份代码，GPU帮你复制出几百上千份，同时执行。是不是很酷？

核函数有几个硬性规定，我列出来：

• 必须用 __global__ 修饰
• 返回值必须是 void
• 不能是类的成员函数（静态成员函数可以）
• 调用时要用特殊的 <<<grid, block>>> 语法

第一个核函数：Hello from GPU

来，直接上代码。这是我当年写的第一个CUDA程序，现在看依然觉得亲切。

#include <stdio.h>

// 定义一个核函数
__global__ void helloFromGPU()
{
    printf("Hello World from GPU! threadId: %d\n", threadIdx.x);
}

int main()
{
    // 调用核函数
    helloFromGPU<<<1, 5>>>();

    // 同步，确保GPU执行完毕
    cudaDeviceSynchronize();

    printf("Hello World from CPU!\n");

    return 0;
}

编译运行一下，你会看到类似这样的输出：

Hello World from GPU! threadId: 0
Hello World from GPU! threadId: 1
Hello World from GPU! threadId: 2
Hello World from GPU! threadId: 3
Hello World from GPU! threadId: 4
Hello World from CPU!

看到了吗？GPU上5个线程同时打印了信息。每个线程都有自己的ID，从0到4。

<<<grid, block>>> 语法详解

这个语法，说实话，刚接触时我也懵过。三个尖括号，里面两个参数，到底啥意思？

我们来拆解一下：

• grid：网格，可以理解成「有多少个线程块」
• block：线程块，可以理解成「每个线程块里有多少个线程」

所以 <<<1, 5>>> 的意思是：1个线程块，每个块里5个线程。总共就是 1 × 5 = 5 个线程。

你想想看，如果改成 <<<2, 3>>>，那就是2个块，每块3个线程，总共6个线程。

我画了一张图，帮你理解这个层次结构：

这张图展示的是 <<<3, 5>>> 的情况。3个Block，每个Block里5个Thread。总共15个线程并行执行。

线程索引：threadIdx 和 blockIdx

每个线程都有自己独特的ID。CUDA提供了内置变量：

变量	含义	示例值
threadIdx.x	线程在块内的索引	0, 1, 2, 3, 4
blockIdx.x	块在网格内的索引	0, 1, 2
blockDim.x	每个块的线程数	5
gridDim.x	网格的块数	3

如果你想计算全局唯一的线程ID，公式是：

int globalThreadId = blockIdx.x * blockDim.x + threadIdx.x;

嗯，这个公式后面会反复用到，建议你记下来。

主机与设备代码分离

CUDA程序天然分成两部分：

• 主机（Host）代码：在CPU上运行，负责数据准备、调用核函数、结果回收
• 设备（Device）代码：在GPU上运行，负责并行计算

我刚开始写CUDA时，经常把这两部分混在一起。后来养成了一个习惯：在代码里用注释明确标出 // Host 和 // Device，清晰很多。

编译和运行

如果你用的是Linux，编译命令很简单：

nvcc hello.cu -o hello
./hello

Windows上如果你装了Visual Studio + CUDA Toolkit，直接在命令行里同样操作就行。

我曾经遇到一个坑：忘了加 -arch=sm_xx 参数，结果在老显卡上跑不了。建议你编译时加上：

nvcc -arch=sm_75 hello.cu -o hello

sm_75 对应Turing架构（RTX 20系列）。具体用哪个，查一下你的显卡算力版本。

小结

今天的内容其实就三件事：

1. 核函数用 __global__ 定义，用 <<<grid, block>>> 调用
2. 线程层次结构：Grid → Block → Thread
3. 主机代码和设备代码是分开的，各司其职

说实话，第一个程序跑通的那一刻，你就已经迈过了CUDA学习中最难的一道坎。剩下的，就是不断加深理解、积累经验了。

想要更多资料，联系可可