40 #ifndef __optix_optix_cuda__internal_h__
41 #define __optix_optix_cuda__internal_h__
45 #include "internal/optix_internal.h"
56 template<
class S>
static __device__ __forceinline__
57 Type make(S s) {
return make_int(s); }
61 template<
class S>
static __device__ __forceinline__
62 Type make(S s) {
return make_int2(s); }
66 template<
class S>
static __device__ __forceinline__
67 Type make(S s) {
return make_int3(s); }
71 template<
class S>
static __device__ __forceinline__
72 Type make(S s) {
return make_int4(s); }
75 typedef unsigned int Type;
76 static __device__ __forceinline__
77 Type make(
unsigned int s) {
return s; }
78 template<
class S>
static __device__ __forceinline__
79 Type make(S s) {
return (
unsigned int)s.x; }
83 template<
class S>
static __device__ __forceinline__
84 Type make(S s) {
return make_uint2(s); }
88 template<
class S>
static __device__ __forceinline__
89 Type make(S s) {
return make_uint3(s); }
93 template<
class S>
static __device__ __forceinline__
94 Type make(S s) {
return make_uint4(s); }
98 template<
class S>
static __device__ __forceinline__
99 Type make(S s) {
return make_float(s); }
103 template<
class S>
static __device__ __forceinline__
104 Type make(S s) {
return make_float2(s); }
108 template<
class S>
static __device__ __forceinline__
109 Type make(S s) {
return make_float3(s); }
113 template<
class S>
static __device__ __forceinline__
114 Type make(S s) {
return make_float4(s); }
117 #if defined(__APPLE__) || defined(__x86_64) || defined(AMD64) || defined(_M_AMD64)
120 static __device__ __forceinline__
121 Type make(
unsigned int s) {
return s; }
122 template<
class S>
static __device__ __forceinline__
123 Type make(S s) {
return (
unsigned int)s.x; }
125 template<>
struct VectorTypes<size_t, 2> {
126 typedef size_t2 Type;
127 template<
class S>
static __device__ __forceinline__
128 Type make(S s) {
return make_size_t2(s); }
130 template<>
struct VectorTypes<size_t, 3> {
131 typedef size_t3 Type;
132 template<
class S>
static __device__ __forceinline__
133 Type make(S s) {
return make_size_t3(s); }
135 template<>
struct VectorTypes<size_t, 4> {
136 typedef size_t4 Type;
137 template<
class S>
static __device__ __forceinline__
138 Type make(S s) {
return make_size_t4(s); }
184 void never_call() { handle = 0; }
241 #define rtDeclareVariable(type, name, semantic, annotation) \
242 namespace rti_internal_typeinfo { \
243 __device__ ::rti_internal_typeinfo::rti_typeinfo name = { ::rti_internal_typeinfo::_OPTIX_VARIABLE, sizeof(type)}; \
245 namespace rti_internal_typename { \
246 __device__ char name[] = #type; \
248 namespace rti_internal_typeenum { \
249 __device__ int name = ::rti_internal_typeinfo::rti_typeenum<type>::m_typeenum; \
251 namespace rti_internal_semantic { \
252 __device__ char name[] = #semantic; \
254 namespace rti_internal_annotation { \
255 __device__ char name[] = #annotation; \
319 #define rtDeclareAnnotation(variable, annotation) \
320 namespace rti_internal_annotation { \
321 __device__ char variable[] = #annotation; \
374 #ifndef RT_USE_TEMPLATED_RTCALLABLEPROGRAM
375 #if (__CUDA_ARCH__ < 200) || (CUDART_VERSION < 4010)
376 #define rtCallableProgram(return_type, function_name, parameter_list) \
377 namespace rti_internal_typeinfo { \
378 __device__ ::rti_internal_typeinfo::rti_typeinfo function_name = { ::rti_internal_typeinfo::_OPTIX_VARIABLE, rtCallableProgramSizeofWrapper<return_type>::value }; \
380 namespace rti_internal_typename { \
381 __device__ char function_name[] = #return_type; \
383 namespace rti_internal_semantic { \
384 __device__ char function_name[] = ""; \
386 namespace rti_internal_annotation { \
387 __device__ char function_name[] = #parameter_list; \
389 __noinline__ __device__ return_type function_name parameter_list { typedef return_type localtype; return localtype(); }
391 #define rtCallableProgram(return_type, function_name, parameter_list) \
392 rtDeclareVariable(optix::boundCallableProgramId<return_type parameter_list>, function_name,,);
400 typedef unsigned int rtPickledLocalPointer;
401 static inline __device__ rtPickledLocalPointer rtPickleLocalPointer(
void *p ) {
402 return optix::rt_pickle_pointer( p );
405 static inline __device__
void * rtUnpickleLocalPointer( rtPickledLocalPointer p ) {
406 return optix::rt_unpickle_pointer( p );
417 template<
typename T,
int Dim = 1>
struct buffer {
421 __device__ __forceinline__ IndexType size()
const {
422 return WrapperType::make(rt_buffer_get_size(
this, Dim,
sizeof(T)));
424 __device__ __forceinline__ T& operator[](IndexType i) {
425 size_t4 c = make_index(i);
426 return *(T*)create(
type<T>(), rt_buffer_get(
this, Dim,
sizeof(T), c.x, c.y, c.z, c.w));
429 __inline__ __device__
static size_t4 make_index(
size_t v0) {
return make_size_t4(v0, 0, 0, 0); }
430 __inline__ __device__
static size_t4 make_index(size_t2 v0) {
return make_size_t4(v0.x, v0.y, 0, 0); }
431 __inline__ __device__
static size_t4 make_index(size_t3 v0) {
return make_size_t4(v0.x, v0.y, v0.z, 0); }
432 __inline__ __device__
static size_t4 make_index(size_t4 v0) {
return make_size_t4(v0.x, v0.y, v0.z, v0.w); }
437 template<
typename T2>
struct type { };
440 template<
typename T2> __device__ __forceinline__
static void* create(
type<T2>,
void* v) {
return v; }
443 template<
typename T2,
int Dim2>
444 __device__ __forceinline__
static void* create(type<
bufferId<T2,Dim2> >,
void* v)
457 template<
typename T,
int Dim = 1>
struct bufferId :
public buffer<T,Dim> {
458 typedef typename buffer<T,Dim>::WrapperType WrapperType;
459 typedef typename buffer<T,Dim>::IndexType IndexType;
462 __device__ __forceinline__ bufferId() {}
464 __device__ __forceinline__ bufferId(
RTbufferidnull nullid) { m_id = (int)nullid; }
466 __device__ __forceinline__
explicit bufferId(
int id) : m_id(id) {}
469 __device__ __forceinline__ bufferId& operator= (
RTbufferidnull nullid) { m_id = nullid;
return *
this; }
473 __device__ __forceinline__ IndexType size()
const {
474 return WrapperType::make(rt_buffer_get_size_id(m_id, Dim,
sizeof(T)));
476 __device__ __forceinline__ T& operator[](IndexType i)
const {
477 size_t4 c = make_index(i);
478 return *(T*)create(
typename buffer<T,Dim>::template type<T>(),
479 rt_buffer_get_id(m_id, Dim,
sizeof(T), c.x, c.y, c.z, c.w));
482 __device__ __forceinline__
int getId()
const {
return m_id; }
484 __device__ __forceinline__
operator bool()
const {
return m_id; }
537 #define rtBuffer __device__ optix::buffer
565 #define rtBufferId optix::bufferId
600 #define rtTextureSampler texture
604 typedef int rtTextureId;
606 #define _OPTIX_TEX_FUNC_DECLARE_(FUNC, SIGNATURE, PARAMS ) \
607 template<> inline __device__ unsigned char FUNC SIGNATURE \
609 uint4 tmp = FUNC <uint4> PARAMS; \
610 return (unsigned char)(tmp.x); \
612 template<> inline __device__ char FUNC SIGNATURE \
614 int4 tmp = FUNC <int4> PARAMS; \
615 return (char)(tmp.x); \
617 template<> inline __device__ unsigned short FUNC SIGNATURE \
619 uint4 tmp = FUNC <uint4> PARAMS; \
620 return (unsigned short)(tmp.x); \
622 template<> inline __device__ short FUNC SIGNATURE \
624 int4 tmp = FUNC <int4> PARAMS; \
625 return (short)(tmp.x); \
627 template<> inline __device__ int FUNC SIGNATURE \
629 int4 tmp = FUNC <int4> PARAMS; \
632 template<> inline __device__ unsigned int FUNC SIGNATURE \
634 uint4 tmp = FUNC <uint4> PARAMS; \
637 template<> inline __device__ uchar1 FUNC SIGNATURE \
639 uint4 tmp = FUNC <uint4> PARAMS; \
640 return make_uchar1(tmp.x); \
642 template<> inline __device__ char1 FUNC SIGNATURE \
644 int4 tmp = FUNC <int4> PARAMS; \
645 return make_char1(tmp.x); \
647 template<> inline __device__ ushort1 FUNC SIGNATURE \
649 uint4 tmp = FUNC <uint4> PARAMS; \
650 return make_ushort1(tmp.x); \
652 template<> inline __device__ short1 FUNC SIGNATURE \
654 int4 tmp = FUNC <int4> PARAMS; \
655 return make_short1(tmp.x); \
657 template<> inline __device__ uint1 FUNC SIGNATURE \
659 uint4 tmp = FUNC <uint4> PARAMS; \
660 return make_uint1(tmp.x); \
662 template<> inline __device__ int1 FUNC SIGNATURE \
664 int4 tmp = FUNC <int4> PARAMS; \
665 return make_int1(tmp.x); \
667 template<> inline __device__ float FUNC SIGNATURE \
669 float4 tmp = FUNC <float4> PARAMS; \
672 template<> inline __device__ uchar2 FUNC SIGNATURE \
674 uint4 tmp = FUNC <uint4> PARAMS; \
675 return make_uchar2(tmp.x, tmp.y); \
677 template<> inline __device__ char2 FUNC SIGNATURE \
679 int4 tmp = FUNC <int4> PARAMS; \
680 return make_char2(tmp.x, tmp.y); \
682 template<> inline __device__ ushort2 FUNC SIGNATURE \
684 uint4 tmp = FUNC <uint4> PARAMS; \
685 return make_ushort2(tmp.x, tmp.y); \
687 template<> inline __device__ short2 FUNC SIGNATURE \
689 int4 tmp = FUNC <int4> PARAMS; \
690 return make_short2(tmp.x, tmp.y); \
692 template<> inline __device__ uint2 FUNC SIGNATURE \
694 uint4 tmp = FUNC <uint4> PARAMS; \
695 return make_uint2(tmp.x, tmp.y); \
697 template<> inline __device__ int2 FUNC SIGNATURE \
699 int4 tmp = FUNC <int4> PARAMS; \
700 return make_int2(tmp.x, tmp.y); \
702 template<> inline __device__ float2 FUNC SIGNATURE \
704 float4 tmp = FUNC <float4> PARAMS; \
705 return ::make_float2(tmp.x, tmp.y); \
707 template<> inline __device__ uchar4 FUNC SIGNATURE \
709 uint4 tmp = FUNC <uint4> PARAMS; \
710 return make_uchar4(tmp.x, tmp.y, tmp.z, tmp.w); \
712 template<> inline __device__ char4 FUNC SIGNATURE \
714 int4 tmp = FUNC <int4> PARAMS; \
715 return make_char4(tmp.x, tmp.y, tmp.z, tmp.w); \
717 template<> inline __device__ ushort4 FUNC SIGNATURE \
719 uint4 tmp = FUNC <uint4> PARAMS; \
720 return make_ushort4(tmp.x, tmp.y, tmp.z, tmp.w); \
722 template<> inline __device__ short4 FUNC SIGNATURE \
724 int4 tmp = FUNC <int4> PARAMS; \
725 return make_short4(tmp.x, tmp.y, tmp.z, tmp.w); \
728 inline __device__ int4 float4AsInt4( float4 f4 ) {
729 return make_int4(__float_as_int(f4.x), __float_as_int(f4.y), __float_as_int(f4.z), __float_as_int(f4.w));
732 inline __device__ uint4 float4AsUInt4( float4 f4 ) {
733 return make_uint4(__float_as_int(f4.x), __float_as_int(f4.y), __float_as_int(f4.z), __float_as_int(f4.w));
784 inline __device__ uint3 rtTexSize(rtTextureId
id)
786 return optix::rt_texture_get_size_id(
id);
790 inline __device__ T rtTex1D(rtTextureId
id,
float x);
791 template<>
inline __device__ float4 rtTex1D(rtTextureId
id,
float x)
793 return optix::rt_texture_get_f_id(
id, 1, x, 0, 0, 0);
795 template<>
inline __device__ int4 rtTex1D(rtTextureId
id,
float x)
797 return optix::rt_texture_get_i_id(
id, 1, x, 0, 0, 0);
799 template<>
inline __device__ uint4 rtTex1D(rtTextureId
id,
float x)
801 return optix::rt_texture_get_u_id(
id, 1, x, 0, 0, 0);
803 _OPTIX_TEX_FUNC_DECLARE_(rtTex1D, (rtTextureId
id,
float x), (
id, x) )
805 inline __device__
void rtTex1D(T* retVal, rtTextureId
id,
float x)
807 T tmp = rtTex1D<T>(id, x);
812 inline __device__ T rtTex1DFetch(rtTextureId
id,
int x);
813 template<>
inline __device__ float4 rtTex1DFetch(rtTextureId
id,
int x)
815 return optix::rt_texture_get_fetch_id(
id, 1, x, 0, 0, 0);
817 template<>
inline __device__ int4 rtTex1DFetch(rtTextureId
id,
int x)
819 return float4AsInt4(optix::rt_texture_get_fetch_id(
id, 1, x, 0, 0, 0));
821 template<>
inline __device__ uint4 rtTex1DFetch(rtTextureId
id,
int x)
823 return float4AsUInt4(optix::rt_texture_get_fetch_id(
id, 1, x, 0, 0, 0));
825 _OPTIX_TEX_FUNC_DECLARE_(rtTex1DFetch, (rtTextureId
id,
int x), (
id, x) )
827 inline __device__
void rtTex1DFetch(T* retVal, rtTextureId
id,
int x)
829 T tmp = rtTex1DFetch<T>(id, x);
834 inline __device__ T rtTex2D(rtTextureId
id,
float x,
float y);
836 inline __device__ float4 rtTex2D(rtTextureId
id,
float x,
float y)
838 return optix::rt_texture_get_f_id(
id, 2, x, y, 0, 0);
841 inline __device__ int4 rtTex2D(rtTextureId
id,
float x,
float y)
843 return optix::rt_texture_get_i_id(
id, 2, x, y, 0, 0);
846 inline __device__ uint4 rtTex2D(rtTextureId
id,
float x,
float y)
848 return optix::rt_texture_get_u_id(
id, 2, x, y, 0, 0);
850 _OPTIX_TEX_FUNC_DECLARE_(rtTex2D, (rtTextureId
id,
float x,
float y), (
id, x, y) )
852 inline __device__
void rtTex2D(T* retVal, rtTextureId
id,
float x,
float y)
854 T tmp = rtTex2D<T>(id, x, y);
859 inline __device__ T rtTex2DFetch(rtTextureId
id,
int x,
int y);
860 template<>
inline __device__ float4 rtTex2DFetch(rtTextureId
id,
int x,
int y)
862 return optix::rt_texture_get_fetch_id(
id, 2, x, y, 0, 0);
864 template<>
inline __device__ int4 rtTex2DFetch(rtTextureId
id,
int x,
int y)
866 return float4AsInt4(optix::rt_texture_get_fetch_id(
id, 2, x, y, 0, 0));
868 template<>
inline __device__ uint4 rtTex2DFetch(rtTextureId
id,
int x,
int y)
870 return float4AsUInt4(optix::rt_texture_get_fetch_id(
id, 2, x, y, 0, 0));
872 _OPTIX_TEX_FUNC_DECLARE_(rtTex2DFetch, (rtTextureId
id,
int x,
int y), (
id, x, y) )
874 inline __device__
void rtTex2DFetch(T* retVal, rtTextureId
id,
int x,
int y)
876 T tmp = rtTex2DFetch<T>(id, x, y);
881 inline __device__ T rtTex3D(rtTextureId
id,
float x,
float y,
float z);
882 template<>
inline __device__ float4 rtTex3D(rtTextureId
id,
float x,
float y,
float z)
884 return optix::rt_texture_get_f_id(
id, 3, x, y, z, 0);
886 template<>
inline __device__ int4 rtTex3D(rtTextureId
id,
float x,
float y,
float z)
888 return optix::rt_texture_get_i_id(
id, 3, x, y, z, 0);
890 template<>
inline __device__ uint4 rtTex3D(rtTextureId
id,
float x,
float y,
float z)
892 return optix::rt_texture_get_u_id(
id, 3, x, y, z, 0);
894 _OPTIX_TEX_FUNC_DECLARE_(rtTex3D, (rtTextureId
id,
float x,
float y,
float z), (
id, x, y, z) )
896 inline __device__
void rtTex3D(T* retVal, rtTextureId
id,
float x,
float y,
float z)
898 T tmp = rtTex3D<T>(id, x, y, z);
903 inline __device__ T rtTex3DFetch(rtTextureId
id,
int x,
int y,
int z);
904 template<>
inline __device__ float4 rtTex3DFetch(rtTextureId
id,
int x,
int y,
int z)
906 return optix::rt_texture_get_fetch_id(
id, 3, x, y, z, 0);
908 template<>
inline __device__ int4 rtTex3DFetch(rtTextureId
id,
int x,
int y,
int z)
910 return float4AsInt4(optix::rt_texture_get_fetch_id(
id, 3, x, y, z, 0));
912 template<>
inline __device__ uint4 rtTex3DFetch(rtTextureId
id,
int x,
int y,
int z)
914 return float4AsUInt4(optix::rt_texture_get_fetch_id(
id, 3, x, y, z, 0));
916 _OPTIX_TEX_FUNC_DECLARE_(rtTex3DFetch, (rtTextureId
id,
int x,
int y,
int z), (
id, x, y, z) )
918 inline __device__
void rtTex3DFetch(T* retVal, rtTextureId
id,
int x,
int y,
int z)
920 T tmp = rtTex3DFetch<T>(id, x, y, z);
925 inline __device__ T rtTex2DGather(rtTextureId
id,
float x,
float y,
int comp = 0);
926 template<>
inline __device__ float4 rtTex2DGather(rtTextureId
id,
float x,
float y,
int comp)
928 return optix::rt_texture_get_gather_id(
id, x, y, comp);
930 template<>
inline __device__ int4 rtTex2DGather(rtTextureId
id,
float x,
float y,
int comp)
932 return float4AsInt4(optix::rt_texture_get_gather_id(
id, x, y, comp));
934 template<>
inline __device__ uint4 rtTex2DGather(rtTextureId
id,
float x,
float y,
int comp)
936 return float4AsUInt4(optix::rt_texture_get_gather_id(
id, x, y, comp));
938 _OPTIX_TEX_FUNC_DECLARE_(rtTex2DGather, (rtTextureId
id,
float x,
float y,
int comp), (
id, x, y, comp) )
940 inline __device__
void rtTex2DGather(T* retVal, rtTextureId
id,
float x,
float y,
int comp = 0)
942 T tmp = rtTex2DGather<T>(id, x, y, comp);
947 inline __device__ T rtTex1DGrad(rtTextureId
id,
float x,
float dPdx,
float dPdy);
948 template<>
inline __device__ float4 rtTex1DGrad(rtTextureId
id,
float x,
float dPdx,
float dPdy)
950 return optix::rt_texture_get_grad_id(
id, TEX_LOOKUP_1D, x, 0, 0, 0, dPdx, 0, 0, dPdy, 0, 0);
952 template<>
inline __device__ int4 rtTex1DGrad(rtTextureId
id,
float x,
float dPdx,
float dPdy)
954 return float4AsInt4(optix::rt_texture_get_grad_id(
id, TEX_LOOKUP_1D, x, 0, 0, 0, dPdx, 0, 0, dPdy, 0, 0));
956 template<>
inline __device__ uint4 rtTex1DGrad(rtTextureId
id,
float x,
float dPdx,
float dPdy)
958 return float4AsUInt4(optix::rt_texture_get_grad_id(
id, TEX_LOOKUP_1D, x, 0, 0, 0, dPdx, 0, 0, dPdy, 0, 0));
960 _OPTIX_TEX_FUNC_DECLARE_(rtTex1DGrad, (rtTextureId
id,
float x,
float dPdx,
float dPdy), (
id, x, dPdx, dPdy) )
962 inline __device__
void rtTex1DGrad(T* retVal, rtTextureId
id,
float x,
float dPdx,
float dPdy)
964 T tmp = rtTex1DGrad<T>(id, x, dPdx, dPdy);
969 inline __device__ T rtTex2DGrad(rtTextureId
id,
float x,
float y, float2 dPdx, float2 dPdy);
970 template<>
inline __device__ float4 rtTex2DGrad(rtTextureId
id,
float x,
float y, float2 dPdx, float2 dPdy)
972 return optix::rt_texture_get_grad_id(
id, TEX_LOOKUP_2D, x, y, 0, 0, dPdx.x, dPdx.y, 0, dPdy.x, dPdy.y, 0);
974 template<>
inline __device__ int4 rtTex2DGrad(rtTextureId
id,
float x,
float y, float2 dPdx, float2 dPdy)
976 return float4AsInt4(optix::rt_texture_get_grad_id(
id, TEX_LOOKUP_2D, x, y, 0, 0, dPdx.x, dPdx.y, 0, dPdy.x, dPdy.y, 0));
978 template<>
inline __device__ uint4 rtTex2DGrad(rtTextureId
id,
float x,
float y, float2 dPdx, float2 dPdy)
980 return float4AsUInt4(optix::rt_texture_get_grad_id(
id, TEX_LOOKUP_2D, x, y, 0, 0, dPdx.x, dPdx.y, 0, dPdy.x, dPdy.y, 0));
982 _OPTIX_TEX_FUNC_DECLARE_(rtTex2DGrad, (rtTextureId
id,
float x,
float y, float2 dPdx, float2 dPdy), (
id, x, y, dPdx, dPdy) )
984 inline __device__
void rtTex2DGrad(T* retVal, rtTextureId
id,
float x,
float y, float2 dPdx, float2 dPdy)
986 T tmp = rtTex2DGrad<T>(id, x, y, dPdx, dPdy);
991 inline __device__ T rtTex3DGrad(rtTextureId
id,
float x,
float y,
float z, float4 dPdx, float4 dPdy);
992 template<>
inline __device__ float4 rtTex3DGrad(rtTextureId
id,
float x,
float y,
float z, float4 dPdx, float4 dPdy)
994 return optix::rt_texture_get_grad_id(
id, TEX_LOOKUP_3D, x, y, z, 0, dPdx.x, dPdx.y, dPdx.z, dPdy.x, dPdy.y, dPdy.z);
996 template<>
inline __device__ int4 rtTex3DGrad(rtTextureId
id,
float x,
float y,
float z, float4 dPdx, float4 dPdy)
998 return float4AsInt4(optix::rt_texture_get_grad_id(
id, TEX_LOOKUP_3D, x, y, z, 0, dPdx.x, dPdx.y, dPdx.z, dPdy.x, dPdy.y, dPdy.z));
1000 template<>
inline __device__ uint4 rtTex3DGrad(rtTextureId
id,
float x,
float y,
float z, float4 dPdx, float4 dPdy)
1002 return float4AsUInt4(optix::rt_texture_get_grad_id(
id, TEX_LOOKUP_3D, x, y, z, 0, dPdx.x, dPdx.y, dPdx.z, dPdy.x, dPdy.y, dPdy.z));
1004 _OPTIX_TEX_FUNC_DECLARE_(rtTex3DGrad, (rtTextureId
id,
float x,
float y,
float z, float4 dPdx, float4 dPdy), (
id, x, y, z, dPdx, dPdy) )
1005 template<
typename T>
1006 inline __device__
void rtTex3DGrad(T* retVal, rtTextureId
id,
float x,
float y,
float z, float4 dPdx, float4 dPdy)
1008 T tmp = rtTex3DGrad<T>(id, x, y, z, dPdx, dPdy);
1012 template<
typename T>
1013 inline __device__ T rtTex1DLayeredGrad(rtTextureId
id,
float x,
int layer,
float dPdx,
float dPdy);
1014 template<>
inline __device__ float4 rtTex1DLayeredGrad(rtTextureId
id,
float x,
int layer,
float dPdx,
float dPdy)
1016 return optix::rt_texture_get_grad_id(
id, TEX_LOOKUP_A1, x, 0, 0, layer, dPdx, 0, 0, dPdy, 0, 0);
1018 template<>
inline __device__ int4 rtTex1DLayeredGrad(rtTextureId
id,
float x,
int layer,
float dPdx,
float dPdy)
1020 return float4AsInt4(optix::rt_texture_get_grad_id(
id, TEX_LOOKUP_A1, x, 0, 0, layer, dPdx, 0, 0, dPdy, 0, 0));
1022 template<>
inline __device__ uint4 rtTex1DLayeredGrad(rtTextureId
id,
float x,
int layer,
float dPdx,
float dPdy)
1024 return float4AsUInt4(optix::rt_texture_get_grad_id(
id, TEX_LOOKUP_A1, x, 0, 0, layer, dPdx, 0, 0, dPdy, 0, 0));
1026 _OPTIX_TEX_FUNC_DECLARE_(rtTex1DLayeredGrad, (rtTextureId
id,
float x,
int layer,
float dPdx,
float dPdy), (
id, x, layer, dPdx, dPdy) )
1027 template<
typename T>
1028 inline __device__
void rtTex1DLayeredGrad(T* retVal, rtTextureId
id,
float x,
int layer,
float dPdx,
float dPdy)
1030 T tmp = rtTex1DLayeredGrad<T>(id, x, layer, dPdx, dPdy);
1034 template<
typename T>
1035 inline __device__ T rtTex2DLayeredGrad(rtTextureId
id,
float x,
float y,
int layer, float2 dPdx, float2 dPdy);
1036 template<>
inline __device__ float4 rtTex2DLayeredGrad(rtTextureId
id,
float x,
float y,
int layer, float2 dPdx, float2 dPdy)
1038 return optix::rt_texture_get_grad_id(
id, TEX_LOOKUP_A2, x, y, 0, layer, dPdx.x, dPdx.y, 0, dPdy.x, dPdy.y, 0);
1040 template<>
inline __device__ int4 rtTex2DLayeredGrad(rtTextureId
id,
float x,
float y,
int layer, float2 dPdx, float2 dPdy)
1042 return float4AsInt4(optix::rt_texture_get_grad_id(
id, TEX_LOOKUP_A2, x, y, 0, layer, dPdx.x, dPdx.y, 0, dPdy.x, dPdy.y, 0));
1044 template<>
inline __device__ uint4 rtTex2DLayeredGrad(rtTextureId
id,
float x,
float y,
int layer, float2 dPdx, float2 dPdy)
1046 return float4AsUInt4(optix::rt_texture_get_grad_id(
id, TEX_LOOKUP_A2, x, y, 0, layer, dPdx.x, dPdx.y, 0, dPdy.x, dPdy.y, 0));
1048 _OPTIX_TEX_FUNC_DECLARE_(rtTex2DLayeredGrad, (rtTextureId
id,
float x,
float y,
int layer, float2 dPdx, float2 dPdy), (
id, x, y, layer, dPdx, dPdy) )
1049 template<
typename T>
1050 inline __device__
void rtTex2DLayeredGrad(T* retVal, rtTextureId
id,
float x,
float y,
int layer, float2 dPdx, float2 dPdy)
1052 T tmp = rtTex2DLayeredGrad<T>(id, x, y, layer, dPdx, dPdy);
1056 template<
typename T>
1057 inline __device__ T rtTex1DLod(rtTextureId
id,
float x,
float level);
1058 template<>
inline __device__ float4 rtTex1DLod(rtTextureId
id,
float x,
float level)
1060 return optix::rt_texture_get_level_id(
id, TEX_LOOKUP_1D, x, 0, 0, 0, level );
1062 template<>
inline __device__ int4 rtTex1DLod(rtTextureId
id,
float x,
float level)
1064 return float4AsInt4(optix::rt_texture_get_level_id(
id, TEX_LOOKUP_1D, x, 0, 0, 0, level ));
1066 template<>
inline __device__ uint4 rtTex1DLod(rtTextureId
id,
float x,
float level)
1068 return float4AsUInt4(optix::rt_texture_get_level_id(
id, TEX_LOOKUP_1D, x, 0, 0, 0, level ));
1070 _OPTIX_TEX_FUNC_DECLARE_(rtTex1DLod, (rtTextureId
id,
float x,
float level), (
id, x, level) )
1071 template<
typename T>
1072 inline __device__
void rtTex1DLod(T* retVal, rtTextureId
id,
float x,
float level)
1074 T tmp = rtTex1DLod<T>(id, x, level);
1078 template<
typename T>
1079 inline __device__ T rtTex2DLod(rtTextureId
id,
float x,
float y,
float level);
1080 template<>
inline __device__ float4 rtTex2DLod(rtTextureId
id,
float x,
float y,
float level)
1082 return optix::rt_texture_get_level_id(
id, TEX_LOOKUP_2D, x, y, 0, 0, level );
1084 template<>
inline __device__ int4 rtTex2DLod(rtTextureId
id,
float x,
float y,
float level)
1086 return float4AsInt4(optix::rt_texture_get_level_id(
id, TEX_LOOKUP_2D, x, y, 0, 0, level ));
1088 template<>
inline __device__ uint4 rtTex2DLod(rtTextureId
id,
float x,
float y,
float level)
1090 return float4AsUInt4(optix::rt_texture_get_level_id(
id, TEX_LOOKUP_2D, x, y, 0, 0, level ));
1092 _OPTIX_TEX_FUNC_DECLARE_(rtTex2DLod, (rtTextureId
id,
float x,
float y,
float level), (
id, x, y, level) )
1093 template<
typename T>
1094 inline __device__
void rtTex2DLod(T* retVal, rtTextureId
id,
float x,
float y,
float level)
1096 T tmp = rtTex2DLod<T>(id, x, y, level);
1100 template<
typename T>
1101 inline __device__ T rtTex3DLod(rtTextureId
id,
float x,
float y,
float z,
float level);
1102 template<>
inline __device__ float4 rtTex3DLod(rtTextureId
id,
float x,
float y,
float z,
float level)
1104 return optix::rt_texture_get_level_id(
id, TEX_LOOKUP_3D, x, y, z, 0, level );
1106 template<>
inline __device__ int4 rtTex3DLod(rtTextureId
id,
float x,
float y,
float z,
float level)
1108 return float4AsInt4(optix::rt_texture_get_level_id(
id, TEX_LOOKUP_3D, x, y, z, 0, level ));
1110 template<>
inline __device__ uint4 rtTex3DLod(rtTextureId
id,
float x,
float y,
float z,
float level)
1112 return float4AsUInt4(optix::rt_texture_get_level_id(
id, TEX_LOOKUP_3D, x, y, z, 0, level ));
1114 _OPTIX_TEX_FUNC_DECLARE_(rtTex3DLod, (rtTextureId
id,
float x,
float y,
float z,
float level), (
id, x, y, z, level) )
1115 template<
typename T>
1116 inline __device__
void rtTex3DLod(T* retVal, rtTextureId
id,
float x,
float y,
float z,
float level)
1118 T tmp = rtTex3DLod<T>(id, x, y, z, level);
1122 template<
typename T>
1123 inline __device__ T rtTex1DLayeredLod(rtTextureId
id,
float x,
int layer,
float level);
1124 template<>
inline __device__ float4 rtTex1DLayeredLod(rtTextureId
id,
float x,
int layer,
float level)
1126 return optix::rt_texture_get_level_id(
id, TEX_LOOKUP_A1, x, 0, 0, layer, level );
1128 template<>
inline __device__ int4 rtTex1DLayeredLod(rtTextureId
id,
float x,
int layer,
float level)
1130 return float4AsInt4(optix::rt_texture_get_level_id(
id, TEX_LOOKUP_A1, x, 0, 0, layer, level ));
1132 template<>
inline __device__ uint4 rtTex1DLayeredLod(rtTextureId
id,
float x,
int layer,
float level)
1134 return float4AsUInt4(optix::rt_texture_get_level_id(
id, TEX_LOOKUP_A1, x, 0, 0, layer, level ));
1136 _OPTIX_TEX_FUNC_DECLARE_(rtTex1DLayeredLod, (rtTextureId
id,
float x,
int layer,
float level), (
id, x, layer, level) )
1137 template<
typename T>
1138 inline __device__
void rtTex1DLayeredLod(T* retVal, rtTextureId
id,
float x,
int layer,
float level)
1140 T tmp = rtTex1DLayeredLod<T>(id, x, layer, level);
1144 template<
typename T>
1145 inline __device__ T rtTex2DLayeredLod(rtTextureId
id,
float x,
float y,
int layer,
float level);
1146 template<>
inline __device__ float4 rtTex2DLayeredLod(rtTextureId
id,
float x,
float y,
int layer,
float level)
1148 return optix::rt_texture_get_level_id(
id, TEX_LOOKUP_A2, x, y, 0, layer, level );
1150 template<>
inline __device__ int4 rtTex2DLayeredLod(rtTextureId
id,
float x,
float y,
int layer,
float level)
1152 return float4AsInt4(optix::rt_texture_get_level_id(
id, TEX_LOOKUP_A2, x, y, 0, layer, level ));
1154 template<>
inline __device__ uint4 rtTex2DLayeredLod(rtTextureId
id,
float x,
float y,
int layer,
float level)
1156 return float4AsUInt4(optix::rt_texture_get_level_id(
id, TEX_LOOKUP_A2, x, y, 0, layer, level ));
1158 _OPTIX_TEX_FUNC_DECLARE_(rtTex2DLayeredLod, (rtTextureId
id,
float x,
float y,
int layer,
float level), (
id, x, y, layer, level) )
1159 template<
typename T>
1160 inline __device__
void rtTex2DLayeredLod(T* retVal, rtTextureId
id,
float x,
float y,
int layer,
float level)
1162 T tmp = rtTex2DLayeredLod<T>(id, x, y, layer, level);
1166 template<
typename T>
1167 inline __device__ T rtTex1DLayered(rtTextureId
id,
float x,
int layer);
1168 template<>
inline __device__ float4 rtTex1DLayered(rtTextureId
id,
float x,
int layer)
1170 return optix::rt_texture_get_base_id(
id, TEX_LOOKUP_A1, x, 0, 0, layer );
1172 template<>
inline __device__ int4 rtTex1DLayered(rtTextureId
id,
float x,
int layer)
1174 return float4AsInt4(optix::rt_texture_get_base_id(
id, TEX_LOOKUP_A1, x, 0, 0, layer ));
1176 template<>
inline __device__ uint4 rtTex1DLayered(rtTextureId
id,
float x,
int layer)
1178 return float4AsUInt4(optix::rt_texture_get_base_id(
id, TEX_LOOKUP_A1, x, 0, 0, layer ));
1180 _OPTIX_TEX_FUNC_DECLARE_(rtTex1DLayered, (rtTextureId
id,
float x,
int layer), (
id, x, layer) )
1181 template<
typename T>
1182 inline __device__
void rtTex1DLayered(T* retVal, rtTextureId
id,
float x,
int layer)
1184 T tmp = rtTex1DLayered<T>(id, x, layer);
1188 template<
typename T>
1189 inline __device__ T rtTex2DLayered(rtTextureId
id,
float x,
float y,
int layer);
1190 template<>
inline __device__ float4 rtTex2DLayered(rtTextureId
id,
float x,
float y,
int layer)
1192 return optix::rt_texture_get_base_id(
id, TEX_LOOKUP_A2, x, y, 0, layer );
1194 template<>
inline __device__ int4 rtTex2DLayered(rtTextureId
id,
float x,
float y,
int layer)
1196 return float4AsInt4(optix::rt_texture_get_base_id(
id, TEX_LOOKUP_A2, x, y, 0, layer ));
1198 template<>
inline __device__ uint4 rtTex2DLayered(rtTextureId
id,
float x,
float y,
int layer)
1200 return float4AsUInt4(optix::rt_texture_get_base_id(
id, TEX_LOOKUP_A2, x, y, 0, layer ));
1202 _OPTIX_TEX_FUNC_DECLARE_(rtTex2DLayered, (rtTextureId
id,
float x,
float y,
int layer), (
id, x, y, layer) )
1203 template<
typename T>
1204 inline __device__
void rtTex2DLayered(T* retVal, rtTextureId
id,
float x,
float y,
int layer)
1206 T tmp = rtTex2DLayered<T>(id, x, y, layer);
1210 template<
typename T>
1211 inline __device__ T rtTexCubemap(rtTextureId
id,
float x,
float y,
float z);
1212 template<>
inline __device__ float4 rtTexCubemap(rtTextureId
id,
float x,
float y,
float z)
1214 return optix::rt_texture_get_base_id(
id, TEX_LOOKUP_CUBE, x, y, z, 0 );
1216 template<>
inline __device__ int4 rtTexCubemap(rtTextureId
id,
float x,
float y,
float z)
1218 return float4AsInt4(optix::rt_texture_get_base_id(
id, TEX_LOOKUP_CUBE, x, y, z, 0 ));
1220 template<>
inline __device__ uint4 rtTexCubemap(rtTextureId
id,
float x,
float y,
float z)
1222 return float4AsUInt4(optix::rt_texture_get_base_id(
id, TEX_LOOKUP_CUBE, x, y, z, 0 ));
1224 _OPTIX_TEX_FUNC_DECLARE_(rtTexCubemap, (rtTextureId
id,
float x,
float y,
float z), (
id, x, y, z) )
1225 template<
typename T>
1226 inline __device__
void rtTexCubemap(T* retVal, rtTextureId
id,
float x,
float y,
float z)
1228 T tmp = rtTexCubemap<T>(id, x, y, z);
1232 template<
typename T>
1233 inline __device__ T rtTexCubemapLayered(rtTextureId
id,
float x,
float y,
float z,
int layer);
1234 template<>
inline __device__ float4 rtTexCubemapLayered(rtTextureId
id,
float x,
float y,
float z,
int layer)
1236 return optix::rt_texture_get_base_id(
id, TEX_LOOKUP_ACUBE, x, y, z, layer );
1238 template<>
inline __device__ int4 rtTexCubemapLayered(rtTextureId
id,
float x,
float y,
float z,
int layer)
1240 return float4AsInt4(optix::rt_texture_get_base_id(
id, TEX_LOOKUP_ACUBE, x, y, z, layer ));
1242 template<>
inline __device__ uint4 rtTexCubemapLayered(rtTextureId
id,
float x,
float y,
float z,
int layer)
1244 return float4AsUInt4(optix::rt_texture_get_base_id(
id, TEX_LOOKUP_ACUBE, x, y, z, layer ));
1246 _OPTIX_TEX_FUNC_DECLARE_(rtTexCubemapLayered, (rtTextureId
id,
float x,
float y,
float z,
int layer), (
id, x, y, z, layer) )
1247 template<
typename T>
1248 inline __device__
void rtTexCubemapLayered(T* retVal, rtTextureId
id,
float x,
float y,
float z,
int layer)
1250 T tmp = rtTexCubemapLayered<T>(id, x, y, z, layer);
1254 template<
typename T>
1255 inline __device__ T rtTexCubemapLod(rtTextureId
id,
float x,
float y,
float z,
float level);
1256 template<>
inline __device__ float4 rtTexCubemapLod(rtTextureId
id,
float x,
float y,
float z,
float level)
1258 return optix::rt_texture_get_level_id(
id, TEX_LOOKUP_CUBE, x, y, z, 0, level );
1260 template<>
inline __device__ int4 rtTexCubemapLod(rtTextureId
id,
float x,
float y,
float z,
float level)
1262 return float4AsInt4(optix::rt_texture_get_level_id(
id, TEX_LOOKUP_CUBE, x, y, z, 0, level ));
1264 template<>
inline __device__ uint4 rtTexCubemapLod(rtTextureId
id,
float x,
float y,
float z,
float level)
1266 return float4AsUInt4(optix::rt_texture_get_level_id(
id, TEX_LOOKUP_CUBE, x, y, z, 0, level ));
1268 _OPTIX_TEX_FUNC_DECLARE_(rtTexCubemapLod, (rtTextureId
id,
float x,
float y,
float z,
float level), (
id, x, y, z, level) )
1269 template<
typename T>
1270 inline __device__
void rtTexCubemapLod(T* retVal, rtTextureId
id,
float x,
float y,
float z,
float level)
1272 T tmp = rtTexCubemapLod<T>(id, x, y, z, level);
1276 template<
typename T>
1277 inline __device__ T rtTexCubemapLayeredLod(rtTextureId
id,
float x,
float y,
float z,
int layer,
float level);
1278 template<>
inline __device__ float4 rtTexCubemapLayeredLod(rtTextureId
id,
float x,
float y,
float z,
int layer,
float level)
1280 return optix::rt_texture_get_level_id(
id, TEX_LOOKUP_ACUBE, x, y, z, layer, level );
1282 template<>
inline __device__ int4 rtTexCubemapLayeredLod(rtTextureId
id,
float x,
float y,
float z,
int layer,
float level)
1284 return float4AsInt4(optix::rt_texture_get_level_id(
id, TEX_LOOKUP_ACUBE, x, y, z, layer, level ));
1286 template<>
inline __device__ uint4 rtTexCubemapLayeredLod(rtTextureId
id,
float x,
float y,
float z,
int layer,
float level)
1288 return float4AsUInt4(optix::rt_texture_get_level_id(
id, TEX_LOOKUP_ACUBE, x, y, z, layer, level ));
1290 _OPTIX_TEX_FUNC_DECLARE_(rtTexCubemapLayeredLod, (rtTextureId
id,
float x,
float y,
float z,
int layer,
float level), (
id, x, y, z, layer, level) )
1291 template<
typename T>
1292 inline __device__
void rtTexCubemapLayeredLod(T* retVal, rtTextureId
id,
float x,
float y,
float z,
int layer,
float level)
1294 T tmp = rtTexCubemapLayeredLod<T>(id, x, y, z, layer, level);
1300 #undef _OPTIX_TEX_FUNC_DECLARE_
1336 #define RT_PROGRAM __global__
1367 #define RT_CALLABLE_PROGRAM __device__ __noinline__
1370 namespace rti_internal_callableprogram {
1382 template<
typename T1>
1388 template<
bool Condition,
typename Dummy =
void >
1390 typedef bool result;
1393 template<
typename IntentionalError>
1395 typedef typename IntentionalError::does_not_exist result;
1402 template <
typename ReturnT
1416 #if (CUDART_VERSION >= 4010)
1432 __device__ __forceinline__ ReturnT operator()()
1444 typedef ReturnT (*funcT)();
1445 funcT call = (funcT)optix::rt_callable_program_from_id(m_id);
1448 __device__ __forceinline__ ReturnT operator()(Arg0T arg0)
1459 typedef ReturnT (*funcT)(Arg0T);
1460 funcT call = (funcT)optix::rt_callable_program_from_id(m_id);
1463 __device__ __forceinline__ ReturnT operator()(Arg0T arg0, Arg1T arg1)
1473 typedef ReturnT (*funcT)(Arg0T,Arg1T);
1474 funcT call = (funcT)optix::rt_callable_program_from_id(m_id);
1475 return call(arg0,arg1);
1477 __device__ __forceinline__ ReturnT operator()(Arg0T arg0, Arg1T arg1, Arg2T arg2)
1486 typedef ReturnT (*funcT)(Arg0T,Arg1T,Arg2T);
1487 funcT call = (funcT)optix::rt_callable_program_from_id(m_id);
1488 return call(arg0,arg1,arg2);
1490 __device__ __forceinline__ ReturnT operator()(Arg0T arg0, Arg1T arg1, Arg2T arg2, Arg3T arg3)
1498 typedef ReturnT (*funcT)(Arg0T,Arg1T,Arg2T,Arg3T);
1499 funcT call = (funcT)optix::rt_callable_program_from_id(m_id);
1500 return call(arg0,arg1,arg2,arg3);
1502 __device__ __forceinline__ ReturnT operator()(Arg0T arg0, Arg1T arg1, Arg2T arg2, Arg3T arg3,
1510 typedef ReturnT (*funcT)(Arg0T,Arg1T,Arg2T,Arg3T,Arg4T);
1511 funcT call = (funcT)optix::rt_callable_program_from_id(m_id);
1512 return call(arg0,arg1,arg2,arg3,arg4);
1514 __device__ __forceinline__ ReturnT operator()(Arg0T arg0, Arg1T arg1, Arg2T arg2, Arg3T arg3,
1515 Arg4T arg4, Arg5T arg5)
1521 typedef ReturnT (*funcT)(Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T);
1522 funcT call = (funcT)optix::rt_callable_program_from_id(m_id);
1523 return call(arg0,arg1,arg2,arg3,arg4,arg5);
1525 __device__ __forceinline__ ReturnT operator()(Arg0T arg0, Arg1T arg1, Arg2T arg2, Arg3T arg3,
1526 Arg4T arg4, Arg5T arg5, Arg6T arg6)
1531 typedef ReturnT (*funcT)(Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T);
1532 funcT call = (funcT)optix::rt_callable_program_from_id(m_id);
1533 return call(arg0,arg1,arg2,arg3,arg4,arg5,arg6);
1535 __device__ __forceinline__ ReturnT operator()(Arg0T arg0, Arg1T arg1, Arg2T arg2, Arg3T arg3,
1536 Arg4T arg4, Arg5T arg5, Arg6T arg6, Arg7T arg7)
1540 typedef ReturnT (*funcT)(Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T,Arg7T);
1541 funcT call = (funcT)optix::rt_callable_program_from_id(m_id);
1542 return call(arg0,arg1,arg2,arg3,arg4,arg5,arg6,arg7);
1544 __device__ __forceinline__ ReturnT operator()(Arg0T arg0, Arg1T arg1, Arg2T arg2, Arg3T arg3,
1545 Arg4T arg4, Arg5T arg5, Arg6T arg6, Arg7T arg7,
1549 typedef ReturnT (*funcT)(Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T,Arg7T,Arg8T);
1550 funcT call = (funcT)optix::rt_callable_program_from_id(m_id);
1551 return call(arg0,arg1,arg2,arg3,arg4,arg5,arg6,arg7,arg8);
1553 __device__ __forceinline__ ReturnT operator()(Arg0T arg0, Arg1T arg1, Arg2T arg2, Arg3T arg3,
1554 Arg4T arg4, Arg5T arg5, Arg6T arg6, Arg7T arg7,
1555 Arg8T arg8, Arg9T arg9)
1557 typedef ReturnT (*funcT)(Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T,Arg7T,Arg8T,Arg9T);
1558 funcT call = (funcT)optix::rt_callable_program_from_id(m_id);
1559 return call(arg0,arg1,arg2,arg3,arg4,arg5,arg6,arg7,arg8,arg9);
1573 #define RT_INTERNAL_CALLABLE_PROGRAM_DEFS(...) public rti_internal_callableprogram::callableProgramIdBase<__VA_ARGS__> \
1577 __device__ __forceinline__ callableProgramId() {} \
1579 __device__ __forceinline__ callableProgramId(RTprogramidnull nullid) \
1580 : rti_internal_callableprogram::callableProgramIdBase<__VA_ARGS__>(nullid) {} \
1582 __device__ __forceinline__ explicit callableProgramId(int id) \
1583 : rti_internal_callableprogram::callableProgramIdBase<__VA_ARGS__>(id) {} \
1585 __device__ __forceinline__ callableProgramId& operator= (RTprogramidnull nullid) \
1586 { this->m_id = nullid; return *this; } \
1588 __device__ __forceinline__ int getId() const { return this->m_id; } \
1590 __device__ __forceinline__ operator bool() const \
1591 { return this->m_id != RT_PROGRAM_ID_NULL; } \
1600 template<
typename T>
1604 template<
typename ReturnT>
1606 template<typename ReturnT, typename Arg0T>
1608 template<typename ReturnT, typename Arg0T, typename Arg1T>
1610 template<typename ReturnT, typename Arg0T, typename Arg1T, typename Arg2T>
1612 template<typename ReturnT, typename Arg0T, typename Arg1T, typename Arg2T, typename Arg3T>
1614 template<typename ReturnT, typename Arg0T, typename Arg1T, typename Arg2T, typename Arg3T,
1617 template<typename ReturnT, typename Arg0T, typename Arg1T, typename Arg2T, typename Arg3T,
1618 typename Arg4T, typename Arg5T>
1619 class
callableProgramId<ReturnT(Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T)>:
RT_INTERNAL_CALLABLE_PROGRAM_DEFS(ReturnT,Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T);
1620 template<typename ReturnT, typename Arg0T, typename Arg1T, typename Arg2T, typename Arg3T,
1621 typename Arg4T, typename Arg5T, typename Arg6T>
1622 class
callableProgramId<ReturnT(Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T)>:
RT_INTERNAL_CALLABLE_PROGRAM_DEFS(ReturnT,Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T);
1623 template<typename ReturnT, typename Arg0T, typename Arg1T, typename Arg2T, typename Arg3T,
1624 typename Arg4T, typename Arg5T, typename Arg6T, typename Arg7T>
1625 class
callableProgramId<ReturnT(Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T,Arg7T)>:
RT_INTERNAL_CALLABLE_PROGRAM_DEFS(ReturnT,Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T,Arg7T);
1626 template<typename ReturnT, typename Arg0T, typename Arg1T, typename Arg2T, typename Arg3T,
1627 typename Arg4T, typename Arg5T, typename Arg6T, typename Arg7T, typename Arg8T>
1628 class
callableProgramId<ReturnT(Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T,Arg7T,Arg8T)>:
RT_INTERNAL_CALLABLE_PROGRAM_DEFS(ReturnT,Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T,Arg7T,Arg8T);
1629 template<typename ReturnT, typename Arg0T, typename Arg1T, typename Arg2T, typename Arg3T,
1630 typename Arg4T, typename Arg5T, typename Arg6T, typename Arg7T, typename Arg8T, typename Arg9T>
1631 class
callableProgramId<ReturnT(Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T,Arg7T,Arg8T,Arg9T)>:
RT_INTERNAL_CALLABLE_PROGRAM_DEFS(ReturnT,Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T,Arg7T,Arg8T,Arg9T);
1647 #if (CUDART_VERSION >= 4010)
1648 #define RT_INTERNAL_BOUND_CALLABLE_PROGRAM_DEFS(...) public rti_internal_callableprogram::callableProgramIdBase<__VA_ARGS__> \
1652 __device__ __forceinline__ boundCallableProgramId() {} \
1655 __device__ __forceinline__ boundCallableProgramId(const boundCallableProgramId& ); \
1656 __device__ __forceinline__ boundCallableProgramId& operator= (const boundCallableProgramId& ); \
1659 #define RT_INTERNAL_BOUND_CALLABLE_PROGRAM_DEFS(...) public rti_internal_callableprogram::callableProgramIdBase<__VA_ARGS__> \
1670 template<typename T>
1674 template<typename ReturnT>
1676 template<typename ReturnT, typename Arg0T>
1678 template<typename ReturnT, typename Arg0T, typename Arg1T>
1679 class
boundCallableProgramId<ReturnT(Arg0T,Arg1T)>: RT_INTERNAL_BOUND_CALLABLE_PROGRAM_DEFS(ReturnT,Arg0T,Arg1T);
1680 template<typename ReturnT, typename Arg0T, typename Arg1T, typename Arg2T>
1681 class
boundCallableProgramId<ReturnT(Arg0T,Arg1T,Arg2T)>: RT_INTERNAL_BOUND_CALLABLE_PROGRAM_DEFS(ReturnT,Arg0T,Arg1T,Arg2T);
1682 template<typename ReturnT, typename Arg0T, typename Arg1T, typename Arg2T, typename Arg3T>
1683 class
boundCallableProgramId<ReturnT(Arg0T,Arg1T,Arg2T,Arg3T)>: RT_INTERNAL_BOUND_CALLABLE_PROGRAM_DEFS(ReturnT,Arg0T,Arg1T,Arg2T,Arg3T);
1684 template<typename ReturnT, typename Arg0T, typename Arg1T, typename Arg2T, typename Arg3T,
1686 class
boundCallableProgramId<ReturnT(Arg0T,Arg1T,Arg2T,Arg3T,Arg4T)>: RT_INTERNAL_BOUND_CALLABLE_PROGRAM_DEFS(ReturnT,Arg0T,Arg1T,Arg2T,Arg3T,Arg4T);
1687 template<typename ReturnT, typename Arg0T, typename Arg1T, typename Arg2T, typename Arg3T,
1688 typename Arg4T, typename Arg5T>
1689 class
boundCallableProgramId<ReturnT(Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T)>: RT_INTERNAL_BOUND_CALLABLE_PROGRAM_DEFS(ReturnT,Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T);
1690 template<typename ReturnT, typename Arg0T, typename Arg1T, typename Arg2T, typename Arg3T,
1691 typename Arg4T, typename Arg5T, typename Arg6T>
1692 class
boundCallableProgramId<ReturnT(Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T)>: RT_INTERNAL_BOUND_CALLABLE_PROGRAM_DEFS(ReturnT,Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T);
1693 template<typename ReturnT, typename Arg0T, typename Arg1T, typename Arg2T, typename Arg3T,
1694 typename Arg4T, typename Arg5T, typename Arg6T, typename Arg7T>
1695 class
boundCallableProgramId<ReturnT(Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T,Arg7T)>: RT_INTERNAL_BOUND_CALLABLE_PROGRAM_DEFS(ReturnT,Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T,Arg7T);
1696 template<typename ReturnT, typename Arg0T, typename Arg1T, typename Arg2T, typename Arg3T,
1697 typename Arg4T, typename Arg5T, typename Arg6T, typename Arg7T, typename Arg8T>
1698 class
boundCallableProgramId<ReturnT(Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T,Arg7T,Arg8T)>: RT_INTERNAL_BOUND_CALLABLE_PROGRAM_DEFS(ReturnT,Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T,Arg7T,Arg8T);
1699 template<typename ReturnT, typename Arg0T, typename Arg1T, typename Arg2T, typename Arg3T,
1700 typename Arg4T, typename Arg5T, typename Arg6T, typename Arg7T, typename Arg8T, typename Arg9T>
1701 class
boundCallableProgramId<ReturnT(Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T,Arg7T,Arg8T,Arg9T)>: RT_INTERNAL_BOUND_CALLABLE_PROGRAM_DEFS(ReturnT,Arg0T,Arg1T,Arg2T,Arg3T,Arg4T,Arg5T,Arg6T,Arg7T,Arg8T,Arg9T);
1705 namespace rti_internal_typeinfo {
1707 template <
typename T>
1710 static const int m_typeenum = _OPTIX_TYPE_ENUM_PROGRAM_ID;
1714 template <
typename T>
1717 static const int m_typeenum = _OPTIX_TYPE_ENUM_PROGRAM_AS_ID;
1751 #define rtCallableProgramId optix::callableProgramId
1787 #define rtCallableProgramX optix::boundCallableProgramId
1788 #ifdef RT_USE_TEMPLATED_RTCALLABLEPROGRAM
1789 # undef rtCallableProgram
1790 # define rtCallableProgram optix::boundCallableProgramId
1827 optix::rt_trace(*(
unsigned int*)&topNode, ray.origin, ray.direction, ray.ray_type, ray.tmin, ray.tmax, &prd,
sizeof(T));
1877 return optix::rt_potential_intersection( tmin );
1906 return optix::rt_report_intersection( material );
1942 optix::rt_ignore_intersection();
1971 optix::rt_terminate_ray();
2011 optix::rt_intersect_child( index );
2051 return optix::rt_transform_point( kind, p );
2092 return optix::rt_transform_vector( kind, v );
2133 return optix::rt_transform_normal( kind, n );
2173 return optix::rt_get_transform( kind, matrix );
2213 static inline __device__
void rtPrintf(
const char* fmt )
2216 optix::rt_print_start(fmt,sz);
2218 template<
typename T1>
2219 static inline __device__
void rtPrintf(
const char* fmt, T1 arg1 )
2222 _RT_PRINTF_ARG_1( arg1 );
2224 _RT_PRINTF_ARG_2( arg1 );
2226 template<
typename T1,
typename T2>
2227 static inline __device__
void rtPrintf(
const char* fmt, T1 arg1, T2 arg2 )
2230 _RT_PRINTF_ARG_1( arg1 );
2231 _RT_PRINTF_ARG_1( arg2 );
2233 _RT_PRINTF_ARG_2( arg1 );
2234 _RT_PRINTF_ARG_2( arg2 );
2236 template<
typename T1,
typename T2,
typename T3>
2237 static inline __device__
void rtPrintf(
const char* fmt, T1 arg1, T2 arg2, T3 arg3 )
2240 _RT_PRINTF_ARG_1( arg1 );
2241 _RT_PRINTF_ARG_1( arg2 );
2242 _RT_PRINTF_ARG_1( arg3 );
2244 _RT_PRINTF_ARG_2( arg1 );
2245 _RT_PRINTF_ARG_2( arg2 );
2246 _RT_PRINTF_ARG_2( arg3 );
2248 template<
typename T1,
typename T2,
typename T3,
typename T4>
2249 static inline __device__
void rtPrintf(
const char* fmt, T1 arg1, T2 arg2, T3 arg3, T4 arg4 )
2252 _RT_PRINTF_ARG_1( arg1 );
2253 _RT_PRINTF_ARG_1( arg2 );
2254 _RT_PRINTF_ARG_1( arg3 );
2255 _RT_PRINTF_ARG_1( arg4 );
2257 _RT_PRINTF_ARG_2( arg1 );
2258 _RT_PRINTF_ARG_2( arg2 );
2259 _RT_PRINTF_ARG_2( arg3 );
2260 _RT_PRINTF_ARG_2( arg4 );
2262 template<
typename T1,
typename T2,
typename T3,
typename T4,
typename T5>
2263 static inline __device__
void rtPrintf(
const char* fmt, T1 arg1, T2 arg2, T3 arg3, T4 arg4, T5 arg5 )
2266 _RT_PRINTF_ARG_1( arg1 );
2267 _RT_PRINTF_ARG_1( arg2 );
2268 _RT_PRINTF_ARG_1( arg3 );
2269 _RT_PRINTF_ARG_1( arg4 );
2270 _RT_PRINTF_ARG_1( arg5 );
2272 _RT_PRINTF_ARG_2( arg1 );
2273 _RT_PRINTF_ARG_2( arg2 );
2274 _RT_PRINTF_ARG_2( arg3 );
2275 _RT_PRINTF_ARG_2( arg4 );
2276 _RT_PRINTF_ARG_2( arg5 );
2278 template<
typename T1,
typename T2,
typename T3,
typename T4,
typename T5,
typename T6>
2279 static inline __device__
void rtPrintf(
const char* fmt, T1 arg1, T2 arg2, T3 arg3, T4 arg4, T5 arg5, T6 arg6 )
2282 _RT_PRINTF_ARG_1( arg1 );
2283 _RT_PRINTF_ARG_1( arg2 );
2284 _RT_PRINTF_ARG_1( arg3 );
2285 _RT_PRINTF_ARG_1( arg4 );
2286 _RT_PRINTF_ARG_1( arg5 );
2287 _RT_PRINTF_ARG_1( arg6 );
2289 _RT_PRINTF_ARG_2( arg1 );
2290 _RT_PRINTF_ARG_2( arg2 );
2291 _RT_PRINTF_ARG_2( arg3 );
2292 _RT_PRINTF_ARG_2( arg4 );
2293 _RT_PRINTF_ARG_2( arg5 );
2294 _RT_PRINTF_ARG_2( arg6 );
2296 template<
typename T1,
typename T2,
typename T3,
typename T4,
typename T5,
typename T6,
typename T7>
2297 static inline __device__
void rtPrintf(
const char* fmt, T1 arg1, T2 arg2, T3 arg3, T4 arg4, T5 arg5, T6 arg6, T7 arg7 )
2300 _RT_PRINTF_ARG_1( arg1 );
2301 _RT_PRINTF_ARG_1( arg2 );
2302 _RT_PRINTF_ARG_1( arg3 );
2303 _RT_PRINTF_ARG_1( arg4 );
2304 _RT_PRINTF_ARG_1( arg5 );
2305 _RT_PRINTF_ARG_1( arg6 );
2306 _RT_PRINTF_ARG_1( arg7 );
2308 _RT_PRINTF_ARG_2( arg1 );
2309 _RT_PRINTF_ARG_2( arg2 );
2310 _RT_PRINTF_ARG_2( arg3 );
2311 _RT_PRINTF_ARG_2( arg4 );
2312 _RT_PRINTF_ARG_2( arg5 );
2313 _RT_PRINTF_ARG_2( arg6 );
2314 _RT_PRINTF_ARG_2( arg7 );
2316 template<
typename T1,
typename T2,
typename T3,
typename T4,
typename T5,
typename T6,
typename T7,
typename T8>
2317 static inline __device__
void rtPrintf(
const char* fmt, T1 arg1, T2 arg2, T3 arg3, T4 arg4, T5 arg5, T6 arg6, T7 arg7, T8 arg8 )
2320 _RT_PRINTF_ARG_1( arg1 );
2321 _RT_PRINTF_ARG_1( arg2 );
2322 _RT_PRINTF_ARG_1( arg3 );
2323 _RT_PRINTF_ARG_1( arg4 );
2324 _RT_PRINTF_ARG_1( arg5 );
2325 _RT_PRINTF_ARG_1( arg6 );
2326 _RT_PRINTF_ARG_1( arg7 );
2327 _RT_PRINTF_ARG_1( arg8 );
2329 _RT_PRINTF_ARG_2( arg1 );
2330 _RT_PRINTF_ARG_2( arg2 );
2331 _RT_PRINTF_ARG_2( arg3 );
2332 _RT_PRINTF_ARG_2( arg4 );
2333 _RT_PRINTF_ARG_2( arg5 );
2334 _RT_PRINTF_ARG_2( arg6 );
2335 _RT_PRINTF_ARG_2( arg7 );
2336 _RT_PRINTF_ARG_2( arg8 );
2338 template<
typename T1,
typename T2,
typename T3,
typename T4,
typename T5,
typename T6,
typename T7,
typename T8,
typename T9>
2339 static inline __device__
void rtPrintf(
const char* fmt, T1 arg1, T2 arg2, T3 arg3, T4 arg4, T5 arg5, T6 arg6, T7 arg7, T8 arg8, T9 arg9 )
2342 _RT_PRINTF_ARG_1( arg1 );
2343 _RT_PRINTF_ARG_1( arg2 );
2344 _RT_PRINTF_ARG_1( arg3 );
2345 _RT_PRINTF_ARG_1( arg4 );
2346 _RT_PRINTF_ARG_1( arg5 );
2347 _RT_PRINTF_ARG_1( arg6 );
2348 _RT_PRINTF_ARG_1( arg7 );
2349 _RT_PRINTF_ARG_1( arg8 );
2350 _RT_PRINTF_ARG_1( arg9 );
2352 _RT_PRINTF_ARG_2( arg1 );
2353 _RT_PRINTF_ARG_2( arg2 );
2354 _RT_PRINTF_ARG_2( arg3 );
2355 _RT_PRINTF_ARG_2( arg4 );
2356 _RT_PRINTF_ARG_2( arg5 );
2357 _RT_PRINTF_ARG_2( arg6 );
2358 _RT_PRINTF_ARG_2( arg7 );
2359 _RT_PRINTF_ARG_2( arg8 );
2360 _RT_PRINTF_ARG_2( arg9 );
2362 template<
typename T1,
typename T2,
typename T3,
typename T4,
typename T5,
typename T6,
typename T7,
typename T8,
typename T9,
typename T10>
2363 static inline __device__
void rtPrintf(
const char* fmt, T1 arg1, T2 arg2, T3 arg3, T4 arg4, T5 arg5, T6 arg6, T7 arg7, T8 arg8, T9 arg9, T10 arg10 )
2366 _RT_PRINTF_ARG_1( arg1 );
2367 _RT_PRINTF_ARG_1( arg2 );
2368 _RT_PRINTF_ARG_1( arg3 );
2369 _RT_PRINTF_ARG_1( arg4 );
2370 _RT_PRINTF_ARG_1( arg5 );
2371 _RT_PRINTF_ARG_1( arg6 );
2372 _RT_PRINTF_ARG_1( arg7 );
2373 _RT_PRINTF_ARG_1( arg8 );
2374 _RT_PRINTF_ARG_1( arg9 );
2375 _RT_PRINTF_ARG_1( arg10 );
2377 _RT_PRINTF_ARG_2( arg1 );
2378 _RT_PRINTF_ARG_2( arg2 );
2379 _RT_PRINTF_ARG_2( arg3 );
2380 _RT_PRINTF_ARG_2( arg4 );
2381 _RT_PRINTF_ARG_2( arg5 );
2382 _RT_PRINTF_ARG_2( arg6 );
2383 _RT_PRINTF_ARG_2( arg7 );
2384 _RT_PRINTF_ARG_2( arg8 );
2385 _RT_PRINTF_ARG_2( arg9 );
2386 _RT_PRINTF_ARG_2( arg10 );
2388 template<
typename T1,
typename T2,
typename T3,
typename T4,
typename T5,
typename T6,
typename T7,
typename T8,
typename T9,
typename T10,
typename T11>
2389 static inline __device__
void rtPrintf(
const char* fmt, T1 arg1, T2 arg2, T3 arg3, T4 arg4, T5 arg5, T6 arg6, T7 arg7, T8 arg8, T9 arg9, T10 arg10, T11 arg11 )
2392 _RT_PRINTF_ARG_1( arg1 );
2393 _RT_PRINTF_ARG_1( arg2 );
2394 _RT_PRINTF_ARG_1( arg3 );
2395 _RT_PRINTF_ARG_1( arg4 );
2396 _RT_PRINTF_ARG_1( arg5 );
2397 _RT_PRINTF_ARG_1( arg6 );
2398 _RT_PRINTF_ARG_1( arg7 );
2399 _RT_PRINTF_ARG_1( arg8 );
2400 _RT_PRINTF_ARG_1( arg9 );
2401 _RT_PRINTF_ARG_1( arg10 );
2402 _RT_PRINTF_ARG_1( arg11 );
2404 _RT_PRINTF_ARG_2( arg1 );
2405 _RT_PRINTF_ARG_2( arg2 );
2406 _RT_PRINTF_ARG_2( arg3 );
2407 _RT_PRINTF_ARG_2( arg4 );
2408 _RT_PRINTF_ARG_2( arg5 );
2409 _RT_PRINTF_ARG_2( arg6 );
2410 _RT_PRINTF_ARG_2( arg7 );
2411 _RT_PRINTF_ARG_2( arg8 );
2412 _RT_PRINTF_ARG_2( arg9 );
2413 _RT_PRINTF_ARG_2( arg10 );
2414 _RT_PRINTF_ARG_2( arg11 );
2416 template<
typename T1,
typename T2,
typename T3,
typename T4,
typename T5,
typename T6,
typename T7,
typename T8,
typename T9,
typename T10,
typename T11,
typename T12>
2417 static inline __device__
void rtPrintf(
const char* fmt, T1 arg1, T2 arg2, T3 arg3, T4 arg4, T5 arg5, T6 arg6, T7 arg7, T8 arg8, T9 arg9, T10 arg10, T11 arg11, T12 arg12 )
2420 _RT_PRINTF_ARG_1( arg1 );
2421 _RT_PRINTF_ARG_1( arg2 );
2422 _RT_PRINTF_ARG_1( arg3 );
2423 _RT_PRINTF_ARG_1( arg4 );
2424 _RT_PRINTF_ARG_1( arg5 );
2425 _RT_PRINTF_ARG_1( arg6 );
2426 _RT_PRINTF_ARG_1( arg7 );
2427 _RT_PRINTF_ARG_1( arg8 );
2428 _RT_PRINTF_ARG_1( arg9 );
2429 _RT_PRINTF_ARG_1( arg10 );
2430 _RT_PRINTF_ARG_1( arg11 );
2431 _RT_PRINTF_ARG_1( arg12 );
2433 _RT_PRINTF_ARG_2( arg1 );
2434 _RT_PRINTF_ARG_2( arg2 );
2435 _RT_PRINTF_ARG_2( arg3 );
2436 _RT_PRINTF_ARG_2( arg4 );
2437 _RT_PRINTF_ARG_2( arg5 );
2438 _RT_PRINTF_ARG_2( arg6 );
2439 _RT_PRINTF_ARG_2( arg7 );
2440 _RT_PRINTF_ARG_2( arg8 );
2441 _RT_PRINTF_ARG_2( arg9 );
2442 _RT_PRINTF_ARG_2( arg10 );
2443 _RT_PRINTF_ARG_2( arg11 );
2444 _RT_PRINTF_ARG_2( arg12 );
2450 #undef _RT_PRINTF_ARG_1
2451 #undef _RT_PRINTF_ARG_2
2454 namespace rti_internal_register {
2455 extern __device__
void* reg_bitness_detector;
2456 extern __device__
volatile unsigned long long reg_exception_64_detail0;
2457 extern __device__
volatile unsigned long long reg_exception_64_detail1;
2458 extern __device__
volatile unsigned long long reg_exception_64_detail2;
2459 extern __device__
volatile unsigned long long reg_exception_64_detail3;
2460 extern __device__
volatile unsigned long long reg_exception_64_detail4;
2461 extern __device__
volatile unsigned long long reg_exception_64_detail5;
2462 extern __device__
volatile unsigned long long reg_exception_64_detail6;
2463 extern __device__
volatile unsigned long long reg_exception_64_detail7;
2464 extern __device__
volatile unsigned long long reg_exception_64_detail8;
2465 extern __device__
volatile unsigned long long reg_exception_64_detail9;
2466 extern __device__
volatile unsigned int reg_exception_detail0;
2467 extern __device__
volatile unsigned int reg_exception_detail1;
2468 extern __device__
volatile unsigned int reg_exception_detail2;
2469 extern __device__
volatile unsigned int reg_exception_detail3;
2470 extern __device__
volatile unsigned int reg_exception_detail4;
2471 extern __device__
volatile unsigned int reg_exception_detail5;
2472 extern __device__
volatile unsigned int reg_exception_detail6;
2473 extern __device__
volatile unsigned int reg_exception_detail7;
2474 extern __device__
volatile unsigned int reg_exception_detail8;
2475 extern __device__
volatile unsigned int reg_exception_detail9;
2476 extern __device__
volatile unsigned int reg_rayIndex_x;
2477 extern __device__
volatile unsigned int reg_rayIndex_y;
2478 extern __device__
volatile unsigned int reg_rayIndex_z;
2516 static inline __device__
void rtThrow(
unsigned int code )
2518 optix::rt_throw( code );
2550 return optix::rt_get_exception_code();
2588 rtPrintf(
"Caught RT_EXCEPTION_STACK_OVERFLOW\n"
2589 " launch index : %d, %d, %d\n",
2590 rti_internal_register::reg_rayIndex_x,
2591 rti_internal_register::reg_rayIndex_y,
2592 rti_internal_register::reg_rayIndex_z
2597 #if defined(__x86_64) || defined(AMD64) || defined(_M_AMD64)
2598 const unsigned int dim = rti_internal_register::reg_exception_detail0;
2600 rtPrintf(
"Caught RT_EXCEPTION_BUFFER_INDEX_OUT_OF_BOUNDS\n"
2601 " launch index : %d, %d, %d\n"
2602 " buffer address : 0x%llX\n"
2603 " dimensionality : %d\n"
2604 " size : %lldx%lldx%lld\n"
2605 " element size : %d\n"
2606 " accessed index : %lld, %lld, %lld\n",
2607 rti_internal_register::reg_rayIndex_x,
2608 rti_internal_register::reg_rayIndex_y,
2609 rti_internal_register::reg_rayIndex_z,
2610 rti_internal_register::reg_exception_64_detail0,
2611 rti_internal_register::reg_exception_detail0,
2612 rti_internal_register::reg_exception_64_detail1,
2613 dim > 1 ? rti_internal_register::reg_exception_64_detail2 : 1,
2614 dim > 2 ? rti_internal_register::reg_exception_64_detail3 : 1,
2615 rti_internal_register::reg_exception_detail1,
2616 rti_internal_register::reg_exception_64_detail4,
2617 rti_internal_register::reg_exception_64_detail5,
2618 rti_internal_register::reg_exception_64_detail6
2621 const unsigned int dim = rti_internal_register::reg_exception_detail1;
2623 rtPrintf(
"Caught RT_EXCEPTION_BUFFER_INDEX_OUT_OF_BOUNDS\n"
2624 " launch index : %d, %d, %d\n"
2625 " buffer address : 0x%X\n"
2626 " dimensionality : %d\n"
2627 " size : %dx%dx%d\n"
2628 " element size : %d\n"
2629 " accessed index : %d, %d, %d\n",
2630 rti_internal_register::reg_rayIndex_x,
2631 rti_internal_register::reg_rayIndex_y,
2632 rti_internal_register::reg_rayIndex_z,
2633 rti_internal_register::reg_exception_detail0,
2634 rti_internal_register::reg_exception_detail1,
2635 rti_internal_register::reg_exception_detail2,
2636 dim > 1 ? rti_internal_register::reg_exception_detail3 : 1,
2637 dim > 2 ? rti_internal_register::reg_exception_detail4 : 1,
2638 rti_internal_register::reg_exception_detail5,
2639 rti_internal_register::reg_exception_detail6,
2640 rti_internal_register::reg_exception_detail7,
2641 rti_internal_register::reg_exception_detail8
2647 rtPrintf(
"Caught RT_EXCEPTION_PROGRAM_ID_INVALID\n");
2648 switch(rti_internal_register::reg_exception_detail1)
2651 rtPrintf(
"\tprogram ID equal to RT_PROGRAM_ID_NULL used\n");
2654 rtPrintf(
"\tprogram ID (%d) is not in the valid range of [1,size)\n", rti_internal_register::reg_exception_detail0);
2657 rtPrintf(
"\tprogram ID of a deleted program used\n");
2663 rtPrintf(
"Caught RT_EXCEPTION_TEXTURE_ID_INVALID\n");
2664 switch(rti_internal_register::reg_exception_detail1)
2667 rtPrintf(
"\ttexture ID (%d) is invalid (0)\n", rti_internal_register::reg_exception_detail0);
2670 rtPrintf(
"\ttexture ID (%d) is not in the valid range of [1,size)\n", rti_internal_register::reg_exception_detail0);
2673 rtPrintf(
"\ttexture ID (%d) is invalid (-1)\n", rti_internal_register::reg_exception_detail0);
2679 rtPrintf(
"Caught RT_EXCEPTION_BUFFER_ID_INVALID\n");
2680 switch(rti_internal_register::reg_exception_detail1)
2683 rtPrintf(
"\tbuffer ID equal to RT_BUFFER_ID_NULL used\n");
2686 rtPrintf(
"\tbuffer ID (%d) is not in the valid range of [1,size)\n", rti_internal_register::reg_exception_detail0);
2689 rtPrintf(
"\tBuffer ID of a deleted buffer used\n");
2695 #if defined(__x86_64) || defined(AMD64) || defined(_M_AMD64)
2696 const unsigned int dim = rti_internal_register::reg_exception_detail0;
2698 rtPrintf(
"Caught RT_EXCEPTION_INDEX_OUT_OF_BOUNDS\n"
2699 " launch index : %d, %d, %d\n"
2700 " buffer address : 0x%llX\n"
2702 " accessed index : %lld\n",
2703 rti_internal_register::reg_rayIndex_x,
2704 rti_internal_register::reg_rayIndex_y,
2705 rti_internal_register::reg_rayIndex_z,
2706 rti_internal_register::reg_exception_64_detail0,
2707 rti_internal_register::reg_exception_64_detail1,
2708 rti_internal_register::reg_exception_64_detail2
2711 const unsigned int dim = rti_internal_register::reg_exception_detail1;
2713 rtPrintf(
"Caught RT_EXCEPTION_INDEX_OUT_OF_BOUNDS\n"
2714 " launch index : %d, %d, %d\n"
2715 " buffer address : 0x%X\n"
2717 " accessed index : %d\n",
2718 rti_internal_register::reg_rayIndex_x,
2719 rti_internal_register::reg_rayIndex_y,
2720 rti_internal_register::reg_rayIndex_z,
2721 rti_internal_register::reg_exception_detail0,
2722 rti_internal_register::reg_exception_detail1,
2723 rti_internal_register::reg_exception_detail2
2729 rtPrintf(
"Caught RT_EXCEPTION_INVALID_RAY\n"
2730 " launch index : %d, %d, %d\n"
2731 " ray origin : %f %f %f\n"
2732 " ray direction : %f %f %f\n"
2736 rti_internal_register::reg_rayIndex_x,
2737 rti_internal_register::reg_rayIndex_y,
2738 rti_internal_register::reg_rayIndex_z,
2739 __int_as_float(rti_internal_register::reg_exception_detail0),
2740 __int_as_float(rti_internal_register::reg_exception_detail1),
2741 __int_as_float(rti_internal_register::reg_exception_detail2),
2742 __int_as_float(rti_internal_register::reg_exception_detail3),
2743 __int_as_float(rti_internal_register::reg_exception_detail4),
2744 __int_as_float(rti_internal_register::reg_exception_detail5),
2745 rti_internal_register::reg_exception_detail6,
2746 __int_as_float(rti_internal_register::reg_exception_detail7),
2747 __int_as_float(rti_internal_register::reg_exception_detail8)
2753 rtPrintf(
"Caught RT_EXCEPTION_INTERNAL_ERROR\n"
2754 " launch index : %d, %d, %d\n"
2756 rti_internal_register::reg_rayIndex_x,
2757 rti_internal_register::reg_rayIndex_y,
2758 rti_internal_register::reg_rayIndex_z,
2759 rti_internal_register::reg_exception_detail0
2764 rtPrintf(
"Caught RT_EXCEPTION_USER+%d\n"
2765 " launch index : %d, %d, %d\n",
2767 rti_internal_register::reg_rayIndex_x,
2768 rti_internal_register::reg_rayIndex_y,
2769 rti_internal_register::reg_rayIndex_z
2775 rtPrintf(
"Caught unknown exception\n"
2776 " launch index : %d, %d, %d\n",
2777 rti_internal_register::reg_rayIndex_x,
2778 rti_internal_register::reg_rayIndex_y,
2779 rti_internal_register::reg_rayIndex_z
Opaque handle to a OptiX object.
Definition: optix_device.h:179
Definition: optix_declarations.h:222
Definition: optix_declarations.h:226
Definition: optix_device.h:53
Definition: optix_device.h:417
static __device__ bool rtPotentialIntersection(float tmin)
Determine whether a computed intersection is potentially valid.
Definition: optix_device.h:1875
Definition: optix_declarations.h:224
Definition: optix_device.h:1389
static __device__ void rtIntersectChild(unsigned int index)
Visit child of selector.
Definition: optix_device.h:2009
Definition: optix_declarations.h:228
RTprogramidnull
Definition: optix_declarations.h:344
bufferId is a host version of the device side bufferId.
Definition: optix_device.h:415
static __device__ float3 rtTransformVector(RTtransformkind kind, const float3 &v)
Apply the current transformation to a vector.
Definition: optix_device.h:2090
static __device__ void rtTrace(rtObject topNode, optix::Ray ray, T &prd)
Traces a ray.
Definition: optix_device.h:1825
static __device__ void rtPrintExceptionDetails()
Print information on a caught exception.
Definition: optix_device.h:2582
Definition: optix_declarations.h:225
static __device__ void rtGetTransform(RTtransformkind kind, float matrix[16])
Get requested transform.
Definition: optix_device.h:2171
Definition: optix_declarations.h:229
Definition: optix_device.h:437
static __device__ void rtPrintf(const char *fmt)
Prints text to the standard output.
Definition: optix_device.h:2213
static __device__ float3 rtTransformNormal(RTtransformkind kind, const float3 &n)
Apply the current transformation to a normal.
Definition: optix_device.h:2131
RTbufferidnull
Definition: optix_declarations.h:341
#define RT_INTERNAL_CALLABLE_PROGRAM_DEFS()
callableProgramId is a host version of the device side callableProgramId.
Definition: optixpp_namespace.h:1770
static __device__ bool rtReportIntersection(unsigned int material)
Report an intersection with the current object and the specified material.
Definition: optix_device.h:1904
Definition: optix_device.h:1414
static __device__ void rtThrow(unsigned int code)
Throw a user exception.
Definition: optix_device.h:2516
Definition: optix_device.h:1381
Definition: optix_defines.h:74
Definition: optix_device.h:1601
static __device__ unsigned int rtGetExceptionCode()
Retrieves the type of a caught exception.
Definition: optix_device.h:2548
Definition: optix_device.h:333
OptiX public API declarations.
Definition: optix_declarations.h:230
Definition: optix_declarations.h:227
Definition: optix_declarations.h:223
Definition: optix_device.h:1383
Definition: optix_device.h:1671
static __device__ float3 rtTransformPoint(RTtransformkind kind, const float3 &p)
Apply the current transformation to a point.
Definition: optix_device.h:2049
static __device__ void rtTerminateRay()
Terminate traversal associated with the current ray.
Definition: optix_device.h:1969
RTtransformkind
Definition: optix_defines.h:45
static __device__ void rtIgnoreIntersection()
Cancels the potential intersection with current ray.
Definition: optix_device.h:1940