Differential D13361 Diff 45286 intern/cycles/kernel/device/optix/compat.h

Changeset View

Standalone View

intern/cycles/kernel/device/optix/compat.h

	Show First 20 Lines • Show All 47 Lines • ▼ Show 20 Lines
	#define ccl_device \			#define ccl_device \
	__device__ __forceinline__ // Function calls are bad for OptiX performance, so inline everything			__device__ __forceinline__ // Function calls are bad for OptiX performance, so inline everything
	#define ccl_device_inline ccl_device			#define ccl_device_inline ccl_device
	#define ccl_device_forceinline ccl_device			#define ccl_device_forceinline ccl_device
	#define ccl_device_inline_method ccl_device			#define ccl_device_inline_method ccl_device
	#define ccl_device_noinline __device__ __noinline__			#define ccl_device_noinline __device__ __noinline__
	#define ccl_device_noinline_cpu ccl_device			#define ccl_device_noinline_cpu ccl_device
	#define ccl_global			#define ccl_global
	#define ccl_static_constant __constant__			#define ccl_inline_constant __constant__
	#define ccl_device_constant __constant__ __device__			#define ccl_device_constant __constant__ __device__
	#define ccl_constant const			#define ccl_constant const
	#define ccl_gpu_shared __shared__			#define ccl_gpu_shared __shared__
	#define ccl_private			#define ccl_private
	#define ccl_may_alias			#define ccl_may_alias
	#define ccl_restrict __restrict__			#define ccl_restrict __restrict__
	#define ccl_loop_no_unroll			#define ccl_loop_no_unroll
	#define ccl_align(n) __align__(n)			#define ccl_align(n) __align__(n)
	Show All 17 Lines
	#define ccl_gpu_global_id_x() (ccl_gpu_block_idx_x * ccl_gpu_block_dim_x + ccl_gpu_thread_idx_x)			#define ccl_gpu_global_id_x() (ccl_gpu_block_idx_x * ccl_gpu_block_dim_x + ccl_gpu_thread_idx_x)
	#define ccl_gpu_global_size_x() (ccl_gpu_grid_dim_x * ccl_gpu_block_dim_x)			#define ccl_gpu_global_size_x() (ccl_gpu_grid_dim_x * ccl_gpu_block_dim_x)

	/* GPU warp synchronization. */			/* GPU warp synchronization. */

	#define ccl_gpu_syncthreads() __syncthreads()			#define ccl_gpu_syncthreads() __syncthreads()
	#define ccl_gpu_ballot(predicate) __ballot_sync(0xFFFFFFFF, predicate)			#define ccl_gpu_ballot(predicate) __ballot_sync(0xFFFFFFFF, predicate)
	#define ccl_gpu_shfl_down_sync(mask, var, detla) __shfl_down_sync(mask, var, detla)			#define ccl_gpu_shfl_down_sync(mask, var, detla) __shfl_down_sync(mask, var, detla)
	#define ccl_gpu_popc(x) __popc(x)

	/* GPU texture objects */			/* GPU texture objects */

	typedef unsigned long long CUtexObject;			typedef unsigned long long CUtexObject;
	typedef CUtexObject ccl_gpu_tex_object;			typedef CUtexObject ccl_gpu_tex_object;

	template<typename T>			template<typename T>
	ccl_device_forceinline T ccl_gpu_tex_object_read_2D(const ccl_gpu_tex_object texobj,			ccl_device_forceinline T ccl_gpu_tex_object_read_2D(const ccl_gpu_tex_object texobj,
	Show All 37 Lines