Hi Jens,
On 17/01/11 11:33, Jens Orthmann wrote:
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.
OK, I saw the inheritance, was just not sure if an osgCuda::Texture2D
could be made from an existing osg::Texture2D. I'll switch our texture
type and do a test.
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); ... }
Thanks, I'll try one or both of these approaches.
Please notice that we currently only support single-threaded osg
applications.
All our apps are single threaded currently, so no problem.
Thanks for your help
rgds
jp
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
--
This message is subject to the CSIR's copyright terms and conditions, e-mail legal notice, and implemented Open Document Format (ODF) standard.
The full disclaimer details can be found at http://www.csir.co.za/disclaimer.html.
This message has been scanned for viruses and dangerous content by MailScanner,
and is believed to be clean. MailScanner thanks Transtec Computers for their support.
_______________________________________________
osg-users mailing list
[email protected]
http://lists.openscenegraph.org/listinfo.cgi/osg-users-openscenegraph.org
