Volume Render : demo in cudasdk2.0 bata, linux

Accelerated Computing CUDA CUDA Programming and Performance
1

In cuda2.0 sdk there’s a demo about volume render:
I’m want to render a multi-layer CT image, the raw data is: 512512220, 16-bit depth.
I cannot understand the source dode, I cannot get informations of some function…

such as: Line59, dbvolumeRender.cu:
cudaExtent volumeSize = make_cudaExtent(32, 32, 32);
where can i get information about the function make_cudaExtent(int, int, int)?
where can i get the fefine of cudaExtent?

Is it possiable to rend all the data at once? my device is 8800gt, 512Mb mem.
The demo is for 8-bit data. may I render 16-bit data? how to modify it?

2

cudaExtent is explained in the CUDA Reference Manual (this is different than the Programmers Guide). The Reference Manual is found in the doc/ directory of the CUDA toolkit after you install it, and lists all of the functions available in CUDA. If you search for cudaExtent, you will find a definition on the page that describes the cudaMemset3D() function.

No description of make_cudaExtent() is given in the manual, but following the pattern for other make_XXX() functions, I assume what it does is return a cudaExtent() with the three struct fields set to the parameters you pass.

3

Yes, you should be able to use the volumeRender sample to load other data sets. 512512220*2 bytes is 110Mbytes, so it should fit in your GPU memory easily. You will need to modify the texture definition and initialization to make it work with 16-bit data.

Note that this sample isn’t really intended as a high performance volume renderer, it’s just an example of using 3D textures.

4

Hi there!

I’m trying to use volume rendering example with a diferent volume example. This example is included in the webpage under “Download new datasets” with the following specification:

Head MRT Angiography 16Bits
(10bits set)
256 x 320 x 128
0.66, 0.66, 0.66

I already try to change the following variables:

char *volumeFilename = “mrt16_angio2.raw”;
cudaExtent volumeSize = make_cudaExtent(256, 320, 128);

But the result is a blur image of the volume.

Image:

800×532 9.3 KB

Is there some change missing in here?? Please help me to get this right.

Thanks in advance,
Pedro Boavida

5

Note that this sample isn’t really intended as a high performance volume renderer,

Good to know; I’ve played as well with this VR sample and was quite upset by its performance. So, can NVIDIA provide the sample intended to demonstrate the capability of GPU/CUDA to perform a high quality volume rendering? I think GPU should outperform CPU in this regards but so far the best CPU volumetric ray-tracing engine dramatically outperforms the best GPU implementations. I’m looking hard to justify investment in CUDA development for HQ VR so far no factual evidence just hype. Please do not reply if you cannot be specific.

Stefan

6

As I mentioned, you need modify the code to use 16-bit data (replace uchar with ushort), it is written assuming 8-bit data.

7

I’m following Pedro’s lead & looking into what Simon Green has advised. When I change the instances of the uchar data to ushort, there is a runtime memory access violation.

The first portion of code with a problem lies within initCuda:

extern “C”

void initCuda(ushort *h_volume, cudaExtent volumeSize)

{

// create 3D array

cudaChannelFormatDesc channelDesc = cudaCreateChannelDesc<ushort>();

cutilSafeCall( cudaMalloc3DArray(&d_volumeArray, &channelDesc, volumeSize) );

// copy data to 3D array

cudaMemcpy3DParms copyParams = {0};

copyParams.srcPtr   = make_cudaPitchedPtr((void*)h_volume, volumeSize.width*sizeof(ushort), volumeSize.width, volumeSize.height);

copyParams.dstArray = d_volumeArray;

copyParams.extent   = volumeSize;

copyParams.kind     = cudaMemcpyHostToDevice;

>>cutilSafeCall( cudaMemcpy3D(&copyParams) );

The cudaMemcpy3D is the point of crash. I’m a novice at memory management within C & have been caught up in trying to figure out how to use the copy params. Why isn’t the correct memory allocated if the volumeSize.width is multiplied by sizeof(ushort)?

Anyone get 10/12/16bit volumetric data working? If anything, I’ll welcome advise into what to look for in the Programming guide & reference material /Doc.

TIA!

EDIT: BTW, I’m using both the 2.3 SDK & example code. I noticed there were a few changes to how GL was handled since 2.0. I’m interested in only a rudimentary example of VR.

8

I used the example to show float volumes from Matlab, make sure that you change the datatype for the texture as well (in the kernel.cu file)

9

Thanks for the reply Wanderine; I think I may have already addressed the texture datatype with something along these lines:

//volumeRender_kernel.cu

typedef unsigned short ushort;

texture<ushort, 3, cudaReadModeNormalizedFloat> tex; // 3D texture

texture<float4, 1, cudaReadModeElementType> transferTex; // 1D transfer function

void initCuda(ushort *h_volume, cudaExtent volumeSize)

{

…

// set texture parameters

tex.normalized = true; // access with normalized texture coordinates

tex.filterMode = cudaFilterModeLinear; // linear interpolation

tex.addressMode[0] = cudaAddressModeClamp; // wrap texture coordinates

tex.addressMode[1] = cudaAddressModeClamp;

…

}

The 3D texture is what you’re referring to, I believe, it is what is bound to the d_VolumeArray:

// bind array to 3D texture

cutilSafeCall(cudaBindTextureToArray(tex, d_volumeArray, channelDesc));

I’ve looked at some other examples and the channelDesc looks okay when it is changed to a ushort:

cudaChannelFormatDesc channelDesc = cudaCreateChannelDesc();

The problem still remains with the host to device memory transaction which uses this source pitched pointer:

//copyParams.srcPtr = make_cudaPitchedPtr((void*)h_volume, volumeSize.width*sizeof(ushort), volumeSize.width, volumeSize.height);

Where the volume dimensions are currently 32 X 32 X 32. The pitch should be 64 using * sizeof(ushort)… not sure what this has to do with anything, just following /DOCs & forum examples :rolleyes:

Thanks for the continued assistance, and also, I’d be happy to see the Matlab example (i’ll look for it here… let you know if I can’t locate it!).

TIA (again!)

10

Did you have any success ?

11

I have tried changing the volume type from unsigned char to unsigned short, both in the kernel.cu and cpp file. In this case nothing is being rendered and the volume render window remains blank.

Is there any other change required.

I made the following changes:

<u>VolumeRenderer.cpp</u>

//char *volumeFilename = "Bucky.raw";

//cudaExtent volumeSize = make_cudaExtent(32, 32, 32);

//typedef unsigned char VolumeType;

char *volumeFilename = "case1_T2.raw";

cudaExtent volumeSize = make_cudaExtent(544, 640, 24);

typedef unsigned short VolumeType;

<u>volumeRender_kernel.cu</u>

//typedef unsigned char VolumeType;

typedef unsigned short VolumeType;

Please let me know if I am missing something. I am able to render a different volume of unsigned char type, facing trouble when trying to render unsigned short type.

Thanks,

Dharaneesh

12

I have tried changing the volume type from unsigned char to unsigned short, both in the kernel.cu and cpp file. In this case nothing is being rendered and the volume render window remains blank.

Is there any other change required.

I made the following changes:

<u>VolumeRenderer.cpp</u>

//char *volumeFilename = "Bucky.raw";

//cudaExtent volumeSize = make_cudaExtent(32, 32, 32);

//typedef unsigned char VolumeType;

char *volumeFilename = "case1_T2.raw";

cudaExtent volumeSize = make_cudaExtent(544, 640, 24);

typedef unsigned short VolumeType;

<u>volumeRender_kernel.cu</u>

//typedef unsigned char VolumeType;

typedef unsigned short VolumeType;

Please let me know if I am missing something. I am able to render a different volume of unsigned char type, facing trouble when trying to render unsigned short type.

Thanks,

Dharaneesh

13

I have tried changing the volume type from unsigned char to unsigned short, both in the kernel.cu and cpp file. In this case nothing is being rendered and the volume render window remains blank.

Is there any other change required.

I made the following changes:

<u>VolumeRenderer.cpp</u>

//char *volumeFilename = "Bucky.raw";

//cudaExtent volumeSize = make_cudaExtent(32, 32, 32);

//typedef unsigned char VolumeType;

char *volumeFilename = "case1_T2.raw";

cudaExtent volumeSize = make_cudaExtent(544, 640, 24);

typedef unsigned short VolumeType;

<u>volumeRender_kernel.cu</u>

//typedef unsigned char VolumeType;

typedef unsigned short VolumeType;

Please let me know if I am missing something. I am able to render a different volume of unsigned char type, facing trouble when trying to render unsigned short type.

Thanks,

Dharaneesh

14

I think this thread has a working 16 bit variant that displays some angio data set

[url=“The Official NVIDIA Forums | NVIDIA”]The Official NVIDIA Forums | NVIDIA

Also a backport to CUDA SDK 2.3 is provided.

I hope this encourages people to use the search function before opening new threads.

15

I think this thread has a working 16 bit variant that displays some angio data set

[url=“http://forums.nvidia.com/index.php?showtopic=166088”]The Official NVIDIA Forums | NVIDIA

Also a backport to CUDA SDK 2.3 is provided.

I hope this encourages people to use the search function before opening new threads.

16

Does this Volume Rendering code represent the best VR NVIDIA may offer? I’ve asked several times in this forum the links to test quality/speed of GPU based VR (preferably CUDA based) to match the best CPU based VR and so far only this “gimmick” sample. Does NVIDIA has a competitive VR engine to render interactively at high quality mid size CT data (1000x512x512 12 bits)? The claims about GPU’s VR supremacy over CPU is everywhere - please provide the testable binaries instead of words. I may download and test Voreen, ImageVis3D, 3D Slicer, etc… and see its quality/speed performance even they provide a totally unsatisfactory performance but it is not just words it is a testable material. I still believe that NVIDIA may provide a way better VR performance but for some reason hides it ;o)

Thank you in advance,

Stefan

17

Does this Volume Rendering code represent the best VR NVIDIA may offer? I’ve asked several times in this forum the links to test quality/speed of GPU based VR (preferably CUDA based) to match the best CPU based VR and so far only this “gimmick” sample. Does NVIDIA has a competitive VR engine to render interactively at high quality mid size CT data (1000x512x512 12 bits)? The claims about GPU’s VR supremacy over CPU is everywhere - please provide the testable binaries instead of words. I may download and test Voreen, ImageVis3D, 3D Slicer, etc… and see its quality/speed performance even they provide a totally unsatisfactory performance but it is not just words it is a testable material. I still believe that NVIDIA may provide a way better VR performance but for some reason hides it ;o)

Thank you in advance,

Stefan