1

考虑以下两个函数模板：

template <typename... Params>
void foo(Params... params)
{
    /* etc etc */
    my_kernel<<<grid_dims, block_dims, shmem_size, stream_id>>>(params...);
}

和：

template <typename... Params> 
void bar(Params... params)
{
    /* etc etc */
    void* arguments_ptrs[sizeof...(Params)];
    auto arg_index = 0;

    for_each_argument(
        [&](auto param) {arguments_ptrs[arg_index++] = &param;},
        params...);

    cudaLaunchKernel<decltype(my_kernel)>(
        &my_kernel, grid_dims, block_dims, argument_ptrs, shmem_size, stream_id);
}

由Sean Parent定义for_each_argument。

问题：

  • foo和的语义是否bar完全相同？
  • 使用其中一个有什么好处吗？（例如，也许第一种形式在引擎盖下进行堆分配或其他东西......）
  • 在第二个函数中使用转发引用是个好主意吗？两个功能？
4

2 回答 2

1

foo 和 bar 的语义是否完全相同？

我还没有签入 CUDA 9，但在此之前，没有。语法被内联<<<>>>扩展为 API 调用和包装函数调用。有趣的是，所使用的内核启动 API早已被弃用。但是包装函数允许在编译时进行显式的参数类型安全检查，这很有帮助。

[编辑：我检查了 CUDA 9.1，它仍然cudaLaunch像所有以前版本的运行时 API 一样使用]

使用其中一个有什么好处吗？（例如，也许第一种形式在引擎盖下进行堆分配或其他东西......）

不是我知道的。

在第二个函数中使用转发引用是个好主意吗？两个功能？

如果内核在与调用代码相同的编译单元范围内编译，则不会。工具链自动发出内核的前向声明。

于 2018-02-01T07:44:20.243 回答
0

请记住，最终，运行时 API 需要进行驱动程序 API 调用（假设它不会进行我们不知道的秘密 APIcuLaunchKernel()调用），所以最终使用的是：

CUresult cuLaunchKernel ( 
    CUfunction f, 
    unsigned int  gridDimX, 
    unsigned int  gridDimY,
    unsigned int  gridDimZ, 
    unsigned int  blockDimX,
    unsigned int  blockDimY, 
    unsigned int  blockDimZ,
    unsigned int  sharedMemBytes,
    CUstream hStream, 
    void** kernelParams, 
    void** extra )

那是一个非模板化的接口，它不关心引用类型等。

当然，有两种方法可以指定启动参数 - usingkernelParams和 using extra。因此，如果您想调整启动内核的方式，您可能只想尝试一下。

于 2022-01-25T16:22:07.893 回答