Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Fixed small details in SPH examples #5

Merged
merged 2 commits into from
Jun 14, 2024
Merged
Show file tree
Hide file tree
Changes from all commits
Commits
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
5 changes: 2 additions & 3 deletions example/Vector/7_SPH_dlb/main.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -303,7 +303,7 @@ inline double Wab(double r)
if (r < 1.0)
return (1.0 - 3.0/2.0*r*r + 3.0/4.0*r*r*r)*a2;
else if (r < 2.0)
return (1.0/4.0*(2.0 - r*r)*(2.0 - r*r)*(2.0 - r*r))*a2;
return (1.0/4.0*(2.0 - r)*(2.0 - r)*(2.0 - r))*a2;
else
return 0.0;
}
Expand Down Expand Up @@ -1031,7 +1031,7 @@ int main(int argc, char* argv[])

// extended boundary around the domain, and the processor domain
Ghost<3,double> g(2*H);

//! \cond [Initialization and parameters] \endcond

/*!
Expand Down Expand Up @@ -1493,4 +1493,3 @@ int main(int argc, char* argv[])
*
*/
}

14 changes: 7 additions & 7 deletions example/Vector/7_SPH_dlb_gpu/main.cu
Original file line number Diff line number Diff line change
Expand Up @@ -45,12 +45,12 @@
*
* my_kernel<<<wthr,thr>>>(arguments ... )
*
* Where wthr is the number of workgroups and thr is the number of threads in a workgroup and arguments... are the arguments to pass to the kernel.
* Where wthr is the number of workgroups and thr is the number of threads in a workgroup and arguments... are the arguments to pass to the kernel.
* Equivalently we can launch a kernel with the macro CUDA_LAUNCH_DIM3(my_kernel,wthr,thr,arguments...) or CUDA_LAUNCH(my_kernel,ite,arguments) where
* ite has been taken using getDomainIteratorGPU. There are several advantage on using CUDA_LAUNCH. The first advantage in using the macro is enabling SE_CLASS1
* all kernel launch become synchronous and an error check is performed before continue to the next kernel making debugging easier. Another feature is the possibility
* to run CUDA code on CPU without a GPU. compiling with "CUDA_ON_CPU=1 make" (Note openfpm must be compiled with GPU support (-g) or with CUDA_ON_CPU support
* (-c "... --enable_cuda_on_cpu"). You can compile this example on CPU. You do not have to change a single line of code for this example. (Check the video to see this
* to run CUDA code on CPU without a GPU. compiling with "CUDA_ON_CPU=1 make" (Note openfpm must be compiled with GPU support (-g) or with CUDA_ON_CPU support
* (-c "... --enable_cuda_on_cpu"). You can compile this example on CPU. You do not have to change a single line of code for this example. (Check the video to see this
* feature in action). All the openfpm GPU example and CUDA example can run on CPU if they use CUDA_LAUNCH as macro. We are planning to support
* AMD GPUs as well using this system.
*
Expand Down Expand Up @@ -223,7 +223,7 @@ inline __device__ __host__ real_number Wab(real_number r)
if (r < 1.0)
return (1.0 - 3.0/2.0*r*r + 3.0/4.0*r*r*r)*a2;
else if (r < 2.0)
return (1.0/4.0*(2.0 - r*r)*(2.0 - r*r)*(2.0 - r*r))*a2;
return (1.0/4.0*(2.0 - r)*(2.0 - r)*(2.0 - r))*a2;
else
return 0.0;
}
Expand Down Expand Up @@ -839,7 +839,7 @@ int main(int argc, char* argv[])

++obstacle_box;
}

vd.map();

// Now that we fill the vector with particles
Expand All @@ -848,7 +848,7 @@ int main(int argc, char* argv[])
vd.addComputationCosts(md);
vd.getDecomposition().decompose();
vd.map();

///////////////////////////

// Ok the initialization is done on CPU on GPU we are doing the main loop, so first we offload all properties on GPU
Expand Down Expand Up @@ -958,7 +958,7 @@ int main(int argc, char* argv[])

openfpm_finalize();
}

#else

int main(int argc, char* argv[])
Expand Down
4 changes: 2 additions & 2 deletions example/Vector/7_SPH_dlb_gpu_more_opt/main.cu
Original file line number Diff line number Diff line change
Expand Up @@ -225,7 +225,7 @@ inline __device__ __host__ real_number Wab(real_number r)
if (r < 1.0)
return (1.0 - 3.0/2.0*r*r + 3.0/4.0*r*r*r)*a2;
else if (r < 2.0)
return (1.0/4.0*(2.0 - r*r)*(2.0 - r*r)*(2.0 - r*r))*a2;
return (1.0/4.0*(2.0 - r)*(2.0 - r)*(2.0 - r))*a2;
else
return 0.0;
}
Expand Down Expand Up @@ -1101,7 +1101,7 @@ int main(int argc, char* argv[])

openfpm_finalize();
}

#else

int main(int argc, char* argv[])
Expand Down
5 changes: 2 additions & 3 deletions example/Vector/7_SPH_dlb_gpu_opt/main.cu
Original file line number Diff line number Diff line change
Expand Up @@ -215,7 +215,7 @@ inline __device__ __host__ real_number Wab(real_number r)
if (r < 1.0)
return (1.0 - 3.0/2.0*r*r + 3.0/4.0*r*r*r)*a2;
else if (r < 2.0)
return (1.0/4.0*(2.0 - r*r)*(2.0 - r*r)*(2.0 - r*r))*a2;
return (1.0/4.0*(2.0 - r)*(2.0 - r)*(2.0 - r))*a2;
else
return 0.0;
}
Expand Down Expand Up @@ -681,7 +681,6 @@ inline void sensor_pressure(Vector & vd,
CUDA_LAUNCH_DIM3(sensor_pressure_gpu,1,1,vd.toKernel(),NN.toKernel(),probe,(real_number *)press_tmp_.toKernel());

vd.template restoreOrder<>(NN);

// move calculated pressure on
press_tmp_.deviceToHost();
press_tmp = *(real_number *)press_tmp_.getPointer();
Expand Down Expand Up @@ -998,7 +997,7 @@ int main(int argc, char* argv[])

openfpm_finalize();
}

#else

int main(int argc, char* argv[])
Expand Down
9 changes: 4 additions & 5 deletions example/Vector/7_SPH_dlb_opt/main.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -383,11 +383,11 @@ inline double Wab(double r)
r /= H;

if (r < 1.0)
{return (1.0 - 3.0/2.0*r*r + 3.0/4.0*r*r*r)*a2;}
return (1.0 - 3.0/2.0*r*r + 3.0/4.0*r*r*r)*a2;
else if (r < 2.0)
{return (1.0/4.0*(2.0 - r*r)*(2.0 - r*r)*(2.0 - r*r))*a2;}
return (1.0/4.0*(2.0 - r)*(2.0 - r)*(2.0 - r))*a2;
else
{return 0.0;}
return 0.0;
}

const double c1 = -3.0/M_PI/H/H/H/H;
Expand Down Expand Up @@ -881,7 +881,7 @@ int main(int argc, char* argv[])
// extended boundary around the domain, and the processor domain
// by the support of the cubic kernel
Ghost<3,double> g(r_gskin);

/*! \cond [skin_calc] \endcond */

// Eliminating the lower part of the ghost
Expand Down Expand Up @@ -1141,4 +1141,3 @@ int main(int argc, char* argv[])

openfpm_finalize();
}

Loading