#include <shrUtils.h>
#include <shrQATest.h>
#include <cutil_inline.h>
#include "ppm.h"

__global__ void fade (unsigned char *in, unsigned char *out, float f, int pixels)
{
  int xIndex = blockIdx.x * blockDim.x + threadIdx.x;
  if (xIndex >= pixels) return;

  unsigned char *t = in + xIndex;
  unsigned int v = *t;
  v = v * f;  
  if (v > 255) v = 255;
  t =  out + xIndex;
  *t = v;
}

int
main( int argc, char** argv) 
{
  float f = 1.5;
  ppm_t *image = ppmread ("Untitled.ppm");
  unsigned char *din, *dout;
  int pixels = image->x * image->y * 3;
  int threads = 256;
 
  int blocks =  (pixels / threads) + ((pixels % threads) ? 1: 0);

  cutilSafeCall( cudaMalloc ((void**) &din, pixels) );
  cutilSafeCall( cudaMalloc ((void**) &dout, pixels) );
  cutilSafeCall( cudaMemcpy (din, image->rgb, pixels, cudaMemcpyHostToDevice) );

  // initialize events
  cudaEvent_t start, stop;
  cutilSafeCall( cudaEventCreate(&start) );
  cutilSafeCall( cudaEventCreate(&stop) );

  cutilSafeCall( cudaEventRecord(start, 0) );
  fade <<<blocks, threads>>> (din, dout, f, pixels);

  cutilSafeCall( cudaEventRecord(stop, 0) );
  cutilSafeCall( cudaEventSynchronize(stop) );

  float outerTime;
  cutilSafeCall( cudaEventElapsedTime(&outerTime, start, stop) );
  cutilSafeCall( cudaMemcpy(image->rgb, dout, pixels, cudaMemcpyDeviceToHost) );

  ppmwrite ("out.ppm", image);
  ppmfree (image);
  cutilSafeCall( cudaFree(din) );
  cutilSafeCall( cudaFree(dout) );

  cutilSafeCall( cudaEventDestroy(start) );
  cutilSafeCall( cudaEventDestroy(stop) );
  
  return 0;
}