线程的索引和它的线程ID以一种简单的方式相互关联，threadIdx、blockIdx、blockDim是内置变量，用于索引线程位置。线程索引方式对于并行编程至关重要，它制约着内核函数中对循环的改写方式。

线程块需要独立执行:必须能够以任何顺序执行它们，并行执行或串行执行。这种独立性要求允许以任意顺序跨任意数量的内核调度线程块。块中的线程可以通过一些共享内存共享数据，并通过同步它们的执行来协调内存访问进行协作。更准确地说，可以通过调用内部函数来指定内核中的同步点;作为一个屏障，在这个屏障上，块中的所有线程都必须等待，然后才允许任何线程继续。

1.一维线程、一维线程块索引

threadIdx是一个三维矢量，包括threadIdx.x、threadIdx.y、threadIdx.z，对于函数主题设置为一维的线程的索引，仅用到threadIdx.x，下面展示一维线程索引。

// Kernel definition
__global__ void VecAdd(float* A, float* B, float* C)
{
    int i = threadIdx.x;
    C[i] = A[i] + B[i];
}

int main()
{
    ...
    // 1个线程网格上的线程块个数，N：一个线程块上的线程个数
    VecAdd<<<1, N>>>(A, B, C);
    ...
}

2.二维线程、二维线程块索引

线程和线程块均按二维排列的情况如下图，blockIdx为block索引的内置变量，是一个三维矢量；blockDim为一个block块包含的线程规模，也是一个三维矢量（例如下图中一个block中包含4列3行线程，因此blockDim.x大小为4，blockDim.y大小为3）。因此图中Thread（0，0）的线程索引为
i=blockIdx.x * blockDim.x+threadIdx.x;
j=blockIdx.y * blockDim.y + threadIdx.y;

下面是一个二维线程、二维线程块索引的示例。

// Kernel definition
__global__ void MatAdd(float A[N][N], float B[N][N],
float C[N][N])
{
    int i = blockIdx.x * blockDim.x + threadIdx.x;
    int j = blockIdx.y * blockDim.y + threadIdx.y;
    if (i < N && j < N)
        C[i][j] = A[i][j] + B[i][j];
}

int main()
{
    ...
    // Kernel invocation
    dim3 threadsPerBlock(16, 16);
    dim3 numBlocks(N / threadsPerBlock.x, N / threadsPerBlock.y);
    MatAdd<<<numBlocks, threadsPerBlock>>>(A, B, C);
    ...
}

对于其它维度的线程块和线程索引，可以依照这样的思想类比得到。这里不再赘述，有疑问的可以在评论区交流。