 texture<short, 2, cudaReadModeElementType> ClassifyPrimaryTexture;
 texture<short, 2, cudaReadModeElementType> ClassifyKeyholeTexture;
 
-#define NUM_THREADS 256
+#define NUM_THREADS 512
+#define MAX_GRID_SIZE 65535
 
 cudaChannelFormatDesc Voxel_Classifier_ChannelDesc = cudaCreateChannelDesc<short>();
 


	//copy information to GPU
	cudaMemcpyToSymbolAsync(info, &information, sizeof(Voxel_Classifier_Information) );

	//translate input onto device
	float* dev_InputData = 0;
	cudaMalloc( (void**) &dev_InputData, 2*sizeof(float)*information.VolumeSize[0]*information.VolumeSize[1]*information.VolumeSize[2] );
	cudaMemcpyAsync(dev_InputData,inputData, 2*sizeof(float)*information.VolumeSize[0]*information.VolumeSize[1]*information.VolumeSize[2], 
					cudaMemcpyHostToDevice, *stream);

	//translate classification textures onto the device

	cudaUnbindTexture( ClassifyKeyholeTexture );
	cudaFreeArray( PrimaryTextureArray );
	cudaFreeArray( KeyholeTextureArray );
	cudaFree( dev_OutputData );

}