Hi J.P.,

Thanks for your interest in using osgCompute. First of all it is possible as 
long as your previous texture is of type osgCuda::TextureXD. There is currently 
no way to use osg::Textures directly. The specialization is necessary as we 
need to keep track of when the texture's apply( State ) method is called within 
the osg pipeline. If you have utilized a osg::Texture2D beforehand it is no 
problem to switch to a osgCuda::Texture2D as it acts like a osg::Texture2D as 
long as no CUDA operation is performed. 
 
If you utilize a osgCuda::Texture2D for example, you can use myTexture->map( 
osgCompute::MAP_DEVICE_SOURCE ) or myTexture->map( osgCompute::MAP_DEVICE_ARRAY 
). However as CUDA maps an OpenGL texture to a cudaArray* we internally copy 
the array's data to writeable device memory with a cudaMemcpyDeviceToDevice in 
case of MAP_DEVICE_SOURCE. Its fast (GPU->GPU) but still it is a copy 
operation. If you want to avoid any copy operation you should use 
MAP_DEVICE_ARRAY. This will return the cudaArray pointer directly. The decision 
which of them to use depends on which type of pointer your CUDA-kernel supports:

MAP_DEVICE_SOURCE:

Code:

...
cudaKernel<<< #blocks,#threads >>>( (float*) 
myTexture->map(osgCompute::MAP_DEVICE_SOURCE ) );
...
__global__ void cudaKernel( float* myPtr ) {
...
float val = myPtr[122];
...
}




MAP_DEVICE_ARRAY:

Code:

...
cudaBindTextureToArray( myTex, 
(cudaArray*)(myTexture->map(osgCompute::MAP_DEVICE_ARRAY ) ) );
...
cudaKernel<<< #blocks,#threads >>>();
...
__global__ void cudaKernel(){ 
...
float val = texture2D(0.3,0.3,myTex);
...
 }




Please notice that we currently only support single-threaded osg applications.


Best regards,
Jens[/code]

------------------
Read this topic online here:
http://forum.openscenegraph.org/viewtopic.php?p=35680#35680





_______________________________________________
osg-users mailing list
[email protected]
http://lists.openscenegraph.org/listinfo.cgi/osg-users-openscenegraph.org

Reply via email to