Parameters to CUDA kernels

Question

When invoking a CUDA kernel for a specific thread configuration, are there any strict rules on which memory space (device/host) kernel parameters should reside in and what type they should be?

Suppose I launch a 1-D grid of threads with

kernel<<<numblocks, threadsperblock >>> (/*parameters*/)

Can I pass an integer parameter int foo which is a host-integer variable, directly to the CUDA kernel? Or should I cudaMalloc memory for a single integer say dev_foo and then cudaMemcpy foo into devfoo and then pass devfoo as a kernel parameter?

Answer 1

The rules for kernel arguments are a logical consequence of C++ parameter passing rules and the fact that device and host memory are physically separate.

CUDA does not allow passing arguments by reference and you must be careful with pointers.

Specifically, you must pass parameters by value. Passing user-defined types requires that the default copy-constructor or your own copy-constructor (if present) does not contain any memory allocations (heap allocations with "new" or "malloc").

In summary pass-by-value works well for integral, floating point or other primitive types, and simple flat user-defined structs or class objects.