 __global__ void kernel(int *a, int *b)
 {
   a[threadIdx.x] = b[threadIdx.x];
+}
+
+#define CudaErrorCheck(ans) { gpuAssert((ans), __FILE__, __LINE__); }
+inline void gpuAssert(cudaError_t code, const char *file, int line, bool abort = true)
+{
+  if (code != cudaSuccess)
+  {
+    fprintf(stderr, "GPUassert: %s %s %d\n", cudaGetErrorString(code), file, line);
+    if (abort) exit(code);
+  }
 }
 
 int cuda_reconstruction(

  kernel<<<dimGrid, dimBlock>>>(d_a, d_b);

  // transfer data from device to host
  cudaMemcpy(h_a, d_a, N * sizeof(int), cudaMemcpyDeviceToHost);
  for (int i = 0; i < N; i++)
  {
    h_outScalar[i] = h_a[i];

  int *d_a, *d_b;
  cudaMalloc((void**)&d_a, N * sizeof(int));
  cudaMalloc((void**)&d_b, N * sizeof(int));
  CudaErrorCheck(cudaMemcpy(d_a, h_a, N * sizeof(int), cudaMemcpyHostToDevice));
  CudaErrorCheck(cudaMemcpy(d_b, h_b, N * sizeof(int), cudaMemcpyHostToDevice));

  // organize threads into blocks and grids
  dim3 dimBlock(N, 1, 1);
