diff --git a/src/gpu/VirtualFluids_GPU/Communication/ExchangeData27.cpp b/src/gpu/VirtualFluids_GPU/Communication/ExchangeData27.cpp index eb8f7aa8139b303efcc49540c7ae3d8255254822..6b651b150c2a4a5da219f440d8efcf3a2d7eb29a 100644 --- a/src/gpu/VirtualFluids_GPU/Communication/ExchangeData27.cpp +++ b/src/gpu/VirtualFluids_GPU/Communication/ExchangeData27.cpp @@ -70,29 +70,34 @@ void startNonBlockingMpiReceive(unsigned int numberOfSendProcessNeighbors, vf::g } -void copyEdgeNodes(std::vector<LBMSimulationParameter::EdgeNodePositions> &edgeNodes, std::vector<ProcessNeighbor27> &recvProcessNeighborHostAllNodes, std::vector<ProcessNeighbor27> &sendProcessNeighborHostAllNodes, - std::vector<ProcessNeighbor27> &sendProcessNeighborHost) +void copyEdgeNodes(std::vector<LBMSimulationParameter::EdgeNodePositions> &edgeNodes, + std::vector<ProcessNeighbor27> &recvProcessNeighborHost, std::vector<ProcessNeighbor27> &sendProcessNeighborHost) { - uint indexInSubdomainRecv = 0; - uint indexInSubdomainSend = 0; - uint numNodesInBufferRecv = 0; - uint numNodesInBufferSend = 0; + + int indexInSubdomainRecv = 0; + int indexInSubdomainSend = 0; + int numNodesInBufferRecv = 0; + int numNodesInBufferSend = 0; + #pragma omp parallel for for (uint i = 0; i < edgeNodes.size(); i++) { indexInSubdomainRecv = edgeNodes[i].indexOfProcessNeighborRecv; indexInSubdomainSend = edgeNodes[i].indexOfProcessNeighborSend; - numNodesInBufferRecv = recvProcessNeighborHostAllNodes[indexInSubdomainRecv].numberOfNodes; - numNodesInBufferSend = sendProcessNeighborHostAllNodes[indexInSubdomainSend].numberOfNodes; - if(edgeNodes[i].indexInSendBuffer >= sendProcessNeighborHost[indexInSubdomainSend].numberOfNodes){ + numNodesInBufferRecv = recvProcessNeighborHost[indexInSubdomainRecv].numberOfNodes; + numNodesInBufferSend = sendProcessNeighborHost[indexInSubdomainSend].numberOfNodes; + if(edgeNodes[i].indexInSendBuffer >= numNodesInBufferSend){ // for reduced communication after fine to coarse: only copy send nodes which are not part of the reduced comm continue; } - for (uint direction = 0; direction <= dirEND; direction++) { - (sendProcessNeighborHostAllNodes[indexInSubdomainSend].f[0] + (direction * numNodesInBufferSend))[edgeNodes[i].indexInSendBuffer] = 1000; - // (recvProcessNeighborHostAllNodes[indexInSubdomainRecv].f[0] + (direction * numNodesInBufferRecv))[edgeNodes[i].indexInRecvBuffer]; + + for (int direction = 0; direction <= (int) dirEND; direction++) { + (sendProcessNeighborHost[indexInSubdomainSend].f[0] + (direction * numNodesInBufferSend))[edgeNodes[i].indexInSendBuffer] = + (recvProcessNeighborHost[indexInSubdomainRecv].f[0] + (direction * numNodesInBufferRecv))[edgeNodes[i].indexInRecvBuffer]; } } } + + //////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////// // X //////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////// diff --git a/src/gpu/VirtualFluids_GPU/Communication/ExchangeData27.h b/src/gpu/VirtualFluids_GPU/Communication/ExchangeData27.h index 9601ace8b1ac37162ad28c09d2b1546246ce8fb2..d57a9dad95d74a77b8e387811ab9d0262072add8 100644 --- a/src/gpu/VirtualFluids_GPU/Communication/ExchangeData27.h +++ b/src/gpu/VirtualFluids_GPU/Communication/ExchangeData27.h @@ -22,8 +22,7 @@ extern "C" void scatterNodesFromRecvBufferGPU(Parameter *para, int level, int st std::vector<ProcessNeighbor27> *recvProcessNeighborDev, unsigned int numberOfRecvProcessNeighbors); extern "C" void copyEdgeNodes(std::vector<LBMSimulationParameter::EdgeNodePositions> &edgeNodes, - std::vector<ProcessNeighbor27> &recvProcessNeighborHostAllNodes, - std::vector<ProcessNeighbor27> &sendProcessNeighborHostAllNodes, + std::vector<ProcessNeighbor27> &recvProcessNeighborHost, std::vector<ProcessNeighbor27> &sendProcessNeighborHost); // x diff --git a/src/gpu/VirtualFluids_GPU/LBM/Simulation.cpp b/src/gpu/VirtualFluids_GPU/LBM/Simulation.cpp index 284b9f2397ca542bf99b1095061d88f4646eefc7..baf268b25e02e86522df749d4820a4e36ee0b99b 100644 --- a/src/gpu/VirtualFluids_GPU/LBM/Simulation.cpp +++ b/src/gpu/VirtualFluids_GPU/LBM/Simulation.cpp @@ -391,14 +391,14 @@ void Simulation::init(SPtr<Parameter> para, SPtr<GridProvider> gridProvider, std //InterfaceDebugWriter::writeInterfaceLinesDebugFC(para.get()); // writers for version with communication hiding - if(para->getNumprocs() > 1 && para->getUseStreams()){ - // InterfaceDebugWriter::writeInterfaceFCC_Send(para.get()); - // InterfaceDebugWriter::writeInterfaceCFC_Recv(para.get()); - // InterfaceDebugWriter::writeSendNodesStream(para.get()); - // InterfaceDebugWriter::writeRecvNodesStream(para.get()); - EdgeNodeDebugWriter::writeEdgeNodesXZ_Send(para); - EdgeNodeDebugWriter::writeEdgeNodesXZ_Recv(para); - } + // if(para->getNumprocs() > 1 && para->getUseStreams()){ + // InterfaceDebugWriter::writeInterfaceFCC_Send(para.get()); + // InterfaceDebugWriter::writeInterfaceCFC_Recv(para.get()); + // InterfaceDebugWriter::writeSendNodesStream(para.get()); + // InterfaceDebugWriter::writeRecvNodesStream(para.get()); + // EdgeNodeDebugWriter::writeEdgeNodesXZ_Send(para); + // EdgeNodeDebugWriter::writeEdgeNodesXZ_Recv(para); + // } } void Simulation::allocNeighborsOffsetsScalesAndBoundaries(SPtr<GridProvider> &gridProvider)