Blender  V3.3
math_float8.h
Go to the documentation of this file.
1 /* SPDX-License-Identifier: Apache-2.0
2  * Copyright 2022 Blender Foundation */
3 
4 #ifndef __UTIL_MATH_FLOAT8_H__
5 #define __UTIL_MATH_FLOAT8_H__
6 
7 #ifndef __UTIL_MATH_H__
8 # error "Do not include this file directly, include util/types.h instead."
9 #endif
10 
12 
13 /*******************************************************************************
14  * Declaration.
15  */
16 
17 ccl_device_inline float8_t operator+(const float8_t a, const float8_t b);
18 ccl_device_inline float8_t operator+(const float8_t a, const float f);
19 ccl_device_inline float8_t operator+(const float f, const float8_t a);
20 
21 ccl_device_inline float8_t operator-(const float8_t a);
22 ccl_device_inline float8_t operator-(const float8_t a, const float8_t b);
23 ccl_device_inline float8_t operator-(const float8_t a, const float f);
24 ccl_device_inline float8_t operator-(const float f, const float8_t a);
25 
26 ccl_device_inline float8_t operator*(const float8_t a, const float8_t b);
27 ccl_device_inline float8_t operator*(const float8_t a, const float f);
28 ccl_device_inline float8_t operator*(const float f, const float8_t a);
29 
30 ccl_device_inline float8_t operator/(const float8_t a, const float8_t b);
31 ccl_device_inline float8_t operator/(const float8_t a, float f);
32 ccl_device_inline float8_t operator/(const float f, const float8_t a);
33 
34 ccl_device_inline float8_t operator+=(float8_t a, const float8_t b);
35 
36 ccl_device_inline float8_t operator*=(float8_t a, const float8_t b);
37 ccl_device_inline float8_t operator*=(float8_t a, float f);
38 
39 ccl_device_inline float8_t operator/=(float8_t a, float f);
40 
41 ccl_device_inline bool operator==(const float8_t a, const float8_t b);
42 
43 ccl_device_inline float8_t rcp(const float8_t a);
44 ccl_device_inline float8_t sqrt(const float8_t a);
45 ccl_device_inline float8_t sqr(const float8_t a);
46 ccl_device_inline bool is_zero(const float8_t a);
47 ccl_device_inline float average(const float8_t a);
48 ccl_device_inline float8_t min(const float8_t a, const float8_t b);
49 ccl_device_inline float8_t max(const float8_t a, const float8_t b);
50 ccl_device_inline float8_t clamp(const float8_t a, const float8_t mn, const float8_t mx);
51 ccl_device_inline float8_t fabs(const float8_t a);
52 ccl_device_inline float8_t mix(const float8_t a, const float8_t b, float t);
53 ccl_device_inline float8_t saturate(const float8_t a);
54 
55 ccl_device_inline float8_t safe_divide(const float8_t a, const float b);
56 ccl_device_inline float8_t safe_divide(const float8_t a, const float8_t b);
57 
58 ccl_device_inline float reduce_min(const float8_t a);
59 ccl_device_inline float reduce_max(const float8_t a);
60 ccl_device_inline float reduce_add(const float8_t a);
61 
62 ccl_device_inline bool isequal(const float8_t a, const float8_t b);
63 
64 /*******************************************************************************
65  * Definition.
66  */
67 
69 {
70 #ifdef __KERNEL_AVX2__
71  return float8_t(_mm256_setzero_ps());
72 #else
73  return make_float8_t(0.0f);
74 #endif
75 }
76 
78 {
79  return make_float8_t(1.0f);
80 }
81 
82 ccl_device_inline float8_t operator+(const float8_t a, const float8_t b)
83 {
84 #ifdef __KERNEL_AVX2__
85  return float8_t(_mm256_add_ps(a.m256, b.m256));
86 #else
87  return make_float8_t(
88  a.a + b.a, a.b + b.b, a.c + b.c, a.d + b.d, a.e + b.e, a.f + b.f, a.g + b.g, a.h + b.h);
89 #endif
90 }
91 
92 ccl_device_inline float8_t operator+(const float8_t a, const float f)
93 {
94  return a + make_float8_t(f);
95 }
96 
97 ccl_device_inline float8_t operator+(const float f, const float8_t a)
98 {
99  return make_float8_t(f) + a;
100 }
101 
102 ccl_device_inline float8_t operator-(const float8_t a)
103 {
104 #ifdef __KERNEL_AVX2__
105  __m256 mask = _mm256_castsi256_ps(_mm256_set1_epi32(0x80000000));
106  return float8_t(_mm256_xor_ps(a.m256, mask));
107 #else
108  return make_float8_t(-a.a, -a.b, -a.c, -a.d, -a.e, -a.f, -a.g, -a.h);
109 #endif
110 }
111 
112 ccl_device_inline float8_t operator-(const float8_t a, const float8_t b)
113 {
114 #ifdef __KERNEL_AVX2__
115  return float8_t(_mm256_sub_ps(a.m256, b.m256));
116 #else
117  return make_float8_t(
118  a.a - b.a, a.b - b.b, a.c - b.c, a.d - b.d, a.e - b.e, a.f - b.f, a.g - b.g, a.h - b.h);
119 #endif
120 }
121 
122 ccl_device_inline float8_t operator-(const float8_t a, const float f)
123 {
124  return a - make_float8_t(f);
125 }
126 
127 ccl_device_inline float8_t operator-(const float f, const float8_t a)
128 {
129  return make_float8_t(f) - a;
130 }
131 
132 ccl_device_inline float8_t operator*(const float8_t a, const float8_t b)
133 {
134 #ifdef __KERNEL_AVX2__
135  return float8_t(_mm256_mul_ps(a.m256, b.m256));
136 #else
137  return make_float8_t(
138  a.a * b.a, a.b * b.b, a.c * b.c, a.d * b.d, a.e * b.e, a.f * b.f, a.g * b.g, a.h * b.h);
139 #endif
140 }
141 
142 ccl_device_inline float8_t operator*(const float8_t a, const float f)
143 {
144  return a * make_float8_t(f);
145 }
146 
147 ccl_device_inline float8_t operator*(const float f, const float8_t a)
148 {
149  return make_float8_t(f) * a;
150 }
151 
152 ccl_device_inline float8_t operator/(const float8_t a, const float8_t b)
153 {
154 #ifdef __KERNEL_AVX2__
155  return float8_t(_mm256_div_ps(a.m256, b.m256));
156 #else
157  return make_float8_t(
158  a.a / b.a, a.b / b.b, a.c / b.c, a.d / b.d, a.e / b.e, a.f / b.f, a.g / b.g, a.h / b.h);
159 #endif
160 }
161 
162 ccl_device_inline float8_t operator/(const float8_t a, const float f)
163 {
164  return a / make_float8_t(f);
165 }
166 
167 ccl_device_inline float8_t operator/(const float f, const float8_t a)
168 {
169  return make_float8_t(f) / a;
170 }
171 
172 ccl_device_inline float8_t operator+=(float8_t a, const float8_t b)
173 {
174  return a = a + b;
175 }
176 
177 ccl_device_inline float8_t operator-=(float8_t a, const float8_t b)
178 {
179  return a = a - b;
180 }
181 
182 ccl_device_inline float8_t operator*=(float8_t a, const float8_t b)
183 {
184  return a = a * b;
185 }
186 
187 ccl_device_inline float8_t operator*=(float8_t a, float f)
188 {
189  return a = a * f;
190 }
191 
192 ccl_device_inline float8_t operator/=(float8_t a, float f)
193 {
194  return a = a / f;
195 }
196 
197 ccl_device_inline bool operator==(const float8_t a, const float8_t b)
198 {
199 #ifdef __KERNEL_AVX2__
200  return (_mm256_movemask_ps(_mm256_castsi256_ps(
201  _mm256_cmpeq_epi32(_mm256_castps_si256(a.m256), _mm256_castps_si256(b.m256)))) &
202  0b11111111) == 0b11111111;
203 #else
204  return (a.a == b.a && a.b == b.b && a.c == b.c && a.d == b.d && a.e == b.e && a.f == b.f &&
205  a.g == b.g && a.h == b.h);
206 #endif
207 }
208 
209 ccl_device_inline float8_t rcp(const float8_t a)
210 {
211 #ifdef __KERNEL_AVX2__
212  return float8_t(_mm256_rcp_ps(a.m256));
213 #else
214  return make_float8_t(1.0f / a.a,
215  1.0f / a.b,
216  1.0f / a.c,
217  1.0f / a.d,
218  1.0f / a.e,
219  1.0f / a.f,
220  1.0f / a.g,
221  1.0f / a.h);
222 #endif
223 }
224 
225 ccl_device_inline float8_t sqrt(const float8_t a)
226 {
227 #ifdef __KERNEL_AVX2__
228  return float8_t(_mm256_sqrt_ps(a.m256));
229 #else
230  return make_float8_t(sqrtf(a.a),
231  sqrtf(a.b),
232  sqrtf(a.c),
233  sqrtf(a.d),
234  sqrtf(a.e),
235  sqrtf(a.f),
236  sqrtf(a.g),
237  sqrtf(a.h));
238 #endif
239 }
240 
241 ccl_device_inline float8_t sqr(const float8_t a)
242 {
243  return a * a;
244 }
245 
246 ccl_device_inline bool is_zero(const float8_t a)
247 {
248  return a == make_float8_t(0.0f);
249 }
250 
251 ccl_device_inline float average(const float8_t a)
252 {
253  return reduce_add(a) / 8.0f;
254 }
255 
256 ccl_device_inline float8_t min(const float8_t a, const float8_t b)
257 {
258 #ifdef __KERNEL_AVX2__
259  return float8_t(_mm256_min_ps(a.m256, b.m256));
260 #else
261  return make_float8_t(min(a.a, b.a),
262  min(a.b, b.b),
263  min(a.c, b.c),
264  min(a.d, b.d),
265  min(a.e, b.e),
266  min(a.f, b.f),
267  min(a.g, b.g),
268  min(a.h, b.h));
269 #endif
270 }
271 
272 ccl_device_inline float8_t max(const float8_t a, const float8_t b)
273 {
274 #ifdef __KERNEL_AVX2__
275  return float8_t(_mm256_max_ps(a.m256, b.m256));
276 #else
277  return make_float8_t(max(a.a, b.a),
278  max(a.b, b.b),
279  max(a.c, b.c),
280  max(a.d, b.d),
281  max(a.e, b.e),
282  max(a.f, b.f),
283  max(a.g, b.g),
284  max(a.h, b.h));
285 #endif
286 }
287 
288 ccl_device_inline float8_t clamp(const float8_t a, const float8_t mn, const float8_t mx)
289 {
290  return min(max(a, mn), mx);
291 }
292 
293 ccl_device_inline float8_t fabs(const float8_t a)
294 {
295 #ifdef __KERNEL_AVX2__
296  return float8_t(_mm256_and_ps(a.m256, _mm256_castsi256_ps(_mm256_set1_epi32(0x7fffffff))));
297 #else
298  return make_float8_t(fabsf(a.a),
299  fabsf(a.b),
300  fabsf(a.c),
301  fabsf(a.d),
302  fabsf(a.e),
303  fabsf(a.f),
304  fabsf(a.g),
305  fabsf(a.h));
306 #endif
307 }
308 
309 ccl_device_inline float8_t mix(const float8_t a, const float8_t b, float t)
310 {
311  return a + t * (b - a);
312 }
313 
314 ccl_device_inline float8_t saturate(const float8_t a)
315 {
316  return clamp(a, make_float8_t(0.0f), make_float8_t(1.0f));
317 }
318 
319 ccl_device_inline float8_t exp(float8_t v)
320 {
321  return make_float8_t(
322  expf(v.a), expf(v.b), expf(v.c), expf(v.d), expf(v.e), expf(v.f), expf(v.g), expf(v.h));
323 }
324 
325 ccl_device_inline float8_t log(float8_t v)
326 {
327  return make_float8_t(
328  logf(v.a), logf(v.b), logf(v.c), logf(v.d), logf(v.e), logf(v.f), logf(v.g), logf(v.h));
329 }
330 
331 ccl_device_inline float dot(const float8_t a, const float8_t b)
332 {
333 #ifdef __KERNEL_AVX2__
334  float8_t t(_mm256_dp_ps(a.m256, b.m256, 0xFF));
335  return t[0] + t[4];
336 #else
337  return (a.a * b.a) + (a.b * b.b) + (a.c * b.c) + (a.d * b.d) + (a.e * b.e) + (a.f * b.f) +
338  (a.g * b.g) + (a.h * b.h);
339 #endif
340 }
341 
342 ccl_device_inline float8_t pow(float8_t v, float e)
343 {
344  return make_float8_t(powf(v.a, e),
345  powf(v.b, e),
346  powf(v.c, e),
347  powf(v.d, e),
348  powf(v.e, e),
349  powf(v.f, e),
350  powf(v.g, e),
351  powf(v.h, e));
352 }
353 
354 ccl_device_inline float reduce_min(const float8_t a)
355 {
356  return min(min(min(a.a, a.b), min(a.c, a.d)), min(min(a.e, a.f), min(a.g, a.h)));
357 }
358 
359 ccl_device_inline float reduce_max(const float8_t a)
360 {
361  return max(max(max(a.a, a.b), max(a.c, a.d)), max(max(a.e, a.f), max(a.g, a.h)));
362 }
363 
364 ccl_device_inline float reduce_add(const float8_t a)
365 {
366 #ifdef __KERNEL_AVX2__
367  float8_t b(_mm256_hadd_ps(a.m256, a.m256));
368  float8_t h(_mm256_hadd_ps(b.m256, b.m256));
369  return h[0] + h[4];
370 #else
371  return a.a + a.b + a.c + a.d + a.e + a.f + a.g + a.h;
372 #endif
373 }
374 
375 ccl_device_inline bool isequal(const float8_t a, const float8_t b)
376 {
377  return a == b;
378 }
379 
380 ccl_device_inline float8_t safe_divide(const float8_t a, const float b)
381 {
382  return (b != 0.0f) ? a / b : make_float8_t(0.0f);
383 }
384 
385 ccl_device_inline float8_t safe_divide(const float8_t a, const float8_t b)
386 {
387  return make_float8_t((b.a != 0.0f) ? a.a / b.a : 0.0f,
388  (b.b != 0.0f) ? a.b / b.b : 0.0f,
389  (b.c != 0.0f) ? a.c / b.c : 0.0f,
390  (b.d != 0.0f) ? a.d / b.d : 0.0f,
391  (b.e != 0.0f) ? a.e / b.e : 0.0f,
392  (b.f != 0.0f) ? a.f / b.f : 0.0f,
393  (b.g != 0.0f) ? a.g / b.g : 0.0f,
394  (b.h != 0.0f) ? a.h / b.h : 0.0f);
395 }
396 
398 {
399  v.a = ensure_finite(v.a);
400  v.b = ensure_finite(v.b);
401  v.c = ensure_finite(v.c);
402  v.d = ensure_finite(v.d);
403  v.e = ensure_finite(v.e);
404  v.f = ensure_finite(v.f);
405  v.g = ensure_finite(v.g);
406  v.h = ensure_finite(v.h);
407 
408  return v;
409 }
410 
412 {
413  return isfinite_safe(v.a) && isfinite_safe(v.b) && isfinite_safe(v.c) && isfinite_safe(v.d) &&
415 }
416 
418 
419 #endif /* __UTIL_MATH_FLOAT8_H__ */
_GL_VOID GLfloat value _GL_VOID_RET _GL_VOID const GLuint GLboolean *residences _GL_BOOL_RET _GL_VOID GLsizei GLfloat GLfloat GLfloat GLfloat const GLubyte *bitmap _GL_VOID_RET _GL_VOID GLenum const void *lists _GL_VOID_RET _GL_VOID const GLdouble *equation _GL_VOID_RET _GL_VOID GLdouble GLdouble blue _GL_VOID_RET _GL_VOID GLfloat GLfloat blue _GL_VOID_RET _GL_VOID GLint GLint blue _GL_VOID_RET _GL_VOID GLshort GLshort blue _GL_VOID_RET _GL_VOID GLubyte GLubyte blue _GL_VOID_RET _GL_VOID GLuint GLuint blue _GL_VOID_RET _GL_VOID GLushort GLushort blue _GL_VOID_RET _GL_VOID GLbyte GLbyte GLbyte alpha _GL_VOID_RET _GL_VOID GLdouble GLdouble GLdouble alpha _GL_VOID_RET _GL_VOID GLfloat GLfloat GLfloat alpha _GL_VOID_RET _GL_VOID GLint GLint GLint alpha _GL_VOID_RET _GL_VOID GLshort GLshort GLshort alpha _GL_VOID_RET _GL_VOID GLubyte GLubyte GLubyte alpha _GL_VOID_RET _GL_VOID GLuint GLuint GLuint alpha _GL_VOID_RET _GL_VOID GLushort GLushort GLushort alpha _GL_VOID_RET _GL_VOID GLenum mode _GL_VOID_RET _GL_VOID GLint GLsizei GLsizei GLenum type _GL_VOID_RET _GL_VOID GLsizei GLenum GLenum const void *pixels _GL_VOID_RET _GL_VOID const void *pointer _GL_VOID_RET _GL_VOID GLdouble v _GL_VOID_RET _GL_VOID GLfloat v _GL_VOID_RET _GL_VOID GLint GLint i2 _GL_VOID_RET _GL_VOID GLint j _GL_VOID_RET _GL_VOID GLfloat param _GL_VOID_RET _GL_VOID GLint param _GL_VOID_RET _GL_VOID GLdouble GLdouble GLdouble GLdouble GLdouble zFar _GL_VOID_RET _GL_UINT GLdouble *equation _GL_VOID_RET _GL_VOID GLenum GLint *params _GL_VOID_RET _GL_VOID GLenum GLfloat *v _GL_VOID_RET _GL_VOID GLenum GLfloat *params _GL_VOID_RET _GL_VOID GLfloat *values _GL_VOID_RET _GL_VOID GLushort *values _GL_VOID_RET _GL_VOID GLenum GLfloat *params _GL_VOID_RET _GL_VOID GLenum GLdouble *params _GL_VOID_RET _GL_VOID GLenum GLint *params _GL_VOID_RET _GL_VOID GLsizei const void *pointer _GL_VOID_RET _GL_VOID GLsizei const void *pointer _GL_VOID_RET _GL_BOOL GLfloat param _GL_VOID_RET _GL_VOID GLint param _GL_VOID_RET _GL_VOID GLenum GLfloat param _GL_VOID_RET _GL_VOID GLenum GLint param _GL_VOID_RET _GL_VOID GLushort pattern _GL_VOID_RET _GL_VOID GLdouble GLdouble GLint GLint const GLdouble *points _GL_VOID_RET _GL_VOID GLdouble GLdouble GLint GLint GLdouble GLdouble GLint GLint const GLdouble *points _GL_VOID_RET _GL_VOID GLdouble GLdouble u2 _GL_VOID_RET _GL_VOID GLdouble GLdouble GLint GLdouble GLdouble v2 _GL_VOID_RET _GL_VOID GLenum GLfloat param _GL_VOID_RET _GL_VOID GLenum GLint param _GL_VOID_RET _GL_VOID GLenum mode _GL_VOID_RET _GL_VOID GLdouble GLdouble nz _GL_VOID_RET _GL_VOID GLfloat GLfloat nz _GL_VOID_RET _GL_VOID GLint GLint nz _GL_VOID_RET _GL_VOID GLshort GLshort nz _GL_VOID_RET _GL_VOID GLsizei const void *pointer _GL_VOID_RET _GL_VOID GLsizei const GLfloat *values _GL_VOID_RET _GL_VOID GLsizei const GLushort *values _GL_VOID_RET _GL_VOID GLint param _GL_VOID_RET _GL_VOID const GLuint const GLclampf *priorities _GL_VOID_RET _GL_VOID GLdouble y _GL_VOID_RET _GL_VOID GLfloat y _GL_VOID_RET _GL_VOID GLint y _GL_VOID_RET _GL_VOID GLshort y _GL_VOID_RET _GL_VOID GLdouble GLdouble z _GL_VOID_RET _GL_VOID GLfloat GLfloat z _GL_VOID_RET _GL_VOID GLint GLint z _GL_VOID_RET _GL_VOID GLshort GLshort z _GL_VOID_RET _GL_VOID GLdouble GLdouble GLdouble w _GL_VOID_RET _GL_VOID GLfloat GLfloat GLfloat w _GL_VOID_RET _GL_VOID GLint GLint GLint w _GL_VOID_RET _GL_VOID GLshort GLshort GLshort w _GL_VOID_RET _GL_VOID GLdouble GLdouble GLdouble y2 _GL_VOID_RET _GL_VOID GLfloat GLfloat GLfloat y2 _GL_VOID_RET _GL_VOID GLint GLint GLint y2 _GL_VOID_RET _GL_VOID GLshort GLshort GLshort y2 _GL_VOID_RET _GL_VOID GLdouble GLdouble GLdouble z _GL_VOID_RET _GL_VOID GLdouble GLdouble z _GL_VOID_RET _GL_VOID GLuint *buffer _GL_VOID_RET _GL_VOID GLdouble t _GL_VOID_RET _GL_VOID GLfloat t _GL_VOID_RET _GL_VOID GLint t _GL_VOID_RET _GL_VOID GLshort t _GL_VOID_RET _GL_VOID GLdouble t
ATTR_WARN_UNUSED_RESULT const BMVert const BMEdge * e
ATTR_WARN_UNUSED_RESULT const BMVert * v
#define logf(x)
Definition: cuda/compat.h:105
#define expf(x)
Definition: cuda/compat.h:106
#define ccl_device_inline
Definition: cuda/compat.h:34
#define powf(x, y)
Definition: cuda/compat.h:103
#define CCL_NAMESPACE_END
Definition: cuda/compat.h:9
ccl_device_inline float4 mask(const int4 &mask, const float4 &a)
Definition: math_float4.h:513
ccl_device_inline float reduce_add(const float8_t a)
Definition: math_float8.h:364
ccl_device_inline float8_t safe_divide(const float8_t a, const float b)
Definition: math_float8.h:380
ccl_device_inline float8_t ensure_finite(float8_t v)
Definition: math_float8.h:397
ccl_device_inline float8_t one_float8_t()
Definition: math_float8.h:77
ccl_device_inline float8_t operator+=(float8_t a, const float8_t b)
Definition: math_float8.h:172
ccl_device_inline float8_t saturate(const float8_t a)
Definition: math_float8.h:314
ccl_device_inline float8_t operator*=(float8_t a, const float8_t b)
Definition: math_float8.h:182
ccl_device_inline float8_t operator/(const float8_t a, const float8_t b)
Definition: math_float8.h:152
ccl_device_inline float8_t log(float8_t v)
Definition: math_float8.h:325
ccl_device_inline float8_t operator*(const float8_t a, const float8_t b)
Definition: math_float8.h:132
ccl_device_inline bool is_zero(const float8_t a)
Definition: math_float8.h:246
ccl_device_inline bool isfinite_safe(float8_t v)
Definition: math_float8.h:411
ccl_device_inline float8_t sqr(const float8_t a)
Definition: math_float8.h:241
ccl_device_inline float8_t max(const float8_t a, const float8_t b)
Definition: math_float8.h:272
ccl_device_inline float average(const float8_t a)
Definition: math_float8.h:251
ccl_device_inline bool isequal(const float8_t a, const float8_t b)
Definition: math_float8.h:375
ccl_device_inline float8_t clamp(const float8_t a, const float8_t mn, const float8_t mx)
Definition: math_float8.h:288
ccl_device_inline float8_t operator-=(float8_t a, const float8_t b)
Definition: math_float8.h:177
ccl_device_inline float reduce_max(const float8_t a)
Definition: math_float8.h:359
ccl_device_inline bool operator==(const float8_t a, const float8_t b)
Definition: math_float8.h:197
ccl_device_inline float8_t fabs(const float8_t a)
Definition: math_float8.h:293
ccl_device_inline float reduce_min(const float8_t a)
Definition: math_float8.h:354
ccl_device_inline float8_t pow(float8_t v, float e)
Definition: math_float8.h:342
ccl_device_inline float8_t mix(const float8_t a, const float8_t b, float t)
Definition: math_float8.h:309
ccl_device_inline float dot(const float8_t a, const float8_t b)
Definition: math_float8.h:331
ccl_device_inline float8_t min(const float8_t a, const float8_t b)
Definition: math_float8.h:256
ccl_device_inline float8_t operator/=(float8_t a, float f)
Definition: math_float8.h:192
ccl_device_inline float8_t zero_float8_t()
Definition: math_float8.h:68
ccl_device_inline float8_t rcp(const float8_t a)
Definition: math_float8.h:209
ccl_device_inline float8_t sqrt(const float8_t a)
Definition: math_float8.h:225
ccl_device_inline float8_t exp(float8_t v)
Definition: math_float8.h:319
CCL_NAMESPACE_BEGIN ccl_device_inline float8_t operator+(const float8_t a, const float8_t b)
Definition: math_float8.h:82
ccl_device_inline float8_t operator-(const float8_t a)
Definition: math_float8.h:102
#define fabsf(x)
Definition: metal/compat.h:219
#define sqrtf(x)
Definition: metal/compat.h:243
static unsigned a[3]
Definition: RandGen.cpp:78
static const pxr::TfToken b("b", pxr::TfToken::Immortal)
struct BMEdge * e
Definition: bmesh_class.h:97
ccl_device_inline float8_t make_float8_t(float f)