     unsigned int ngpu = gpu_partition.getNumActiveGPUs();
     for (int idev = ngpu - 1; idev >= 0; --idev)
         {
-        auto range = gpu_partition.getRangeAndSetGPU(idev);
+        gpu_partition.getRangeAndSetGPU(idev);
 
         cudaMemcpyToSymbol(GPU_rho_coeff, &(CPU_rho_coeff[0]), order * (2*order+1) * sizeof(Scalar));
         }


        unsigned int nwork = range.second - range.first;
        unsigned int n_blocks = nwork/run_block_size+1;

        // force mesh includes ghost cells
        unsigned int num_cells = grid_dim.x*grid_dim.y*grid_dim.z;
        inv_fourier_mesh_tex_x.normalized = false;
        inv_fourier_mesh_tex_x.filterMode = cudaFilterModePoint;
        inv_fourier_mesh_tex_y.normalized = false;
        inv_fourier_mesh_tex_y.filterMode = cudaFilterModePoint;
        inv_fourier_mesh_tex_z.normalized = false;
        inv_fourier_mesh_tex_z.filterMode = cudaFilterModePoint;

        cudaBindTexture(0, inv_fourier_mesh_tex_x, d_inv_fourier_mesh_x, sizeof(cufftComplex)*num_cells);
        cudaBindTexture(0, inv_fourier_mesh_tex_y, d_inv_fourier_mesh_y, sizeof(cufftComplex)*num_cells);
        cudaBindTexture(0, inv_fourier_mesh_tex_z, d_inv_fourier_mesh_z, sizeof(cufftComplex)*num_cells);

        gpu_compute_forces_kernel<<<n_blocks,run_block_size>>>(nwork,
                 d_postype,

                                                      NNN);
    }

__global__ void gpu_compute_forces_kernel(const unsigned int work_size,
                                          const Scalar4 *d_postype,
                                          Scalar4 *d_force,
