diff --git a/.clang-format b/.clang-format index b3bb8d132..6f7fef667 100644 --- a/.clang-format +++ b/.clang-format @@ -108,6 +108,10 @@ IncludeCategories: Priority: 4 - Regex: '^"checkpoint\/.*\.h"' Priority: 4 + - Regex: '^"kernels\/.*\.hpp"' + Priority: 4 + - Regex: '^"kernels\/.*\.h"' + Priority: 4 - Regex: '^"output\/.*\.h"' Priority: 4 - Regex: '^"archetypes\/.*\.h"' diff --git a/.gitattributes b/.gitattributes new file mode 100644 index 000000000..f8ff2b5de --- /dev/null +++ b/.gitattributes @@ -0,0 +1 @@ +*.mp4 filter=lfs diff=lfs merge=lfs -text diff --git a/.github/workflows/cpuarch.yml b/.github/workflows/cpuarch.yml index 56184e035..12345ccc7 100644 --- a/.github/workflows/cpuarch.yml +++ b/.github/workflows/cpuarch.yml @@ -63,8 +63,10 @@ jobs: cmake --version - name: Install OpenMPI 5 run: | - sudo apt-get install -y libopenmpi-dev openmpi-bin - mpirun --version + if [ "${{ matrix.mpi }}" = "ON" ]; then + sudo apt-get install -y libopenmpi-dev openmpi-bin + mpirun --version + fi - name: Configure run: | if [ "${{ matrix.mpi }}" = "ON" ]; then @@ -87,7 +89,7 @@ jobs: strategy: fail-fast: false matrix: - pgen: [streaming, turbulence, reconnection, shock, magnetosphere, accretion, wald] + pgen: [streaming, turbulence, reconnection, shock, magnetosphere, accretion, wald, examples/custom_emission] steps: - name: Checkout uses: actions/checkout@v4 diff --git a/.gitignore b/.gitignore index a544e9717..62248c605 100644 --- a/.gitignore +++ b/.gitignore @@ -40,6 +40,7 @@ venv/ *.mov *.mp4 !pgens/**/*.png +!pgens/**/*.mp4 # Accidental files *.xc diff --git a/dependencies.py b/dependencies.py index 270b32bfb..bbb986f8d 100755 --- a/dependencies.py +++ b/dependencies.py @@ -143,7 +143,7 @@ def InstallKokkosScriptModfile(settings: Settings) -> tuple[str, str]: f"Kokkos install path {install_path} already exists and overwrite is disabled" ) - extra_flags = "-D ".join(settings.extra_kokkos_flags) + extra_flags = " ".join(["-D " + kf for kf in settings.extra_kokkos_flags]) cxx_standard = 20 if tuple(map(int, version.split("."))) >= (5, 0, 0) else 17 if arch == "": @@ -212,7 +212,7 @@ def InstallAdios2Script(settings: Settings) -> tuple[str, str]: f"Adios2 install path {install_path} already exists and overwrite is disabled" ) - extra_flags = "-D ".join(settings.extra_adios2_flags) + extra_flags = " ".join(["-D " + af for af in settings.extra_adios2_flags]) cxx_standard = ( 20 if tuple(map(int, settings.kokkos_version.split("."))) >= (5, 0, 0) @@ -301,7 +301,32 @@ def InstallNt2pyScript(settings: Settings) -> str: "adios2_mpi": "mpi", }, "stellar": {"module_loads": []}, - "perlmutter": {"module_loads": []}, + "perlmutter": { + "module_loads": ["gpu/1.0"], + "kokkos_backend": "cuda", + "kokkos_arch": "AMPERE80", + "extra_kokkos_flags": [ + "Kokkos_ENABLE_IMPL_CUDA_MALLOC_ASYNC=OFF", + "CMAKE_CXX_COMPILER=CC", + ], + "extra_adios2_flags": [ + "LIBFABRIC_ROOT=/opt/cray/libfabric/1.15.2.0/", + "MPI_ROOT=/opt/cray/pe/craype/2.7.30", + ], + }, + "lumi": { + "module_loads": ["PrgEnv-cray", "cray-mpich", "craype-accel-amd-gfx90a", "rocm"], + "kokkos_backend": "hip", + "kokkos_arch": "AMD_GFX90A", + "extra_kokkos_flags": [ + "CMAKE_CXX_COMPILER=hipcc", + "AMDGPU_TARGETS=gfx90a", + ], + "extra_adios2_flags": [ + "CMAKE_CXX_COMPILER=CC", + "CMAKE_C_COMPILER=cc" + ] + }, "frontier": {"module_loads": []}, "aurora": {"module_loads": []}, } @@ -314,6 +339,8 @@ def apply_preset(s: Settings, name: str) -> None: s.apps["Kokkos"] = True s.apps["adios2"] = True s.apps["nt2py"] = False + s.write_modulefiles = True + s.overwrite = True s.module_loads = cluster_preset.get("module_loads", []) s.kokkos_backend = cluster_preset.get("kokkos_backend", "cpu") s.kokkos_arch = cluster_preset.get("kokkos_arch", "NATIVE") @@ -1068,24 +1095,14 @@ def toggle(k: str): def menu_cluster(self) -> Tuple[str, str, List[MenuItem]]: def choose(name: str): + print ("CALLING:", name) apply_preset(self.s, name) self.push("custom") return ( "cluster-specific", "pick a preset:", - [ - MenuItem("rusty", "apply preset", on_enter=lambda: choose("rusty")), - MenuItem("stellar", "apply preset", on_enter=lambda: choose("stellar")), - MenuItem( - "perlmutter", "apply preset", on_enter=lambda: choose("perlmutter") - ), - MenuItem( - "frontier", "apply preset", on_enter=lambda: choose("frontier") - ), - MenuItem("aurora", "apply preset", on_enter=lambda: choose("aurora")), - MenuItem("back", "", on_enter=self.pop), - ], + [MenuItem(cluster, "apply preset", on_enter=lambda c=cluster: choose(c)) for cluster in list(PRESETS.keys())] + [MenuItem("back", "", on_enter=self.pop)], ) def get_menu(self) -> Tuple[str, str, List[MenuItem]]: diff --git a/pgens/examples/README.md b/pgens/examples/README.md new file mode 100644 index 000000000..cc9ebdfc5 --- /dev/null +++ b/pgens/examples/README.md @@ -0,0 +1,7 @@ +# Examples + +Problem generations in this directory are just examples demonstrating how to use some of the features of the Entity. + +- `custom_emission`: simple example where two particles are initialized on gyrating trajectories probabilistically emitting photons while the total energy is conserved. + + https://github.com/user-attachments/assets/6c5c399a-bbab-4b93-be5f-35fed1959fcc diff --git a/pgens/examples/custom_emission/custom_emission.mp4 b/pgens/examples/custom_emission/custom_emission.mp4 new file mode 100644 index 000000000..2efaca769 --- /dev/null +++ b/pgens/examples/custom_emission/custom_emission.mp4 @@ -0,0 +1,3 @@ +version https://git-lfs.github.com/spec/v1 +oid sha256:320b85667465db5f2537a67f95462aeca96caa6cff2f80ae3fb57abfa09c95d4 +size 2896072 diff --git a/pgens/examples/custom_emission/custom_emission.py b/pgens/examples/custom_emission/custom_emission.py new file mode 100644 index 000000000..ae5d9b3be --- /dev/null +++ b/pgens/examples/custom_emission/custom_emission.py @@ -0,0 +1,51 @@ +import nt2 +import matplotlib.pyplot as plt +import numpy as np + +plt.rcParams["text.usetex"] = True + +data = nt2.Data("custom_emission") + + +def plot(t, data): + fig = plt.figure(figsize=(8, 4), dpi=150) + ax1 = fig.add_subplot(121) + prtls = data.particles.sel(t=slice(t - 0.5, t)).load() + prtls.plot.scatter( + ax=ax1, + x="x", + y="y", + color=["r" if sp == 1 else "b" for sp in prtls["sp"]], + ec=None, + alpha=(1 - (t - prtls["t"]) / 0.5) ** 2, + s=(1 - (t - prtls["t"]) / 0.5) * 5, + ) + ax1.set(xlabel=r"$x$", ylabel=r"$y$", xlim=(-1, 1), ylim=(-1, 1), aspect=1) + + ax2 = fig.add_subplot(122) + prtls1 = data.particles.sel(sp=1, t=slice(None, t)).load() + prtls2 = data.particles.sel(sp=2, t=slice(None, t)).load() + e1 = ( + prtls1.assign( + e=np.sqrt(1 + prtls1["ux"] ** 2 + prtls1["uy"] ** 2 + prtls1["uz"] ** 2) + ) + .groupby("t", as_index=False)["e"] + .sum() + ) + e2 = ( + prtls2.assign( + e=np.sqrt(prtls2["ux"] ** 2 + prtls2["uy"] ** 2 + prtls2["uz"] ** 2) + ) + .groupby("t", as_index=False)["e"] + .sum() + ) + e = e1.merge(e2, on="t", how="outer", suffixes=("_1", "_2")).fillna(0) + ax2.plot(e.t, e.e_1, label="emitters", c="r") + ax2.plot(e.t, e.e_2, label="emitted", c="b") + ax2.plot(e.t, e.e_1 + e.e_2, label="total", c="k") + ax2.set(xlabel=r"$t$", ylabel=r"total energy", xlim=(0, 5), ylim=(0, 3)) + ax2.axvline(t, color="gray", ls="--") + ax2.legend(loc="center left") + + +data.makeMovie(plot, framerate=30) diff --git a/pgens/examples/custom_emission/custom_emission.toml b/pgens/examples/custom_emission/custom_emission.toml new file mode 100644 index 000000000..fd4cd135e --- /dev/null +++ b/pgens/examples/custom_emission/custom_emission.toml @@ -0,0 +1,68 @@ +[simulation] + name = "custom_emission" + engine = "srpic" + runtime = 5.0 + +[grid] + resolution = [64, 64] + extent = [[-1.0, 1.0], [-1.0, 1.0]] + + [grid.metric] + metric = "minkowski" + + [grid.boundaries] + fields = [["PERIODIC"], ["PERIODIC"]] + particles = [["PERIODIC"], ["PERIODIC"]] + +[scales] + larmor0 = 0.25 + skindepth0 = 1.0 + +[algorithms] + + [algorithms.deposit] + enable = false + + [algorithms.timestep] + CFL = 0.5 + +[particles] + ppc0 = 1.0 + + [[particles.species]] + label = "p" + mass = 1.0 + charge = 1.0 + maxnpart = 1e2 + emission = "custom" + + [[particles.species]] + label = "ph" + mass = 0.0 + charge = 0.0 + maxnpart = 1e4 + +[setup] + x1_arr = [0.0, 0.0] + x2_arr = [0.0, 0.0] + x3_arr = [0.0, 0.0] + ux1_arr = [0.0, 0.0] + ux2_arr = [1.0, -1.0] + ux3_arr = [0.0, 0.0] + emission_probability = 0.05 + +[output] + interval = 1 + + [output.fields] + quantities = ["N_1", "N_2", "B"] + + [output.particles] + enable = true + stride = 1 + + [output.spectra] + enable = false + +[checkpoint] + keep = 0 diff --git a/pgens/examples/custom_emission/pgen.hpp b/pgens/examples/custom_emission/pgen.hpp new file mode 100644 index 000000000..e1d200bab --- /dev/null +++ b/pgens/examples/custom_emission/pgen.hpp @@ -0,0 +1,240 @@ +#ifndef PROBLEM_GENERATOR_H +#define PROBLEM_GENERATOR_H + +#include "enums.h" +#include "global.h" + +#include "arch/kokkos_aliases.h" + +#include "archetypes/particle_injector.h" +#include "archetypes/problem_generator.h" +#include "archetypes/traits.h" +#include "framework/domain/metadomain.h" +#include "kernels/emission/traits.h" +#include "kernels/injectors.hpp" + +#include + +#include +#include +#include + +namespace user { + using namespace ntt; + + template + struct InitFields { + Inline auto bx3(const coord_t&) const -> real_t { + return ONE; + } + }; + + template + struct RandomEmission { + struct Payload { + real_t photon_energy { ZERO }; + }; + + random_number_pool_t random_pool; + const real_t probability; + + array_t inj_idx { "inj_idx" }; + const npart_t inj_offset; + + array_t inj_i1, inj_i2, inj_i3; + array_t inj_dx1, inj_dx2, inj_dx3; + array_t inj_ux1, inj_ux2, inj_ux3; + array_t inj_phi; + array_t inj_weight; + array_t inj_tag; + array_t inj_pld_i; + + RandomEmission(random_number_pool_t& random_pool, + real_t probability, + npart_t inj_offset, + array_t& inj_i1, + array_t& inj_i2, + array_t& inj_i3, + array_t& inj_dx1, + array_t& inj_dx2, + array_t& inj_dx3, + array_t& inj_ux1, + array_t& inj_ux2, + array_t& inj_ux3, + array_t& inj_phi, + array_t& inj_weight, + array_t& inj_tag, + array_t& inj_pld_i) + : random_pool { random_pool } + , probability { probability } + , inj_offset { inj_offset } + , inj_i1 { inj_i1 } + , inj_i2 { inj_i2 } + , inj_i3 { inj_i3 } + , inj_dx1 { inj_dx1 } + , inj_dx2 { inj_dx2 } + , inj_dx3 { inj_dx3 } + , inj_ux1 { inj_ux1 } + , inj_ux2 { inj_ux2 } + , inj_ux3 { inj_ux3 } + , inj_phi { inj_phi } + , inj_weight { inj_weight } + , inj_tag { inj_tag } + , inj_pld_i { inj_pld_i } {} + + Inline auto shouldEmit(const coord_t&, + const coord_t&, + const vec_t& u_Ph, + const vec_t&, + const vec_t&, + vec_t& delta_u_Ph, + Payload& payload) const -> Kokkos::pair { + auto generator = random_pool.get_state(); + const auto rnd = Random(generator); + random_pool.free_state(generator); + if (rnd < probability) { + delta_u_Ph[0] = -0.1 * u_Ph[0]; + delta_u_Ph[1] = -0.1 * u_Ph[1]; + delta_u_Ph[2] = -0.1 * u_Ph[2]; + + const auto uSqr = NORM_SQR(u_Ph[0], u_Ph[1], u_Ph[2]); + const auto gammaSqr = ONE + uSqr; + const auto delta_uSqr = NORM_SQR(delta_u_Ph[0], + delta_u_Ph[1], + delta_u_Ph[2]); + const auto u_dot_delta_u = DOT(u_Ph[0], + u_Ph[1], + u_Ph[2], + delta_u_Ph[0], + delta_u_Ph[1], + delta_u_Ph[2]); + payload.photon_energy = math::sqrt(gammaSqr) * + (math::sqrt(ONE + delta_uSqr / gammaSqr + + TWO * u_dot_delta_u / gammaSqr) - + ONE); + return { true, true }; + } + return { false, false }; + } + + Inline auto emit(const tuple_t& xi_Cd, + const tuple_t& dxi_Cd, + const vec_t& direction, + real_t, + real_t, + const Payload& payload) const -> void { + const auto inj_index = Kokkos::atomic_fetch_add(&inj_idx(), 1); + kernel::InjectParticle( + inj_offset + inj_index, + inj_i1, + inj_i2, + inj_i3, + inj_dx1, + inj_dx2, + inj_dx3, + inj_ux1, + inj_ux2, + inj_ux3, + inj_phi, + inj_weight, + inj_tag, + inj_pld_i, + xi_Cd, + dxi_Cd, + { payload.photon_energy * direction[0], + payload.photon_energy * direction[1], + payload.photon_energy * direction[2] }); + } + + auto emitted_species_indices() const -> std::vector { + return { 2u }; + } + + auto numbers_injected() const -> std::vector { + auto inj_idx_h = Kokkos::create_mirror_view(inj_idx); + Kokkos::deep_copy(inj_idx_h, inj_idx); + return { inj_idx_h() }; + } + }; + + template + struct PGen : public arch::ProblemGenerator { + static constexpr auto engines { + arch::traits::pgen::compatible_with::value + }; + static constexpr auto metrics { + arch::traits::pgen::compatible_with::value + }; + static constexpr auto dimensions { + arch::traits::pgen::compatible_with::value + }; + + using arch::ProblemGenerator::D; + using arch::ProblemGenerator::C; + using arch::ProblemGenerator::params; + + const Metadomain& metadomain; + + const real_t emission_probability; + + InitFields init_flds {}; + + inline PGen(const SimulationParams& p, const Metadomain& metadomain) + : arch::ProblemGenerator { p } + , metadomain { metadomain } + , emission_probability { params.template get( + "setup.emission_probability") } { + static_assert(kernel::traits::emission::IsValid, M>, "RandomEmission does not satisfy the requirements of an emission policy"); + } + + inline auto EmissionPolicy(simtime_t, + spidx_t, + Domain& domain) const -> RandomEmission { + return RandomEmission { + domain.random_pool(), emission_probability, + domain.species[1].npart(), domain.species[1].i1, + domain.species[1].i2, domain.species[1].i3, + domain.species[1].dx1, domain.species[1].dx2, + domain.species[1].dx3, domain.species[1].ux1, + domain.species[1].ux2, domain.species[1].ux3, + domain.species[1].phi, domain.species[1].weight, + domain.species[1].tag, domain.species[1].pld_i + }; + } + + inline void InitPrtls(Domain& domain) { + const auto empty = std::vector {}; + const auto x1_arr = params.template get>( + "setup.x1_arr", + empty); + const auto x2_arr = params.template get>( + "setup.x2_arr", + empty); + const auto x3_arr = params.template get>( + "setup.x3_arr", + empty); + const auto ux1_arr = params.template get>( + "setup.ux1_arr", + empty); + const auto ux2_arr = params.template get>( + "setup.ux2_arr", + empty); + const auto ux3_arr = params.template get>( + "setup.ux3_arr", + empty); + + std::map> data_arr { + { "x1", x1_arr }, + { "x2", x2_arr }, + { "x3", x3_arr }, + { "ux1", ux1_arr }, + { "ux2", ux2_arr }, + { "ux3", ux3_arr } + }; + arch::InjectGlobally(metadomain, domain, 1u, data_arr); + } + }; + +} // namespace user + +#endif diff --git a/pgens/reconnection/pgen.hpp b/pgens/reconnection/pgen.hpp index bc6f7c75e..4711aec22 100644 --- a/pgens/reconnection/pgen.hpp +++ b/pgens/reconnection/pgen.hpp @@ -14,7 +14,6 @@ #include "archetypes/traits.h" #include "archetypes/utils.h" #include "framework/domain/metadomain.h" - #include "kernels/particle_moments.hpp" namespace user { diff --git a/src/archetypes/energy_dist.h b/src/archetypes/energy_dist.h index 230e4a932..a493d7d85 100644 --- a/src/archetypes/energy_dist.h +++ b/src/archetypes/energy_dist.h @@ -250,7 +250,7 @@ namespace arch { } } - Inline void operator()(const coord_t& x_Code, vec_t& v) const { + Inline void operator()(const coord_t&, vec_t& v) const { if (cmp::AlmostZero(temperature)) { v[0] = ZERO; v[1] = ZERO; diff --git a/src/archetypes/field_setter.h b/src/archetypes/field_setter.h index d4fab7a3b..a9e45d2d6 100644 --- a/src/archetypes/field_setter.h +++ b/src/archetypes/field_setter.h @@ -36,23 +36,23 @@ namespace arch { template requires metric::traits::HasD && - ((S == SimEngine::SRPIC && metric::traits::HasConvert && + (((S == SimEngine::SRPIC) && metric::traits::HasConvert && metric::traits::HasTransform_i) || - (S == SimEngine::GRPIC && metric::traits::HasConvert_i)) && - (S == SimEngine::SRPIC && - (::traits::fieldsetter::HasEx1 || - ::traits::fieldsetter::HasEx2 || - ::traits::fieldsetter::HasEx3 || - ::traits::fieldsetter::HasBx1 || - ::traits::fieldsetter::HasBx2 || - ::traits::fieldsetter::HasBx3) || - (S == SimEngine::GRPIC && - (::traits::fieldsetter::HasDx1 && - ::traits::fieldsetter::HasDx2 && - ::traits::fieldsetter::HasDx3) || - (::traits::fieldsetter::HasBx1 && - ::traits::fieldsetter::HasBx2 && - ::traits::fieldsetter::HasBx3))) + ((S == SimEngine::GRPIC) && metric::traits::HasConvert_i)) && + (((S == SimEngine::SRPIC) && + (::traits::fieldsetter::HasEx1 || + ::traits::fieldsetter::HasEx2 || + ::traits::fieldsetter::HasEx3 || + ::traits::fieldsetter::HasBx1 || + ::traits::fieldsetter::HasBx2 || + ::traits::fieldsetter::HasBx3)) || + ((S == SimEngine::GRPIC) && + ((::traits::fieldsetter::HasDx1 && + ::traits::fieldsetter::HasDx2 && + ::traits::fieldsetter::HasDx3) || + (::traits::fieldsetter::HasBx1 && + ::traits::fieldsetter::HasBx2 && + ::traits::fieldsetter::HasBx3)))) class SetEMFields_kernel { static constexpr Dimension D = M::Dim; diff --git a/src/archetypes/particle_injector.h b/src/archetypes/particle_injector.h index d41996393..e1ca1b7ca 100644 --- a/src/archetypes/particle_injector.h +++ b/src/archetypes/particle_injector.h @@ -26,7 +26,6 @@ #include "framework/domain/domain.h" #include "framework/domain/metadomain.h" - #include "kernels/injectors.hpp" #include diff --git a/src/archetypes/tests/pgen.cpp b/src/archetypes/tests/pgen.cpp index a0e619901..ad5174894 100644 --- a/src/archetypes/tests/pgen.cpp +++ b/src/archetypes/tests/pgen.cpp @@ -27,6 +27,14 @@ struct ExtForce { Inline auto fx1(spidx_t, simtime_t, const coord_t&) const -> real_t { return ZERO; } + + Inline auto ex1(spidx_t, simtime_t, const coord_t&) const -> real_t { + return ZERO; + } + + Inline auto bx3(spidx_t, simtime_t, const coord_t&) const -> real_t { + return ZERO; + } }; template @@ -142,6 +150,22 @@ auto main(int argc, char* argv[]) -> int { Domain>>) { throw std::runtime_error("CustomPgen should have CustomStat"); } + if constexpr ( + not ::traits::external::HasFx1) { + throw std::runtime_error("CustomPgen's ext_fields should have fx1"); + } + if constexpr ( + not ::traits::external::HasEx1) { + throw std::runtime_error("CustomPgen's ext_fields should have ex1"); + } + if constexpr ( + ::traits::external::HasBx1) { + throw std::runtime_error("CustomPgen's ext_fields should not have bx1"); + } + if constexpr ( + not ::traits::external::HasBx3) { + throw std::runtime_error("CustomPgen's ext_current should have bx3"); + } } catch (std::exception& e) { std::cerr << e.what() << std::endl; diff --git a/src/archetypes/traits.h b/src/archetypes/traits.h index 522d3f24b..5d11ba3ab 100644 --- a/src/archetypes/traits.h +++ b/src/archetypes/traits.h @@ -77,6 +77,14 @@ namespace arch { template concept HasInitFlds = requires(const PG& pgen) { pgen.init_flds; }; + template + concept HasEmissionPolicy = requires(const PG& pgen, + simtime_t time, + spidx_t sp, + D& domain) { + pgen.EmissionPolicy(time, sp, domain); + }; + template concept HasInitPrtls = requires(PG& pgen, D& domain) { { pgen.InitPrtls(domain) } -> std::same_as; diff --git a/src/engines/grpic.hpp b/src/engines/grpic.hpp index d204ad74b..7952153ba 100644 --- a/src/engines/grpic.hpp +++ b/src/engines/grpic.hpp @@ -22,8 +22,6 @@ #include "framework/domain/domain.h" #include "framework/parameters/parameters.h" - -#include "engines/engine.hpp" #include "kernels/ampere_gr.hpp" #include "kernels/aux_fields_gr.hpp" #include "kernels/currents_deposit.hpp" @@ -31,6 +29,8 @@ #include "kernels/faraday_gr.hpp" #include "kernels/fields_bcs.hpp" #include "kernels/particle_pusher_gr.hpp" + +#include "engines/engine.hpp" #include "pgen.hpp" #include diff --git a/src/engines/srpic/currents.h b/src/engines/srpic/currents.h index 479ba0bc6..5733875b4 100644 --- a/src/engines/srpic/currents.h +++ b/src/engines/srpic/currents.h @@ -13,7 +13,6 @@ #include "engines/srpic/utils.h" #include "framework/domain/domain.h" #include "framework/domain/metadomain.h" - #include "kernels/currents_deposit.hpp" #include "kernels/digital_filter.hpp" diff --git a/src/engines/srpic/fields_bcs.h b/src/engines/srpic/fields_bcs.h index 07792002b..cf4c44dc8 100644 --- a/src/engines/srpic/fields_bcs.h +++ b/src/engines/srpic/fields_bcs.h @@ -13,7 +13,6 @@ #include "engines/srpic/utils.h" #include "framework/domain/domain.h" #include "framework/parameters/parameters.h" - #include "kernels/fields_bcs.hpp" namespace ntt { diff --git a/src/engines/srpic/fieldsolvers.h b/src/engines/srpic/fieldsolvers.h index 04bca2744..bc68cee3a 100644 --- a/src/engines/srpic/fieldsolvers.h +++ b/src/engines/srpic/fieldsolvers.h @@ -12,7 +12,6 @@ #include "engines/srpic/utils.h" #include "framework/domain/domain.h" #include "framework/parameters/parameters.h" - #include "kernels/ampere_mink.hpp" #include "kernels/ampere_sr.hpp" #include "kernels/faraday_mink.hpp" diff --git a/src/engines/srpic/particle_pusher.h b/src/engines/srpic/particle_pusher.h index 3ba120ef3..2a0848de4 100644 --- a/src/engines/srpic/particle_pusher.h +++ b/src/engines/srpic/particle_pusher.h @@ -16,16 +16,16 @@ #include "framework/domain/domain.h" #include "framework/domain/grid.h" #include "framework/parameters/parameters.h" - #include "kernels/emission/compton.hpp" #include "kernels/emission/emission.hpp" #include "kernels/emission/synchrotron.hpp" +#include "kernels/emission/traits.h" #include "kernels/particle_pusher_sr.hpp" namespace ntt { namespace srpic { - template + template void CallPusher(Domain& domain, const SimulationParams& params, const kernel::sr::PusherParams& pusher_params, @@ -34,6 +34,7 @@ namespace ntt { const range_t& range, const ndfield_t& EB, const M& metric, + const PG& pgen, const F& external_fields) { if (emission_policy_flag == EmissionType::NONE) { const auto no_emission = kernel::NoEmissionPolicy_t {}; @@ -62,13 +63,17 @@ namespace ntt { HERE); const auto emission_policy = kernel::emission::Synchrotron( emitted_species, + photon_species, pusher_params.mass, pusher_params.charge, pusher_params.radiative_drag_flags, - pusher_params.dt, domain.index(), params, domain.random_pool()); + static_assert( + kernel::traits::emission::IsValid, M>, + "Synchrotron emission policy does not satisfy the required " + "interface"); Kokkos::parallel_for( "ParticlePusher", range, @@ -79,11 +84,14 @@ namespace ntt { metric, external_fields, emission_policy)); - const auto n_inj = emission_policy.number_injected(); + const auto n_inj = emission_policy.numbers_injected(); + raise::ErrorIf(n_inj.size() != 1, + "Synchrotron emission should only inject one species", + HERE); domain.species[photon_species - 1].set_npart( - emitted_species.npart() + n_inj); + emitted_species.npart() + n_inj[0]); domain.species[photon_species - 1].set_counter( - emitted_species.counter() + n_inj); + emitted_species.counter() + n_inj[0]); } else if (emission_policy_flag == EmissionType::COMPTON) { const auto photon_species = params.get( "radiation.emission.compton.photon_species"); @@ -99,13 +107,16 @@ namespace ntt { HERE); const auto emission_policy = kernel::emission::Compton( emitted_species, + photon_species, pusher_params.mass, pusher_params.charge, pusher_params.radiative_drag_flags, - pusher_params.dt, domain.index(), params, domain.random_pool()); + static_assert( + kernel::traits::emission::IsValid, M>, + "Compton emission policy does not satisfy the required interface"); Kokkos::parallel_for( "ParticlePusher", range, @@ -116,11 +127,58 @@ namespace ntt { metric, external_fields, emission_policy)); - const auto n_inj = emission_policy.number_injected(); + const auto n_inj = emission_policy.numbers_injected(); + raise::ErrorIf(n_inj.size() != 1, + "Compton emission should only inject one species", + HERE); domain.species[photon_species - 1].set_npart( - emitted_species.npart() + n_inj); + emitted_species.npart() + n_inj[0]); domain.species[photon_species - 1].set_counter( - emitted_species.counter() + n_inj); + emitted_species.counter() + n_inj[0]); + } else if (emission_policy_flag == EmissionType::CUSTOM) { + if constexpr ( + arch::traits::pgen::HasEmissionPolicy) { + const auto emission_policy = pgen.EmissionPolicy(pusher_params.time, + pusher_params.species_index, + domain); + static_assert( + kernel::traits::emission::IsValid, + "Custom emission policy does not satisfy the required " + "interface"); + Kokkos::parallel_for( + "ParticlePusher", + range, + kernel::sr::Pusher_kernel( + pusher_params, + pusher_arrays, + EB, + metric, + external_fields, + emission_policy)); + const auto emitted_species = emission_policy.emitted_species_indices(); + const auto n_inj = emission_policy.numbers_injected(); + raise::ErrorIf(emitted_species.size() != n_inj.size(), + "Emission policy emitted_species_indices and " + "numbers_injected must have the same size", + HERE); + for (auto i = 0u; i < emitted_species.size(); ++i) { + const auto sp_idx = emitted_species[i]; + raise::ErrorIf(sp_idx > domain.species.size(), + "Invalid emitted species index from custom " + "emission policy", + HERE); + domain.species[sp_idx - 1].set_npart( + domain.species[sp_idx - 1].npart() + n_inj[i]); + domain.species[sp_idx - 1].set_counter( + domain.species[sp_idx - 1].counter() + n_inj[i]); + } + } else { + raise::Error("Custom emission policy flag is set but problem " + "generator does not define an emission policy", + HERE); + } + } else { + raise::Error("Unrecognized emission policy flag", HERE); } } @@ -184,6 +242,7 @@ namespace ntt { HERE); kernel::sr::PusherParams pusher_params {}; + pusher_params.species_index = species.index(); pusher_params.pusher_flags = species.pusher(); pusher_params.radiative_drag_flags = species.radiative_drag_flags(); pusher_params.mass = species.mass(); @@ -226,26 +285,7 @@ namespace ntt { params.template get("radiation.drag.compton.gamma_rad")); } - kernel::sr::PusherArrays pusher_arrays {}; - pusher_arrays.sp = species.index(); - pusher_arrays.i1 = species.i1; - pusher_arrays.i2 = species.i2; - pusher_arrays.i3 = species.i3; - pusher_arrays.i1_prev = species.i1_prev; - pusher_arrays.i2_prev = species.i2_prev; - pusher_arrays.i3_prev = species.i3_prev; - pusher_arrays.dx1 = species.dx1; - pusher_arrays.dx2 = species.dx2; - pusher_arrays.dx3 = species.dx3; - pusher_arrays.dx1_prev = species.dx1_prev; - pusher_arrays.dx2_prev = species.dx2_prev; - pusher_arrays.dx3_prev = species.dx3_prev; - pusher_arrays.ux1 = species.ux1; - pusher_arrays.ux2 = species.ux2; - pusher_arrays.ux3 = species.ux3; - pusher_arrays.phi = species.phi; - pusher_arrays.weight = species.weight; - pusher_arrays.tag = species.tag; + auto pusher_arrays = species.PusherKernelArrays(); // toggle to indicate whether pgen defines the external force bool has_extfields = false; @@ -261,7 +301,7 @@ namespace ntt { } if (not has_atmosphere and not has_extfields) { - CallPusher( + CallPusher( domain, params, pusher_params, @@ -270,9 +310,10 @@ namespace ntt { species.rangeActiveParticles(), domain.fields.em, domain.mesh.metric, + pgen, kernel::sr::NoField_t {}); } else if (has_atmosphere and not has_extfields) { - CallPusher( + CallPusher( domain, params, pusher_params, @@ -281,10 +322,11 @@ namespace ntt { species.rangeActiveParticles(), domain.fields.em, domain.mesh.metric, + pgen, kernel::sr::NoField_t {}); } else if (not has_atmosphere and has_extfields) { if constexpr (arch::traits::pgen::HasExtFields) { - CallPusher( + CallPusher( domain, params, pusher_params, @@ -293,13 +335,14 @@ namespace ntt { species.rangeActiveParticles(), domain.fields.em, domain.mesh.metric, + pgen, pgen.ext_fields); } else { raise::Error("External fields not implemented", HERE); } } else { // has_atmosphere and has_extforce if constexpr (arch::traits::pgen::HasExtFields) { - CallPusher( + CallPusher( domain, params, pusher_params, @@ -308,6 +351,7 @@ namespace ntt { species.rangeActiveParticles(), domain.fields.em, domain.mesh.metric, + pgen, pgen.ext_fields); } else { raise::Error("External fields not implemented", HERE); diff --git a/src/engines/srpic/particles_bcs.h b/src/engines/srpic/particles_bcs.h index 1f921f033..5d444385f 100644 --- a/src/engines/srpic/particles_bcs.h +++ b/src/engines/srpic/particles_bcs.h @@ -16,7 +16,6 @@ #include "framework/domain/domain.h" #include "framework/domain/metadomain.h" #include "framework/parameters/parameters.h" - #include "kernels/particle_moments.hpp" namespace ntt { diff --git a/src/engines/srpic/srpic.hpp b/src/engines/srpic/srpic.hpp index 3d39951f4..246f7f056 100644 --- a/src/engines/srpic/srpic.hpp +++ b/src/engines/srpic/srpic.hpp @@ -69,8 +69,6 @@ namespace ntt { "algorithms.fieldsolver.enable"); const auto deposit_enabled = m_params.template get( "algorithms.deposit.enable"); - const auto clear_interval = m_params.template get( - "particles.clear_interval"); if (step == 0) { // communicate fields and apply BCs on the first timestep diff --git a/src/framework/containers/particles.cpp b/src/framework/containers/particles.cpp index 31b08a40c..7682718a3 100644 --- a/src/framework/containers/particles.cpp +++ b/src/framework/containers/particles.cpp @@ -84,6 +84,31 @@ namespace ntt { } } + template + auto Particles::PusherKernelArrays() -> kernel::sr::PusherArrays { + kernel::sr::PusherArrays pusher_arrays {}; + pusher_arrays.sp = index(); + pusher_arrays.i1 = i1; + pusher_arrays.i2 = i2; + pusher_arrays.i3 = i3; + pusher_arrays.i1_prev = i1_prev; + pusher_arrays.i2_prev = i2_prev; + pusher_arrays.i3_prev = i3_prev; + pusher_arrays.dx1 = dx1; + pusher_arrays.dx2 = dx2; + pusher_arrays.dx3 = dx3; + pusher_arrays.dx1_prev = dx1_prev; + pusher_arrays.dx2_prev = dx2_prev; + pusher_arrays.dx3_prev = dx3_prev; + pusher_arrays.ux1 = ux1; + pusher_arrays.ux2 = ux2; + pusher_arrays.ux3 = ux3; + pusher_arrays.phi = phi; + pusher_arrays.weight = weight; + pusher_arrays.tag = tag; + return pusher_arrays; + } + template struct Particles; template struct Particles; template struct Particles; diff --git a/src/framework/containers/particles.h b/src/framework/containers/particles.h index c7ba5acc5..c3ee1c1b5 100644 --- a/src/framework/containers/particles.h +++ b/src/framework/containers/particles.h @@ -25,6 +25,7 @@ #include "framework/containers/species.h" #include "framework/domain/grid.h" +#include "kernels/particle_pusher_sr.hpp" #include @@ -270,6 +271,12 @@ namespace ntt { */ void SyncHostDevice(); + /** + * @brief Get the arrays required for the particle pusher kernel + * @returns The struct of arrays for the particle pusher kernel + */ + auto PusherKernelArrays() -> kernel::sr::PusherArrays; + #if defined(MPI_ENABLED) /** * @brief Communicate particles across neighboring meshblocks diff --git a/src/framework/containers/particles_comm.cpp b/src/framework/containers/particles_comm.cpp index 4d6d67118..1cf17efef 100644 --- a/src/framework/containers/particles_comm.cpp +++ b/src/framework/containers/particles_comm.cpp @@ -10,7 +10,6 @@ #include "utils/log.h" #include "framework/containers/particles.h" - #include "kernels/comm.hpp" #include diff --git a/src/framework/containers/particles_io.cpp b/src/framework/containers/particles_io.cpp index d35ec59f5..fb72ccc5f 100644 --- a/src/framework/containers/particles_io.cpp +++ b/src/framework/containers/particles_io.cpp @@ -7,11 +7,10 @@ #include "framework/containers/particles.h" #include "framework/specialization_registry.h" +#include "kernels/prtls_to_phys.hpp" #include "output/utils/readers.h" #include "output/utils/writers.h" -#include "kernels/prtls_to_phys.hpp" - #include #include @@ -123,7 +122,10 @@ namespace ntt { npart_t nout_offset = 0; npart_t nout_total = nout; -#if defined(MPI_ENABLED) +#if !defined(MPI_ENABLED) + (void)domains_total; + (void)domains_offset; +#else auto nout_total_vec = std::vector(domains_total); MPI_Allgather(&nout, 1, @@ -133,7 +135,7 @@ namespace ntt { mpi::get_type(), MPI_COMM_WORLD); nout_total = 0; - for (auto r = 0; r < domains_total; ++r) { + for (auto r = 0u; r < domains_total; ++r) { if (r < domains_offset) { nout_offset += nout_total_vec[r]; } @@ -429,7 +431,9 @@ namespace ntt { domains_offset); set_npart(npart_read); -#if defined(MPI_ENABLED) +#if !defined(MPI_ENABLED) + (void)domains_total; +#else { const auto npart_send = npart(); std::vector glob_nparts(domains_total); @@ -616,7 +620,7 @@ namespace ntt { mpi::get_type(), MPI_COMM_WORLD); npart_total = 0u; - for (auto r = 0; r < domains_total; ++r) { + for (auto r = 0u; r < domains_total; ++r) { if (r < domains_offset) { npart_offset += glob_nparts[r]; } diff --git a/src/framework/containers/species.h b/src/framework/containers/species.h index f3e4f369d..5fcec5878 100644 --- a/src/framework/containers/species.h +++ b/src/framework/containers/species.h @@ -32,10 +32,10 @@ namespace ntt { const float m_charge; // Max number of allocated particles for the species npart_t m_maxnpart; + // Clearing interval for the species (0 means no clearing) + const timestep_t m_clearing_interval; // Spatial sorting interval for the species (0 means no sorting) const timestep_t m_spatial_sorting_interval; - // Spatial sorting interval for the species (0 means no clearing) - const timestep_t m_clearing_interval; // Pusher assigned for the species const ParticlePusherFlags m_particle_pusher_flags; diff --git a/src/framework/domain/metadomain_chckpt.cpp b/src/framework/domain/metadomain_chckpt.cpp index 693cafc1a..bd6b87cfd 100644 --- a/src/framework/domain/metadomain_chckpt.cpp +++ b/src/framework/domain/metadomain_chckpt.cpp @@ -1,5 +1,3 @@ -#include "output/checkpoint.h" - #include "enums.h" #include "global.h" @@ -10,6 +8,7 @@ #include "framework/domain/metadomain.h" #include "framework/parameters/parameters.h" #include "framework/specialization_registry.h" +#include "output/checkpoint.h" namespace ntt { diff --git a/src/framework/domain/metadomain_io.cpp b/src/framework/domain/metadomain_io.cpp index 1d29c4102..b9f351a0f 100644 --- a/src/framework/domain/metadomain_io.cpp +++ b/src/framework/domain/metadomain_io.cpp @@ -11,7 +11,6 @@ #include "framework/domain/metadomain.h" #include "framework/parameters/parameters.h" #include "framework/specialization_registry.h" - #include "kernels/divergences.hpp" #include "kernels/fields_to_phys.hpp" #include "kernels/particle_moments.hpp" @@ -81,10 +80,6 @@ namespace ntt { "output.particles.species"); g_writer.defineFieldOutputs(S, all_fields_to_write); - Dimension dim = M::PrtlDim; - if constexpr (M::CoordType != Coord::Cart) { - dim = Dim::_3D; - } g_writer.clearSpeciesIndex(); for (const auto& s : species_to_write) { g_writer.addSpeciesIndex(s); @@ -281,7 +276,7 @@ namespace ntt { for (auto nr1 { 0u }; nr1 < nranks_x1; ++nr1) { const auto rank_send = rank_send_pre + nr1; const auto rank_recv = rank_recv_pre + nr1; - if (local_domain->mpi_rank() == rank_send) { + if (static_cast(local_domain->mpi_rank()) == rank_send) { array_t aphi_r { "Aphi_r", nx1 }; Kokkos::deep_copy( aphi_r, @@ -295,7 +290,8 @@ namespace ntt { rank_recv, 0, MPI_COMM_WORLD); - } else if (local_domain->mpi_rank() == rank_recv) { + } else if (static_cast(local_domain->mpi_rank()) == + rank_recv) { array_t aphi_r { "Aphi_r", nx1 }; MPI_Recv(aphi_r.data(), nx1, diff --git a/src/framework/domain/metadomain_sort.cpp b/src/framework/domain/metadomain_sort.cpp index 68ee28cfa..c529ca738 100644 --- a/src/framework/domain/metadomain_sort.cpp +++ b/src/framework/domain/metadomain_sort.cpp @@ -11,9 +11,9 @@ namespace ntt { template requires IsCompatibleWithMetadomain void Metadomain::SortParticles(simtime_t, - timestep_t step, - const SimulationParams& params, - Domain& domain) const { + timestep_t step, + const SimulationParams&, + Domain& domain) const { for (auto& species : domain.species) { const auto clearing_interval = species.clearing_interval(); if ((clearing_interval > 0u) and (step % clearing_interval == 0u) and diff --git a/src/framework/domain/metadomain_stats.cpp b/src/framework/domain/metadomain_stats.cpp index 811b35111..21176a836 100644 --- a/src/framework/domain/metadomain_stats.cpp +++ b/src/framework/domain/metadomain_stats.cpp @@ -11,7 +11,6 @@ #include "framework/domain/metadomain.h" #include "framework/parameters/parameters.h" #include "framework/specialization_registry.h" - #include "kernels/reduced_stats.hpp" #include @@ -190,8 +189,8 @@ namespace ntt { timestep_t finished_step, simtime_t current_time, simtime_t finished_time, - std::function&)> - CustomStat) -> bool { + std::function&)> CustomStat) + -> bool { if (not(params.template get("output.stats.enable") and g_stats_writer.shouldWrite(finished_step, finished_time))) { return false; @@ -281,17 +280,18 @@ namespace ntt { return true; } -#define METADOMAIN_STATS(S, M, D) \ - template void Metadomain>::InitStatsWriter(const SimulationParams&, \ - bool); \ - template auto Metadomain>::WriteStats( \ - const SimulationParams&, \ - timestep_t, \ - timestep_t, \ - simtime_t, \ - simtime_t, \ - std::function< \ - real_t(const std::string&, timestep_t, simtime_t, const Domain>&)>) -> bool; +#define METADOMAIN_STATS(S, M, D) \ + template void Metadomain>::InitStatsWriter(const SimulationParams&, \ + bool); \ + template auto Metadomain>::WriteStats( \ + const SimulationParams&, \ + timestep_t, \ + timestep_t, \ + simtime_t, \ + simtime_t, \ + std::function< \ + real_t(const std::string&, timestep_t, simtime_t, const Domain>&)>) \ + -> bool; NTT_FOREACH_SPECIALIZATION(METADOMAIN_STATS) diff --git a/src/framework/parameters/algorithms.cpp b/src/framework/parameters/algorithms.cpp index 944c323f8..856d05fc7 100644 --- a/src/framework/parameters/algorithms.cpp +++ b/src/framework/parameters/algorithms.cpp @@ -4,10 +4,11 @@ #include "global.h" #include "utils/numeric.h" -#include #include "framework/parameters/parameters.h" +#include + namespace ntt { namespace params { diff --git a/src/framework/parameters/algorithms.h b/src/framework/parameters/algorithms.h index fb41d731c..a496cb7b6 100644 --- a/src/framework/parameters/algorithms.h +++ b/src/framework/parameters/algorithms.h @@ -14,10 +14,10 @@ #include "global.h" -#include - #include "framework/parameters/parameters.h" +#include + #include #include diff --git a/src/framework/parameters/extra.h b/src/framework/parameters/extra.h index cec6f619b..cf5c0849c 100644 --- a/src/framework/parameters/extra.h +++ b/src/framework/parameters/extra.h @@ -14,10 +14,10 @@ #include "global.h" -#include - #include "framework/parameters/parameters.h" +#include + #include #include diff --git a/src/framework/parameters/grid.cpp b/src/framework/parameters/grid.cpp index 78072680f..9e5d39df5 100644 --- a/src/framework/parameters/grid.cpp +++ b/src/framework/parameters/grid.cpp @@ -6,7 +6,6 @@ #include "utils/error.h" #include "utils/formatting.h" #include "utils/numeric.h" -#include #include "metrics/kerr_schild.h" #include "metrics/kerr_schild_0.h" @@ -17,6 +16,8 @@ #include "framework/parameters/parameters.h" +#include + #include #include #include diff --git a/src/framework/parameters/grid.h b/src/framework/parameters/grid.h index bc73063d7..978b7f329 100644 --- a/src/framework/parameters/grid.h +++ b/src/framework/parameters/grid.h @@ -15,10 +15,10 @@ #include "enums.h" #include "global.h" -#include - #include "framework/parameters/parameters.h" +#include + #include #include #include diff --git a/src/framework/parameters/output.cpp b/src/framework/parameters/output.cpp index 3d06ff054..547800971 100644 --- a/src/framework/parameters/output.cpp +++ b/src/framework/parameters/output.cpp @@ -5,6 +5,7 @@ #include "utils/error.h" #include "utils/log.h" + #include namespace ntt { diff --git a/src/framework/parameters/output.h b/src/framework/parameters/output.h index 08b3377d2..e301fc745 100644 --- a/src/framework/parameters/output.h +++ b/src/framework/parameters/output.h @@ -14,10 +14,10 @@ #include "global.h" -#include - #include "framework/parameters/parameters.h" +#include + #include #include #include diff --git a/src/framework/parameters/particles.cpp b/src/framework/parameters/particles.cpp index 0b15e20b4..afc795a9a 100644 --- a/src/framework/parameters/particles.cpp +++ b/src/framework/parameters/particles.cpp @@ -86,6 +86,8 @@ namespace ntt { return EmissionType::SYNCHROTRON; } else if (fmt::toLower(emission_policy_str) == "compton") { return EmissionType::COMPTON; + } else if (fmt::toLower(emission_policy_str) == "custom") { + return EmissionType::CUSTOM; } else { raise::Error(fmt::format("Invalid emission_policy value: %s", emission_policy_str.c_str()), diff --git a/src/framework/simulation.cpp b/src/framework/simulation.cpp index db133c0f8..4e0da3eda 100644 --- a/src/framework/simulation.cpp +++ b/src/framework/simulation.cpp @@ -9,6 +9,7 @@ #include "utils/formatting.h" #include "utils/log.h" #include "utils/plog.h" + #include #include diff --git a/src/framework/simulation.h b/src/framework/simulation.h index 745c76db4..7387d62dd 100644 --- a/src/framework/simulation.h +++ b/src/framework/simulation.h @@ -17,11 +17,12 @@ #include "enums.h" #include "utils/error.h" -#include #include "engines/traits.h" #include "framework/parameters/parameters.h" +#include + namespace ntt { class Simulation { diff --git a/src/framework/tests/comm-mpi.cpp b/src/framework/tests/comm-mpi.cpp index 487976f73..8da0fa54b 100644 --- a/src/framework/tests/comm-mpi.cpp +++ b/src/framework/tests/comm-mpi.cpp @@ -1,5 +1,3 @@ -#include "framework/domain/comm_mpi.hpp" - #include "enums.h" #include "global.h" @@ -8,6 +6,8 @@ #include "utils/error.h" #include "utils/numeric.h" +#include "framework/domain/comm_mpi.hpp" + #include #include diff --git a/src/framework/tests/comm-nompi.cpp b/src/framework/tests/comm-nompi.cpp index c7646ef03..059c21cdb 100644 --- a/src/framework/tests/comm-nompi.cpp +++ b/src/framework/tests/comm-nompi.cpp @@ -1,5 +1,3 @@ -#include "framework/domain/comm_nompi.hpp" - #include "enums.h" #include "global.h" @@ -7,6 +5,8 @@ #include "arch/kokkos_aliases.h" #include "utils/numeric.h" +#include "framework/domain/comm_nompi.hpp" + #include #include diff --git a/src/global/arch/directions.h b/src/global/arch/directions.h index 850bc130d..5f8281ed3 100644 --- a/src/global/arch/directions.h +++ b/src/global/arch/directions.h @@ -132,8 +132,8 @@ namespace dir { using dirs_t = std::vector>; template - inline auto operator<<(std::ostream& os, - const direction_t& dir) -> std::ostream& { + inline auto operator<<(std::ostream& os, const direction_t& dir) + -> std::ostream& { for (auto& d : dir) { os << std::setw(2) << std::left; if (d > 0) { diff --git a/src/global/arch/kokkos_aliases.cpp b/src/global/arch/kokkos_aliases.cpp index e81d41280..25397af8e 100644 --- a/src/global/arch/kokkos_aliases.cpp +++ b/src/global/arch/kokkos_aliases.cpp @@ -9,18 +9,18 @@ auto CreateParticleRangePolicy(npart_t p1, npart_t p2) -> range_t { } template <> -auto CreateRangePolicy( - const tuple_t& i1, - const tuple_t& i2) -> range_t { +auto CreateRangePolicy(const tuple_t& i1, + const tuple_t& i2) + -> range_t { index_t i1min = i1[0]; index_t i1max = i2[0]; return Kokkos::RangePolicy(i1min, i1max); } template <> -auto CreateRangePolicy( - const tuple_t& i1, - const tuple_t& i2) -> range_t { +auto CreateRangePolicy(const tuple_t& i1, + const tuple_t& i2) + -> range_t { index_t i1min = i1[0]; index_t i1max = i2[0]; index_t i2min = i1[1]; @@ -31,9 +31,9 @@ auto CreateRangePolicy( } template <> -auto CreateRangePolicy( - const tuple_t& i1, - const tuple_t& i2) -> range_t { +auto CreateRangePolicy(const tuple_t& i1, + const tuple_t& i2) + -> range_t { index_t i1min = i1[0]; index_t i1max = i2[0]; index_t i2min = i1[1]; @@ -46,18 +46,18 @@ auto CreateRangePolicy( } template <> -auto CreateRangePolicyOnHost( - const tuple_t& i1, - const tuple_t& i2) -> range_h_t { +auto CreateRangePolicyOnHost(const tuple_t& i1, + const tuple_t& i2) + -> range_h_t { index_t i1min = i1[0]; index_t i1max = i2[0]; return Kokkos::RangePolicy(i1min, i1max); } template <> -auto CreateRangePolicyOnHost( - const tuple_t& i1, - const tuple_t& i2) -> range_h_t { +auto CreateRangePolicyOnHost(const tuple_t& i1, + const tuple_t& i2) + -> range_h_t { index_t i1min = i1[0]; index_t i1max = i2[0]; index_t i2min = i1[1]; @@ -68,9 +68,9 @@ auto CreateRangePolicyOnHost( } template <> -auto CreateRangePolicyOnHost( - const tuple_t& i1, - const tuple_t& i2) -> range_h_t { +auto CreateRangePolicyOnHost(const tuple_t& i1, + const tuple_t& i2) + -> range_h_t { index_t i1min = i1[0]; index_t i1max = i2[0]; index_t i2min = i1[1]; diff --git a/src/global/arch/mpi_tags.h b/src/global/arch/mpi_tags.h index aaf38a8f4..775670614 100644 --- a/src/global/arch/mpi_tags.h +++ b/src/global/arch/mpi_tags.h @@ -190,13 +190,8 @@ namespace mpi { tag; } - Inline auto SendTag(short tag, - bool im1, - bool ip1, - bool jm1, - bool jp1, - bool km1, - bool kp1) -> short { + Inline auto SendTag(short tag, bool im1, bool ip1, bool jm1, bool jp1, bool km1, bool kp1) + -> short { return ((im1 && jm1 && km1) * (PrtlSendTag::im1_jm1_km1 - 1) + (im1 && jm1 && kp1) * (PrtlSendTag::im1_jm1_kp1 - 1) + (im1 && jp1 && km1) * (PrtlSendTag::im1_jp1_km1 - 1) + diff --git a/src/global/arch/traits.h b/src/global/arch/traits.h index be3b377c5..a2d40adac 100644 --- a/src/global/arch/traits.h +++ b/src/global/arch/traits.h @@ -169,45 +169,6 @@ namespace traits { } // namespace external - namespace emission { - - template - concept HasPayloadType = requires { typename E::Payload; }; - - template - concept HasShouldEmit = HasPayloadType and requires( - const E& e, - const coord_t& x_Cd, - const coord_t& x_Ph, - const vec_t& u_Ph, - const vec_t& ep, - const vec_t& bp, - vec_t& delta_u_Ph, - typename E::Payload& payload) { - { - e.shouldEmit(x_Cd, x_Ph, u_Ph, ep, bp, delta_u_Ph, payload) - } -> std::convertible_to>; - }; - - template - concept HasEmit = HasPayloadType and - requires(const E& e, - const tuple_t& xi_Cd, - const tuple_t& dxi_Cd, - const vec_t& direction, - real_t weight, - real_t phi, - const typename E::Payload& payload) { - { - e.emit(xi_Cd, dxi_Cd, direction, weight, phi, payload) - } -> std::same_as; - }; - - template - concept IsValidEmissionPolicy = HasShouldEmit and HasEmit; - - } // namespace emission - template