From 9c9161adf913477f7aad49b939078ac5abacafd5 Mon Sep 17 00:00:00 2001
From: Rafael Ravedutti <rafael.r.ravedutti@fau.de>
Date: Mon, 12 Feb 2024 22:41:00 +0100
Subject: [PATCH] Small fixes
Signed-off-by: Rafael Ravedutti <rafael.r.ravedutti@fau.de>
---
runtime/domain/ParticleDataHanding.hpp | 27 +++++++++++++++-----------
runtime/domain/block_forest.cpp | 2 +-
runtime/domain/block_forest.hpp | 11 +++++++++++
runtime/interfaces/md.hpp | 27 ++++++++++++++++++++++++++
runtime/pairs.hpp | 2 ++
src/pairs/sim/domain_partitioning.py | 20 +++++++++----------
6 files changed, 67 insertions(+), 22 deletions(-)
create mode 100644 runtime/interfaces/md.hpp
diff --git a/runtime/domain/ParticleDataHanding.hpp b/runtime/domain/ParticleDataHanding.hpp
index b36cd0e..6ee2710 100644
--- a/runtime/domain/ParticleDataHanding.hpp
+++ b/runtime/domain/ParticleDataHanding.hpp
@@ -25,7 +25,7 @@ inline bool operator==(const ParticleDeleter& lhs, const ParticleDeleter& rhs) {
class ParticleDataHandling : public blockforest::BlockDataHandling<internal::ParticleDeleter>{
public:
- ParticleDataHandling() {}
+ ParticleDataHandling(PairsSimulation *_ps) { ps = _ps; }
virtual ~ParticleDataHandling() {}
virtual internal::ParticleDeleter *initialize(IBlock *const block) WALBERLA_OVERRIDE {
@@ -71,6 +71,8 @@ public:
}
private:
+ PairsSimulation *ps;
+
void serializeImpl(Block *const block, const BlockDataID&, mpi::SendBuffer& buffer, const uint_t child, bool check_child) {
auto ptr = buffer.allocate<uint_t>();
double aabb_check[6];
@@ -95,7 +97,7 @@ private:
}
for(auto& p: ps->getNonVolatileProperties()) {
- ps->copyPropertyToHost(p, ReadOnly);
+ ps->copyPropertyToHost(p, WriteAfterRead);
}
auto position = ps->getPropertyByName("position");
@@ -167,6 +169,8 @@ private:
void deserializeImpl(IBlock *const, const BlockDataID&, mpi::RecvBuffer& buffer) {
int nlocal = ps->getNumberOfLocalParticles();
+ real_t real_tmp;
+ int int_tmp;
uint_t nrecv;
buffer >> nrecv;
@@ -184,28 +188,33 @@ private:
constexpr int nelems = 3;
for(int e = 0; e < nelems; e++) {
- buffer >> vector_ptr(nlocal + i, e);
+ buffer >> real_tmp;
+ vector_ptr(nlocal + i, e) = real_tmp;
}
} else if(prop_type == Prop_Matrix) {
auto matrix_ptr = ps->getAsMatrixProperty(prop);
constexpr int nelems = 9;
for(int e = 0; e < nelems; e++) {
- buffer >> matrix_ptr(nlocal + i, e);
+ buffer >> real_tmp;
+ matrix_ptr(nlocal + i, e) = real_tmp;
}
} else if(prop_type == Prop_Quaternion) {
auto quat_ptr = ps->getAsQuaternionProperty(prop);
constexpr int nelems = 4;
for(int e = 0; e < nelems; e++) {
- buffer >> quat_ptr(nlocal + i, e);
+ buffer >> real_tmp;
+ quat_ptr(nlocal + i, e) = real_tmp;
}
} else if(prop_type == Prop_Integer) {
auto int_ptr = ps->getAsIntegerProperty(prop);
- buffer >> int_ptr(nlocal + i);
+ buffer >> int_tmp;
+ int_ptr(nlocal + i) = int_tmp;
} else if(prop_type == Prop_Real) {
auto float_ptr = ps->getAsFloatProperty(prop);
- buffer >> float_ptr(nlocal + i);
+ buffer >> real_tmp;
+ float_ptr(nlocal + i) = real_tmp;
} else {
std::cerr << "deserializeImpl(): Invalid property type!" << std::endl;
return 0;
@@ -213,10 +222,6 @@ private:
}
}
- for(auto& p: ps->getNonVolatileProperties()) {
- ps->clearDeviceFlags(p);
- }
-
ps->setNumberOfLocalParticles(nlocal + nrecv);
}
};
diff --git a/runtime/domain/block_forest.cpp b/runtime/domain/block_forest.cpp
index 02a8fca..54aba1d 100644
--- a/runtime/domain/block_forest.cpp
+++ b/runtime/domain/block_forest.cpp
@@ -308,7 +308,7 @@ void BlockForest::initializeWorkloadBalancer() {
forest->setRefreshPhantomBlockMigrationPreparationFunction(prepFunc);
}
- forest->addBlockData(make_shared<ParticleDataHandling>(), "Interface");
+ forest->addBlockData(make_shared<ParticleDataHandling>(ps), "Interface");
}
void BlockForest::finalize() {
diff --git a/runtime/domain/block_forest.hpp b/runtime/domain/block_forest.hpp
index 6facb7d..22b7fe1 100644
--- a/runtime/domain/block_forest.hpp
+++ b/runtime/domain/block_forest.hpp
@@ -1,5 +1,16 @@
#include <map>
//---
+#include <blockforest/BlockForest.h>
+#include <blockforest/Initialization.h>
+#include <blockforest/loadbalancing/DynamicCurve.h>
+#include <blockforest/loadbalancing/DynamicDiffusive.h>
+#include <blockforest/loadbalancing/DynamicParMetis.h>
+#include <blockforest/loadbalancing/InfoCollection.h>
+#include <blockforest/loadbalancing/PODPhantomData.h>
+#include <pe/amr/level_determination/MinMaxLevelDetermination.h>
+#include <pe/amr/weight_assignment/MetisAssignmentFunctor.h>
+#include <pe/amr/weight_assignment/WeightAssignmentFunctor.h>
+//---
#include "../pairs_common.hpp"
#include "domain_partitioning.hpp"
diff --git a/runtime/interfaces/md.hpp b/runtime/interfaces/md.hpp
new file mode 100644
index 0000000..c8bb6b2
--- /dev/null
+++ b/runtime/interfaces/md.hpp
@@ -0,0 +1,27 @@
+#include "../pairs.hpp"
+
+namespace pairs_host_interface {
+
+int get_uid(int *uid, int i) { return uid[i]; }
+int get_shape(int *shape, int i) { return shape[i]; }
+int get_flags(int *flags, int i) { return flags[i]; }
+double get_position(double *position, int i, int j, int capacity) { return position[i * 3 + j]; }
+double get_mass(double *mass, int i) { return mass[i]; }
+double get_linear_velocity(double *linear_velocity, int i, int j, int capacity) { return linear_velocity[i * 3 + j]; }
+double get_force(double *force, int i, int j, int capacity) { return force[i * 3 + j]; }
+int get_type(int *type, int i) { return type[i]; }
+
+}
+
+namespace pairs_cuda_interface {
+
+__inline__ __device__ int get_uid(int *uid, int i) { return uid[i]; }
+__inline__ __device__ int get_shape(int *shape, int i) { return shape[i]; }
+__inline__ __device__ int get_flags(int *flags, int i) { return flags[i]; }
+__inline__ __device__ double get_position(double *position, int i, int j, int capacity) { return position[i * 3 + j]; }
+__inline__ __device__ double get_mass(double *mass, int i) { return mass[i]; }
+__inline__ __device__ double get_linear_velocity(double *linear_velocity, int i, int j, int capacity) { return linear_velocity[i * 3 + j]; }
+__inline__ __device__ double get_force(double *force, int i, int j, int capacity) { return force[i * 3 + j]; }
+__inline__ __device__ int get_type(int *type, int i) { return type[i]; }
+
+}
diff --git a/runtime/pairs.hpp b/runtime/pairs.hpp
index 93f4583..f47e714 100644
--- a/runtime/pairs.hpp
+++ b/runtime/pairs.hpp
@@ -70,7 +70,9 @@ public:
return RuntimeVar<T>(h_ptr);
}
+ void setNumberOfLocalParticles(int n) { *nlocal = n; }
const int getNumberOfLocalParticles() { return *nlocal; }
+ void setNumberOfGhostParticles(int n) { *nghost = n; }
const int getNumberOfGhostParticles() { return *nghost; }
// Arrays
diff --git a/src/pairs/sim/domain_partitioning.py b/src/pairs/sim/domain_partitioning.py
index dae2d28..93822fe 100644
--- a/src/pairs/sim/domain_partitioning.py
+++ b/src/pairs/sim/domain_partitioning.py
@@ -75,12 +75,13 @@ class BlockForest:
self.sim = sim
self.nranks = sim.add_var('nranks', Types.Int32)
self.nranks_capacity = sim.add_var('nranks_capacity', Types.Int32)
+ self.ntotal_aabbs = sim.add_var('ntotal_aabbs', Types.Int32)
self.aabb_capacity = sim.add_var('aabb_capacity', Types.Int32)
- self.ranks = sim.add_static_array('ranks', [self.nranks_capacity], Types.Int32)
- self.naabbs = sim.add_static_array('naabbs', [self.nranks_capacity], Types.Int32)
- self.aabb_offsets = sim.add_static_array('aabb_offsets', [self.nranks_capacity], Types.Int32)
- self.aabbs = sim.add_static_array('aabbs', [self.aabb_capacity, 6], Types.Real)
- self.subdom = sim.add_static_array('subdom', [sim.ndims() * 2], Types.Real)
+ self.ranks = sim.add_array('ranks', [self.nranks_capacity], Types.Int32)
+ self.naabbs = sim.add_array('naabbs', [self.nranks_capacity], Types.Int32)
+ self.aabb_offsets = sim.add_array('aabb_offsets', [self.nranks_capacity], Types.Int32)
+ self.aabbs = sim.add_array('aabbs', [self.aabb_capacity, 6], Types.Real)
+ self.subdom = sim.add_array('subdom', [sim.ndims() * 2], Types.Real)
def min(self, dim):
return self.subdom[dim * 2 + 0]
@@ -99,7 +100,7 @@ class BlockForest:
Call_Void(self.sim, "pairs->initDomain", [param for delim in grid_array for param in delim])
Assign(self.sim, self.nranks, Call_Int(self.sim, "pairs->getNumberOfNeighborRanks", []))
- Assign(self.sim, self.naabbs, Call_Int(self.sim, "pairs->getNumberOfNeighborAABBs", []))
+ Assign(self.sim, self.ntotal_aabbs, Call_Int(self.sim, "pairs->getNumberOfNeighborAABBs", []))
for _ in Filter(self.sim, self.nranks_capacity < self.nranks):
Assign(self.sim, self.nranks_capacity, self.nranks + 10)
@@ -107,15 +108,14 @@ class BlockForest:
self.naabbs.realloc()
self.aabb_offsets.realloc()
- for _ in Filter(self.sim, self.aabb_capacity < self.naabbs):
- Assign(self.sim, self.aabb_capacity, self.naabbs + 20)
- self.pbc.realloc()
+ for _ in Filter(self.sim, self.aabb_capacity < self.ntotal_aabbs):
+ Assign(self.sim, self.aabb_capacity, self.ntotal_aabbs + 20)
self.aabbs.realloc()
Call_Void(self.sim, "pairs->copyRuntimeArray", ['ranks', self.ranks, self.nranks])
Call_Void(self.sim, "pairs->copyRuntimeArray", ['naabbs', self.naabbs, self.nranks])
Call_Void(self.sim, "pairs->copyRuntimeArray", ['aabb_offsets', self.aabb_offsets, self.nranks])
- Call_Void(self.sim, "pairs->copyRuntimeArray", ['aabbs', self.aabbs, self.naabbs * 6])
+ Call_Void(self.sim, "pairs->copyRuntimeArray", ['aabbs', self.aabbs, self.ntotal_aabbs * 6])
Call_Void(self.sim, "pairs->copyRuntimeArray", ['subdom', self.subdom, self.sim.ndims() * 2])
def ghost_particles(self, step, position, offset=0.0):
--
GitLab