                   int nx, int ny, int nxf, int nyf, float dx, float dy,
                   float * gamma_up_d, float rho, float gamma,
                   int * matching_indices_d, int ng, int rank, float * qc_comp) {
+    /*
+    Prolong coarse grid data to fine grid
+
+    Parameters
+    ----------
+    kernels, threads, blocks : dim3 *
+        number of kernels, threads and blocks for each process/kernel
+    cumulative_kernels : int *
+        cumulative number of kernels in mpi processes of r < rank
+    q_cd, q_fd : float *
+        coarse and fine grids of state vectors
+    nx, ny : int
+        dimensions of coarse grid
+    nxf, nyf : int
+        dimensions of fine grid
+    dx, dy : float
+        coarse grid cell spacings
+    gamma_up_d : float *
+        spatial metric
+    rho, gamma : float
+        density and adiabatic index
+    matching_indices_d : int *
+        position of fine grid wrt coarse grid
+    ng : int
+        number of ghost cells
+    rank : int
+        rank of MPI process
+    qc_comp : float *
+        grid of compressible variables on coarse grid
+    */
 
     int kx_offset = 0;
     int ky_offset = (kernels[0].y * blocks[0].y * threads[0].y - 2*ng) * rank;

    Using Brent's method, return the root of a function or functor func known
    to lie between x1 and x2. The root will be regined until its accuracy is
    tol.
    */

    const int ITMAX = 300;


__device__ float p_from_swe(float * q, float * gamma_up, float rho,
                            float gamma, float W) {
    /*
    Calculates p and returns using SWE conserved variables

    Parameters
    ----------
    q : float *
        state vector
    gamma_up : float *
        spatial metric
    rho : float
        density
    gamma : float
        adiabatic index
    W : float
        Lorentz factor
    */

    float ph = q[0] / W;

