parallel-processing - CUDA内核如何启动？

Question

我创建了一个简单的CUDA应用程序以添加两个矩阵。它编译正常。我想知道内核将如何通过所有线程启动，以及流将在CUDA内部发生什么？我的意思是，每个线程将以哪种方式执行矩阵的每个元素。

我知道这是一个非常基本的概念，但我不知道。我对流程感到困惑。

Answer 1

您将启动一个网格块。

块是不可分割地分配给多处理器的(其中，多处理器上的块数决定了可用共享内存的数量)。

块进一步分解为扭曲。对于32个线程的Fermi GPU，它们要么执行相同的指令，要么处于非事件状态(因为它们分支了，例如，比同一个warp中的邻居更早退出循环，或者不采用自己执行的if)。在Fermi GPU上，一次最多可以在一个多处理器上运行两个扭曲。

每当有延迟(即执行停顿以完成内存访问或完成数据依赖性)时，都会运行另一种warp(适合一个多处理器的warp数量(相同或不同的块)取决于每个处理器使用的寄存器数量)。线程和一个或多个块使用的共享内存量)。

此调度透明地进行。也就是说，您不必考虑太多。
但是，您可能要使用预定义的整数矢量threadIdx(块中我的线程在哪里？)，blockDim(一个块有多大？)，blockIdx(网格中我的块在哪里？)和gridDim(我的块在网格中多少？)网格？)以在线程之间分配工作(读取:输入和输出)。您可能还想了解如何有效访问不同类型的内存(以便可以在单个事务中为多个线程提供服务)-但这是主题。

NSight提供了一个图形调试器，使您可以轻松了解行话丛林中设备上正在发生的事情。对于您在调试器中看不到的那些东西(例如，停顿原因或内存压力)，其探查器也是如此。

您可以通过另一个内核启动来同步网格中的所有线程(所有线程)。
对于不重叠的顺序内核执行，不需要进一步的同步。

一个网格(或一个内核运行-但是您要调用它)中的线程可以使用原子操作(用于算术)或适当的内存隔离(用于加载或存储访问)通过全局内存进行通信。

您可以将一个块内的所有线程与内部指令__syncthreads()同步(之后所有线程将处于事件状态-尽管像往常一样，在Fermi GPU上最多可以运行两个扭曲)。一个块内的线程可以使用原子操作(用于算术)或适当的内存屏障(用于加载或存储访问)，通过共享或全局内存进行通信。

如前所述，扭曲中的所有线程总是“同步的”，尽管有些可能是不事件的。他们可以通过共享或全局内存进行通信(或在即将发布的具有计算能力3的硬件上进行“ channel 交换”)。您可以使用原子运算(用于算术运算)和限定volatile的共享或全局变量(在同一warp中顺序发生的加载或存储访问)。 volatile 限定符告诉编译器始终访问内存，而从不访问其他线程无法查看其状态的寄存器。

此外，还有整个经纱范围内的投票功能，可以帮助您做出分支决策或计算整数(前缀)总和。

好的，基本上就是这样。希望能有所帮助。写得很好:-)。