diff --git a/example/particle/3d/cube.xml b/example/particle/3d/cube.xml
new file mode 100644
index 000000000..d0739c874
--- /dev/null
+++ b/example/particle/3d/cube.xml
@@ -0,0 +1,26 @@
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
diff --git a/example/particle/3d/roll_lb.xml b/example/particle/3d/roll_lb.xml
new file mode 100644
index 000000000..748b4c06b
--- /dev/null
+++ b/example/particle/3d/roll_lb.xml
@@ -0,0 +1,22 @@
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
+
diff --git a/example/particle/3d/roll_sp.xml b/example/particle/3d/roll_sp.xml
new file mode 100644
index 000000000..1e4035306
--- /dev/null
+++ b/example/particle/3d/roll_sp.xml
@@ -0,0 +1,5 @@
+
+
+
+
+
diff --git a/src/CartLatticeAccess.hpp.Rt b/src/CartLatticeAccess.hpp.Rt
index 7b8911922..af0d74d93 100644
--- a/src/CartLatticeAccess.hpp.Rt
+++ b/src/CartLatticeAccess.hpp.Rt
@@ -15,6 +15,10 @@
#include "CartLatticeContainer.h"
#include "StorageConversions.h"
+#ifndef NODE_SYMZ
+ #define NODE_SYMZ 0
+#endif
+
/// Push all densities
template < class PARENT >
template
@@ -407,12 +413,12 @@ CudaDeviceFunction real_t SymmetryAccess< PARENT >::%s paste0(this_fun, f$nice
{
if ( > range_int<0>()) {
- if ((this->getNodeType() & NODE_SYM%s ch[i] ?>) == NODE_Symmetry%s ch[i] ?>_plus) {
- return %s paste0(sig, next_fun, sf$nicename) ?>();
+ if ((this->getNodeType() & NODE_SYM%s ch[i] ?>) == NODE_%s autosym_name ?>%s ch[i] ?>_plus) {
+ return %s paste0(sig, next_fun, sf$nicename) ?>();
}
} else if ( < range_int<0>()) {
- if ((this->getNodeType() & NODE_SYM%s ch[i] ?>) == NODE_Symmetry%s ch[i] ?>_minus) {
- return %s paste0(sig, next_fun, sf$nicename) ?>();
+ if ((this->getNodeType() & NODE_SYM%s ch[i] ?>) == NODE_%s autosym_name ?>%s ch[i] ?>_minus) {
+ return %s paste0(sig, next_fun, sf$nicename) ?>();
}
}
@@ -425,9 +431,27 @@ CudaDeviceFunction real_t SymmetryAccess< PARENT >::%s paste0(this_fun, f$nice
template < class PARENT >
template
CudaDeviceFunction void SymmetryAccess< PARENT >::pop%s s$suffix ?>(N & node) const
-{
- parent::pop%s s$suffix ?>(node);
-
+{
+ parent::pop%s s$suffix ?>(node);
+ parent::pop%s s$suffix ?>(node);
+ if (this->getNodeType() & (NODE_SYMX | NODE_SYMY | NODE_SYMZ)) {
+ %s paste("node",d$name,sep=".") ?> = load_%s f$nicename ?>(range_int< %d dp[1] ?> >(),range_int< %d dp[2] ?> >(),range_int< %d dp[3] ?> >());
+ %s paste("node",d$name,sep=".") ?> = %f d$default ?>;
+ } else {
+ parent::pop%s s$suffix ?>(node);
+ }
}
diff --git a/src/Handlers/GenericOptimizer.cpp b/src/Handlers/GenericOptimizer.cpp
index ad78d632a..ccb97ac49 100644
--- a/src/Handlers/GenericOptimizer.cpp
+++ b/src/Handlers/GenericOptimizer.cpp
@@ -6,7 +6,7 @@ int GenericOptimizer::Init () {
int ret;
DEBUG_M;
ret = OptimizerInit();
- MPI_Bcast( &ret, 1, MPI_INT, 0, MPI_COMM_WORLD );
+ MPI_Bcast( &ret, 1, MPI_INT, 0, solver->mpi_comm );
if (ret) {
ERROR("Failed to initialize Optimizer");
return -1;
diff --git a/src/Handlers/OptimalControl.cpp b/src/Handlers/OptimalControl.cpp
index 92ad09185..039b020bd 100644
--- a/src/Handlers/OptimalControl.cpp
+++ b/src/Handlers/OptimalControl.cpp
@@ -88,7 +88,7 @@ int OptimalControl::Parameters (int type, double * tab) {
return 0;
case PAR_SET:
output("Setting the params in the zone\n");
- MPI_Bcast(tab, Pars, MPI_DOUBLE, 0, MPI_COMM_WORLD);
+ MPI_Bcast(tab, Pars, MPI_DOUBLE, 0, solver->mpi_comm);
if (f != NULL) {
fprintf(f,"SET");
for (int i=0;ilattice->zSet.get_grad(par_index, zone_number, tmptab);
- MPI_Reduce(tmptab, tab, Pars, MPI_DOUBLE, MPI_SUM, 0, MPI_COMM_WORLD);
+ MPI_Reduce(tmptab, tab, Pars, MPI_DOUBLE, MPI_SUM, 0, solver->mpi_comm);
if (f != NULL) {
fprintf(f,"GRAD");
for (int i=0;impi_comm );
return gsum;
}
diff --git a/src/Handlers/acRemoteForceInterface.cpp b/src/Handlers/acRemoteForceInterface.cpp
index 993fa127c..a3b26d9dd 100644
--- a/src/Handlers/acRemoteForceInterface.cpp
+++ b/src/Handlers/acRemoteForceInterface.cpp
@@ -2,6 +2,8 @@
std::string acRemoteForceInterface::xmlname = "RemoteForceInterface";
#include "../HandlerFactory.h"
+#include
+
int acRemoteForceInterface::Init () {
Action::Init();
pugi::xml_attribute attr = node.attribute("integrator");
@@ -22,6 +24,26 @@ int acRemoteForceInterface::ConnectRemoteForceInterface(std::string integrator_)
solver->lattice->RFI.setUnits(units[0],units[1],units[2]);
solver->lattice->RFI.CanCopeWithUnits(false);
+ solver->lattice->RFI.setVar("output", solver->info.outpath);
+
+
+ std::string element_content;
+ int node_children = 0;
+ for (pugi::xml_node par = node.first_child(); par; par = par.next_sibling()) {
+ node_children ++;
+ if (node_children > 1) {
+ ERROR("Only a single element/CDATA allowed inside of a RemoteForceInterface xml element\n");
+ return -1;
+ }
+ if ((par.type() == pugi::node_pcdata) || (par.type() == pugi::node_cdata)) {
+ element_content = par.value();
+ } else {
+ std::stringstream ss;
+ par.print(ss);
+ element_content = ss.str();
+ }
+ }
+ if (node_children > 0) solver->lattice->RFI.setVar("content", element_content);
bool stats = false;
std::string stats_prefix = solver->outpath;
stats_prefix = stats_prefix + "_RFI";
@@ -56,7 +78,7 @@ int acRemoteForceInterface::ConnectRemoteForceInterface(std::string integrator_)
bool use_box = true;
attr = node.attribute("use_box");
if (attr) use_box = attr.as_bool();
-
+
if (use_box) {
lbRegion reg = lattice->getLocalRegion();
double px = lattice->px;
@@ -70,6 +92,12 @@ int acRemoteForceInterface::ConnectRemoteForceInterface(std::string integrator_)
pz + reg.dz - PART_MAR_BOX,
pz + reg.dz + reg.nz + PART_MAR_BOX);
}
+
+ attr = node.attribute("omega");
+ if (attr) solver->lattice->RFI_omega = attr.as_bool();
+ attr = node.attribute("torque");
+ if (attr) solver->lattice->RFI_torque = attr.as_bool();
+
MPI_Barrier(MPMD.local);
lattice->RFI.Connect(MPMD.work,inter.work);
diff --git a/src/Handlers/acSolve.cpp b/src/Handlers/acSolve.cpp
index f66f841c6..030a316c4 100644
--- a/src/Handlers/acSolve.cpp
+++ b/src/Handlers/acSolve.cpp
@@ -7,7 +7,8 @@ int acSolve::Init () {
if (GenericAction::ExecuteInternal()) return -1;
int stop=0;
do {
- int next_it = Next(solver->iter);
+ int my_next_it = Next(solver->iter);
+ int next_it = my_next_it;
for (size_t i=0; ihands.size(); i++) {
int it = solver->hands[i].Next(solver->iter);
if ((it > 0) && (it < next_it)) next_it = it;
@@ -15,7 +16,9 @@ int acSolve::Init () {
solver->steps = next_it;
MPI_Bcast(&solver->steps, 1, MPI_INT, 0, MPMD.local);
solver->iter += solver->steps;
- solver->lattice->Iterate(solver->steps, solver->iter_type);
+ int iter_type = solver->iter_type;
+ if (solver->steps == my_next_it) iter_type |= ITER_LASTGLOB;
+ solver->lattice->Iterate(solver->steps, iter_type);
CudaDeviceSynchronize();
MPI_Barrier(MPMD.local);
for (size_t i=0; ihands.size(); i++) {
diff --git a/src/Handlers/cbPID.cpp b/src/Handlers/cbPID.cpp
index 6b6fc05bb..013dc1482 100644
--- a/src/Handlers/cbPID.cpp
+++ b/src/Handlers/cbPID.cpp
@@ -111,7 +111,7 @@ int cbPID::DoIt () {
old_err = err;
}
- MPI_Bcast(&control, 1, MPI_DOUBLE, 0, MPI_COMM_WORLD);
+ MPI_Bcast(&control, 1, MPI_DOUBLE, 0, solver->mpi_comm);
double nval;
/* if (zone_number < 0) {
sval = solver->lattice->zSet.get(setting, 0, (size_t) 0);
diff --git a/src/Handlers/cbRunR.cpp b/src/Handlers/cbRunR.cpp
index b73eb39e9..42ba5d6b0 100644
--- a/src/Handlers/cbRunR.cpp
+++ b/src/Handlers/cbRunR.cpp
@@ -742,6 +742,8 @@ int cbRunR::Init() {
output("%s\n",source.c_str());
output("--------------------------------\n");
}
+ old_iter_type = solver->iter_type;
+ solver->iter_type |= ITER_LASTGLOB;
return 0;
}
@@ -788,6 +790,11 @@ int cbRunR::DoIt() {
return 0;
}
+int cbRunR::Finish () {
+ solver->iter_type = old_iter_type;
+ return Callback::Finish();
+}
+
#endif // WITH_R
diff --git a/src/Handlers/cbRunR.h b/src/Handlers/cbRunR.h
index a62fa8b3a..d9330a653 100644
--- a/src/Handlers/cbRunR.h
+++ b/src/Handlers/cbRunR.h
@@ -29,9 +29,11 @@ class cbRunR : public Callback {
bool python;
static int s_tag;
int tag;
+ int old_iter_type;
public:
int Init ();
int DoIt ();
+ int Finish ();
};
#endif // WITH_R
diff --git a/src/Lattice.h.Rt b/src/Lattice.h.Rt
index 6cdde4882..220b997a5 100644
--- a/src/Lattice.h.Rt
+++ b/src/Lattice.h.Rt
@@ -74,6 +74,7 @@ public:
real_t px, py, pz;
MPIInfo mpi; ///< MPI information
rfi_t RFI;
+ bool RFI_omega, RFI_torque;
solidcontainer_t SC;
size_t particle_data_size_max;
char snapFileName[STRING_LEN];
diff --git a/src/LatticeBase.cpp.Rt b/src/LatticeBase.cpp.Rt
index ef3e075c3..0460f2904 100644
--- a/src/LatticeBase.cpp.Rt
+++ b/src/LatticeBase.cpp.Rt
@@ -14,6 +14,9 @@ LatticeBase::LatticeBase(int zonesettings, int zones, int num_snaps_, const Unit
data.particle_data_size = 0;
SC.balls = &RFI;
RFI.name = "TCLB";
+ RFI_omega = true;
+ RFI_torque = true;
+
CudaStreamCreate(&kernelStream);
CudaStreamCreate(&inStream);
CudaStreamCreate(&outStream);
@@ -107,12 +110,11 @@ void LatticeBase::CopyInParticles() {
CudaMalloc(&data.particle_data, RFI.mem_size());
}
data.particle_data_size = RFI.size();
- for (size_t i = 0; i < RFI.size(); i++) {
- RFI.RawData(i, RFI_DATA_FORCE + 0) = 0;
- RFI.RawData(i, RFI_DATA_FORCE + 1) = 0;
- RFI.RawData(i, RFI_DATA_FORCE + 2) = 0;
+ for (int j=0; j<6; j++) for (size_t i=0; i 0) {
+ CudaMemcpyAsync(data.particle_data, RFI.Particles(), RFI.mem_size(), CudaMemcpyHostToDevice, kernelStream);
}
- if (RFI.mem_size() > 0) { CudaMemcpyAsync(data.particle_data, RFI.Particles(), RFI.mem_size(), CudaMemcpyHostToDevice, kernelStream); }
DEBUG_PROF_PUSH("Tree Build");
SC.Build();
DEBUG_PROF_POP();
@@ -120,16 +122,22 @@ void LatticeBase::CopyInParticles() {
}
void LatticeBase::CopyOutParticles() {
- if (RFI.mem_size() > 0) { CudaMemcpyAsync(RFI.Particles(), data.particle_data, RFI.mem_size(), CudaMemcpyDeviceToHost, kernelStream); }
+ if (RFI.mem_size() > 0) {
+ CudaMemcpyAsync(RFI.Particles(), data.particle_data, RFI.mem_size(), CudaMemcpyDeviceToHost, kernelStream);
+ }
CudaStreamSynchronize(kernelStream);
DEBUG_PROF_PUSH("Testing particles for NaNs");
int nans = 0;
- for (size_t i = 0; i < RFI.size(); i++) {
- for (int j = 0; j < 3; j++) {
- if (!isfinite(RFI.RawData(i, RFI_DATA_FORCE + j))) {
- nans++;
- RFI.RawData(i, RFI_DATA_FORCE + j) = 0.0;
+ for (int j=0; j<6; j++) {
+ if (RFI_torque || (j<3)) {
+ for (size_t i=0; i 0) notice("%d NANs in particle forces (overwritten with 0.0)\n", nans);
diff --git a/src/LatticeBase.hpp b/src/LatticeBase.hpp
index 8604790f5..275d3ec70 100644
--- a/src/LatticeBase.hpp
+++ b/src/LatticeBase.hpp
@@ -64,6 +64,7 @@ class LatticeBase {
solidcontainer_t SC; ///<
size_t particle_data_size_max = 0; ///<
rfi_t RFI; ///<
+ bool RFI_omega, RFI_torque;
SyntheticTurbulence ST; ///<
std::string snapFileName;
diff --git a/src/Particle.hpp b/src/Particle.hpp
index 6285aebe0..91b70841c 100644
--- a/src/Particle.hpp
+++ b/src/Particle.hpp
@@ -30,9 +30,9 @@ struct ParticleI : Particle {
angvel.x = particle_data[i*RFI_DATA_SIZE+RFI_DATA_ANGVEL+0];
angvel.y = particle_data[i*RFI_DATA_SIZE+RFI_DATA_ANGVEL+1];
angvel.z = particle_data[i*RFI_DATA_SIZE+RFI_DATA_ANGVEL+2];
- diff.x = pos.x - node[0];
- diff.y = pos.y - node[1];
- diff.z = pos.z - node[2];
+ diff.x = node[0] - pos.x;
+ diff.y = node[1] - pos.y;
+ diff.z = node[2] - pos.z;
dist = sqrt(diff.x*diff.x + diff.y*diff.y + diff.z*diff.z);
cvel.x = vel.x + angvel.y*diff.z - angvel.z*diff.y;
cvel.y = vel.y + angvel.z*diff.x - angvel.x*diff.z;
diff --git a/src/RemoteForceInterface.h b/src/RemoteForceInterface.h
index af70e81ce..1dc109122 100644
--- a/src/RemoteForceInterface.h
+++ b/src/RemoteForceInterface.h
@@ -6,6 +6,7 @@
#include
#include
#include
+#include