CUDA Bug? nested structs / compiler assertion failure short and simple example, would take you 5 sec

very short and simple, could use some insight.

trying to compile:

struct S


	float2 s[1];


__global__ void f(float t, S s)



void g()


	float t = 0;

	S s;

	f<<<1,1>>>( t , s );


fails with nvcc 3.0, windows 7, 64 bit.

builds with nvcc 2.3, windows xp, 32 bit

those are the only machines i have for testing.

nothing necessary for the error has been elided.

gives a compilation error “error PRJ0019: A tool returned an error code from …”

more specifically nvcc seems to give up with “compiler error in file … during data layout phase: unexpected field mtype

most interesting are ways around the error:

  • switch float2 to float

  • switch the order of the t and s parameters

successfully compiles with either of those changes.

any thoughts? is this a bug? am i doing something stupid? anyone willing to try building it on their machine? any help is appreciated.

googling “unexpected field mtype” shows that apparently i am the only one with this error…

Have you tried nvcc 3.1b? Seems to work for me with a recent compiler:

[codebox]C:\dev>nvcc warning: variable “s” is used before its value is set


tmpxft_00000f98_00000000-8_test.cudafe2.gpu warning: variable “s” is used before its value is set



c:\dev\ : warning C4700: uninitialized local variable ‘s’ used

LIBCMT.lib(crt0.obj) : error LNK2019: unresolved external symbol _main reference

a.exe : fatal error LNK1120: 1 unresolved externals

C:\dev>nvcc --version

nvcc: NVIDIA ® Cuda compiler driver

Copyright © 2005-2010 NVIDIA Corporation

Built on Wed_Jun__9_21:21:54_PDT_2010

Cuda compilation tools, release 3.1, V0.2.1221[/codebox]

stuff like this is the reason I never upgraded to 3.0. Still waiting for the “service pack”. ;)