正如您所写的,该内核是完全串行的。启动执行它的每个线程都将执行相同的工作。
CUDA(以及 OpenCL 和其他类似的“单程序、多数据”类型编程模型)背后的主要思想是采用“数据并行”操作 - 因此必须多次执行相同的、很大程度上独立的操作 - 并且编写一个执行该操作的内核。然后启动大量(半)自治线程来跨输入数据集执行该操作。
在数组加法示例中,数据并行操作是
C[k] = A[k] + B[k];
对于 0 到 128 * 1024 之间的所有 k。每个加法操作完全独立,没有顺序要求,因此可以由不同的线程执行。为了在 CUDA 中表达这一点,可以这样编写内核:
__global__ void mAdd(float* A, float* B, float* C, int n)
{
int k = threadIdx.x + blockIdx.x * blockDim.x;
if (k < n)
C[k] = A[k] + B[k];
}
[免责声明:代码在浏览器中编写,未经测试,使用风险自负]
在这里,串行代码中的内部和外部循环被每个操作一个 CUDA 线程取代,并且我在代码中添加了限制检查,以便在启动的线程多于所需操作的情况下,不会发生缓冲区溢出。如果内核是这样启动的:
const int n = 128 * 1024;
int blocksize = 512; // value usually chosen by tuning and hardware constraints
int nblocks = n / blocksize; // value determine by block size and total work
madd<<<nblocks,blocksize>>>mAdd(A,B,C,n);
然后,256 个块(每个块包含 512 个线程)将被启动到 GPU 硬件上,以并行执行数组加法操作。请注意,如果输入数据大小无法表示为块大小的整数倍,则需要对块数进行舍入以覆盖完整的输入数据集。
以上所有内容都是针对非常琐碎的操作的 CUDA 范式的极大简化的概述,但也许它为您提供了足够的洞察力,让您可以继续自己的工作。如今 CUDA 已经相当成熟,网络上有很多优质的免费教育材料,您可以使用它来进一步阐明我在这个答案中掩盖的编程模型的许多方面。