bug report, static global CUDA function?

class ZNCudaTemplateTest_ApiObj_V0

{

public:

	static __device__ void testFunc(int &x)

	{

  ++x;

	}

};

template <class ApiObj>

static __global__ void testTemplate()

{

	int x = 10;

	ApiObj::testFunc(x);

}

int main(int argc, const char **argv)

{

	testTemplate<ZNCudaTemplateTest_ApiObj_V0><<<256, 16>>>();

	return 0;

}

A pretty simple test to kill the compiler…the message is:

...

/home/sdk/ZNCudaTemplateTest.cu(27): internal error: 

          assertion failed: get_scope_for_routine_definition: scope is NULL

          (/home/buildmeister/build/rel/gpgpu/toolkit/r2.0/compiler/edg/EDG_3.9

          /src/c_gen_be.c, line 8878)

...

But if I remove the “static” keyword for the global function, it works. I know most SDK examples uses static keyword for device function but none of them use “static” for global function. What does that static keyword mean exactly? (Sorry maybe it sounds more like a C/C++ language problem, but nvcc treats it differently)

Thanks!

How are you attempting to build this code? Also, please provide the output from “nvcc -V”.

thanks,
Lonni