-#include <stdio.h>
-#include <assert.h>
+#include <cstdio>
+#include <cassert>
+#include <algorithm>
+#include <thrust/device_vector.h>
+#include <thrust/host_vector.h>
 #include <agency/cuda/execution_policy.hpp>
-#include <algorithm>
-
-// Convenience function for checking CUDA runtime API results
-// can be wrapped around any runtime API call. No-op in release builds.
-inline
-cudaError_t checkCuda(cudaError_t result)
-{
-#if defined(DEBUG) || defined(_DEBUG)
-  if (result != cudaSuccess) {
-    fprintf(stderr, "CUDA Runtime Error: %s\n", cudaGetErrorString(result));
-    assert(result == cudaSuccess);
-  }
-#endif
-  return result;
-}
 
 const int TILE_DIM = 32;
 const int BLOCK_ROWS = 8;
 const int NUM_REPS = 100;
 
 // Check errors and print GB/s
-void postprocess(const std::vector<float>& ref, const std::vector<float>& res, float ms)
+void postprocess(const thrust::host_vector<float>& ref, const thrust::host_vector<float>& res, float ms)
 {
   auto mismatch = std::mismatch(ref.begin(), ref.end(), res.begin());
   if(mismatch.first != ref.end())


  cuda_timer(cudaStream_t stream = 0) : stream_(stream), start_{0}, end_{0}
  {
    checkCuda(cudaEventCreate(&start_));
    checkCuda(cudaEventCreate(&end_));
    reset();
  }

  void reset()
  {
    checkCuda(cudaEventRecord(start_, stream_));
  }

  float elapsed_milliseconds() const
  {
    checkCuda(cudaEventRecord(end_, stream_));
    checkCuda(cudaEventSynchronize(end_));

    float result = 0;
    checkCuda(cudaEventElapsedTime(&result, start_, end_));
    return result;
  }

  ~cuda_timer()
  {
    checkCuda(cudaEventDestroy(start_));
    checkCuda(cudaEventDestroy(end_));
  }
};


    for (int i = 0; i < nx; i++)
      gold[j*nx + i] = h_idata[i*nx + j];
  
  // copy input to device
  d_idata = h_idata;
  
  float ms;
  cuda_timer timer;
