                                          , i_to);
     }
     cudaDeviceSynchronize();
-    check_error();
+    check_error("after kernel loop");
     // get partial results from GPU
     std::vector<unsigned int> partial_pops(n_rows*n_radii);
     cudaMemcpy(partial_pops.data()

             , sizeof(unsigned int) * n_rows * n_radii);
    cudaMalloc((void**) &d_rad2
             , sizeof(float) * n_radii);
    cudaMemset(d_pops
             , 0
             , sizeof(unsigned int) * n_rows * n_radii);
    cudaMemcpy(d_coords
             , coords
             , sizeof(float) * n_rows * n_cols

#include "lts_cuda_kernels.cuh"

// for pops
//#define BSIZE_POPS 128
#define BSIZE_POPS 1024

// for neighborhood search
#define BSIZE_NH 128
