diff --git a/src/grid/grid3D.h b/src/grid/grid3D.h index 7b6a7ae59..01635fdf0 100644 --- a/src/grid/grid3D.h +++ b/src/grid/grid3D.h @@ -845,13 +845,13 @@ class Grid3D void Advance_Particles_KDK_Step2_GPU(); void Set_Particles_Boundary_GPU(int dir, int side); void Set_Particles_Density_Boundaries_Periodic_GPU(int direction, int side); + int Load_Particles_Density_Boundary_to_Buffer_GPU(int direction, int side, Real *buffer); + void Unload_Particles_Density_Boundary_From_Buffer_GPU(int direction, int side, Real *buffer); #endif // PARTICLES_GPU #ifdef GRAVITY_GPU void Copy_Potential_From_GPU(); void Copy_Particles_Density_to_GPU(); void Copy_Particles_Density_GPU(); - int Load_Particles_Density_Boundary_to_Buffer_GPU(int direction, int side, Real *buffer); - void Unload_Particles_Density_Boundary_From_Buffer_GPU(int direction, int side, Real *buffer); #endif // GRAVITY_GPU #endif // PARTICLES diff --git a/src/particles/density_CIC.cpp b/src/particles/density_CIC.cpp index 428a0e864..f576aafa4 100644 --- a/src/particles/density_CIC.cpp +++ b/src/particles/density_CIC.cpp @@ -70,7 +70,7 @@ void Grid3D::Copy_Particles_Density() Copy_Particles_Density_to_GPU(); #endif Copy_Particles_Density_GPU(); - #else + #elif defined(GRAVITY) #ifndef PARALLEL_OMP Copy_Particles_Density_function(0, Grav.nz_local); @@ -90,11 +90,12 @@ void Grid3D::Copy_Particles_Density() } #endif // PARALLEL_OMP - #endif // GRAVITY_GPU + #endif // GRAVITY_GPU and GRAVITY } void Grid3D::Copy_Particles_Density_function(int g_start, int g_end) { + #ifdef GRAVITY int nx_part, ny_part, nz_part, nGHST; nGHST = Particles.G.n_ghost_particles_grid; nx_part = Particles.G.nx_local + 2 * nGHST; @@ -116,6 +117,7 @@ void Grid3D::Copy_Particles_Density_function(int g_start, int g_end) } } } + #endif // GRAVITY } // Clear the density array: density=0 diff --git a/src/particles/density_boundaries_gpu.cu b/src/particles/density_boundaries_gpu.cu index fd5c4ddca..d14069e68 100644 --- a/src/particles/density_boundaries_gpu.cu +++ b/src/particles/density_boundaries_gpu.cu @@ -1,9 +1,10 @@ -#if defined(PARTICLES_GPU) && defined(GRAVITY_GPU) +#if defined(PARTICLES_GPU) #include #include "../grid/grid3D.h" #include "../io/io.h" + #include "../utils/error_handling.h" #include "particles_3D.h" __global__ void Set_Particles_Density_Boundaries_Periodic_kernel(int direction, int side, int n_i, int n_j, int nx, @@ -68,6 +69,9 @@ __global__ void Set_Particles_Density_Boundaries_Periodic_kernel(int direction, void Grid3D::Set_Particles_Density_Boundaries_Periodic_GPU(int direction, int side) { + #ifndef GRAVITY_GPU + CHOLLA_ERROR("This function should not be invoked when compiled without GPU-Gravity"); + #endif GRAVITY_GPU int n_ghost, nx_g, ny_g, nz_g, size, ngrid, n_i, n_j; n_ghost = Particles.G.n_ghost_particles_grid; nx_g = Particles.G.nx_local + 2 * n_ghost; @@ -148,6 +152,9 @@ __global__ void Load_Particles_Density_Boundary_to_Buffer_kernel(int direction, int Grid3D::Load_Particles_Density_Boundary_to_Buffer_GPU(int direction, int side, Real *buffer) { + #ifndef GRAVITY_GPU + CHOLLA_ERROR("This function should not be invoked when compiled without GPU-Gravity"); + #endif GRAVITY_GPU int n_ghost, nx_g, ny_g, nz_g, size_buffer, ngrid, n_i, n_j; n_ghost = Particles.G.n_ghost_particles_grid; nx_g = Particles.G.nx_local + 2 * n_ghost; @@ -236,6 +243,9 @@ __global__ void Unload_Particles_Density_Boundary_to_Buffer_kernel(int direction void Grid3D::Unload_Particles_Density_Boundary_From_Buffer_GPU(int direction, int side, Real *buffer) { + #ifndef GRAVITY_GPU + CHOLLA_ERROR("This function should not be invoked when compiled without GPU-Gravity"); + #endif GRAVITY_GPU int n_ghost, nx_g, ny_g, nz_g, size_buffer, ngrid, n_i, n_j; n_ghost = Particles.G.n_ghost_particles_grid; nx_g = Particles.G.nx_local + 2 * n_ghost; @@ -276,4 +286,4 @@ void Grid3D::Unload_Particles_Density_Boundary_From_Buffer_GPU(int direction, in #endif // MPI_CHOLLA -#endif // PARTICLES_GPU & GRAVITY_GPU +#endif // PARTICLES_GPU diff --git a/src/particles/gravity_CIC.cpp b/src/particles/gravity_CIC.cpp index 495e7cf33..c6bc5368d 100644 --- a/src/particles/gravity_CIC.cpp +++ b/src/particles/gravity_CIC.cpp @@ -46,7 +46,7 @@ void Grid3D::Get_Gravity_Field_Particles() #endif // PARALLEL_OMP #endif // PARTICLES_CPU - #ifdef PARTICLES_GPU + #if defined(PARTICLES_GPU) && defined(GRAVITY_GPU) Particles.Get_Gravity_Field_Particles_GPU(Grav.F.potential_h); #endif } diff --git a/src/particles/gravity_CIC_gpu.cu b/src/particles/gravity_CIC_gpu.cu index 4711b1a32..62e30555f 100644 --- a/src/particles/gravity_CIC_gpu.cu +++ b/src/particles/gravity_CIC_gpu.cu @@ -108,9 +108,10 @@ void Particles3D::Get_Gravity_Field_Particles_GPU_function(int nx_local, int ny_ Real *gravity_x_dev, Real *gravity_y_dev, Real *gravity_z_dev) { - #ifndef GRAVITY_GPU + #ifdef GRAVITY + #ifndef GRAVITY_GPU Copy_Potential_To_GPU(potential_host, potential_dev, n_cells_potential); - #endif + #endif int nx_g, ny_g, nz_g; nx_g = nx_local + 2 * N_GHOST_POTENTIAL; @@ -133,6 +134,7 @@ void Particles3D::Get_Gravity_Field_Particles_GPU_function(int nx_local, int ny_ gravity_y_dev, gravity_z_dev, nx_local, ny_local, nz_local, n_ghost_particles_grid, N_GHOST_POTENTIAL, dx, dy, dz); GPU_Error_Check(); + #endif // GRAVITY } // Get CIC indexes from the particles positions diff --git a/src/particles/particles_3D.h b/src/particles/particles_3D.h index dbda79be5..d64e35e34 100644 --- a/src/particles/particles_3D.h +++ b/src/particles/particles_3D.h @@ -17,6 +17,8 @@ #include "../grid/spatial_domain_props.h" #ifdef PARTICLES_GPU + #include "../utils/gpu.hpp" // cudaFree + #define TPB_PARTICLES 1024 // #define PRINT_GPU_MEMORY #define PRINT_MAX_MEMORY_USAGE diff --git a/src/particles/particles_boundaries.cpp b/src/particles/particles_boundaries.cpp index 96e4f110e..5310a847d 100644 --- a/src/particles/particles_boundaries.cpp +++ b/src/particles/particles_boundaries.cpp @@ -96,9 +96,11 @@ void Grid3D::Unload_Particles_From_Buffers_BLOCK(int index, int *flags) if (H.TRANSFER_HYDRO_BOUNDARIES) { return; } + #ifdef GRAVITY if (Grav.TRANSFER_POTENTIAL_BOUNDARIES) { return; } + #endif if (index == 0) { Unload_Particles_from_Buffer_X0(flags);