/**
 *This is the source code for Mars, a MapReduce framework on graphics
 *processors.
 *Author: Wenbin Fang (HKUST), Bingsheng He (HKUST)
 *Mentor: Naga K. Govindaraju (Microsoft Corp.), Qiong Luo (HKUST), Tuyong
 *Wang (Sina.com).
 *If you have any question on the code, please contact us at {saven,
 *wenbin, luo}@cse.ust.hk.
 *The copyright is held by HKUST. Mars is provided "as is" without any 
 *guarantees of any kind.
 */

//Function forward declarations
#ifndef __MARSFUNC_H__
#define __MARSFUNC_H__

//--------------------------------------------------------
//MarsUtils
//--------------------------------------------------------
extern "C"
Spec_t *GetDefaultSpec();
extern "C"
void AddMapInputRecord(Spec_t*		spec, 
					   void*		key, 
					   void*		val,
					   size_t		keySize, 
					   size_t		valSize);
extern "C"
void InitMapReduce(Spec_t *spec);
extern "C"
void MapReduce(Spec_t *spec);
extern "C"
void FinishMapReduce(Spec_t *spec);

extern "C"
void ResetInput(Spec_t *spec);
extern "C"
void ResetInter(Spec_t *spec);
extern "C"
void ResetOutput(Spec_t *spec);

extern "C"
void FreeChunk(ChunkInfo_t *chunk);
extern "C"
void FlushInputToDisk(Spec_t *spec);
extern "C"
bool ReadChunkFromFile(ChunkInfo_t *chunk, FileName_t *file, 
					  size_t totalRec, size_t cursor, size_t chunkSize);
extern "C"
bool ReadGroupChunkFromFile(ChunkInfo_t *chunk, FileName_t *file, 
					  size_t totalRec, size_t cursor, size_t chunkSize);
extern "C"
bool WriteChunkToFile(ChunkInfo_t *chunk, FileName_t *file, size_t *totalRecCount, char mode);

extern "C"
RecIterator_t *InitIterator(ChunkInfo_t *chunk, FileName_t *file, size_t totalRec, Spec_t *spec);
extern "C"
void FinishIterator(RecIterator_t *it);
extern "C"
char NextRec(RecIterator_t *it, Record_t *rec);

extern "C"
GroupIterator_t *InitGroupIterator(ChunkInfo_t *chunk, FileName_t *file, size_t totalGroup, Spec_t *spec);
extern "C"
char NextGroup(GroupIterator_t *it, Group_t *rec);
extern "C"
char NextRecInGroup(Group_t *group, Record_t *rec);
extern "C"
void FinishGroupIterator(GroupIterator_t *it);

#define INT	0x00
#define CHAR 0X01
#define FLOAT 0x02
#define STRING 0x03
extern "C"
void GenRecords(Spec_t *spec, char keyType, char valType, int maxRec);
extern "C"
void PrintRecords(ChunkInfo_t *chunk, FileName_t *file, size_t totalRec, Spec_t *spec,
				 char keyType, char valType, int maxRec);
void PrintGroup(ChunkInfo_t *chunk, FileName_t *file, size_t totalGroup, Spec_t *spec,
				 char keyType, char valType, int maxGroup, int maxRec);
//--------------------------------------------------------
//MarsLib
//--------------------------------------------------------
extern "C"
void InitMapReduce(Spec_t *spec);
extern "C"
char ScheduleSingleMem(Spec_t *spec, char phase);
extern "C"
char ScheduleCoprocessMem(Spec_t *spec, char phase);
extern "C"
char ScheduleSingleFile(Spec_t *spec, char phase);
extern "C"
char ScheduleCoprocessFile(Spec_t *spec, char phase);

//--------------------------------------------------------
//MarsCPULib
//--------------------------------------------------------
extern "C"
char *cpuGetVal(void *vals, int4 *index, size_t valStartIndex,
				size_t valCount, size_t i);

extern "C"
size_t cpuGetValSize(void *vals, int4 *index, size_t valStartIndex,
				size_t valCount, size_t i);

extern "C"
void StartCPUMap(Schedule_t *sched, char mode);
extern "C"
void StartCPUSort(Schedule_t *sched, char mode);
extern "C"
void StartCPUReduce(Schedule_t *sched, char mode);

extern "C"
void cpuEmitInterCount(size_t	keySize,
					   size_t	valSize,
					   size_t*	interKeysSizePerTask,
					   size_t*	interValsSizePerTask,
					   size_t*	interCountPerTask,
					   int index);
extern "C"
void cpuEmitIntermediate(char*	key, 
				 char*		val, 
				 size_t		keySize, 
				 size_t		valSize,
				 size_t*	psKeySizes,
				 size_t*	psValSizes,
				 size_t*	psCounts,
				 int2*		keyValOffsets,
				 char*		interKeys,
				 char*		interVals,
				 int4*		interOffsetSizes,
				 size_t*	curIndex, int index);

extern "C"
void cpuEmitCount(size_t		keySize,
	  size_t		valSize,
	  size_t*		outputKeysSizePerTask,
	  size_t*		outputValsSizePerTask,
	  size_t*		outputCountPerTask,
	  int			index);

extern "C"
void cpuEmit  (char*		key, 
	   char*		val, 
	   size_t		keySize, 
	   size_t		valSize,
	   size_t*		psKeySizes, 
	   size_t*		psValSizes, 
	   size_t*		psCounts, 
	   int2*		keyValOffsets, 
	   char*		outputKeys,
	   char*		outputVals,
	   int4*		outputOffsetSizes,
	   size_t*		curIndex,
       int		index);

//--------------------------------------------------------
//MarsGPULib
//--------------------------------------------------------
extern "C"
__device__ char *gpuGetVal(void *vals, int4 *index, size_t valStartIndex,
				size_t valCount, size_t i);

extern "C"
__device__ size_t gpuGetValSize(void *vals, int4 *index, size_t valStartIndex,
				size_t valCount, size_t i);

extern "C"
void StartGPUMap(Schedule_t *sched, char mode);
extern "C"
void StartGPUSort(Schedule_t *sched, char mode);
extern "C"
void StartGPUSort_cpu(Schedule_t *sched, char mode);
extern "C"
void StartGPUReduce(Schedule_t *sched, char mode);

extern "C"
__device__ void gpuEmitInterCount(size_t		keySize,
						  size_t		valSize,
						  size_t*		outputKeysSizePerTask,
						  size_t*		outputValsSizePerTask,
						  size_t*		outputCountPerTask);

extern "C"
__device__ void gpuEmitCount(size_t		keySize,
	  size_t		valSize,
	  size_t*		outputKeysSizePerTask,
	  size_t*		outputValsSizePerTask,
	  size_t*		outputCountPerTask);

extern "C"
__device__ void gpuEmit  (char*		key, 
	   char*		val, 
	   size_t		keySize, 
	   size_t		valSize,
	   size_t*		psKeySizes, 
	   size_t*		psValSizes, 
	   size_t*		psCounts, 
	   int2*		keyValOffsets, 
	   char*		outputKeys,
	   char*		outputVals,
	   int4*		outputOffsetSizes,
	   size_t*		curIndex);

//--------------------------------------------------------
//MarsScan
//--------------------------------------------------------
extern "C"
void saven_initialPrefixSum(unsigned int maxNumElements);
extern "C"
int prefexSum( int* d_inArr, int* d_outArr, int numRecords );
extern "C"
void prescanArray(int *outArray, int *inArray, int numElements);

//--------------------------------------------------------
//MarsSort
//--------------------------------------------------------
extern "C"
void QuickSortMem(ChunkInfo_t *chunk);
extern "C"
void InitGroupMem(ChunkInfo_t *chunk);
extern "C"
void GroupByMem(ChunkInfo_t *chunk);
extern "C"
int cmp_wrap(const void *arg1, const void *arg2);
extern "C"
void MergeSortFile(FileName_t *file, FileName_t *tmpfile, SortInfo_t *sortInfo, 
				   size_t chunkRecCount, size_t *totalRecCount, size_t *totalGroupCount);
extern "C"
void RearrangeKeyVal(ChunkInfo_t *chunk);
extern "C"
int GPUBitonicSortMem (void * d_inputKeyArray, int totalKeySize, void * d_inputValArray, int totalValueSize, 
		  cmp_type_t * d_inputPointerArray, int rLen, 
		  void * d_outputKeyArray, void * d_outputValArray, 
		  cmp_type_t * d_outputPointerArray, int2 ** h_outputKeyListRange
		  );

//--------------------------------------------------------
//MarsSort
//--------------------------------------------------------
extern "C"
__device__ void gpuEmitKeyFloat(float		key, 
								size_t		keySize, 
								size_t*	psKeySizes,
								size_t*	psCounts,
								int2*		keyValOffsets,
								char*		interKeys,
								int4*		interOffsetSizes,
								size_t*	curIndex);
extern "C"
__device__ void gpuEmitValFloat(float		val, 
								size_t		valSize,
								size_t*	psValSizes,
								size_t*	psCounts,
								int2*		keyValOffsets,
								char*		interVals,
								int4*		interOffsetSizes,
								size_t*	curIndex);
extern "C"
__device__ void gpuEmitKeyFloat2(float2		key, 
								size_t		keySize, 
								size_t*	psKeySizes,
								size_t*	psCounts,
								int2*		keyValOffsets,
								char*		interKeys,
								int4*		interOffsetSizes,
								size_t*	curIndex);
extern "C"
__device__ void gpuEmitValFloat2(float2		val, 
								size_t		valSize,
								size_t*	psValSizes,
								size_t*	psCounts,
								int2*		keyValOffsets,
								char*		interVals,
								int4*		interOffsetSizes,
								size_t*	curIndex);
extern "C"
__device__ void gpuEmitKeyInt(int		key, 
								size_t		keySize, 
								size_t*	psKeySizes,
								size_t*	psCounts,
								int2*		keyValOffsets,
								char*		interKeys,
								int4*		interOffsetSizes,
								size_t*	curIndex);
extern "C"
__device__ void gpuEmitValInt(int		val, 
								size_t		valSize,
								size_t*	psValSizes,
								size_t*	psCounts,
								int2*		keyValOffsets,
								char*		interVals,
								int4*		interOffsetSizes,
								size_t*	curIndex);
extern "C"
__device__ void gpuEmitKeyInt2(int2		key, 
							  size_t		keySize, 
							  size_t*	psKeySizes,
							  size_t*	psCounts,
							  int2*		keyValOffsets,
							  char*		interKeys,
							  int4*		interOffsetSizes,
							  size_t*	curIndex);
extern "C"
__device__ void gpuEmitValInt2(int2		val, 
							  size_t		valSize,
							  size_t*	psValSizes,
							  size_t*	psCounts,
							  int2*		keyValOffsets,
							  char*		interVals,
							  int4*		interOffsetSizes,
							  size_t*	curIndex);
__device__ void gpuEmitKeyInt5(int5		key, 
							   size_t		keySize, 
							   size_t*	psKeySizes,
							   size_t*	psCounts,
							   int2*		keyValOffsets,
							   char*		interKeys,
							   int4*		interOffsetSizes,
							   size_t*	curIndex);
extern "C"
__device__ void gpuEmitValInt5(int5		val, 
							   size_t		valSize,
							   size_t*	psValSizes,
							   size_t*	psCounts,
							   int2*		keyValOffsets,
							   char*		interVals,
							   int4*		interOffsetSizes,
							   size_t*	curIndex);
extern "C"
__device__ void gpuEmitKeyPointer(char	*key, 
								  size_t		keySize, 
								  size_t*	psKeySizes,
								  size_t*	psCounts,
								  int2*		keyValOffsets,
								  char*		interKeys,
								  int4*		interOffsetSizes,
								  size_t*	curIndex);

extern "C"
__device__ void gpuEmitValPointer(char		*val, 
								  size_t		valSize,
								  size_t*	psValSizes,
								  size_t*	psCounts,
								  int2*		keyValOffsets,
								  char*		interVals,
								  int4*		interOffsetSizes,
								  size_t*	curIndex);

extern "C"
__device__ void gpuEmitIntermediate(char*		key, 
									char*		val, 
									size_t		keySize, 
									size_t		valSize,
									size_t*	psKeySizes,
									size_t*	psValSizes,
									size_t*	psCounts,
									int2*		keyValOffsets,
									char*		interKeys,
									char*		interVals,
									int4*		interOffsetSizes,
									size_t*	curIndex);
#endif //__MARSFUNC_H__