[Driver API] Kernel parameters

Accelerated Computing CUDA CUDA Programming and Performance
1

Hello there,

I have a weird problem with kernel parameters.

Here’s the kernel prototype :

__global__ void projPlanGPU_kernel(float * gOut, 

													 float dt, 

													 float2 dtache,

													 unsigned int wima,

													 unsigned int hima)

I’m setting the parameters this way :

int offset = 0;

	

		void* ptr = (void*)(size_t)d_output;

	ALIGN_UP(offset, __alignof(ptr));

	cuParamSetv(projPlanGPU, offset, &ptr, sizeof(ptr));

		offset += sizeof(ptr);

	ALIGN_UP(offset, __alignof(dt));

	cuParamSetf(projPlanGPU, offset, dt);

	offset += sizeof(dt);

	ALIGN_UP(offset, 8);

	cuParamSetv(projPlanGPU, offset, &dtache, sizeof(dtache));

	offset += sizeof(dtache);

	ALIGN_UP(offset, __alignof(wima));

	cuParamSeti(projPlanGPU, offset, wima);

		offset += sizeof(wima);

	

	ALIGN_UP(offset, __alignof(hima));

	cuParamSeti(projPlanGPU, offset, hima);

		offset += sizeof(hima);

	

	cuParamSetSize( projPlanGPU, offset );

In this configuration, i have an error after kernel launch when i do a cuCtxSynchronize().

1 - If i remove ‘dtache’ parameter, i get the same error.

2 - If i remove ‘dt’ parameter, i don’t have the error anymore.

At the begining i thought of a maximum number or size for parameters, but it seems i’m far from it (256 bytes). Moreover i’m surprised by the behaviour when i remove one parameter…

Any idea would be of great help because i cannot see where i missed a thing !

Thanks in advance =)

/******************************/

While writing the post i thought of an alignment problem… And actually, if my offset is modulo 8, everything’s running fine. When i’m modulo 4, i get the error.

Maybe i missed a thing in programming guide, but i still wanted to post my problem just in case someone have the same.

Problem resovled then, but still, if anyone can confirm this behaviour, i’d be interested !

Thanks again !

2

Well, it seems i was tired yesterday. I don’t have alignment problems. My only problem was that i did not declare my kernel as extern “C”. I cannot really explain why i had those problems but i resolved a second problem i had which was the kernel entry name.

Each time i changed the number of parameters, the kernel entry changed in ptx, so i had to copy it each time. It looked like that : _Z18projPlanGPU_kernelPfjjjff6float2jj

Now everything seems fine.

Good day =)