warning: inline qualifier ignored for "global" function

Question

I have a templated function in a header file and getting some warnings when compiling with CUDA.

warning: inline qualifier ignored for "global" function

template<typename A,typename B>
__global__ inline void functionA(ar1, arg2)
{}

Can someone help on this please?

Answer 1

When a function is "inlined" it means that the code to be executed inside the function is promoted into the calling function, thereby avoiding the overhead of saving registers, jumping into the function, and restoring the registers afterwards (search online for "ABI" for details on this).

It's not possible to inline a kernel call, it makes no sense because the processor executing the kernel code (the GPU) is not the same as the processor launching the kernel (the CPU).

Even with dynamic parallelism it makes no sense since the semantics mean that the child kernel can be run anywhere, not necessarily on the same SM.

Answer 2

Well, for me it was a huge problem because I wanted to have a common global function for many different files, and leaving it in a hpp file would lead to either a linking error (multiple definitions) or to this warning, if I added the inline keyword.

Solution: Add a inline device function instead, and create an auxiliary global function inside each desired file that simply calls this device function. That makes the trick.