Channels ▼
RSS

C/C++

CUDA, Supercomputing for the Masses: Part 18


A Simple CUDA Kernel

The first example kernel draws a time-varying sinusoidal surface using a simple adaptation of the NVIDIA kernel provided in the simpleGL SDK example. Essentially, each height value stored in pos is updated based on position in the mesh, a frequency and the animation time passed via the parameter time.


    // calculate simple sine wave pattern
    float freq = 4.0f;
    float w = sinf(u*freq + time) * cosf(v*freq + time) * 0.5f;

The position and height information is stored in the float4 position array:


    // write output vertex
    pos[y*width+x] = make_float4(u, w, v, 1.0f);

Similarly, the colors are calculated based on position in the mesh and the animation time:


    // write the color
    colorPos[y*width+x].w = 0;
    colorPos[y*width+x].x = 255.f *0.5*(1.f+sinf(w+x));
    colorPos[y*width+x].y = 255.f *0.5*(1.f+sinf(x)*cosf(y));
    colorPos[y*width+x].z = 255.f *0.5*(1.f+sinf(w+time/10.f));

The complete listing for kernelVBO is as follows:


// kernelVBO.cpp (Rob Farber)
// Simple kernel to modify vertex positions in sine wave pattern
__global__ void kernel(float4* pos, uchar4 *colorPos,
		       unsigned int width, unsigned int height, float time)
{
    unsigned int x = blockIdx.x*blockDim.x + threadIdx.x;
    unsigned int y = blockIdx.y*blockDim.y + threadIdx.y;

    // calculate uv coordinates
    float u = x / (float) width;
    float v = y / (float) height;
    u = u*2.0f - 1.0f;
    v = v*2.0f - 1.0f;

    // calculate simple sine wave pattern
    float freq = 4.0f;
    float w = sinf(u*freq + time) * cosf(v*freq + time) * 0.5f;

    // write output vertex
    pos[y*width+x] = make_float4(u, w, v, 1.0f);
    colorPos[y*width+x].w = 0;
    colorPos[y*width+x].x = 255.f *0.5*(1.f+sinf(w+x));
    colorPos[y*width+x].y = 255.f *0.5*(1.f+sinf(x)*cosf(y));
    colorPos[y*width+x].z = 255.f *0.5*(1.f+sinf(w+time/10.f));
}

// Wrapper for the __global__ call that sets up the kernel call
extern "C" void launch_kernel(float4* pos, uchar4* colorPos,
			      unsigned int mesh_width, unsigned int mesh_height, float time)
{
  // execute the kernel
  dim3 block(8, 8, 1);
  dim3 grid(mesh_width / block.x, mesh_height / block.y, 1);
  kernel<<< grid, block>>>(pos, colorPos, mesh_width, mesh_height, time);
}

The program can be compiled under Linux with the following after the complete source of each of the files has been saved in correctly named files:


#/bin/bash
SDK_PATH= PATH_TO_NVIDIA_GPU_Computing_SDK

nvcc -O3 -L $SDK_PATH/C/lib -I $SDK_PATH/C/common/inc simpleGLmain.cpp simpleVBO.cpp callbacksVBO.cpp kernelVBO.cu -lglut -lGLEW -lcutil_x86_64 -o testGL

Microsoft users should consult the GPUcomputing.net website for information about building the examples with Microsoft Visual Studio.

Figures 10, 11, and 12 are images created with this program. You will likely see something similar, but the colors and shape of the surface do dynamically change with time. Also, the orientation of the surface in three-space also affects what is shown on the screen.

Figure 10: Example surface sinusoidal surface.

Figure 11: Wire-frame image.

Figure 12: Points drawn in 3D.


Related Reading


More Insights






Currently we allow the following HTML tags in comments:

Single tags

These tags can be used alone and don't need an ending tag.

<br> Defines a single line break

<hr> Defines a horizontal line

Matching tags

These require an ending tag - e.g. <i>italic text</i>

<a> Defines an anchor

<b> Defines bold text

<big> Defines big text

<blockquote> Defines a long quotation

<caption> Defines a table caption

<cite> Defines a citation

<code> Defines computer code text

<em> Defines emphasized text

<fieldset> Defines a border around elements in a form

<h1> This is heading 1

<h2> This is heading 2

<h3> This is heading 3

<h4> This is heading 4

<h5> This is heading 5

<h6> This is heading 6

<i> Defines italic text

<p> Defines a paragraph

<pre> Defines preformatted text

<q> Defines a short quotation

<samp> Defines sample computer code text

<small> Defines small text

<span> Defines a section in a document

<s> Defines strikethrough text

<strike> Defines strikethrough text

<strong> Defines strong text

<sub> Defines subscripted text

<sup> Defines superscripted text

<u> Defines underlined text

Dr. Dobb's encourages readers to engage in spirited, healthy debate, including taking us to task. However, Dr. Dobb's moderates all comments posted to our site, and reserves the right to modify or remove any content that it determines to be derogatory, offensive, inflammatory, vulgar, irrelevant/off-topic, racist or obvious marketing or spam. Dr. Dobb's further reserves the right to disable the profile of any commenter participating in said activities.

 
Disqus Tips To upload an avatar photo, first complete your Disqus profile. | View the list of supported HTML tags you can use to style comments. | Please read our commenting policy.
 

Video