Is this correct way to code function pointers?

Accelerated Computing CUDA CUDA Programming and Performance
1

Hi,

See the below code…

// functions…
device void Add(int* a, int* B );
device void Subtract(int* a, int* B );
device void Division(int* a, int* B );

//kernal function
global void FunctionPointers(int* result);

//function pointer
void (* ptrToFun)(int* x, int* y);

/************************************************************
/
/ ADD() /
/************************************************
**********/
device void Add(int a, int B )
{
*a = *a + *b;
printf(“\nAdd result is %2d\n”, *a);
}

/************************************************************
/
/ Subtract() /
/************************************************
**********/
device void Subtract(int a, int B )
{
*a = *a - *b;
printf(“\mSubtract result is %2d\n”, *a);
}

/************************************************************
/
/ Division() /
/************************************************
**********/
device void Division(int a, int B )
{
if( *b == 0 )
return;

*a = (int)(*a)/(*B );
printf("\nDivision result is %2d\n", *a);

}

/************************************************************
/
/ Example /
/************************************************
***********/
global static void FunctionPointers(int result)
{
ptrToFun = &Division;
int a=2,b=5;
ptrToFun(&a,&b);
*result = a;
}

/************************************************************
/
/ HelloCUDA /
/************************************************
***********/
int main(int argc, char argv)
{
CUT_DEVICE_INIT(argc, argv);

int* device_result = 0;
CUDA_SAFE_CALL( cudaMalloc((void**) &device_result, sizeof(int)));

unsigned int timer = 0;
CUT_SAFE_CALL( cutCreateTimer( &timer));
CUT_SAFE_CALL( cutStartTimer( timer));

FunctionPointers<<<1, 1, 0>>>(device_result);
CUT_CHECK_ERROR("Kernel execution failed\n");

CUDA_SAFE_CALL( cudaThreadSynchronize() );
CUT_SAFE_CALL( cutStopTimer( timer));
printf("Processing time: %f (ms)\n", cutGetTimerValue( timer));
CUT_SAFE_CALL( cutDeleteTimer( timer));

int host_result;
CUDA_SAFE_CALL( cudaMemcpy(&host_result, device_result, sizeof(int), cudaMemcpyDeviceToHost));

if( host_result )
	printf("\nFunction Pointer concept checked successfully\n");
else
	printf("\nFunction Pointer concept checking failed\n");

CUDA_SAFE_CALL( cudaFree(device_result));
CUT_EXIT(argc, argv);

return 0;

}

now My questions are…
1) The handling of function pointers is correct in the above program??
2) Can I call printf() function from device function?

Please answer my questions.

Thanks
Manjunath

2

There is no correct way to handle function pointers. They are not supported on the device. This is documented in the programming guide.

Yes, but only if you are compiling in the device emulation mode.

3

Yes, I just have read the docs. it had given that “device functions cannot have their address taken; function pointers to global functions, on the other hand, are supported”

but When I use the above code, compile and run it, there is no crash [all working fine] not at least a warning. why?

4

Then may b u r using “deviceemu”

5

you mean I compiled and run it in EmuDebug mode???