Assigning function pointers from another file

Accelerated Computing CUDA CUDA Programming and Performance
1

Hi,
I want to use device function pointers for a plugin interface. Unfortunately it seem so, that assigning function pointer has to be done in the same *.cu file as the kernel is called. Why?
When I use this code and call the kernel everything works.

SOMKernel.cu:

typedef float (*pDistanceFu) (float, float);
__device__ pDistanceFu pBubble 		= ANN::fcn_bubble_nhood; 
__device__ pDistanceFu pGaussian 	= ANN::fcn_gaussian_nhood; 
__device__ pDistanceFu pCutGauss 	= ANN::fcn_cutgaussian_nhood; 
__device__ pDistanceFu pMexican 	= ANN::fcn_mexican_nhood; 
__device__ pDistanceFu pEpanech 	= ANN::fcn_epanechicov_nhood;

void AssignDistanceFunction(ANN::DistFunction &DistFunc) {
	pDistanceFu hBubble; 
	pDistanceFu hGaussian; 
	pDistanceFu hCutGauss; 
	pDistanceFu hMexican; 
	pDistanceFu hEpanech;

	cudaMemcpyFromSymbol(&hBubble, pBubble, sizeof(pDistanceFu) );
	cudaMemcpyFromSymbol(&hGaussian, pGaussian, sizeof(pDistanceFu) );
	cudaMemcpyFromSymbol(&hCutGauss, pCutGauss, sizeof(pDistanceFu) );
	cudaMemcpyFromSymbol(&hMexican, pMexican, sizeof(pDistanceFu) );
	cudaMemcpyFromSymbol(&hEpanech, pEpanech, sizeof(pDistanceFu) );

	if (strcmp (DistFunc.name, "gaussian") == 0) {
		DistFunc.distance = hGaussian;
	} else if (strcmp (DistFunc.name, "mexican") == 0) {
		DistFunc.distance = hMexican;
	} else if (strcmp (DistFunc.name, "bubble") == 0) {
		DistFunc.distance = hBubble;
	} else if (strcmp (DistFunc.name, "cutgaussian") == 0) {
		DistFunc.distance = hCutGauss;
	} else if (strcmp (DistFunc.name, "epanechicov") == 0) {
		DistFunc.distance = hEpanech;
	} else {
		printf("No preimplemented function recognized. No assignment done.\n");
	}
	printf("Preimplemented function recognized. Assignment done.\n");
}

AssignDistanceFunction(&DistFunc);
// .. kernel call with DistFunc.distance
// .. Works

When I try to make the assignment in another file from outside, not…
This behavior also happens, when I just implement the Assignment-Functions in another *.cu file.
So it seems to be impossible to create something like plugin-system with the help of function pointers.
I have not to mention, that real (non CUDA) function pointers are working perfectly. So how can I assign function pointers from outside?

main.cu

__device__ static float
fcn_own_nhood(float r, float t) {
	return t*sqrt(r);
}

typedef float (*pDistanceFu) (float, float);

__device__ pDistanceFu pOwn = fcn_own_nhood; 

DistFunction ownDistFunction = {
	(char*)"own",
	fcn_bubble_nhood,
	fcn_rad_decay,
	fcn_lrate_decay
};

int main(int argc, char *argv[]) {
        // ..

	SOMNetGPU gpu;
	gpu.CreateSOM(3, 1, w1,w1);
	gpu.SetTrainingSet(input);
	gpu.SetConscienceRate(0.1);
	
	pDistanceFu hOwn;
	cudaMemcpyFromSymbol(&hOwn, pOwn, sizeof(pDistanceFu) );
	
	DistFunction distFn = ownDistFunction;
	//distFn.rad_decay = fcn_nearest_nhood;
	distFn.distance = hOwn;
	gpu.SetDistFunction(distFn);

// what():  unspecified launch failure