From 1f3bfff9739227bb61f4ac0123653061df102a03 Mon Sep 17 00:00:00 2001 From: Weiqun Zhang Date: Sat, 10 Feb 2024 04:22:01 -0800 Subject: [PATCH 1/8] ROCm workaround: Use ParallelFor instead of Reduce (#2749) Assuming the failure is not often, we can use ParallelFor with atomicAdd to obtain the number of failures. With this change, the ROCm memory issue seems to be gone. --- Source/reactions/Castro_react.cpp | 85 ++++++++++++++++++++----------- 1 file changed, 55 insertions(+), 30 deletions(-) diff --git a/Source/reactions/Castro_react.cpp b/Source/reactions/Castro_react.cpp index 9a53e398d3..f49efc9247 100644 --- a/Source/reactions/Castro_react.cpp +++ b/Source/reactions/Castro_react.cpp @@ -186,12 +186,14 @@ Castro::react_state(MultiFab& s, MultiFab& r, Real time, Real dt, const int stra MultiFab tmp_mask_mf; const MultiFab& mask_mf = mask_covered_zones ? getLevel(level+1).build_fine_mask() : tmp_mask_mf; - ReduceOps reduce_op; - ReduceData reduce_data(reduce_op); - using ReduceTuple = typename decltype(reduce_data)::Type; +#if defined(AMREX_USE_GPU) + Gpu::Buffer d_num_failed({0}); + auto* p_num_failed = d_num_failed.data(); +#endif + int num_failed = 0; #ifdef _OPENMP -#pragma omp parallel +#pragma omp parallel reduction(+:num_failed) #endif for (MFIter mfi(s, TilingIfNotGPU()); mfi.isValid(); ++mfi) { @@ -208,8 +210,11 @@ Castro::react_state(MultiFab& s, MultiFab& r, Real time, Real dt, const int stra const auto problo = geom.ProbLoArray(); #endif - reduce_op.eval(bx, reduce_data, - [=] AMREX_GPU_HOST_DEVICE (int i, int j, int k) -> ReduceTuple +#if defined(AMREX_USE_GPU) + ParallelFor(bx, [=] AMREX_GPU_DEVICE (int i, int j, int k) +#else + LoopOnCpu(bx, [&] (int i, int j, int k) mutable +#endif { burn_t burn_state; @@ -230,7 +235,7 @@ Castro::react_state(MultiFab& s, MultiFab& r, Real time, Real dt, const int stra bool do_burn = true; burn_state.success = true; - Real burn_failed = 0.0_rt; + int burn_failed = 0; // Don't burn on zones inside shock regions, if the relevant option is set. @@ -329,7 +334,7 @@ Castro::react_state(MultiFab& s, MultiFab& r, Real time, Real dt, const int stra // If we were unsuccessful, update the failure count. if (!burn_state.success) { - burn_failed = 1.0_rt; + burn_failed = 1; } // Add burning rates to reactions MultiFab, but be @@ -399,19 +404,25 @@ Castro::react_state(MultiFab& s, MultiFab& r, Real time, Real dt, const int stra } - - return {burn_failed}; - +#if defined(AMREX_USE_GPU) + if (burn_failed) { + Gpu::Atomic::Add(p_num_failed, burn_failed); + } +#else + num_failed += burn_failed; +#endif }); +#if defined(AMREX_USE_HIP) + Gpu::streamSynchronize(); // otherwise HIP may faile to allocate the necessary resources. +#endif } - ReduceTuple hv = reduce_data.value(); - Real burn_failed = amrex::get<0>(hv); +#if defined(AMREX_USE_GPU) + num_failed = *(d_num_failed.copyToHost()); +#endif - if (burn_failed != 0.0) { - burn_success = 0; - } + burn_success = !num_failed; ParallelDescriptor::ReduceIntMin(burn_success); @@ -516,11 +527,13 @@ Castro::react_state(Real time, Real dt) int burn_success = 1; - ReduceOps reduce_op; - ReduceData reduce_data(reduce_op); - - using ReduceTuple = typename decltype(reduce_data)::Type; +#if defined(AMREX_USE_GPU) + Gpu::Buffer d_num_failed({0}); + auto* p_num_failed = d_num_failed.data(); +#endif + int num_failed = 0; + // why no omp here? for (MFIter mfi(S_new, TilingIfNotGPU()); mfi.isValid(); ++mfi) { const Box& bx = mfi.growntilebox(ng); @@ -542,8 +555,11 @@ Castro::react_state(Real time, Real dt) const auto dx = geom.CellSizeArray(); const auto problo = geom.ProbLoArray(); - reduce_op.eval(bx, reduce_data, - [=] AMREX_GPU_HOST_DEVICE (int i, int j, int k) -> ReduceTuple +#if defined(AMREX_USE_GPU) + ParallelFor(bx, [=] AMREX_GPU_DEVICE (int i, int j, int k) +#else + LoopOnCpu(bx, [&] (int i, int j, int k) mutable +#endif { burn_t burn_state; @@ -563,7 +579,7 @@ Castro::react_state(Real time, Real dt) bool do_burn = true; burn_state.success = true; - Real burn_failed = 0.0_rt; + int burn_failed = 0; // Don't burn on zones inside shock regions, if the // relevant option is set. @@ -687,7 +703,7 @@ Castro::react_state(Real time, Real dt) // If we were unsuccessful, update the failure count. if (!burn_state.success) { - burn_failed = 1.0_rt; + burn_failed = 1; } // update the state data. @@ -780,16 +796,25 @@ Castro::react_state(Real time, Real dt) } } - return {burn_failed}; +#if defined(AMREX_USE_GPU) + if (burn_failed) { + Gpu::Atomic::Add(p_num_failed, burn_failed); + } +#else + num_failed += burn_failed; +#endif }); + +#if defined(AMREX_USE_HIP) + Gpu::streamSynchronize(); // otherwise HIP may faile to allocate the necessary resources. +#endif } - ReduceTuple hv = reduce_data.value(); - Real burn_failed = amrex::get<0>(hv); +#if defined(AMREX_USE_GPU) + num_failed = *(d_num_failed.copyToHost()); +#endif - if (burn_failed != 0.0) { - burn_success = 0; - } + burn_success = !num_failed; ParallelDescriptor::ReduceIntMin(burn_success); From 8d4459fdb1f17cf607a0f8a01fce13c59563f343 Mon Sep 17 00:00:00 2001 From: Michael Zingale Date: Sat, 10 Feb 2024 16:14:02 -0500 Subject: [PATCH 2/8] codespell fix (#2751) --- Source/reactions/Castro_react.cpp | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/Source/reactions/Castro_react.cpp b/Source/reactions/Castro_react.cpp index 6859b891a5..73c5a3c873 100644 --- a/Source/reactions/Castro_react.cpp +++ b/Source/reactions/Castro_react.cpp @@ -414,7 +414,7 @@ Castro::react_state(MultiFab& s, MultiFab& r, Real time, Real dt, const int stra }); #if defined(AMREX_USE_HIP) - Gpu::streamSynchronize(); // otherwise HIP may faile to allocate the necessary resources. + Gpu::streamSynchronize(); // otherwise HIP may fail to allocate the necessary resources. #endif } @@ -806,7 +806,7 @@ Castro::react_state(Real time, Real dt) }); #if defined(AMREX_USE_HIP) - Gpu::streamSynchronize(); // otherwise HIP may faile to allocate the necessary resources. + Gpu::streamSynchronize(); // otherwise HIP may fail to allocate the necessary resources. #endif } From 9bbf9dc13b949eaaaadf89c7e048c0cc0e58f88a Mon Sep 17 00:00:00 2001 From: Michael Zingale Date: Sun, 11 Feb 2024 10:49:57 -0500 Subject: [PATCH 3/8] add some checks on dynamic_casts in gravity (#2737) --- Source/gravity/Gravity.cpp | 25 +++++++++++++++++++------ 1 file changed, 19 insertions(+), 6 deletions(-) diff --git a/Source/gravity/Gravity.cpp b/Source/gravity/Gravity.cpp index 8ef59f0e46..35fab221d3 100644 --- a/Source/gravity/Gravity.cpp +++ b/Source/gravity/Gravity.cpp @@ -1844,8 +1844,13 @@ Gravity::fill_multipole_BCs(int crse_level, int fine_level, const Vector(&(parent->getLevel(lev+1)))->build_fine_mask(); - MultiFab::Multiply(source, mask, 0, 0, 1, 0); + auto *castro_level = dynamic_cast(&(parent->getLevel(lev+1))); + if (castro_level != nullptr) { + const MultiFab& mask = castro_level->build_fine_mask(); + MultiFab::Multiply(source, mask, 0, 0, 1, 0); + } else { + amrex::Abort("unable to access mask"); + } } // Loop through the grids and compute the individual contributions @@ -2968,7 +2973,11 @@ Gravity::set_mass_offset (Real time, bool multi_level) { for (int lev = 0; lev <= parent->finestLevel(); lev++) { auto* cs = dynamic_cast(&parent->getLevel(lev)); - mass_offset += cs->volWgtSum("density", time); + if (cs != nullptr) { + mass_offset += cs->volWgtSum("density", time); + } else { + amrex::Abort("unable to access volWgtSum"); + } } } else @@ -3132,9 +3141,13 @@ Gravity::make_radial_gravity(int level, Real time, RealVector& radial_grav) if (lev < level) { auto* fine_level = dynamic_cast(&(parent->getLevel(lev+1))); - const MultiFab& mask = fine_level->build_fine_mask(); - for (int n = 0; n < NUM_STATE; ++n) { - MultiFab::Multiply(S, mask, 0, n, 1, 0); + if (fine_level != nullptr) { + const MultiFab& mask = fine_level->build_fine_mask(); + for (int n = 0; n < NUM_STATE; ++n) { + MultiFab::Multiply(S, mask, 0, n, 1, 0); + } + } else { + amrex::Abort("unable to create mask"); } } From 08d58b44f90b946b857fbc1cc572614c7bc69adf Mon Sep 17 00:00:00 2001 From: Khanak Bhargava <56127611+khanakbhargava@users.noreply.github.com> Date: Sun, 11 Feb 2024 17:29:07 -0500 Subject: [PATCH 4/8] Moving enuc to derived variables in io.rst (#2752) --- Docs/source/io.rst | 5 ++--- 1 file changed, 2 insertions(+), 3 deletions(-) diff --git a/Docs/source/io.rst b/Docs/source/io.rst index 1a29b56988..5f244f2776 100644 --- a/Docs/source/io.rst +++ b/Docs/source/io.rst @@ -271,8 +271,6 @@ radiation quantities). | (where X is any of the species | :math:`\omegadot_k = DX_k/Dt` | | | defined in the network) | | | +-----------------------------------+---------------------------------------------------+--------------------------------------+ -| ``enuc`` | Nuclear energy generation rate / gram | :math:`{\rm erg~g^{-1}~s^{-1}}` | -+-----------------------------------+---------------------------------------------------+--------------------------------------+ | ``rho_enuc`` | Nuclear energy generation rate density | :math:`{\rm erg~cm^{-3}~s^{-1}}` | +-----------------------------------+---------------------------------------------------+--------------------------------------+ | ``phiGrav`` | Gravitational potential | :math:`{\rm erg~g^{-1}}` | @@ -393,7 +391,8 @@ Derived variables | ``y_velocity``, | :math:`\ub = (\rho \ub)/\rho` | | | | ``z_velocity`` | | | | +-----------------------------------+---------------------------------------------------+-----------------------------+-----------------------------------------+ - +| ``enuc`` | Nuclear energy generation rate / gram | ``derenuc`` | :math:`{\rm erg~g^{-1}~s^{-1}}` | ++-----------------------------------+---------------------------------------------------+-----------------------------+-----------------------------------------+ problem-specific plotfile variables ^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^ From fe6ff480b6c88c9ca16d281740373c864893aec3 Mon Sep 17 00:00:00 2001 From: Khanak Bhargava <56127611+khanakbhargava@users.noreply.github.com> Date: Mon, 12 Feb 2024 14:04:00 -0500 Subject: [PATCH 5/8] Fixed a character in the table (#2753) I think the table stuff might just be a warning and not caught by the CI --- Docs/source/io.rst | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/Docs/source/io.rst b/Docs/source/io.rst index 5f244f2776..2054eb4e1c 100644 --- a/Docs/source/io.rst +++ b/Docs/source/io.rst @@ -391,7 +391,7 @@ Derived variables | ``y_velocity``, | :math:`\ub = (\rho \ub)/\rho` | | | | ``z_velocity`` | | | | +-----------------------------------+---------------------------------------------------+-----------------------------+-----------------------------------------+ -| ``enuc`` | Nuclear energy generation rate / gram | ``derenuc`` | :math:`{\rm erg~g^{-1}~s^{-1}}` | +| ``enuc`` | Nuclear energy generation rate / gram | ``derenuc`` | :math:`{\rm erg~g^{-1}~s^{-1}}` | +-----------------------------------+---------------------------------------------------+-----------------------------+-----------------------------------------+ problem-specific plotfile variables From 62193e94e15339132028443bb960f7bea0545846 Mon Sep 17 00:00:00 2001 From: "Eric T. Johnson" Date: Mon, 12 Feb 2024 19:41:09 -0500 Subject: [PATCH 6/8] Turn Sphinx warnings into errors in CI (#2754) --- .github/workflows/docs-test.yml | 9 ++++++++- .gitignore | 2 ++ Docs/rp.py | 10 +++++----- Docs/source/FlowChart.rst | 2 +- Docs/source/docutils.conf | 2 ++ Docs/source/index.rst | 2 +- 6 files changed, 19 insertions(+), 8 deletions(-) create mode 100644 Docs/source/docutils.conf diff --git a/.github/workflows/docs-test.yml b/.github/workflows/docs-test.yml index be80c197fc..562a305b7a 100644 --- a/.github/workflows/docs-test.yml +++ b/.github/workflows/docs-test.yml @@ -9,6 +9,10 @@ on: branches: - development +env: + # enable color output from Sphinx + FORCE_COLOR: "1" + jobs: docs: runs-on: ubuntu-latest @@ -33,7 +37,10 @@ jobs: - name: Build docs run: | cd Docs/ - make SPHINXOPTS=-v NO_DOXYGEN=TRUE html + # remove missing TOC entries that would otherwise be generated by + # doxygen to avoid warnings about missing references + sed -i -e 's/^ filelist$//; s/^ classlist$//' source/index.rst + make SPHINXOPTS='-v -W --keep-going' NO_DOXYGEN=TRUE html - name: Check links run: | diff --git a/.gitignore b/.gitignore index 6f29568fcb..648030bee4 100644 --- a/.gitignore +++ b/.gitignore @@ -72,6 +72,8 @@ Docs/source/namespacelist.rst Docs/source/runtime_parameters.rst Docs/source/*_files.rst Docs/source/preprocessed_files +Docs/source/yt_example.rst +Docs/source/yt_example_files/ amr_diag.out diff --git a/Docs/rp.py b/Docs/rp.py index 343d1f6dc6..5bb7f43301 100755 --- a/Docs/rp.py +++ b/Docs/rp.py @@ -7,17 +7,17 @@ from more_itertools import unique_everseen MAIN_HEADER = """ -+--------------------------------------------+-------------------------------------------------------------+---------------+ -| parameter | description | default value | -+============================================+=============================================================+===============+ ++--------------------------------------------+-------------------------------------------------------------+-----------------------------+ +| parameter | description | default value | ++============================================+=============================================================+=============================+ """ SEPARATOR = """ -+--------------------------------------------+-------------------------------------------------------------+---------------+ ++--------------------------------------------+-------------------------------------------------------------+-----------------------------+ """ ENTRY = """ -| {:42} | {:59} | {:13} | +| {:42} | {:59} | {:27} | """ WRAP_LEN = 59 diff --git a/Docs/source/FlowChart.rst b/Docs/source/FlowChart.rst index f24c4cfa7b..a47b22ad55 100644 --- a/Docs/source/FlowChart.rst +++ b/Docs/source/FlowChart.rst @@ -309,7 +309,7 @@ In the code, the objective is to evolve the state from the old time, A. Create ``Sborder``, initialized from ``S_old`` B. Call ``clean_state()`` to make sure the thermodynamics are in - sync, in particular, compute the temperature. + sync, in particular, compute the temperature. C. [``SHOCK_VAR``] zero out the shock flag. diff --git a/Docs/source/docutils.conf b/Docs/source/docutils.conf new file mode 100644 index 0000000000..3b4d0981a5 --- /dev/null +++ b/Docs/source/docutils.conf @@ -0,0 +1,2 @@ +[parsers] +line_length_limit = 1000000 diff --git a/Docs/source/index.rst b/Docs/source/index.rst index 14b646d378..0500e0bd62 100644 --- a/Docs/source/index.rst +++ b/Docs/source/index.rst @@ -60,7 +60,7 @@ https://github.com/amrex-astro/Castro filelist classlist - .. namespacelist +.. namespacelist .. toctree:: :caption: References From 1bdc65076bb13ee9279f1c4fdf7b0b34648adcab Mon Sep 17 00:00:00 2001 From: KiranEiden <32306975+KiranEiden@users.noreply.github.com> Date: Tue, 13 Feb 2024 14:44:55 -0800 Subject: [PATCH 7/8] Update documentation regarding species sources and normalization (#2742) This PR updates the Hydrodynamics section documentation to highlight that sources for passive variables in the conserved state are off by default. It also removes the reference to castro.normalize_species, which does not appear to be a valid runtime parameter any longer. I am happy to introduce additional documentation regarding the species normalization if that would be helpful -- my understanding is that it is on by default now. --- Docs/source/Hydrodynamics.rst | 11 ++++++++--- 1 file changed, 8 insertions(+), 3 deletions(-) diff --git a/Docs/source/Hydrodynamics.rst b/Docs/source/Hydrodynamics.rst index 8d76bd5fdf..44cdf3e553 100644 --- a/Docs/source/Hydrodynamics.rst +++ b/Docs/source/Hydrodynamics.rst @@ -342,6 +342,14 @@ accounted for in **Steps 1** and **6**. The source terms are: S_{{\rm ext},\rho Y_k} \end{array}\right)^n. +.. index:: USE_SPECIES_SOURCES + +.. note:: To reduce memory usage, we do not include source terms for the + advected quantities, species, and auxiliary variables in the conserved + state vector by default. If your application needs external source terms for + these variables, set `USE_SPECIES_SOURCES=TRUE` when compiling so that space + will be allocated for them. + Primitive Forms =============== @@ -585,9 +593,6 @@ runtime parameters for hydrodynamics: See :ref:`sponge_section` for more details on the sponge. -- ``castro.normalize_species``: enforce that :math:`\sum_i X_i = 1` - (0 or 1; default: 0) - .. index:: castro.small_dens, castro.small_temp, castro.small_pres Several floors are imposed on the thermodynamic quantities to prevet unphysical From 73aacc436dd06fddded5b02cc8d9418b6f896f91 Mon Sep 17 00:00:00 2001 From: KiranEiden <32306975+KiranEiden@users.noreply.github.com> Date: Wed, 14 Feb 2024 05:13:15 -0800 Subject: [PATCH 8/8] Use double-backticks for compilation parameter in docs (#2756) --- Docs/source/Hydrodynamics.rst | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/Docs/source/Hydrodynamics.rst b/Docs/source/Hydrodynamics.rst index 44cdf3e553..edaaf8accd 100644 --- a/Docs/source/Hydrodynamics.rst +++ b/Docs/source/Hydrodynamics.rst @@ -347,7 +347,7 @@ accounted for in **Steps 1** and **6**. The source terms are: .. note:: To reduce memory usage, we do not include source terms for the advected quantities, species, and auxiliary variables in the conserved state vector by default. If your application needs external source terms for - these variables, set `USE_SPECIES_SOURCES=TRUE` when compiling so that space + these variables, set ``USE_SPECIES_SOURCES=TRUE`` when compiling so that space will be allocated for them. Primitive Forms