了解CUDA网格维度，块维度和线程组织（简单说明）

线程如何组织成由GPU执行？

例如，如果一个GPU设备有4个多处理单元，并且每个单元可以运行768个线程，那么在给定的时刻，不会有超过4 * 768个线程真正并行运行（如果您计划了更多的线程，它们将等待轮到他们）。

线程按块组织。块由多处理单元执行。一个块的线程可以使用1Dimension（x），2Dimension（x，y）或者3Dim索引（x，y，z）来识别（索引），但是对于我们的例子，在任何情况下x y z <= 768到x，y，z，看指南和您的设备能力）。

显然，如果你需要比4 * 768多的线程，你需要多于4个块。块也可以索引为1D，2D或3D。有一排等待进入GPU的块（因为在我们的例子中，GPU有4个多处理器，同时只有4块正在执行）。

假设我们想要一个线程来处理一个像素（i，j）。

我们可以使用每个64个线程的块。那么我们需要512 * 512/64 = 4096块（所以512 * 512线程= 4096 * 64）

组织（使图像索引更容易）blockDim = 8 x 8（每块64个线程）的2D块中的线程是很常见的。我更喜欢把它称为threadsPerBlock。

dim3 threadsPerBlock(8, 8); // 64 threads

2D gridDim = 64×64块（需要4096块）。我更喜欢把它叫做numBlocks。

 dim3 numBlocks(imageWidth/threadsPerBlock.x, /* for instance 512/8 = 64*/ imageHeight/threadsPerBlock.y);

内核是这样启动的：

 myKernel <<<numBlocks,threadsPerBlock>>>( /* params for the kernel function */ );

最后：会有类似“4096个块的队列”，其中一个块正在等待GPU的多处理器之一分配，以执行其64个线程。

在内核中，由线程处理的像素（i，j）是这样计算的：

 uint i = (blockIdx.x * blockDim.x) + threadIdx.x; uint j = (blockIdx.y * blockDim.y) + threadIdx.y;

假设一个9800GT GPU：14个多处理器，每个处理器有8个线程处理器，warpsize是32，这意味着每个线程处理器最多处理32个线程。 14 * 8 * 32 = 3584是实际cuncurrent线程的最大数量。

如果你用超过3584个线程来执行这个内核（比如说4000线程，你怎么定义块和网格并不重要，gpu会像对待它们一样）：

 func1(); __syncthreads(); func2(); __syncthreads();

那么这两个函数的执行顺序如下：

1.func1被执行的第一个3584线程

2.func2被执行的第一个3584线程

3.为其余线程执行func1

4.为其余线程执行func2

CUDA编程指南应该是一个很好的开始。我也build议从这里查看CUDA的介绍。