Merge branch 'AMReX-Codes:development' into development

.github/workflows/intel.yml

@@ -24,7 +24,11 @@ jobs:
         restore-keys: |
              ccache-${{ github.workflow }}-${{ github.job }}-git-
     - name: Build & Install
-      env: {CXXFLAGS: "-fno-operator-names -Werror -Wall -Wextra -Wpedantic -Wnull-dereference -Wfloat-conversion -Wshadow -Woverloaded-virtual -Wextra-semi -Wunreachable-code -Wnon-virtual-dtor"}
+      # /tmp/icpx-2d34de0e47/global_vars-header-4390fb.h:25:36: error: zero size arrays are an extension [-Werror,-Wzero-length-array]
+      #    25 | const char* const kernel_names[] = {
+      #       |                                    ^
+      # 1 error generated.
+      env: {CXXFLAGS: "-fno-operator-names -Werror -Wall -Wextra -Wpedantic -Wnull-dereference -Wfloat-conversion -Wshadow -Woverloaded-virtual -Wextra-semi -Wunreachable-code -Wnon-virtual-dtor -Wno-zero-length-array"}
       run: |
         export CCACHE_COMPRESS=1
         export CCACHE_COMPRESSLEVEL=10
@@ -68,7 +72,11 @@ jobs:
         restore-keys: |
              ccache-${{ github.workflow }}-${{ github.job }}-git-
     - name: Build & Install
-      env: {CXXFLAGS: "-fno-operator-names -Werror -Wall -Wextra -Wpedantic -Wnull-dereference -Wfloat-conversion -Wshadow -Woverloaded-virtual -Wextra-semi -Wunreachable-code -Wnon-virtual-dtor"}
+      # /tmp/icpx-2d34de0e47/global_vars-header-4390fb.h:25:36: error: zero size arrays are an extension [-Werror,-Wzero-length-array]
+      #    25 | const char* const kernel_names[] = {
+      #       |                                    ^
+      # 1 error generated.
+      env: {CXXFLAGS: "-fno-operator-names -Werror -Wall -Wextra -Wpedantic -Wnull-dereference -Wfloat-conversion -Wshadow -Woverloaded-virtual -Wextra-semi -Wunreachable-code -Wnon-virtual-dtor -Wno-zero-length-array"}
       run: |
         export CCACHE_COMPRESS=1
         export CCACHE_COMPRESSLEVEL=10

Docs/sphinx_documentation/source/Basics.rst

@@ -419,6 +419,56 @@ will become
     foo.a = 2
     foo.b = 2
 
+Enum Class
+----------
+
+.. versionadded:: 24.09
+   Enum class support in :cpp:`ParmParse`.
+
+AMReX provides a macro :cpp:`AMREX_ENUM` for defining :cpp:`enum class` that
+supports reflection. For example,
+
+.. highlight:: c++
+
+::
+
+   AMREX_ENUM(MyColor, red, green, blue);
+
+   void f ()
+   {
+       MyColor color = amrex::getEnum<MyColor>("red"); // MyColor::red
+       std::string name = amrex::getEnumNameString(MyColor::blue); // "blue"
+       std::vector<std::string> names = amrex::getEnumNameStrings<MyColor>();
+       // names = {"red", "green", "blue"};
+       std::string class_name = amrex::getEnumClassName<MyColor>(); // "MyColor"
+   }
+
+This allows us to read :cpp:`ParmParse` parameters into enum class objects.
+
+.. highlight:: python
+
+::
+
+   color1 = red
+   color2 = BLue
+
+The following code shows how to query the enumerators.
+
+.. highlight:: c++
+
+::
+
+   AMREX_ENUM(MyColor, none, red, green, blue);
+
+   void f (MyColor& c1, MyColor& c2)
+   {
+       ParmParse pp;
+       pp.query("color1", c1); // c1 becomes MyColor::red
+       pp.query_enum_case_insensitive("color2", c2); // c2 becomes MyColor::blue
+       MyColor default_color; // MyColor::none
+       pp.query("color3", default_color); // Still MyColor::none
+   }
+
 Overriding Parameters with Command-Line Arguments
 -------------------------------------------------
 

Src/AmrCore/AMReX_AmrMesh.cpp

@@ -471,6 +471,9 @@ AmrMesh::ChopGrids (int lev, BoxArray& ba, int target_size) const
     IntVect chunk = max_grid_size[lev];
     chunk.min(Geom(lev).Domain().length());
 
+    // Note that ba already satisfies the max_grid_size requirement and it's
+    // coarsenable if it's a fine level BoxArray.
+
     while (ba.size() < target_size)
     {
         IntVect chunk_prev = chunk;
@@ -485,11 +488,24 @@ AmrMesh::ChopGrids (int lev, BoxArray& ba, int target_size) const
             int idim = chunk_dir[idx].second;
             if (refine_grid_layout_dims[idim]) {
                 int new_chunk_size = chunk[idim] / 2;
+                int rr = (lev > 0) ? ref_ratio[lev-1][idim] : 1;
+                if (rr > 1) {
+                    new_chunk_size = (new_chunk_size/rr) * rr;
+                }
                 if (new_chunk_size != 0 &&
                     new_chunk_size%blocking_factor[lev][idim] == 0)
                 {
                     chunk[idim] = new_chunk_size;
-                    ba.maxSize(chunk);
+                    if (rr == 1) {
+                        ba.maxSize(chunk);
+                    } else {
+                        IntVect bf(1);
+                        bf[idim] = rr;
+                        // Note that only idim-direction will be chopped by
+                        // minmaxSize because the sizes in other directions
+                        // are already smaller than chunk.
+                        ba.minmaxSize(bf, chunk);
+                    }
                     break;
                 }
             }

Src/Base/AMReX_BLBackTrace.cpp

@@ -13,12 +13,13 @@
 #include <AMReX_TinyProfiler.H>
 #endif
 
+#include <csignal>
+#include <cstdio>
+#include <cstdlib>
+#include <cstring>
+#include <fstream>
 #include <iostream>
 #include <sstream>
-#include <fstream>
-#include <cstring>
-#include <cstdio>
-#include <csignal>
 
 #if !(defined(_MSC_VER) && defined(__CUDACC__))
 //MSVC can't pre-processor cfenv with `Zc:preprocessor`
@@ -177,6 +178,18 @@ namespace {
         }
         return r;
     }
+
+#ifdef __linux__
+    bool command_exists(std::string const &cmd)
+    {
+        // command -v is part of POSIX so should be available
+        std::string check_command = "command -v " + cmd + " > /dev/null 2>&1";
+        int r = std::system(check_command.c_str());
+        // return value of std::system is implementation defined and can be
+        // decoded using WEXITSTATUS but it should be 0 on success
+        return r == 0;
+    }
+#endif
 }
 #endif
 
@@ -209,19 +222,32 @@ BLBackTrace::print_backtrace_info (FILE* f)
         int have_addr2line = 0;
         std::string eu_cmd;
         {
-            have_eu_addr2line = file_exists("/usr/bin/eu-addr2line");
+            if (command_exists("eu-addr2line")) {
+                have_eu_addr2line = 1;
+                eu_cmd = "eu-addr2line";
+            } else {
+                std::string eu_fallback_path = "/usr/bin/eu-addr2line";
+                have_eu_addr2line = file_exists(eu_fallback_path.c_str());
+                eu_cmd = std::move(eu_fallback_path);
+            }
             if (have_eu_addr2line) {
                 const pid_t pid = getpid();
                 // cmd = "/usr/bin/eu-addr2line -C -f -i --pretty-print -p "
-                eu_cmd = "/usr/bin/eu-addr2line -C -f -i -p "
-                    + std::to_string(pid);
+                eu_cmd += " -C -f -i -p " + std::to_string(pid);
             }
         }
         std::string cmd;
         {
-            have_addr2line = file_exists("/usr/bin/addr2line");
+            if (command_exists("addr2line")) {
+                have_addr2line = 1;
+                cmd = "addr2line";
+            } else {
+                std::string fallback_path = "/usr/bin/addr2line";
+                have_addr2line = file_exists(fallback_path.c_str());
+                cmd = std::move(fallback_path);
+            }
             if (have_addr2line) {
-                cmd = "/usr/bin/addr2line -Cpfie " + amrex::system::exename;
+                cmd += " -Cpfie " + amrex::system::exename;
             }
         }
 

Src/Base/AMReX_BoxArray.H

@@ -615,6 +615,11 @@ public:
 
     BoxArray& maxSize (const IntVect& block_size);
 
+    //! Forces each Box in BoxArray to have sizes >= min_size and <=
+    //! max_size. It's the caller's responsibility to make sure both the
+    //! BoxArray and max_size are coarsenable by min_size.
+    BoxArray& minmaxSize (const IntVect& min_size, const IntVect& max_size);
+
     //! Refine each Box in the BoxArray to the specified ratio.
     BoxArray& refine (int refinement_ratio);
 

Src/Base/AMReX_BoxArray.cpp

@@ -555,12 +555,27 @@ BoxArray::maxSize (const IntVect& block_size)
     blst.maxSize(block_size);
     const int N = static_cast<int>(blst.size());
     if (size() != N) { // If size doesn't change, do nothing.
-        BoxList bak = (m_simplified_list) ? *m_simplified_list : BoxList();
+        std::shared_ptr<BoxList> bak;
+        bak.swap(m_simplified_list);
         define(std::move(blst));
-        if (bak.isNotEmpty()) {
-            m_simplified_list = std::make_shared<BoxList>(std::move(bak));
-        }
+        m_simplified_list = std::move(bak);
+    }
+    return *this;
+}
+
+BoxArray&
+BoxArray::minmaxSize (const IntVect& min_size, const IntVect& max_size)
+{
+    AMREX_ASSERT(this->coarsenable(min_size) &&
+                 (max_size/min_size)*min_size == max_size);
+    std::shared_ptr<BoxList> bak;
+    if (m_bat.is_simple() && crseRatio() == IntVect::TheUnitVector()) {
+        bak.swap(m_simplified_list);
     }
+    this->coarsen(min_size);
+    this->maxSize(max_size/min_size);
+    this->refine(min_size);
+    m_simplified_list = std::move(bak);
     return *this;
 }
 

Src/Base/AMReX_Enum.H

@@ -0,0 +1,81 @@
+#ifndef AMREX_ENUM_H_
+#define AMREX_ENUM_H_
+
+#include <AMReX_String.H>
+
+#include <algorithm>
+#include <array>
+#include <stdexcept>
+#include <string>
+#include <string_view>
+#include <type_traits>
+#include <utility>
+#include <vector>
+
+template <typename T>
+using amrex_enum_traits = decltype(amrex_get_enum_traits(std::declval<T>()));
+
+namespace amrex {
+    template <typename T, typename ET = amrex_enum_traits<T>,
+              std::enable_if_t<ET::value,int> = 0>
+    T getEnum (std::string_view const& s)
+    {
+        auto pos = ET::enum_names.find(s);
+        if (pos == std::string_view::npos) {
+            std::string error_msg("amrex::getEnum: Unknown enum: ");
+            error_msg.append(s).append(" in AMREX_ENUM(").append(ET::class_name)
+                .append(", ").append(ET::enum_names).append(").");
+            throw std::runtime_error(error_msg);
+        }
+        auto count = std::count(ET::enum_names.begin(),
+                                ET::enum_names.begin()+pos, ',');
+        return static_cast<T>(count);
+    }
+
+    template <typename T, typename ET = amrex_enum_traits<T>,
+              std::enable_if_t<ET::value,int> = 0>
+    std::string getEnumNameString (T const& v)
+    {
+        auto n = static_cast<int>(v);
+        std::size_t pos = 0;
+        for (int i = 0; i < n; ++i) {
+            pos = ET::enum_names.find(',', pos);
+            if (pos == std::string::npos) {
+                std::string error_msg("amrex::getEnum: Unknown enum value: ");
+                error_msg.append(std::to_string(n)).append(" in AMREX_ENUM(")
+                    .append(ET::class_name).append(", ").append(ET::enum_names)
+                    .append(").");
+                throw std::runtime_error(error_msg);
+            }
+            ++pos;
+        }
+        auto pos2 = ET::enum_names.find(',', pos);
+        return amrex::trim(std::string(ET::enum_names.substr(pos,pos2-pos)));
+    }
+
+    template <typename T, typename ET = amrex_enum_traits<T>,
+              std::enable_if_t<ET::value,int> = 0>
+    std::vector<std::string> getEnumNameStrings ()
+    {
+        return amrex::split(std::string(ET::enum_names), ", ");
+    }
+
+    template <typename T, typename ET = amrex_enum_traits<T>,
+              std::enable_if_t<ET::value,int> = 0>
+    std::string getEnumClassName ()
+    {
+        return std::string(ET::class_name);
+    }
+}
+
+#define AMREX_ENUM(CLASS, ...) \
+    enum class CLASS : int { __VA_ARGS__ }; \
+    struct CLASS##_EnumTraits { \
+        using enum_class_t = CLASS; \
+        static constexpr bool value = true; \
+        static constexpr std::string_view class_name{#CLASS}; \
+        static constexpr std::string_view enum_names{#__VA_ARGS__}; \
+    }; \
+    CLASS##_EnumTraits amrex_get_enum_traits(CLASS)
+
+#endif

Src/Base/AMReX_GpuDevice.H

@@ -14,6 +14,7 @@
 #include <algorithm>
 #include <array>
 #include <cstdlib>
+#include <cstring>
 #include <memory>
 
 #define AMREX_GPU_MAX_STREAMS 8
@@ -318,6 +319,52 @@ dtod_memcpy (void* p_d_dst, const void* p_d_src, const std::size_t sz) noexcept
 void hypreSynchronize ();
 #endif
 
+//! Copy `nbytes` bytes from host to device global variable. `offset` is the
+//! offset in bytes from the start of the device global variable.
+template <typename T>
+void memcpy_from_host_to_device_global_async (T& dg, const void* src,
+                                              std::size_t nbytes,
+                                              std::size_t offset = 0)
+{
+#if defined(AMREX_USE_CUDA)
+    AMREX_CUDA_SAFE_CALL(cudaMemcpyToSymbolAsync(dg, src, nbytes, offset,
+                                                 cudaMemcpyHostToDevice,
+                                                 Device::gpuStream()));
+#elif defined(AMREX_USE_HIP)
+    AMREX_HIP_SAFE_CALL(hipMemcpyToSymbolAsync(dg, src, nbytes, offset,
+                                               hipMemcpyHostToDevice,
+                                               Device::gpuStream()));
+#elif defined(AMREX_USE_SYCL)
+    Device::streamQueue().memcpy(dg, src, nbytes, offset);
+#else
+    auto* p = (char*)(&dg);
+    std::memcpy(p+offset, src, nbytes);
+#endif
+}
+
+//! Copy `nbytes` bytes from device global variable to host. `offset` is the
+//! offset in bytes from the start of the device global variable.
+template <typename T>
+void memcpy_from_device_global_to_host_async (void* dst, T const& dg,
+                                              std::size_t nbytes,
+                                              std::size_t offset = 0)
+{
+#if defined(AMREX_USE_CUDA)
+    AMREX_CUDA_SAFE_CALL(cudaMemcpyFromSymbolAsync(dst, dg, nbytes, offset,
+                                                   cudaMemcpyDeviceToHost,
+                                                   Device::gpuStream()));
+#elif defined(AMREX_USE_HIP)
+    AMREX_HIP_SAFE_CALL(hipMemcpyFromSymbolAsync(dst, dg, nbytes, offset,
+                                                  hipMemcpyDeviceToHost,
+                                                  Device::gpuStream()));
+#elif defined(AMREX_USE_SYCL)
+    Device::streamQueue().memcpy(dst, dg, nbytes, offset);
+#else
+    auto const* p = (char const*)(&dg);
+    std::memcpy(dst, p+offset, nbytes);
+#endif
+}
+
 }
 
 #endif