重载 cuda 核函数

Question

在 CUDA 中使用重载的内核函数时遇到问题。

我可以理解 CUDA 可以通过参数启动重载函数。

但是，如果我想使用cudaOccupancyMaxPotentialBlockSize() 来计算最大占用率的块大小，请参阅doc。

__global__ void foo_cuda_kernel(int a)
{
  /*implementation 1*/
}

//overloaded kernel function
__global__ void foo_cuda_kernel(int a, int b)
{
  /*implementation 2*/
}

void foo_cuda()
{
  int min_grid_size, grid_size, block_size;
  cudaOccupancyMaxPotentialBlockSize
  (
    &min_grid_size, &block_size, 
    foo_cuda_kernel, //how does it distinguish overloaded functions?
    0, thread_num
  );
  grid_size = (thread_num + block_size - 1) / block_size;
  
  //I can understand compiler can distinguish the launched function by its arguments
  foo_cuda_kernel<<<grid_size, block_size>>>((int)1);
  cudaDeviceSynchronize();
}

如何让它发挥作用？ cudaOccupancyMaxPotentialBlockSize() 如何区分重载函数？

Answer 1

如 cmets 中所述，您可以将函数强制转换为指向正确特化的指针：

auto foo_ii = static_cast<void (*)(int, int)>(&foo_cuda_kernel);
auto foo_i = static_cast<void (*)(int)>(&foo_cuda_kernel);

然后您将foo_i 或foo_ii 传递给cudaOccupancyMaxPotentialBlockSize，具体取决于您需要的函数版本。

这将起作用，因为工具链静默发出主机样板函数，这些函数包装底层运行时 API 调用以运行内核并强制内核参数类型检查。宿主编译器将这些包装器视为任何其他宿主函数（因为它们是），并自动选择匹配的版本。