如何将设备函数作为输入参数传递给主机端函数? [英] How to pass device function as an input argument to host-side function?

问题描述

我只想将设备函数作为宿主函数的参数传递，原因是宿主函数可以使用该设备端函数启动某些内核.

I just want to pass device function as argument of a host function, of cause, the host function then can launch some kernels with this device side function.

我尝试了普通的C ++方式(通过指针/引用传递)，CUDA调试器告诉我内核无法启动.

I tried the usual C++ way (pass by pointer/reference) and the CUDA debugger told me the kernel cannot launch.

更新:

我想做的是:

__host__ void hostfunction(int a, int (*DeviceFunction)(int))
{
   /...do something.../
   somekernel<<<blocks, threads>>>(int * in, DeviceFunction);
}

并使用以下命令启动主机:

And launch the host with:

hostfunction(x, &SomeDeviceFunctionTemplate<int>);

推荐答案

此示例可能很有趣:

$ cat t237.cu
#include <stdio.h>


__device__ int f1(){ printf("dev f1\n"); return 0;}
__device__ int f2(){ printf("dev f2\n"); return 0;}
__device__ int f3(){ printf("dev f3\n"); return 0;}

__device__ int *fptrf1 = (int *)f1;
__device__ int *fptrf2 = (int *)f2;
__device__ int *fptrf3 = (int *)f3;


__global__ void mykernel(int (*fptr)()){

  fptr();
  printf("executed\n");
}

int main(){

  int *hf1, *hf2, *hf3;
  cudaMemcpyFromSymbol(&hf1, fptrf1, sizeof(int *));
  cudaMemcpyFromSymbol(&hf2, fptrf2, sizeof(int *));
  cudaMemcpyFromSymbol(&hf3, fptrf3, sizeof(int *));
  mykernel<<<1,1>>>((int (*)())hf1);
  cudaDeviceSynchronize();
  mykernel<<<1,1>>>((int (*)())hf2);
  cudaDeviceSynchronize();
  mykernel<<<1,1>>>((int (*)())hf3);
  cudaDeviceSynchronize();
  return 0;
}
$ nvcc -arch=sm_20 -O3 -o t237 t237.cu
$ ./t237
dev f1
executed
dev f2
executed
dev f3
executed
[bob@cluster1 misc]$

我认为这大致符合Jared的建议. 正如他提到的那样，这在主机代码中是不可能的:

I think this is roughly along the lines of what Jared was suggesting. As he mentioned, this will not be possible in host code:

&SomeDeviceFunctionTemplate<int>

假设SomeDeviceFunctionTemplate表示__device__函数.

这篇关于如何将设备函数作为输入参数传递给主机端函数?的文章就介绍到这了，希望我们推荐的答案对大家有所帮助，也希望大家多多支持IT屋！