Using OptiX with CUDA generated obj files

For background, I am working on a very large application that has a library of preexisting CUDA code. This code has been kept fairly up to date, and is now compiled into libraries using the linking features that are included in the CUDA SDK.

I have a scenario where I would like to call one of those functions from one of the OptiX programs that I am using (specifically the ray generation program, but I have a feeling it will come up in others as well). I believe it is possible as I am pretty sure calls to cudart are available, but I am not sure what the best way of doing this is.

I’ve tried to create small examples of calling a compiled sample into a .dll and then linking against it, but I get unhandled exception errors. I am wondering if I am missing something straight forward. I’ve included a small toy example I created that generates the same errors as the larger code base.

I guess there are two questions.

  1. Is it possible to create CUDA based libraries that can be called from inside of the programs passed to OptiX?
  2. If so, what is the best way of doing that?

Main Program:

#include "optix_world.h"

int main( void ) {

	optix::Context context = optix::Context::create();

	context->setEntryPointCount( 1 );

	optix::Program hitProgram =
		context->createProgramFromPTXFile( "cuda_compile_ptx_generated_PtxFunction.cu.ptx", "nearestHit" );

	context->setRayGenerationProgram( 0, rayGenerationProgram );*/

	return 0;
}

PtxFunction.cu

#include <optix_world.h>
#include "CudaFunction.h"

RT_PROGRAM void nearestHit() {
	optix::float3 color;
	setColorValue( color, shading_normal );
}

CudaFunction.cu

#include <optix_world.h>
#include <cuda.h>

__device__ void setColorValue( optix::float3& color, optix::float3 normal ) {
	color.x = normal.x;
	color.y = normal.y;
	color.z = normal.z;
}

The cmake file I am using looks like this:

PROJECT( CudaOptiXTest )

cmake_minimum_required( VERSION 2.8 )

find_package(CUDA REQUIRED)
find_package(OPTIX REQUIRED)

set(BUILD_SHARED_LIBS ON)

INCLUDE_DIRECTORIES( ${OptiX_INCLUDE} )
INCLUDE_DIRECTORIES( ${CudaOptixTest_SOURCE_DIR} )

set( CudaSourceFiles
	CudaFunction.cu 
)

set( PtxSourceFiles
	PtxFunction.cu
)

set( SourceFiles
	main.cpp
)

cuda_compile_ptx( cuda_ptx_files ${PtxSourceFiles} )

cuda_add_library( cudalib  ${CudaSourceFiles} )

cuda_add_executable( testDemo ${SourceFiles} ${cuda_ptx_files} )

target_link_libraries( testDemo cudalib ${optix_LIBRARY} )

if(WIN32)
  if(CMAKE_SIZEOF_VOID_P EQUAL 8 AND NOT APPLE)
    set(bit_dest "64")
  else()
    set(bit_dest "")
  endif()
  foreach(config ${CMAKE_CONFIGURATION_TYPES})
    get_target_property(loc testDemo ${config}_LOCATION)
    if(loc)
      # A little helper function
      function(copy_dll lib)
        get_filename_component(path ${loc} PATH)
        get_filename_component(name ${lib} NAME)
        #message("${CMAKE_COMMAND} -E copy_if_different ${lib} ${path}/${name}")
        execute_process(COMMAND ${CMAKE_COMMAND} -E copy_if_different ${lib} ${path}/${name})
      endfunction()
	  
	  function(copy_later lib)
	    get_filename_component(path ${loc} PATH)
        get_filename_component(name ${lib} NAME)
        #message("${CMAKE_COMMAND} -E copy_if_different ${lib} ${path}/${name}")
        add_custom_command(TARGET testDemo POST_BUILD COMMAND ${CMAKE_COMMAND} -E copy_if_different ${lib} ${path}/${name})
	  endfunction()
      
      # Copy the binary directory into the build
      file(GLOB dlls "${OptiX_INSTALL_DIR}/bin${bit_dest}/*.dll")
      foreach(file ${dlls})
        copy_later("${file}")
      endforeach()
    else()
      message(WARNING "Unable to find location to copy DLLs into the build")
    endif()
	
	foreach( file ${cuda_ptx_files} )
	  copy_dll("${file}")
	endforeach()
  endforeach()
endif(WIN32)

I’ve used functions from cuRAND in my OptiX programs, so I don’t see why you shouldn’t be able to use some CUDA library functions. Keep in mind that some CUDA functions like __syncthreads() are not allowed, however. Also, the functions that you use will be inlined with your OptiX programs by the compiler.

It sounds like you want to use callable programs. They act like the other OptiX programs that you write, but they can be called from any OptiX program, and they are not inlined.