CUDA 父内核能否启动一个线程数多于父内核的子内核？

Question

我正在尝试学习如何使用 CUDA 动态并行。

我有一个简单的 CUDA 内核，可以创建一些工作，然后启动新内核来执行该工作。假设我启动了只有 1 个线程块的父内核，如下所示：

int nItems = 100
parentKernel<<<1,1>>>(nItems);

现在，我在父内核中创建工作，然后启动子内核，如下所示：

__global__ void parentKernel(int nItems)
{
    // create some work
    
    // invoke child kernel
    childKernel<<<2, 256>>>();
}

请注意，启动子内核时使用的线程和块 (2x256) 多于给定的父内核 (1x1)。

子内核实际上运行 512个线程并行吗？还是父内核必须将其线程分配给子内核？

Answer 1

子内核启动就像您从主机代码启动了一个新内核一样。没有基于管理子内核网格大小的父内核的限制。

您的子内核将运行有 2 个块，每个块有 256 个线程，就像您从主机代码启动它一样。这些都将运行在我能想到的任何 GPU 上“并行”。

Can a CUDA parent kernel launch a child kernel with more threads than the parent?