Extras/CUDA/btCudaDefines.h

   1 /*
   2 Bullet Continuous Collision Detection and Physics Library, http://bulletphysics.org
   3 Copyright (C) 2006, 2007 Sony Computer Entertainment Inc.
   4
   5 This software is provided 'as-is', without any express or implied warranty.
   6 In no event will the authors be held liable for any damages arising from the use of this software.
   7 Permission is granted to anyone to use this software for any purpose,
   8 including commercial applications, and to alter it and redistribute it freely,
   9 subject to the following restrictions:
  10
  11 1. The origin of this software must not be misrepresented; you must not claim that you wrote the original software. If you use this software in a product, an acknowledgment in the product documentation would be appreciated but is not required.
  12 2. Altered source versions must be plainly marked as such, and must not be misrepresented as being the original software.
  13 3. This notice may not be removed or altered from any source distribution.
  14 */
  15
  16
  17
  18 // Common preprocessor definitions for CUDA compiler
  19
  20
  21
  22 #ifndef BTCUDADEFINES_H
  23 #define BTCUDADEFINES_H
  24
  25
  26
  27 #ifdef __DEVICE_EMULATION__
  28         #define B_CUDA_USE_TEX 0
  29 #else
  30         #define B_CUDA_USE_TEX 1
  31 #endif
  32
  33
  34 #if B_CUDA_USE_TEX
  35         #define BT_GPU_FETCH(t, i) tex_fetch3F1U(tex1Dfetch(t##Tex, i))
  36         #define BT_GPU_FETCH4(t, i) tex1Dfetch(t##Tex, i)
  37 #else
  38         #define BT_GPU_FETCH(t, i) t[i]
  39         #define BT_GPU_FETCH4(t, i) t[i]
  40 #endif
  41
  42
  43
  44 #define BT_GPU___device__ __device__
  45 #define BT_GPU___devdata__ __device__
  46 #define BT_GPU___constant__ __constant__
  47 #define BT_GPU_max(a, b) max(a, b)
  48 #define BT_GPU_min(a, b) min(a, b)
  49 #define BT_GPU_params params
  50 #define BT_GPU___mul24(a, b) __mul24(a, b)
  51 #define BT_GPU___global__ __global__
  52 #define BT_GPU___shared__ __shared__
  53 #define BT_GPU___syncthreads() __syncthreads()
  54 #define BT_GPU_make_uint2(x, y) make_uint2(x, y)
  55 #define BT_GPU_make_int3(x, y, z) make_int3(x, y, z)
  56 #define BT_GPU_make_float3(x, y, z) make_float3(x, y, z)
  57 #define BT_GPU_make_float34(x) make_float3(x)
  58 #define BT_GPU_make_float31(x) make_float3(x)
  59 #define BT_GPU_make_float42(a, b) make_float4(a, b)
  60 #define BT_GPU_make_float44(a, b, c, d) make_float4(a, b, c, d)
  61 #define BT_GPU_PREF(func) btCuda_##func
  62 #define BT_GPU_Memset cudaMemset
  63 #define BT_GPU_MemcpyToSymbol(a, b, c) cudaMemcpyToSymbol(a, b, c)
  64 #define BT_GPU_blockIdx blockIdx
  65 #define BT_GPU_blockDim blockDim
  66 #define BT_GPU_threadIdx threadIdx
  67 #define BT_GPU_dot(a, b) dot(a, b)
  68 #define BT_GPU_dot4(a, b) dot(a, b)
  69 #define BT_GPU_cross(a, b) cross(a, b)
  70 #define BT_GPU_BindTexture(a, b, c, d) cudaBindTexture(a, b, c, d)
  71 #define BT_GPU_UnbindTexture(a) cudaUnbindTexture(a)
  72 #define BT_GPU_EXECKERNEL(numb, numt, kfunc, args) kfunc<<<numb, numt>>>args
  73
  74
  75
  76 //! Check for CUDA error
  77 #define BT_GPU_CHECK_ERROR(errorMessage)                                                                        \
  78         do                                                                                                                                              \
  79         {                                                                                                                                               \
  80                 cudaError_t err = cudaGetLastError();                                                           \
  81                 if(err != cudaSuccess)                                                                                          \
  82                 {                                                                                                                                       \
  83                         fprintf(stderr,"Cuda error: %s in file '%s' in line %i : %s.\n",\
  84                                 errorMessage, __FILE__, __LINE__, cudaGetErrorString( err));\
  85                         btCuda_exit(EXIT_FAILURE);                                      \
  86                 }                                                                   \
  87                 err = cudaThreadSynchronize();                                      \
  88                 if(err != cudaSuccess)                                                                                          \
  89                 {                                                                                                                                       \
  90                         fprintf(stderr,"Cuda error: %s in file '%s' in line %i : %s.\n",\
  91                                 errorMessage, __FILE__, __LINE__, cudaGetErrorString( err));\
  92                         btCuda_exit(EXIT_FAILURE);                                                                              \
  93                 }                                                                                                                                       \
  94         }                                                                                                                                               \
  95         while(0)
  96
  97
  98
  99 #define BT_GPU_SAFE_CALL_NO_SYNC(call)                                                                          \
 100         do                                                                                                                                              \
 101         {                                                                                                                                               \
 102                 cudaError err = call;                                                                                           \
 103                 if(err != cudaSuccess)                                                                                          \
 104                 {                                                                                                                                       \
 105                         fprintf(stderr, "Cuda error in file '%s' in line %i : %s.\n",   \
 106                                 __FILE__, __LINE__, cudaGetErrorString( err) );             \
 107                         btCuda_exit(EXIT_FAILURE);                                                                              \
 108                 }                                                                                                                                       \
 109         }                                                                                                                                               \
 110         while(0)
 111
 112
 113
 114 #define BT_GPU_SAFE_CALL(call)                                                                                          \
 115         do                                                                                                                                              \
 116         {                                                                                                                                               \
 117                 BT_GPU_SAFE_CALL_NO_SYNC(call);                                                                         \
 118                 cudaError err = cudaThreadSynchronize();                                                        \
 119                 if(err != cudaSuccess)                                                                                          \
 120                 {                                                                                                                                       \
 121                         fprintf(stderr,"Cuda errorSync in file '%s' in line %i : %s.\n",\
 122                                 __FILE__, __LINE__, cudaGetErrorString( err) );                         \
 123                         btCuda_exit(EXIT_FAILURE);                                                                              \
 124                 }                                                                                                                                       \
 125         } while (0)
 126
 127
 128
 129 extern "C" void btCuda_exit(int val);
 130
 131
 132
 133 #endif // BTCUDADEFINES_H
 134
 135
 136
 137
 138