Device Function Pointer Type Checking Bug

I would expect the following piece of code to either generate a compile-time type checking error or run correctly. Currently, it does neither and instead results in a segmentation fault. If the CUDA compiler is going to enforce the “host” and “device” type qualifiers I expect it to be able to reason about function pointers correctly as well. I should either be able to take a function pointer to a device function on the host side and be able to execute it on the device by passing it as an argument to a global function, or I shouldn’t even be able to reference device functions on the host side. Perhaps mandating type qualifiers on function pointers is an acceptable solution.

[codebox]device

void printThread()

{

printf(“This is thread %d in block %d”, threadIdx.x, blockIdx.x);

}

global

void runKernel(void (*function)())

{

(*function)();

}

host

int main()

{

void (*func)() = printThread;

runKernel<<<4,128>>>(func);

cudaThreadSynchronize();

return 0;

}[/codebox]