2 Bullet Continuous Collision Detection and Physics Library, Copyright (c) 2007 Erwin Coumans
4 This software is provided 'as-is', without any express or implied warranty.
5 In no event will the authors be held liable for any damages arising from the use of this software.
6 Permission is granted to anyone to use this software for any purpose,
7 including commercial applications, and to alter it and redistribute it freely,
8 subject to the following restrictions:
10 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.
11 2. Altered source versions must be plainly marked as such, and must not be misrepresented as being the original software.
12 3. This notice may not be removed or altered from any source distribution.
18 #include "LinearMath/btScalar.h"
20 #include "MiniCL/cl.h"
26 #define get_global_id(a) __guid_arg
27 #define get_local_id(a) ((__guid_arg) % gMiniCLNumOutstandingTasks)
28 #define get_local_size(a) (gMiniCLNumOutstandingTasks)
29 #define get_group_id(a) ((__guid_arg) / gMiniCLNumOutstandingTasks)
31 static unsigned int as_uint(float val) { return *((unsigned int*)&val); }
34 #define CLK_LOCAL_MEM_FENCE 0x01
35 #define CLK_GLOBAL_MEM_FENCE 0x02
37 static void barrier(unsigned int a)
42 //ATTRIBUTE_ALIGNED16(struct) float8
56 s0=s1=s2=s3=s4=s5=s6=s7=scalar;
61 float select( float arg0, float arg1, bool select)
75 float3& operator+=(const float3& other)
83 float3& operator-=(const float3& other)
93 static float dot(const float3&a ,const float3& b)
99 return tmp.x+tmp.y+tmp.z;
102 static float3 operator-(const float3& a,const float3& b)
111 static float3 operator*(const float& scalar,const float3& b)
114 tmp.x = scalar * b.x;
115 tmp.y = scalar * b.y;
116 tmp.z = scalar * b.z;
120 static float3 operator*(const float3& a,const float& scalar)
123 tmp.x = a.x * scalar;
124 tmp.y = a.y * scalar;
125 tmp.z = a.z * scalar;
130 static float3 operator*(const float3& a,const float3& b)
140 //ATTRIBUTE_ALIGNED16(struct) float4
156 float4(float v0, float v1, float v2, float v3)
164 float4(float3 xyz, float scalarW)
176 float4 operator*(const float4& other)
188 float4 operator*(const float& other)
200 float4& operator+=(const float4& other)
209 float4& operator-=(const float4& other)
218 float4& operator *=(float scalar)
233 static float4 fabs(const float4& a)
236 tmp.x = a.x < 0.f ? 0.f : a.x;
237 tmp.y = a.y < 0.f ? 0.f : a.y;
238 tmp.z = a.z < 0.f ? 0.f : a.z;
239 tmp.w = a.w < 0.f ? 0.f : a.w;
242 static float4 operator+(const float4& a,const float4& b)
253 static float8 operator+(const float8& a,const float8& b)
256 tmp.s0 = a.s0 + b.s0;
257 tmp.s1 = a.s1 + b.s1;
258 tmp.s2 = a.s2 + b.s2;
259 tmp.s3 = a.s3 + b.s3;
260 tmp.s4 = a.s4 + b.s4;
261 tmp.s5 = a.s5 + b.s5;
262 tmp.s6 = a.s6 + b.s6;
263 tmp.s7 = a.s7 + b.s7;
268 static float4 operator-(const float4& a,const float4& b)
278 static float8 operator-(const float8& a,const float8& b)
281 tmp.s0 = a.s0 - b.s0;
282 tmp.s1 = a.s1 - b.s1;
283 tmp.s2 = a.s2 - b.s2;
284 tmp.s3 = a.s3 - b.s3;
285 tmp.s4 = a.s4 - b.s4;
286 tmp.s5 = a.s5 - b.s5;
287 tmp.s6 = a.s6 - b.s6;
288 tmp.s7 = a.s7 - b.s7;
292 static float4 operator*(float a,const float4& b)
302 static float4 operator/(const float4& b,float a)
316 static float dot(const float4&a ,const float4& b)
323 return tmp.x+tmp.y+tmp.z+tmp.w;
326 static float length(const float4&a)
328 float l = sqrtf(a.x*a.x+a.y*a.y+a.z*a.z);
332 static float4 normalize(const float4&a)
342 static float4 cross(const float4&a ,const float4& b)
345 tmp.x = a.y*b.z - a.z*b.y;
346 tmp.y = -a.x*b.z + a.z*b.x;
347 tmp.z = a.x*b.y - a.y*b.x;
352 static float max(float a, float b)
354 return (a >= b) ? a : b;
358 static float min(float a, float b)
360 return (a <= b) ? a : b;
363 static float fmax(float a, float b)
365 return (a >= b) ? a : b;
368 static float fmin(float a, float b)
370 return (a <= b) ? a : b;
383 //typedef int2 uint2;
385 typedef unsigned int uint;
394 unsigned int x,y,z,w;
396 uint4(uint val) { x = y = z = w = val; }
397 uint4& operator+=(const uint4& other)
406 static uint4 operator+(const uint4& a,const uint4& b)
415 static uint4 operator-(const uint4& a,const uint4& b)
425 #define native_sqrt sqrtf
426 #define native_sin sinf
427 #define native_cos cosf
428 #define native_powr powf
430 #define GUID_ARG ,int __guid_arg
431 #define GUID_ARG_VAL ,__guid_arg
434 #define as_int(a) (*((int*)&(a)))
436 extern "C" int gMiniCLNumOutstandingTasks;
437 // extern "C" void __kernel_func();