CUDA C Programming

线程管理

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

关系：

1. 由一个内核启动所产生的所有线程统称为一个网格。同一个网格中的所有线程共享相同的内存空间。
2. 一个网格由多个线程块构成，一个线程块包含一组线程，同一个线程块内的线程协作通过：“同步”，“共享内存”的方式来实现，不同块内的线程不能协作。
3. 线程依靠两个坐标变量来区分彼此

• blockIdx : 线程块在线程网格内的索引
• threadIdx : 块内的线程索引

4. CUDA可以组织三维的网格和块。网格和块的维度由两个内置变量指定。

• gridDim : 线程格的维度，用每个线程格中的线程数来表示
• blockDim : 线程块的维度，用每个线程块中的线程数来表示

5. 一个线程格会被组织成线程块的二维数组形式，一个线程块会被组织成线程的三维数组形式。

#include <cuda_runtime.h>
#include <stdio.h>

__global__ void CheckIndex()
{
    // 核函数中，每个线程都输出自己的线程索引，块索引，块维度，网格维度
    printf("threadIdx: (%d, %d, %d) \tblockIdx: (%d, %d, %d) \t\
            blockDim: (%d, %d, %d) \tgridDim: (%d, %d, %d)\n",
            threadIdx.x, threadIdx.y, threadIdx.z, 
            blockIdx.x, blockIdx.y, blockIdx.z, 
            blockDim.x, blockDim.y, blockDim.z, 
            gridDim.x, gridDim.y, gridDim.z);
}

int main()
{
    int n_element = 6;

    dim3 block = 3;
    dim3 grid = ((n_element + block.x - 1) / block.x);

    // 主机端，检查块和网格维度
    printf("block.x = %d, block.y = %d, block.z = %d\n", block.x, block.y, block.z);
    printf("grid.x =  %d, grid.y =  %d, grid.z =  %d\n", grid.x, grid.y, grid.z);

    CheckIndex<<<grid, block>>>();
    cudaDeviceReset();

    return 0;
}

block.x = 3, block.y = 1, block.z = 1
grid.x =  2, grid.y =  1, grid.z =  1
threadIdx: (0, 0, 0)  blockIdx: (0, 0, 0)               blockDim: (3, 1, 1)   gridDim: (2, 1, 1)
threadIdx: (1, 0, 0)  blockIdx: (0, 0, 0)               blockDim: (3, 1, 1)   gridDim: (2, 1, 1)
threadIdx: (2, 0, 0)  blockIdx: (0, 0, 0)               blockDim: (3, 1, 1)   gridDim: (2, 1, 1)
threadIdx: (0, 0, 0)  blockIdx: (1, 0, 0)               blockDim: (3, 1, 1)   gridDim: (2, 1, 1)
threadIdx: (1, 0, 0)  blockIdx: (1, 0, 0)               blockDim: (3, 1, 1)   gridDim: (2, 1, 1)
threadIdx: (2, 0, 0)  blockIdx: (1, 0, 0)               blockDim: (3, 1, 1)   gridDim: (2, 1, 1)