
#include "Image_Filters.h"

/*TEXTURES*/
texture<uchar4, 2, cudaReadModeNormalizedFloat> texImage_rgb;

cudaChannelFormatDesc uchar4tex = cudaCreateChannelDesc<uchar4>();

cudaArray *cu_image;

/**********************************/
/**********************************/
extern "C"
void CUDA_CreateMemoryArray(int imageW,int imageH){
	cudaMallocArray(&cu_image, &uchar4tex, imageW, imageH);
}

/**********************************/
/**********************************/
extern "C"
void CUDA_BindTextureToArray(){
	cudaBindTextureToArray(texImage_rgb,cu_image);
}

/**********************************/
/**********************************/
extern "C"
void CUDA_FreeArrays(){
	cudaFreeArray(cu_image);
}

/**********************************/
/**********************************/
extern "C"
void CUDA_MemcpyToArray(uchar4 *src,int imageW,int imageH){
	cudaMemcpyToArray( cu_image, 0, 0,src, imageW * imageH * sizeof(uchar4), cudaMemcpyDeviceToDevice);
}

/**********************************/
/**********************************/
/*Fill the filter matrix*/
__constant__ float GaussianKernel[9] = {/*HERE*/};



/**********************************/
/**********************************/

/*Image Filters*/
__global__ void GaussianFilter(uchar4 *Image_dev,int w,int h){
    
    // write index for the pixels (it is similar to "idx")
	const int ix = /*HERE*/;
    const int iy = /*HERE*/;
    const float x=(float)ix+0.5f;
    const float y=(float)iy+0.5f;

	int win_W=1;

	if(ix < w && iy < h){
		float4 pixval;
		float3 sum;
		sum.x=0.0f;
		sum.y=0.0f;
		sum.z=0.0f;
		int k=0;
		for(int ii=-win_W;ii<=win_W;ii++){
			for(int jj=-win_W;jj<=win_W;jj++){
				pixval=tex2D(texImage_rgb,x+ii,y+jj);
                // Write the convolution with the filter matrix
				/*HERE*/ // for the first channel
                /*HERE*/ // for the second channel
                /*HERE*/ // for the third channel
				k++;
			}
		}
		Image_dev[w*iy+ix].x=(unsigned char)((sum.x/16)*255);
		Image_dev[w*iy+ix].y=(unsigned char)((sum.y/16)*255);
		Image_dev[w*iy+ix].z=(unsigned char)((sum.z/16)*255);
	}
}


/**********************************/
/**********************************/
extern "C"
void CUDA_GaussianFilter(uchar4 *Image_dev,int imageW,int imageH,dim3 grid,dim3 threads){
	
    // Call the GaussianFilter
    /*HERE*/
    
    }

