CUDA线程管理

核函数在主机端启动时，执行会转移到设备上，并且将控制权转移回主机。当核函数在GPU上运行时，主机可以运行其他函数。因此，主机与核函数是异步的。

此时，设备端也就是GPU上会产生大量的线程，并且每个线程都执行由核函数指定的语句。

CUDA是如何组织线程的

CUDA明确了线程层次抽象的概念，是一个两层的线程层次组织结构，由线程块和网格组成。

在主机上调用一次核函数，会启动多个线程，线程的组织方式像上图中所示。一个内核对应启动一个网格，每个网格中包含多个线程块，每个线程块中启动多个线程。

网格中组织线程块的维度和线程的维度由gridDim与blockDim确定。

gridDim：每个网格中，如何组织线程块
blockDim：每个线程块中，如何组织线程

gridDim与blockDim都是dim3数据类型。

每个线程块在网格中的索引与每个线程在线程块中的索引由blockIdx与threadIdx获取。

例如，

__global__ void show_index() {
    printf("grid dim: %d, %d, %d\n", gridDim.x, gridDim.y, gridDim.z);
    printf("block dim: %d, %d, %d\n", blockDim.x, blockDim.y, blockDim.z);
    printf("block index: %d, %d, %d\n", blockIdx.x, blockIdx.y, blockIdx.z);
    printf("thread index: %d, %d, %d\n", threadIdx.x, threadIdx.y, threadIdx.z);
}

int main()
{
    using namespace std;
    
    dim3 grid_dim(2, 2, 1);
    dim3 block_dim(2, 4, 2);

    show_index << <grid_dim,block_dim > >> ();
    cudaDeviceReset();
    return 0;
}

运行结果得到

grid dim: 2, 2, 1
block dim: 2, 4, 2

...
block index: 0, 1, 0
block index: 1, 0, 0
...
block index: 1, 1, 0
block index: 1, 1, 0
...
block index: 0, 0, 0
...

thread index: 0, 0, 0
thread index: 1, 0, 0
thread index: 0, 1, 0
thread index: 1, 1, 0
thread index: 0, 2, 0
thread index: 1, 2, 0
thread index: 0, 3, 0
thread index: 1, 3, 0
thread index: 0, 0, 1
thread index: 1, 0, 1
thread index: 0, 1, 1
thread index: 1, 1, 1
thread index: 0, 2, 1
thread index: 1, 2, 1
thread index: 0, 3, 1
thread index: 1, 3, 1
thread index: 0, 0, 0
thread index: 1, 0, 0
thread index: 0, 1, 0
thread index: 1, 1, 0
thread index: 0, 2, 0
thread index: 1, 2, 0
thread index: 0, 3, 0
thread index: 1, 3, 0
thread index: 0, 0, 1
thread index: 1, 0, 1
thread index: 0, 1, 1
thread index: 1, 1, 1
thread index: 0, 2, 1
thread index: 1, 2, 1
thread index: 0, 3, 1
thread index: 1, 3, 1
thread index: 0, 0, 0
thread index: 1, 0, 0
thread index: 0, 1, 0
thread index: 1, 1, 0
thread index: 0, 2, 0
thread index: 1, 2, 0
thread index: 0, 3, 0
thread index: 1, 3, 0
thread index: 0, 0, 1
thread index: 1, 0, 1
thread index: 0, 1, 1
thread index: 1, 1, 1
thread index: 0, 2, 1
thread index: 1, 2, 1
thread index: 0, 3, 1
thread index: 1, 3, 1