// © Copyright 2008 Владимир Фролов
#define MGML_MATH_CUDA_SIMD_GUARDIAN

//#ifndef MGML_GLOBALS
//	#include "MGML_GLOBALS.h"
//#endif

#ifndef MGML_MATH_GUARDIAN
	#include "MGML_MATH.h"
#endif


#ifdef __CUDACC__

//typedef MGML_MATH::VECTOR<4,float> vec4f;
/*
inline __device__ vec4f make_vec4f(float x, float y, float z, float w, vec4f& v)
{
		v[0] = x;
		v[1] = y;
		v[2] = z;
		v[3] = w;
		return v;
}


__device__ vec4f _mm_div_ps(const vec4f& a, const vec4f& b)
{
	vec4f res = make_vec4f(a[0]/b[0], a[1]/b[1], a[2]/b[2], a[3]/b[3], res);
	return res;
}

__device__ vec4f _mm_mul_ps(const vec4f& a, const vec4f& b)
{
	vec4f res = make_vec4f(a[0]*b[0], a[1]*b[1], a[2]*b[2], a[3]*b[3], res);
	return res;
}

__device__ vec4f _mm_sub_ps(const vec4f& a, const vec4f& b)
{
	return a-b;
}

__device__ int _mm_movemask_ps(const vec4f& rhs)
{
	return (signbit(rhs[3]) << 3) | 
		   (signbit(rhs[2]) << 2) |
		   (signbit(rhs[1]) << 1) |
		   (signbit(rhs[0]) << 0);
}


__device__ vec4f _mm_cmpge_ps(const vec4f& a, const vec4f& b)
{
	vec4f res = make_vec4f( (a[0] >= b[0]) ? __int_as_float(0xffffffff) : __int_as_float(0x0),
							(a[1] >= b[1]) ? __int_as_float(0xffffffff) : __int_as_float(0x0),
							(a[2] >= b[2]) ? __int_as_float(0xffffffff) : __int_as_float(0x0),
							(a[3] >= b[3]) ? __int_as_float(0xffffffff) : __int_as_float(0x0), res);
	return res;
}

__device__ vec4f _mm_cmpgt_ps(const vec4f& a, const vec4f& b)
{
	vec4f res = make_vec4f( (a[0] > b[0]) ? __int_as_float(0xffffffff) : __int_as_float(0x0),
							(a[1] > b[1]) ? __int_as_float(0xffffffff) : __int_as_float(0x0),
							(a[2] > b[2]) ? __int_as_float(0xffffffff) : __int_as_float(0x0),
							(a[3] > b[3]) ? __int_as_float(0xffffffff) : __int_as_float(0x0), res);
	return res;
}

__device__ vec4f _mm_cmple_ps(const vec4f& a, const vec4f& b)
{
	vec4f res = make_vec4f( (a[0] <= b[0]) ? __int_as_float(0xffffffff) : __int_as_float(0x0),
							(a[1] <= b[1]) ? __int_as_float(0xffffffff) : __int_as_float(0x0),
							(a[2] <= b[2]) ? __int_as_float(0xffffffff) : __int_as_float(0x0),
							(a[3] <= b[3]) ? __int_as_float(0xffffffff) : __int_as_float(0x0), res);
	return res;
}

__device__ vec4f _mm_cmplt_ps(const vec4f& a, const vec4f& b)
{
	vec4f res = make_vec4f( (a[0] < b[0]) ? __int_as_float(0xffffffff) : __int_as_float(0x0),
							(a[1] < b[1]) ? __int_as_float(0xffffffff) : __int_as_float(0x0),
							(a[2] < b[2]) ? __int_as_float(0xffffffff) : __int_as_float(0x0),
							(a[3] < b[3]) ? __int_as_float(0xffffffff) : __int_as_float(0x0), res);
	return res;
}


__device__ vec4f _mm_or_ps(const vec4f& a, const vec4f& b)
{
	unsigned int x = __float_as_int(a[0]) | __float_as_int(b[0]);
	unsigned int y = __float_as_int(a[1]) | __float_as_int(b[1]);
	unsigned int z = __float_as_int(a[2]) | __float_as_int(b[2]);
	unsigned int w = __float_as_int(a[3]) | __float_as_int(b[3]);

	vec4f res = make_vec4f(	__int_as_float(x),
							__int_as_float(y),
							__int_as_float(z),
							__int_as_float(w), res);
	return res;
}


__device__ vec4f _mm_and_ps(const vec4f& a, const vec4f& b)
{
	unsigned int x = __float_as_int(a[0]) & __float_as_int(b[0]);
	unsigned int y = __float_as_int(a[1]) & __float_as_int(b[1]);
	unsigned int z = __float_as_int(a[2]) & __float_as_int(b[2]);
	unsigned int w = __float_as_int(a[3]) & __float_as_int(b[3]);

	vec4f res = make_vec4f(	__int_as_float(x),
							__int_as_float(y),
							__int_as_float(z),
							__int_as_float(w), res);
	return res;
}

__device__ vec4f _mm_andnot_ps(const vec4f& a, const vec4f& b)
{
	unsigned int x = ~__float_as_int(a[0]) & __float_as_int(b[0]);
	unsigned int y = ~__float_as_int(a[1]) & __float_as_int(b[1]);
	unsigned int z = ~__float_as_int(a[2]) & __float_as_int(b[2]);
	unsigned int w = ~__float_as_int(a[3]) & __float_as_int(b[3]);

	vec4f res = make_vec4f(	__int_as_float(x),
							__int_as_float(y),
							__int_as_float(z),
							__int_as_float(w), res);
	return res;
}
*/

#endif