Hi,
I have the following code in a single main.cu
file:
namespace my
{
using Arr = float[5];
void __global__ foo(Arr const){};
}
int main(int argc, char *[])
{
my::foo<<<1, 1>>>(nullptr);
return 0;
}
I compile it with the command nvcc main.cu
. Also, I tried to build it with the CMake and set different properties, like architecture, optimization, standard version, etc. On my machine (Ubuntu 18.04) I have the following error:
nvcc main.cu
In file included from tmpxft_00004f21_00000000-6_main.cudafe1.stub.c:1:0:
main.cu:5:21: error: ‘_ZN2my3ArrE’ does not name a type
void __global__ foo(Arr const){};
^~~~~~~~~~~
main.cu:5:41: error: ‘void my::foo(int)’ should have been declared inside ‘my’
void __global__ foo(Arr const){};
^
In file included from tmpxft_00004f21_00000000-6_main.cudafe1.stub.c:1:0:
main.cu: In function ‘void my::foo(int)’:
main.cu:5:40: error: invalid conversion from ‘int’ to ‘const float*’ [-fpermissive]
void __global__ foo(Arr const){};
^
In file included from tmpxft_00004f21_00000000-6_main.cudafe1.stub.c:1:0:
/tmp/tmpxft_00004f21_00000000-6_main.cudafe1.stub.c:13:6: note: initializing argument 1 of ‘void __device_stub__ZN2my3fooEPKf(const float*)’
void __device_stub__ZN2my3fooEPKf(const float *__par0){__cudaLaunchPrologue(1);__cudaSetupArgSimple(__par0, 0UL);__cudaLaunch(((char *)((void ( *)(const float *))my::foo)));}
It complains about the Arr const
argument of the kernel foo
. If I remove the const
keyword, the code compiles. Also, if I move the using
directive to the global namespace, it compiles even with the const
. I tried NVCC 11.4 and 10.2. Also, I tried GCC 7.5 and 7.3 as the host compiler.
The same code compiles fine on godbolt and this confuses me a lot. Is it the problem in NVCC? Or OS? Or the host compiler?