第一个CUDA程序：编写Hello World核函数

好，咱们正式开始写代码了。

说实话，我当年学CUDA的时候，最兴奋的就是看到屏幕上打出"Hello World"的那一刻。虽然只是几个字，但那是GPU算出来的，感觉完全不一样。

这一章，我们就来写你的第一个CUDA核函数。别怕，代码很短，逻辑也很直白。

什么是核函数？

核函数，说白了就是在GPU上跑的函数。CPU上的函数叫"普通函数"，GPU上的函数就叫"核函数"。

你想想看，CPU和GPU是两套不同的硬件。CPU负责串行逻辑，GPU负责并行计算。要让GPU干活，你得把任务写成核函数，然后告诉GPU："嘿，帮我跑这个函数，跑1000次！"

核函数有几个特点：

• 前面必须加 __global__ 修饰符
• 返回类型必须是 void
• 调用时要用特殊的 <<<...>>> 语法

第一个核函数：Hello World

咱们直接上代码。我习惯先写一个最简单的版本，跑通了再慢慢加东西。

#include <stdio.h>

// 定义一个核函数
__global__ void helloFromGPU()
{
    printf("Hello World from GPU!\n");
}

int main()
{
    // 调用核函数，使用 <<<1, 1>>> 表示1个线程块，每个块1个线程
    helloFromGPU<<<1, 1>>>();

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

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

    return 0;
}

这段代码干了什么呢？

1. 定义了一个 __global__ 函数 helloFromGPU
2. 在 main 函数里用 <<<1, 1>>> 调用它
3. GPU打印一行字，CPU也打印一行字

编译运行一下：

nvcc hello.cu -o hello
./hello

输出应该是：

Hello World from GPU!
Hello World from CPU!

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

这个语法是CUDA里最核心的东西。我刚开始学的时候也觉得它有点怪，但用多了就习惯了。

<<<grid, block>>> 其实是在告诉GPU两件事：

• grid：你要启动多少个线程块（Block）
• block：每个线程块里有多少个线程（Thread）

举个例子：

helloFromGPU<<<2, 4>>>();

这表示：启动2个线程块，每个块里有4个线程。总共就是 2 × 4 = 8 个线程在跑。

每个线程都有一个唯一的ID，可以通过 threadIdx.x 和 blockIdx.x 拿到。咱们改一下代码看看：

__global__ void helloFromGPU()
{
    printf("Block %d, Thread %d: Hello World!\n", blockIdx.x, threadIdx.x);
}

int main()
{
    helloFromGPU<<<2, 4>>>();
    cudaDeviceSynchronize();
    return 0;
}

输出大概是这样（顺序可能不同）：

Block 0, Thread 0: Hello World!
Block 0, Thread 1: Hello World!
Block 0, Thread 2: Hello World!
Block 0, Thread 3: Hello World!
Block 1, Thread 0: Hello World!
Block 1, Thread 1: Hello World!
Block 1, Thread 2: Hello World!
Block 1, Thread 3: Hello World!

一维、二维、三维的线程组织

线程块和网格都可以是一维、二维或三维的。为什么需要多维？因为很多数据天然就是多维的，比如图像是二维的，体素数据是三维的。

定义方式如下：

// 一维：1个块，每个块8个线程
helloFromGPU<<<1, 8>>>();

// 二维：2x2个块，每个块4x4个线程
dim3 grid(2, 2);
dim3 block(4, 4);
helloFromGPU<<<grid, block>>>();

// 三维：2x2x2个块，每个块2x2x2个线程
dim3 grid3(2, 2, 2);
dim3 block3(2, 2, 2);
helloFromGPU<<<grid3, block3>>>();

对应的线程索引：

维度	块索引	线程索引
一维	blockIdx.x	threadIdx.x
二维	blockIdx.x, blockIdx.y	threadIdx.x, threadIdx.y
三维	blockIdx.x, blockIdx.y, blockIdx.z	threadIdx.x, threadIdx.y, threadIdx.z

避坑指南：我踩过的几个坑

写第一个CUDA程序，有几个地方特别容易出错。我当年都踩过：

• 忘记加 __global__：编译报错，说找不到核函数。检查一下函数前面有没有这个修饰符。
• 忘记 cudaDeviceSynchronize()：GPU的输出没打印出来，程序就退出了。尤其是用 printf 的时候，一定要加。
• 线程块大小设置太大：每个线程块最多1024个线程（不同GPU可能不同）。设成2048直接报错。
• 用 return 返回值：核函数必须是 void 类型，不能有返回值。想传数据？用指针参数。

本章知识结构图

下面这张图帮你理清这一章的核心逻辑：

这张图展示了核函数的三个核心要素：__global__ 修饰符、void 返回类型、以及 <<<grid, block>>> 语法。记住这三样，你的第一个CUDA程序就能跑起来了。

总结

这一章我们做了三件事：

• 写了一个最简单的Hello World核函数
• 理解了 <<<grid, block>>> 的含义
• 学会了如何获取线程ID

嗯，到这里你已经迈出了CUDA编程的第一步。别小看这个Hello World，它背后是整个GPU并行计算的模型。后面我们会在这个基础上，一步步加上数据传递、并行计算、性能优化等内容。

想要更多资料，联系可可