CUDA中grid、block、thread的设置简介

本文参考https://blog.csdn.net/hujingshuang/article/details/53097222上的内容

不同在于，本文 将对原文进行补充，同时澄清一些易混淆的概念

CUDA中

grid、block、thread只要在最大值范围内，可以随意设置，最后系统会根据内部算法重新分配。因此grid、block、thread的设置其实只以你的“目的”有关。也就是说这几个值的索引如何方便如何来，要处理的矩阵就用二维，要处理的向量就用一维，要处理的张量就用三维

在设置上，变量是遵循向下设置的， 即

dim3 grid(1, 1, 1)表示的是一个grid由一个block组成

dim3 grid(4, 1, 1)表示的是一个grid由四个一维block组成，不要混淆为有4个grid

block(8, 1, 1)表示的是一个block由1维thread组成，1乘8的一维分布，不要混淆为有8个block

block(4, 2, 1)表示的是一个block由2维thread组成，4乘2的二维分布，不要混淆为有4个长度为2的block

这就是为什么没有threadDim.x，threadDim.y，threadDim.z的原因，如果你在window下配置CUDA，用VS编写的时候会发现没有threadDim变量 

 threadIdx.x，blockIdx.x，gridIdx.x下标都是从0开始 

因为本质上是在网格索引，所以从x方向开始查找和从y方向开始查找结果是一样的。对于一个block(4, 2, 1)而言

int tid = threadIdx.x*blockDim.y + threadIdx.y;

int tid = threadIdx.y*blockDim.x + threadIdx.x;

以上两个索引结果等价。

在索引的时候要计算，不要自己套 公式，例如：

dim3 grid(8, 1, 1), block(2, 1, 1); 的情况，索引就是： 

int tid = blockIdx.x * blockDim.x + threadIdx.x;而不是原文中的blockIdx.x * gridDim.x + threadIdx.x， 原文只是恰巧gridDim.x = blockDim.x  = 4 而已

在dim3 grid(2, 2, 1), block(2, 2, 1);   这种block和thread都是二维分布的情况下：

    int tid = (blockIdx.y * gridDim.x + blockIdx.x)*(blockDim.x * blockDim.y) 
              + threadIdx.y*blockDim.x + threadIdx.x;

    int tid = (blockIdx.x * gridDim.y + blockIdx.y)*(blockDim.x * blockDim.y) 
              + threadIdx.x*blockDim.y + threadIdx.y;

以上是等价 的

在三维情况下，更是多种等价的查询，比如可以走列切片，或是走横切片，或是走高度切片，其实意义不大，只要按顺序索引就行。