#include "Tools.cuh"
#include <math_functions.h>

// operator of Vector3D
__device__ __host__ inline 
Vector3D operator+(const Vector3D& param1, const Vector3D& param2){
	Vector3D result;
	result.x = param1.x + param2.x;
	result.y = param1.y + param2.y;
	result.z = param1.z + param2.z;
	return result;
}

__device__ __host__ inline 
Vector3D operator-(const Vector3D& param1, const Vector3D& param2){
	Vector3D result;
	result.x = param1.x - param2.x;
	result.y = param1.y - param2.y;
	result.z = param1.z - param2.z;
	return result;
}

__device__ __host__ inline
Vector3D operator*(const float param1,const Vector3D& param2){
	Vector3D result;
	result.x = param1 * param2.x;
	result.y = param1 * param2.y;
	result.z = param1 * param2.z;
	return result;
}

__device__ __host__ inline 
Vector3D operator*(const Vector3D& param1,const float param2){
	Vector3D result;
	result.x = param1.x * param2;
	result.y = param1.y * param2;
	result.z = param1.z * param2;
	return result;
}
__device__ __host__ inline 
Vector3D operator/(const Vector3D& param1,const float param2){
	Vector3D result;
	result.x = param1.x / param2;
	result.y = param1.y / param2;
	result.z = param1.z / param2;
	return result;
}

//__global__ Vector3D& operator=(const Vector3D &param);
__device__ __host__ inline 
float Exponentiation(const Vector3D &param){
	float result;
	result = param.x * param.x + param.y * param.y + param.z * param.z;
	return result;
}

__device__ __host__	inline
float Norm(const Vector3D &param){
		float result;
		result = sqrt( Exponentiation( param ));
		return result;
}

__device__ __host__	inline 
float operator*(const Vector3D &param1,const Vector3D &param2){
	float result;
	result = param1.x * param2.x + param1.y * param2.y + param1.z * param2.z;
	return result;
}

__device__ __host__	inline 
Vector3D CrossProduct(const Vector3D &param1, const Vector3D &param2){
	Vector3D result;
	result.x = param1.y * param2.z - param1.z * param2.y;
	result.y = param1.z * param2.x - param1.x * param2.z;
	result.z = param1.x * param2.y - param1.y * param2.x;	
	return result;
}

__device__ __host__	inline 
Vector3D Normalize(const Vector3D &param){
	Vector3D result;
	float norm = Norm(param);
	result.x = param.x / norm;
	result.y = param.y / norm;
	result.z = param.z / norm;
	return result;
}


__device__ __host__ inline 
float Distance(const Point3D &param1,const Point3D &param2){
	return ( sqrtf( (param1.x - param2.x ) * ( param1.x - param2.x ) 
		+ ( param1.y - param2.y ) * ( param1.y - param2.y )
		+ ( param1.z - param2.z ) * ( param1.z - param2.z ) ));
}
/*
// operator of Point
__device__ __host__ inline 
Point3D operator+(const Point3D &param1, const Vector3D& param2){
	Point3D result;
	result.x = param1.x + param2.x;
	result.y = param1.y + param2.y;
	result.z = param1.z + param2.z;
	return result;
}
__device__ __host__ inline 
Point3D operator-(const Point3D &param1, const Vector3D& param2){
	Point3D result;
	result.x = param1.x - param2.x;
	result.y = param1.y - param2.y;
	result.z = param1.z - param2.z;
	return result;
}
__device__ __host__ inline 
Vector3D operator-(const Point3D& param1, const Point3D& param2){
	Vector3D result;
	result.x = param1.x - param2.x;
	result.y = param1.y - param2.y;
	result.z = param1.z - param2.z;
	return result;
}
*/
/*
__device__ __host__ inline 
RGBColor operator+(const RGBColor& param1, const RGBColor &param2){
	RGBColor result;
	result.x = param1.x + param2.x;
	result.y = param1.y + param2.y;
	result.z = param1.z + param2.z;
	return result;
}

__device__ __host__ inline 
RGBColor operator-(const RGBColor& param1, const RGBColor &param2){
	RGBColor result;
	result.x = param1.x - param2.x;
	result.y = param1.y - param2.y;
	int tempZ = param1.z - param2.z;
	
	return result;
}
*//*
__device__ __host__ inline 
RGBColor operator*(const RGBColor& param1, const float &param2){
	RGBColor result;
	result.x = param1.x * param2;
	result.y = param1.y * param2;
	result.z = param1.z * param2;

	return result;
}
*/
__device__ __host__ inline 
RGBColor powc(const RGBColor& param1, const RGBColor &param2){
	RGBColor result;
	result.x = param1.x * param2.x;
	result.y = param1.y * param2.y;
	result.z = param1.z * param2.z;
	return result;
}
/*
__device__ __host__ inline 
RGBColor operator/(const RGBColor& param1, const float &param2){
	RGBColor result;
	result.x = param1.x / param2;
	result.y = param1.y / param2;
	result.z = param1.z / param2;

	return result;
}
*/
__device__ __host__ inline 
uchar3 MapToUchar(const RGBColor& param){
	uchar3 result;
	result.x = param.x * 255;
	result.y = param.y * 255;
	result.z = param.z * 255;

	if( param.x > 1.0 ){
		result.x = 255;
	}else if( param.x < 0.0 ){
		result.x = 0;
	}
	if( param.y > 1.0 ){
		result.y = 255;
	}else if( param.x < 0.0 ){
		result.y = 0;
	}
	if( param.z > 1.0 ){
		result.z = 255;
	}else if( param.x < 0.0 ){
		result.z = 0;
	}
	return result;
}
/* pseudo random number generator */
/*  linear congruential  */
__device__
static unsigned int prng_seed = 0xcdcdcdcd;

__device__ 
void rand_seed(unsigned int seed){
	prng_seed = seed;
}

__device__ inline
unsigned int uintRand(){
	prng_seed = ( prng_seed * 0xababa ) % ( MAX_RANDOM_NUM );
	return prng_seed;
}
__device__ inline
float floatRand(){
	prng_seed = ( prng_seed * 0xababa ) % ( MAX_RANDOM_NUM );

	float result = float(prng_seed) / float(MAX_RANDOM_NUM);

	return result;
}
	.file	"tmpxft_002fc142_00000000-6_Tools.cudafe1.cpp"
	.text
#APP

#NO_APP
	.type	_ZL26__cudaUnregisterBinaryUtilv, @function
_ZL26__cudaUnregisterBinaryUtilv:
.LFB2026:
	.cfi_startproc
	endbr64
	movq	_ZL20__cudaFatCubinHandle(%rip), %rdi
	jmp	__cudaUnregisterFatBinary@PLT
	.cfi_endproc
.LFE2026:
	.size	_ZL26__cudaUnregisterBinaryUtilv, .-_ZL26__cudaUnregisterBinaryUtilv
	.globl	_Z9rand_seedj
	.type	_Z9rand_seedj, @function
_Z9rand_seedj:
.LFB2021:
	.cfi_startproc
	endbr64
	pushq	%rax
	.cfi_def_cfa_offset 16
	popq	%rax
	.cfi_def_cfa_offset 8
	subq	$24, %rsp
	.cfi_def_cfa_offset 32
	movl	$1, 12(%rsp)
	movl	12(%rsp), %edi
	call	exit@PLT
	.cfi_endproc
.LFE2021:
	.size	_Z9rand_seedj, .-_Z9rand_seedj
	.section	.rodata.str1.1,"aMS",@progbits,1
.LC0:
	.string	"prng_seed"
	.section	.text.startup,"ax",@progbits
	.type	_ZL24__sti____cudaRegisterAllv, @function
_ZL24__sti____cudaRegisterAllv:
.LFB2049:
	.cfi_startproc
	endbr64
	pushq	%rax
	.cfi_def_cfa_offset 16
	leaq	_ZL15__fatDeviceText(%rip), %rdi
	call	__cudaRegisterFatBinary@PLT
	pushq	$0
	.cfi_def_cfa_offset 24
	movl	$4, %r9d
	xorl	%r8d, %r8d
	pushq	$0
	.cfi_def_cfa_offset 32
	leaq	.LC0(%rip), %rdx
	movq	%rax, %rdi
	leaq	_ZL9prng_seed(%rip), %rsi
	movq	%rdx, %rcx
	movq	%rax, _ZL20__cudaFatCubinHandle(%rip)
	call	__cudaRegisterVar@PLT
	movq	_ZL20__cudaFatCubinHandle(%rip), %rdi
	popq	%rdx
	.cfi_def_cfa_offset 24
	popq	%rcx
	.cfi_def_cfa_offset 16
	call	__cudaRegisterFatBinaryEnd@PLT
	leaq	_ZL26__cudaUnregisterBinaryUtilv(%rip), %rdi
	popq	%rsi
	.cfi_def_cfa_offset 8
	jmp	atexit@PLT
	.cfi_endproc
.LFE2049:
	.size	_ZL24__sti____cudaRegisterAllv, .-_ZL24__sti____cudaRegisterAllv
	.section	.init_array,"aw"
	.align 8
	.quad	_ZL24__sti____cudaRegisterAllv
	.section	.nvFatBinSegment,"aw"
	.align 8
	.type	_ZL15__fatDeviceText, @object
	.size	_ZL15__fatDeviceText, 24
_ZL15__fatDeviceText:
	.long	1180844977
	.long	1
	.quad	fatbinData
	.quad	0
	.local	_ZL20__cudaFatCubinHandle
	.comm	_ZL20__cudaFatCubinHandle,8,8
	.local	_ZL9prng_seed
	.comm	_ZL9prng_seed,4,4
	.ident	"GCC: (Ubuntu 13.3.0-6ubuntu2~24.04) 13.3.0"
	.section	.note.GNU-stack,"",@progbits
	.section	.note.gnu.property,"a"
	.align 8
	.long	1f - 0f
	.long	4f - 1f
	.long	5
0:
	.string	"GNU"
1:
	.align 8
	.long	0xc0000002
	.long	3f - 2f
2:
	.long	0x3
3:
	.align 8
4:


#include <hip/hip_runtime.h>
#if !defined __TOOLS_CUH__
#define __TOOLS__CUH__

typedef unsigned char Byte;

#define hugeValue  3.402823466e+38F
#define PI         3.14f
#define invPI      0.31831f

typedef float2		Point2D;
typedef float3		Point3D;
typedef float3		Vector3D;
typedef float3		Normal;
typedef float3		RGBColor;

#define Vector3D(x,y,z)		make_float3(x,y,z)
#define Point3D(x,y,z)		make_float3(x,y,z)
#define Normal(x,y,z)		make_float3(x,y,z)
#define RGBColor(r,g,b)		make_float3(r,g,b)
#define Point2D(x,y)		make_float2(x,y)

#define white	make_float3(1.0,1.0,1.0)
#define black	make_float3(0,0,0)
#define red		make_float3(1.0,0,0)
#define green   make_float3(0,1.0,0)
#define blue	make_float3(0,0,1.0)
#define yellow	make_float3(1.0,1.0,0)
// operator of Vector3D

extern __device__ __host__ inline Vector3D operator+(const Vector3D& param1, const Vector3D& param2);
extern __device__ __host__ inline Vector3D operator-(const Vector3D& param1, const Vector3D& param2);
extern __device__ __host__ inline Vector3D operator*(const float param1,const Vector3D& param2);
extern __device__ __host__ inline Vector3D operator*(const Vector3D& param1,const float param2);
extern __device__ __host__ inline Vector3D operator/(const Vector3D& param1,const float param2);
//__global__ Vector3D& operator=(const Vector3D &param);
extern __device__ __host__ inline float Exponentiation(const Vector3D &param);
extern __device__ __host__ inline float Norm(const Vector3D &param);
extern __device__ __host__ inline float operator*(const Vector3D &param1,const Vector3D &param2);
extern __device__ __host__ inline Vector3D CrossProduct(const Vector3D &param1, const Vector3D &param2);
extern __device__ __host__ inline Vector3D Normalize(const Vector3D &param);

extern __device__ __host__ inline float Distance(const Point3D &param1,const Point3D &param2);
// operator of Point3D
//__device__ __host__ inline Point3D operator+(const Point3D &param1, const Vector3D& param2);
//__device__ __host__ inline Point3D operator-(const Point3D &param1, const Vector3D& param2);
//__device__ __host__ inline Vector3D operator-(const Point3D& param1, const Point3D& param2);
//operator of RGBColor
//extern __device__ __host__ inline RGBColor operator+(const RGBColor& param1, const RGBColor &param2);
//extern __device__ __host__ inline RGBColor operator-(const RGBColor& param1, const RGBColor &param2);
//extern __device__ __host__ inline RGBColor operator*(const RGBColor& param1, const float &param2);
extern __device__ __host__ inline RGBColor powc(const RGBColor& param1, const RGBColor &param2);
//extern __device__ __host__ inline RGBColor operator/(const RGBColor& param1, const float &param2);

extern __device__ __host__ inline uchar3 MapToUchar(const RGBColor& param);
/***** some utility function *****/

/* pseudo random number generator */
/*  linear congruential  */
#define MAX_RANDOM_NUM	0x7fffffff

/*
__device__ __host__
void rand_seed(unsigned int seed);
*/

extern __device__ __host__ inline
unsigned int uintRand();

extern __device__ __host__ inline
float floatRand();

#define STACK_MAX	10

#endif
	.text
	.file	"Tools.hip"
	.type	__hip_cuid_,@object             # @__hip_cuid_
	.bss
	.globl	__hip_cuid_
__hip_cuid_:
	.byte	0                               # 0x0
	.size	__hip_cuid_, 1

	.ident	"AMD clang version 18.0.0git (https://github.com/RadeonOpenCompute/llvm-project roc-6.3.2 25012 e5bf7e55c91490b07c49d8960fa7983d864936c4)"
	.section	".note.GNU-stack","",@progbits
	.addrsig
	.addrsig_sym __hip_cuid_

	.text
	.p2alignl 7, 3214868480
	.fill 96, 4, 3214868480
	.type	__hip_cuid_,@object             ; @__hip_cuid_
	.section	.bss,"aw",@nobits
	.globl	__hip_cuid_
__hip_cuid_:
	.byte	0                               ; 0x0
	.size	__hip_cuid_, 1

	.ident	"AMD clang version 18.0.0git (https://github.com/RadeonOpenCompute/llvm-project roc-6.3.2 25012 e5bf7e55c91490b07c49d8960fa7983d864936c4)"
	.section	".note.GNU-stack","",@progbits
	.addrsig
	.addrsig_sym __hip_cuid_
	.amdgcn_target "amdgcn-amd-amdhsa--gfx1100"
	.amdgpu_metadata
---
amdhsa.kernels:  []
amdhsa.target:   amdgcn-amd-amdhsa--gfx1100
amdhsa.version:
  - 1
  - 2
...

	.end_amdgpu_metadata
