CUDA & GLM

Apparently their is a kind of incompatibility between CUDA and the template used into GLM.

I don’t achieve to compile a simple kernel which use an array of glm::vec3.

Typical kind of error that I obtain :
core/func_common.hpp(205): error: expected a “)”
core/func_common.hpp(214): error: expected an identifier
gtx/compatibility.hpp(58): error: expected an identifier
gtx/compatibility.hpp(58): error: expected a “)”

I use the 0.9b2 version of GLM and the binary is compiled with nvcc (GCC 4.3).

Is it a known bug or have I made a mistake ?

Thank you in advance.

Oh! BTW, great lib, I love it :slight_smile: Thank Groovounet !

Ahhh… look like some sort of incompatibility. Could you give me a full log so that I could understand what’s going on as I have no clue.

Thanks! :slight_smile:

The output of nvcc for the the GLM files :


/home/derousie/.local/include/Extern/./core/func_common.hpp(205): error: expected an identifier

/home/derousie/.local/include/Extern/./core/func_common.hpp(205): error: expected a ")"

/home/derousie/.local/include/Extern/./core/func_common.hpp(214): error: expected an identifier

/home/derousie/.local/include/Extern/./core/func_common.hpp(214): error: expected a ")"

/home/derousie/.local/include/Extern/./core/func_common.inl(1200): error: expected an identifier

/home/derousie/.local/include/Extern/./core/func_common.inl(1200): error: expected a ")"

/home/derousie/.local/include/Extern/./core/func_common.inl(1215): error: expected an identifier

/home/derousie/.local/include/Extern/./core/func_common.inl(1215): error: expected a ")"

/home/derousie/.local/include/Extern/./core/func_common.inl(1226): error: expected an identifier

/home/derousie/.local/include/Extern/./core/func_common.inl(1226): error: expected a ")"

/home/derousie/.local/include/Extern/./core/func_common.inl(1238): error: expected an identifier

/home/derousie/.local/include/Extern/./core/func_common.inl(1238): error: expected a ")"

/home/derousie/.local/include/Extern/./core/func_common.inl(1251): error: expected an identifier

/home/derousie/.local/include/Extern/./core/func_common.inl(1251): error: expected a ")"

/home/derousie/.local/include/Extern/./core/func_common.inl(1266): error: expected an identifier

/home/derousie/.local/include/Extern/./core/func_common.inl(1266): error: expected a ")"

/home/derousie/.local/include/Extern/./core/func_common.inl(1277): error: expected an identifier

/home/derousie/.local/include/Extern/./core/func_common.inl(1277): error: expected a ")"

/home/derousie/.local/include/Extern/./core/func_common.inl(1289): error: expected an identifier

/home/derousie/.local/include/Extern/./core/func_common.inl(1289): error: expected a ")"

/home/derousie/.local/include/Extern/./gtc/half_float.inl(28): warning: pointless comparison of unsigned integer with zero

/home/derousie/.local/include/Extern/./gtc/half_float.inl(34): warning: pointless comparison of unsigned integer with zero

/home/derousie/.local/include/Extern/./gtc/half_float.inl(299): warning: pointless comparison of unsigned integer with zero

/home/derousie/.local/include/Extern/./gtc/half_float.inl(309): warning: pointless comparison of unsigned integer with zero

/home/derousie/.local/include/Extern/./gtc/half_float.inl(611): warning: pointless comparison of unsigned integer with zero

/home/derousie/.local/include/Extern/./gtc/half_float.inl(621): warning: pointless comparison of unsigned integer with zero

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(58): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(58): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(59): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(59): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(60): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(60): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(61): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(61): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(63): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(63): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(64): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(64): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(65): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(65): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(66): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(66): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(68): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(68): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(69): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(69): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(70): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(70): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(71): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.hpp(71): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(16): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(16): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(27): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(27): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(36): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(36): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(46): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(46): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(58): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(58): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(69): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(69): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(78): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(78): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(88): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(88): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(100): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(100): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(110): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(110): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(119): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(119): error: expected a ")"

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(129): error: expected an identifier

/home/derousie/.local/include/Extern/./gtx/compatibility.inl(129): error: expected a ")"

Thank you in advance :slight_smile:

Hi,

Are you using using namespace glm; somewhere?
I really don’t see from where is come and that’s the only explanation I have so far.

Cuda and GLM might have similar names which involve name collisions.

Yop,

Usually I rename glm namespace into Math. But I don’t think the troubles come from here.

Anyway, I’ve written a simple cuda example :


// nvcc main.cu -o test

//------------------------------------------------------------------------------
// Includes
//------------------------------------------------------------------------------
#include <cstdio>
#include <cassert>
#include <cuda.h>

#include "glm.hpp"
//typedef glm::vec3 	VectorType;
typedef glm::float3 	VectorType;
//typedef float3	VectorType;

VectorType * h_A;
VectorType * h_B;

VectorType * d_A;
VectorType * d_B;

float* h_dot;
float* d_dot;

//------------------------------------------------------------------------------
// Kernel
//------------------------------------------------------------------------------
__global__ void Megadot(	VectorType* _A,
							VectorType* _B,
							float* 		_dot,
							int 		_n)
{
	int index	= blockIdx.x * blockDim.x + threadIdx.x;
	_dot[index] = _A[index].x*_B[index].x + _A[index].y*_B[index].y + _A[index].z*_B[index].z;
}

//------------------------------------------------------------------------------
// Main
//------------------------------------------------------------------------------
int main()
{
	int nElts = 100000;

	h_A   = (VectorType *)malloc(sizeof(VectorType)*nElts);
	h_B   = (VectorType *)malloc(sizeof(VectorType)*nElts);
	h_dot = (float *)malloc(sizeof(float)*nElts);

	cudaMalloc((void **) &d_A, sizeof(VectorType)*nElts);
	cudaMalloc((void **) &d_B, sizeof(VectorType)*nElts);
	cudaMalloc((void **) &d_dot, sizeof(float)*nElts);

	for(int i=0;i<nElts;++i)
	{
		h_A[i].x = 1.f;
		h_A[i].y = 2.f;
		h_A[i].z = 3.f;

		h_B[i].x = 3.f;
		h_B[i].y = 2.f;
		h_B[i].z = 1.f;
	}

	cudaMemcpy(d_A, h_A, nElts*sizeof(VectorType), cudaMemcpyHostToDevice);
	cudaMemcpy(d_B, h_B, nElts*sizeof(VectorType), cudaMemcpyHostToDevice);

	int blocksize = 16;
	int dimBlock  = blocksize;
	int dimGrid( ceil( nElts / (float)blocksize) );
	Megadot <<< dimGrid, dimBlock >>> (	d_A, d_B, d_dot, nElts);
	cudaError_t error = cudaGetLastError();
	if(error!=cudaSuccess)
	{
		printf("CUDA ERROR : %s 
",cudaGetErrorString(error));
		assert(false);
	}
	cudaThreadSynchronize();

	cudaMemcpy(h_dot, d_dot, nElts*sizeof(float), cudaMemcpyDeviceToHost);

	for(int i=0;i<nElts;++i)
	{
		printf("%f
",h_dot[i]);
	}

	cudaFree(d_A);
	cudaFree(d_B);

	free(h_A);
	free(h_B);
	return 0;
}

If you try to compile with float3 (defined by Cuda), it works well, but if you modify the typedef to glm::vec3, you get the following output :


./core/func_common.hpp(205): error: expected an identifier

./core/func_common.hpp(205): error: expected a ")"

./core/func_common.hpp(214): error: expected an identifier

./core/func_common.hpp(214): error: expected a ")"

./core/func_common.inl(1200): error: expected an identifier

./core/func_common.inl(1200): error: expected a ")"

./core/func_common.inl(1215): error: expected an identifier

./core/func_common.inl(1215): error: expected a ")"

./core/func_common.inl(1226): error: expected an identifier

./core/func_common.inl(1226): error: expected a ")"

./core/func_common.inl(1238): error: expected an identifier

./core/func_common.inl(1238): error: expected a ")"

./core/func_common.inl(1251): error: expected an identifier

./core/func_common.inl(1251): error: expected a ")"

./core/func_common.inl(1266): error: expected an identifier

./core/func_common.inl(1266): error: expected a ")"

./core/func_common.inl(1277): error: expected an identifier

./core/func_common.inl(1277): error: expected a ")"

./core/func_common.inl(1289): error: expected an identifier

./core/func_common.inl(1289): error: expected a ")"

main.cu(12): error: namespace "glm" has no member "float3"

21 errors detected in the compilation of "/tmp/tmpxft_00002aac_00000000-4_main.cpp1.ii".

I hope this can help you.

“C for Cuda” is built on top of C, not C++. I don’t think you are allowed to use GLM inside “C for Cuda”.

So I guess, you can use GLM inside your main function but not into your kernel.

I’m not sure about that… but there is probably a clue here.

I think you’re right but not completely. As you said, I should be able to use GLM into the main function but I get an error from the include (and thus before the kernel parsing/compilation). Unless you mean “no GLM into a ‘.cu’ file” (and in this case you’re right, it works perfectly).

According to the specifications :

The front end of the compiler processes CUDA source files according to C++ syntax rules. Full C++ is supported for the host code. However, only a subset of C++ is fully supported for the device code as described in detail in Appendix D. As a consequence of the use of C++ syntax rules, void pointers (e.g., returned by malloc()) cannot be assigned to non-void pointers without a typecast. nvcc also support specific keywords and directives detailed in Appendix E.

Appendix D :
CUDA supports the following C++ language constructs for device code:
 Polymorphism
 Default Parameters
 Operator Overloading
 Namespaces
 Function Templates
 Classes for devices of compute capability 2.0
These C++ constructs are implemented as specified in “The C++ Programming Langue” reference. It is valid to use any of these constructs in .cu CUDA files for host, device, and kernel (global) functions. Any restrictions detailed in previous parts of this programming guide, like the lack of support for recursion, still apply.
The following subsections provide examples of the various constructs.

Thus there is nothing about “class template”. Maybe the problem comes from here.

Anyway, I will use an other library for CUDA. Thank you a lot for this checking !

Welcome!