CUDA中的线程层次

线程层次

为方便起见，threadIdx 是一个 3 分量向量，因此可以使用一维、二维或三维的线程索引来识别线程，形成一个一维、二维或三维的线程块，称为block。这提供了一种跨域的元素（例如向量、矩阵或体积）调用计算的方法。

线程的索引和它的线程 ID 以一种直接的方式相互关联：对于一维块，它们是相同的；对于大小为(Dx, Dy)的二维块，索引为(x, y)的线程的线程ID为(x + y*Dx)；对于大小为 (Dx, Dy, Dz) 的三维块，索引为 (x, y, z) 的线程的线程 ID 为 (x + y*Dx + z*Dx*Dy)。

例如，下面的代码将两个大小为NxN的矩阵A和B相加，并将结果存储到矩阵C中:

// Kernel definition
__global__ void MatAdd(float A[N][N], float B[N][N],float C[N][N])
{int i = threadIdx.x;int j = threadIdx.y;C[i][j] = A[i][j] + B[i][j];
}int main()
{...// Kernel invocation with one block of N * N * 1 threadsint numBlocks = 1;dim3 threadsPerBlock(N, N);MatAdd<<<numBlocks, threadsPerBlock>>>(A, B, C);...
}

每个块的线程数量是有限制的，因为一个块的所有线程都应该驻留在同一个处理器核心上，并且必须共享该核心有限的内存资源。在当前的gpu上，一个线程块可能包含多达1024个线程。

但是，一个内核可以由多个形状相同的线程块执行，因此线程总数等于每个块的线程数乘以块数。

块被组织成一维、二维或三维的线程块网格(grid)，如下图所示。网格中的线程块数量通常由正在处理的数据的大小决定，通常超过系统中的处理器数量。

<<<...>>> 语法中指定的每个块的线程数和每个网格的块数可以是 int或 dim3 类型。如上例所示，可以指定二维块或网格。

网格中的每个块都可以由一个一维、二维或三维的惟一索引标识，该索引可以通过内置的blockIdx变量在内核中访问。线程块的维度可以通过内置的blockDim变量在内核中访问。

扩展前面的MatAdd()示例来处理多个块，代码如下所示。

// 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 invocationdim3 threadsPerBlock(16, 16);dim3 numBlocks(N / threadsPerBlock.x, N / threadsPerBlock.y);MatAdd<<<numBlocks, threadsPerBlock>>>(A, B, C);...
}

线程块大小为16x16(256个线程)，尽管在本例中是任意更改的，但这是一种常见的选择。网格是用足够的块创建的，这样每个矩阵元素就有一个线程来处理。为简单起见，本例假设每个维度中每个网格的线程数可以被该维度中每个块的线程数整除，尽管事实并非如此。

程块需要独立执行：必须可以以任何顺序执行它们，并行或串行。这种独立性要求允许跨任意数量的内核以任意顺序调度线程块，如下图所示，使程序员能够编写随内核数量扩展的代码。

块内的线程可以通过一些共享内存共享数据并通过同步它们的执行来协调内存访问来进行协作。更准确地说，可以通过调用 __syncthreads() 内部函数来指定内核中的同步点； __syncthreads() 充当屏障，块中的所有线程必须等待，然后才能继续。 Shared Memory 给出了一个使用共享内存的例子。除了 __syncthreads() 之外，Cooperative Groups API 还提供了一组丰富的线程同步示例。

为了高效协作，共享内存是每个处理器内核附近的低延迟内存（很像 L1 缓存），并且 __syncthreads() 是轻量级的。

CUDA中的线程层次相关推荐

深入理解CUDA线程层次以及关于设置线程数的思考
深入理解CUDA线程层次以及关于设置线程数的思考 2015-09-16 08:45 215人阅读评论(0) 收藏举报分类: cuda(24) GPU线程以网格(grid)的方式组织,而每个网格中 ...
GPU（CUDA）学习日记（十一）------ 深入理解CUDA线程层次以及关于设置线程数的思考
GPU(CUDA)学习日记(十一)------ 深入理解CUDA线程层次以及关于设置线程数的思考标签: cuda存储线程结构网格 2012-12-07 16:30 6298人阅读评论(4)收藏举 ...
CUDA中SM对线程块的调度
sm流处理器簇对blocks的调度策略在cuda中,GPU中的SM(比如GTX650有两个SM处理器)被CPU调度器把线程块逐个分配到SM上,每个SM同时处理这个被分配的线程块,但是每次每个时刻只能 ...
CUDA中grid、block、thread、warp与SM、SP的关系
首先概括一下这几个概念.其中SM(Streaming Multiprocessor)和SP(streaming Processor)是硬件层次的,其中一个SM可以包含多个SP.thread是一个线程, ...
[原]CUDA中grid、block、thread、warp与SM、SP的关系
[原]CUDA中grid.block.thread.warp与SM.SP的关系 2015-3-27阅读209 评论0 首先概括一下这几个概念.其中SM(Streaming Multiprocessor ...
CUDA中的一些基本概念
线程线程是CUDA中并行程序的基本构建,一个线程就是程序中国的一个单一的执行流,就像一件衣服上的一块棉,一块块棉交织在一起组成衣服,同样一个个线程组成成并行程序. 随着处理器的核越来越多,硬件可以 ...
c语言 cuda核函数,CUDA核函数与线程配置
CUDA核函数在GPU上调用的函数成为CUDA核函数(Kernel function),核函数会被GPU上的多个线程执行.每个线程都会执行核函数里的代码,当然由于线程编号的不同,执行的代码路径可能会 ...
java如何给一个链表定义和传值_如何在CUDA中为Transformer编写一个PyTorch自定义层...
如今,深度学习模型处于持续的演进中,它们正变得庞大而复杂.研究者们通常通过组合现有的 TensorFlow 或 PyTorch 操作符来发现新的架构.然而,有时候,我们可能需要通过自定义的操作符来实现 ...
CUDA性能优化----线程配置
CUDA性能优化----线程配置 2017-01-12 14:19:29| 分类: HPC&CUDA优化 | 标签:cuda gpu hpc |举报 |字号订阅下载LOFTER ...

CUDA中的线程层次

线程层次

CUDA中的线程层次相关推荐

最新文章

热门文章