Category Archives

2 Articles

Venturing into OptiX – cudaMemcpy breaks everything

I’ve been looking into using Nvidia Optix for my Fractal lookdev tool, as I need realtime feedback while traversing the landscape.

I really like the API so far, it even comes with a Julia example demo in the SDK! I’ve modified it to render the Mandelbulb instead, with little trouble (and with refraction).

12318072_10208255826582355_1853094794_o

However, I needed to read the buffer data to write .exr’s to disk (this is a vital feature, better to make sure it’s viable at the beginning of the project). That shouldn’t be a problem, as I can easily grab the buffer’s device pointer and then cudaMemcpy it over.

All I need to do is include and , no big deal right?

#include <cuda.h>
#include <cuda_runtime.h>
 
#include <optixu/optixpp_namespace.h>
#include <optixu/optixu_math_namespace.h>
 
using namespace optix;
Linking CXX executable ../bin/julia
CMakeFiles/julia.dir/julia.cpp.o: In function `AnimCamera::apply(optix::Handle<optix::ContextObj>)':
/home/tom/src/optix/SDK/julia/julia.cpp:151: undefined reference to `PinholeCamera::PinholeCamera(float3, float3, float3, float, float, PinholeCamera::AspectRatioMode)'
/home/tom/src/optix/SDK/julia/julia.cpp:153: undefined reference to `PinholeCamera::getEyeUVW(float3&, float3&, float3&, float3&)'
collect2: error: ld returned 1 exit status

Oh.

Well maybe it’s because I included them before OptiX.

#include <optixu/optixpp_namespace.h>
#include <optixu/optixu_math_namespace.h>
 
#include <cuda.h>
#include <cuda_runtime.h>
 
using namespace optix;
/usr/local/cuda-7.0/include/cuda_runtime_api.h:257:17: error: ‘cudaError_t’ does not name a type
 extern __host__ cudaError_t CUDARTAPI cudaDeviceReset(void);
                 ^
/usr/local/cuda-7.0/include/cuda_runtime_api.h:274:36: error: ‘cudaError_t’ does not name a type
 extern __host__ __cudart_builtin__ cudaError_t CUDARTAPI cudaDeviceSynchronize(void);
                                    ^
/usr/local/cuda-7.0/include/cuda_runtime_api.h:349:17: error: ‘cudaError_t’ does not name a type
 extern __host__ cudaError_t CUDARTAPI cudaDeviceSetLimit(enum cudaLimit limit, size_t value);
                 ^
 
...
 
/home/tom/src/optix/SDK/julia/julia.cpp:299:95: error: ‘cudaMemcpy’ was not declared in this scope
   cudaMemcpy( (void*)h_ptr, (void*)d_ptr, sizeof(float) * totalPixels, cudaMemcpyDeviceToHost );

Oh.

This is dire, could it be related to the optix namespace somehow?

#include <optixu/optixpp_namespace.h>
#include <optixu/optixu_math_namespace.h>
 
using namespace optix;
 
#include <cuda.h>
#include <cuda_runtime.h>
Linking CXX executable ../bin/julia
[100%] Built target tutorial

Success! For some reason the optix namespace interferes with Cuda, setting it up like I did above somehow fixes that. Maybe this is mentioned in the documentation somewhere but I never found any reference to it.

So if you’re getting undefined references when you need cudaMemcpy in an OptiX project, check out your include order and namespace setup (and don’t forget to add ${CUDA_LIBRARIES} to target_link_libraries in the cmake configuration if it’s not there already).

I haven’t had any other strange problems besides this and so far OptiX seems pretty solid, I hope I’ll have fun with it.

Update

Turns out in this particular instance I was better off using the map() and unmap() functions within the Optix Buffer class instead of cudaMemcpy(), they remove the dependency of including the CUDA headers. However, I would no doubt have run into the same problems using something like the NVRTC headers.

Playing with Fractals in WebGL

Playing with Fractals in WebGL

As an excuse to play with WebGL I decided to finally got around to playing with the Mandelbrot and Julia set. There’s a live demo at the end of the post, the zoom control/mouse movement needs some work but the basic idea works great.

Links

A couple of useful places I used as reference :
Ozone3D – The Mandelbrot Set: Colors of Infinity
Fractals: Useful Beauty
dat.GUI – A great JavaScript GUI

The Shader Code

Vertex Shader

Really simple, all we need is a fullscreen quad with smooth UV’s.

attribute vec3 vtx_pos;
attribute vec2 vtx_uv;
 
varying vec2 texcoord;
 
void main(void)
{
	texcoord = vtx_uv;
	gl_Position = vec4(vtx_pos, 1.0);
}
Mandelbrot Set Fragment Shader

This one is very similar to the code in Ozone3D’s tutorial, I basically copied it for the most part (there’s only so many ways to skin a cat anyway).

The most notable aspect is MAX_ITER, which I replace with the actual iteration value before I compile the shader code. This was to get around the limitation of most webgl implementations which disallow you from using non-constant loop conditions, as they will be unrolled. This isn’t very convenient for real time display of any possible iteration size, so I simply recompile the shader when the iteration size changes.

Note that the “uniform vec2 c” uniform isn’t used here, it’s just there to keep the uniform layout identical for both shaders (I’m lazy and it’s just a demo).

Finally, I use a 1D texture (passed in through a 2D sampler, but the actual texture is only 1 pixel tall) to lookup the expected colour based on the iteration level.

#define iterations MAX_ITER
 
precision highp float;
 
varying vec2 texcoord;
 
uniform sampler2D gradientSampler;
uniform vec2 center;
uniform float zoom;
uniform vec2 c;
 
void main(void)
{
	// Mandelbrot
	vec2 C = vec2(1.0, 0.75) * (texcoord - vec2(0.5, 0.5)) * vec2(zoom, zoom) - center;
	vec2 z = C;
 
	gl_FragColor = vec4(0,0,0,1);
 
	for(float i = 0.0; i < float(iterations); i += 1.0)
        {
                z = vec2( z.x*z.x - z.y*z.y, 2.0 * z.x * z.y) + C;
                if(dot(z,z) > 4.0)
		{
			gl_FragColor = texture2D(gradientSampler, vec2(i / float(iterations), 0.5));
			break;
		}
	}
}

 

Julia Set Fragment Shader

The base code is identical to the Mandelbrot shader, except now we can pass in the desired complex parameter through c.

I can’t explain the maths as well as the pages I referenced earlier; but the main change is now the smooth UVs of the fullscreen quad are now assigned to Z instead of C, as the complex parameter is now a user defined constant and Z varies.

#define iterations MAX_ITER
 
precision highp float;
 
varying vec2 texcoord;
 
uniform sampler2D gradientSampler;
uniform vec2 center;
uniform float zoom;
uniform vec2 c;
 
void main(void)
{
	// Julia
	vec2 z = vec2(1.0, 0.75) * (texcoord - vec2(0.5, 0.5)) * vec2(zoom, zoom) - center;
 
	gl_FragColor = vec4(0,0,0,1);
 
	for(float i = 0.0; i < float(iterations); i += 1.0)
        {
                z = vec2( z.x*z.x - z.y*z.y, 2.0 * z.x * z.y) + c;
                if(dot(z,z) > 4.0)
		{
			gl_FragColor = texture2D(gradientSampler, vec2(i / float(iterations), 0.5));
			break;
		}
	}
}

The Live Demo (full page demo)

  • Toggle between the Mandelbrot/Julia set.
  • Use the zoom control to zoom in (this isn’t ideal).
  • Expect slow down on high iterations.
  • If you want to tinker around with the code then I apologise for how messy it is, it’s my very first adventure into JavaScript and I kept bolting more stuff onto it until I ended up with this.

ToDo

The Ozone3D page mentions that “a better approach is to use a multipass algorithm.”, which is next in line for this project.

It would also be cool to choose a complex parameter for the Julia set by clicking on the Mandelbrot itself, as both sets are connected.