       cudaFree(d_in_radius[s]);
       cudaFree(d_coords_ref[s]);
     }
+    return pops;
   }
 
   Pops

               , coords_ref.data()
               , sizeof(float) * n_cols
               , cudaMemcpyHostToDevice);
      //TODO honor boxlimits (-> rng & offset)
      in_radius <<< rng
                  , BSIZE
                  , 0

    float* d_radii2;
    cudaMalloc((void**) &d_radii2
             , sizeof(float) * n_radii);
    std::vector<float> radii2(radii);
    for (float& r: radii2) {
      r *= r;
    }
    cudaMemcpy(d_radii2
             , radii2.data()
             , sizeof(float) * n_radii
             , cudaMemcpyHostToDevice);
    // tmp buffer for in/out info & reference coords (per stream)
