intern/cycles/kernel/kernel_compat_opencl.h

   1 /*
   2  * Copyright 2011-2013 Blender Foundation
   3  *
   4  * Licensed under the Apache License, Version 2.0 (the "License");
   5  * you may not use this file except in compliance with the License.
   6  * You may obtain a copy of the License at
   7  *
   8  * http://www.apache.org/licenses/LICENSE-2.0
   9  *
  10  * Unless required by applicable law or agreed to in writing, software
  11  * distributed under the License is distributed on an "AS IS" BASIS,
  12  * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
  13  * See the License for the specific language governing permissions and
  14  * limitations under the License.
  15  */
  16
  17 #ifndef __KERNEL_COMPAT_OPENCL_H__
  18 #define __KERNEL_COMPAT_OPENCL_H__
  19
  20 #define __KERNEL_GPU__
  21 #define __KERNEL_OPENCL__
  22
  23 /* no namespaces in opencl */
  24 #define CCL_NAMESPACE_BEGIN
  25 #define CCL_NAMESPACE_END
  26
  27 #ifdef __CL_NOINLINE__
  28 #  define ccl_noinline __attribute__((noinline))
  29 #else
  30 #  define ccl_noinline
  31 #endif
  32
  33 /* in opencl all functions are device functions, so leave this empty */
  34 #define ccl_device
  35 #define ccl_device_inline ccl_device
  36 #define ccl_device_forceinline ccl_device
  37 #define ccl_device_noinline ccl_device ccl_noinline
  38 #define ccl_may_alias
  39 #define ccl_constant __constant
  40 #define ccl_global __global
  41 #define ccl_local __local
  42 #define ccl_local_param __local
  43 #define ccl_private __private
  44 #define ccl_restrict restrict
  45 #define ccl_align(n) __attribute__((aligned(n)))
  46
  47 #ifdef __SPLIT_KERNEL__
  48 #  define ccl_addr_space __global
  49 #else
  50 #  define ccl_addr_space
  51 #endif
  52
  53 #define ccl_local_id(d) get_local_id(d)
  54 #define ccl_global_id(d) get_global_id(d)
  55
  56 #define ccl_local_size(d) get_local_size(d)
  57 #define ccl_global_size(d) get_global_size(d)
  58
  59 #define ccl_group_id(d) get_group_id(d)
  60 #define ccl_num_groups(d) get_num_groups(d)
  61
  62 /* Selective nodes compilation. */
  63 #ifndef __NODES_MAX_GROUP__
  64 #  define __NODES_MAX_GROUP__ NODE_GROUP_LEVEL_MAX
  65 #endif
  66 #ifndef __NODES_FEATURES__
  67 #  define __NODES_FEATURES__ NODE_FEATURE_ALL
  68 #endif
  69
  70 /* no assert in opencl */
  71 #define kernel_assert(cond)
  72
  73 /* make_type definitions with opencl style element initializers */
  74 #ifdef make_float2
  75 #  undef make_float2
  76 #endif
  77 #ifdef make_float3
  78 #  undef make_float3
  79 #endif
  80 #ifdef make_float4
  81 #  undef make_float4
  82 #endif
  83 #ifdef make_int2
  84 #  undef make_int2
  85 #endif
  86 #ifdef make_int3
  87 #  undef make_int3
  88 #endif
  89 #ifdef make_int4
  90 #  undef make_int4
  91 #endif
  92 #ifdef make_uchar4
  93 #  undef make_uchar4
  94 #endif
  95
  96 #define make_float2(x, y) ((float2)(x, y))
  97 #define make_float3(x, y, z) ((float3)(x, y, z))
  98 #define make_float4(x, y, z, w) ((float4)(x, y, z, w))
  99 #define make_int2(x, y) ((int2)(x, y))
 100 #define make_int3(x, y, z) ((int3)(x, y, z))
 101 #define make_int4(x, y, z, w) ((int4)(x, y, z, w))
 102 #define make_uchar4(x, y, z, w) ((uchar4)(x, y, z, w))
 103
 104 /* math functions */
 105 #define __uint_as_float(x) as_float(x)
 106 #define __float_as_uint(x) as_uint(x)
 107 #define __int_as_float(x) as_float(x)
 108 #define __float_as_int(x) as_int(x)
 109 #define powf(x, y) pow(((float)(x)), ((float)(y)))
 110 #define fabsf(x) fabs(((float)(x)))
 111 #define copysignf(x, y) copysign(((float)(x)), ((float)(y)))
 112 #define asinf(x) asin(((float)(x)))
 113 #define acosf(x) acos(((float)(x)))
 114 #define atanf(x) atan(((float)(x)))
 115 #define floorf(x) floor(((float)(x)))
 116 #define ceilf(x) ceil(((float)(x)))
 117 #define hypotf(x, y) hypot(((float)(x)), ((float)(y)))
 118 #define atan2f(x, y) atan2(((float)(x)), ((float)(y)))
 119 #define fmaxf(x, y) fmax(((float)(x)), ((float)(y)))
 120 #define fminf(x, y) fmin(((float)(x)), ((float)(y)))
 121 #define fmodf(x, y) fmod((float)(x), (float)(y))
 122
 123 #ifndef __CL_USE_NATIVE__
 124 #  define sinf(x) native_sin(((float)(x)))
 125 #  define cosf(x) native_cos(((float)(x)))
 126 #  define tanf(x) native_tan(((float)(x)))
 127 #  define expf(x) native_exp(((float)(x)))
 128 #  define sqrtf(x) native_sqrt(((float)(x)))
 129 #  define logf(x) native_log(((float)(x)))
 130 #else
 131 #  define sinf(x) sin(((float)(x)))
 132 #  define cosf(x) cos(((float)(x)))
 133 #  define tanf(x) tan(((float)(x)))
 134 #  define expf(x) exp(((float)(x)))
 135 #  define sqrtf(x) sqrt(((float)(x)))
 136 #  define logf(x) log(((float)(x)))
 137 #endif
 138
 139 /* data lookup defines */
 140 #define kernel_data (*kg->data)
 141 #define kernel_tex_fetch(t, index) kg->t[index]
 142
 143 /* define NULL */
 144 #define NULL 0
 145
 146 #include "util/util_half.h"
 147 #include "util/util_types.h"
 148
 149 #endif /* __KERNEL_COMPAT_OPENCL_H__ */
 150