From 4d0bc1c78f26b63b8427981ec190a25da7ca7409 Mon Sep 17 00:00:00 2001
From: "TESLA03\\Master" <a.wellmann@tu-bs.de>
Date: Wed, 21 Jul 2021 14:29:19 +0200
Subject: [PATCH] use ParH and ParD (LBMSimulationParameter) instead of class
NodeIndicesMultiGPU
---
apps/gpu/LBM/MusselOyster/MusselOyster.cpp | 16 ++++-----
src/gpu/GridGenerator/grid/Grid.h | 7 ++--
.../grid/GridBuilder/GridBuilder.h | 3 ++
.../grid/GridBuilder/LevelGridBuilder.cpp | 7 ++++
.../grid/GridBuilder/LevelGridBuilder.h | 3 +-
.../grid/GridBuilder/MultipleGridBuilder.cpp | 21 +-----------
.../grid/GridBuilder/MultipleGridBuilder.h | 2 --
src/gpu/GridGenerator/grid/GridImp.cu | 34 ++++++++-----------
src/gpu/GridGenerator/grid/GridImp.h | 12 +++----
.../DataStructureInitializer/GridProvider.cpp | 6 ++++
.../DataStructureInitializer/GridProvider.h | 3 ++
.../GridReaderFiles/GridReader.cpp | 4 +++
.../GridReaderFiles/GridReader.h | 1 +
.../GridReaderGenerator/GridGenerator.cpp | 10 ++++++
.../GridReaderGenerator/GridGenerator.h | 1 +
.../GPU/CudaMemoryManager.cpp | 21 +++++++++++-
.../VirtualFluids_GPU/GPU/CudaMemoryManager.h | 4 ++-
src/gpu/VirtualFluids_GPU/LBM/Simulation.cpp | 1 +
.../Parameter/NodeIndicesMultiGPU.cpp | 18 ----------
.../Parameter/NodeIndicesMultiGPU.h | 20 -----------
.../VirtualFluids_GPU/Parameter/Parameter.cpp | 11 ------
.../VirtualFluids_GPU/Parameter/Parameter.h | 11 ++----
22 files changed, 96 insertions(+), 120 deletions(-)
delete mode 100644 src/gpu/VirtualFluids_GPU/Parameter/NodeIndicesMultiGPU.cpp
delete mode 100644 src/gpu/VirtualFluids_GPU/Parameter/NodeIndicesMultiGPU.h
diff --git a/apps/gpu/LBM/MusselOyster/MusselOyster.cpp b/apps/gpu/LBM/MusselOyster/MusselOyster.cpp
index c23e4d5df..551e3ffd9 100644
--- a/apps/gpu/LBM/MusselOyster/MusselOyster.cpp
+++ b/apps/gpu/LBM/MusselOyster/MusselOyster.cpp
@@ -56,7 +56,6 @@
#include "utilities/communication.h"
-#include "VirtualFluids_GPU/Parameter/NodeIndicesMultiGPU.h"
////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////
////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////
@@ -110,7 +109,7 @@ void multipleLevel(const std::string& configPath)
////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////
bool useGridGenerator = true;
- bool useMultiGPU = true;
+ bool useMultiGPU = false;
////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////
////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////
@@ -145,8 +144,8 @@ void multipleLevel(const std::string& configPath)
// para->setMaxLevel(2);
- // para->setMainKernel("CumulantK15Comp");
- para->setMainKernel("CumulantK17CompChimSparse");
+ para->setMainKernel("CumulantK17CompChim");
+ //para->setMainKernel("CumulantK17CompChimSparse");
if (useMultiGPU) {
para->setDevices(std::vector<uint>{ (uint)0, (uint)1 });
@@ -241,14 +240,11 @@ void multipleLevel(const std::string& configPath)
gridBuilder->setVelocityBoundaryCondition(SideType::GEOMETRY, 0.0, 0.0, 0.0);
//////////////////////////////////////////////////////////////////////////
- gridBuilder->writeGridsToVtk(path + "/" + bivalveType + "/grid/part" + std::to_string(generatePart) + "_");
+ //gridBuilder->writeGridsToVtk(path + "/" + bivalveType + "/grid/part" + std::to_string(generatePart) + "_");
//gridBuilder->writeGridsToVtk(path + "/" + bivalveType + "/" + std::to_string(generatePart) + "/grid/");
//gridBuilder->writeArrows(path + "/" + bivalveType + "/" + std::to_string(generatePart) + " /arrow");
gridBuilder->findGeoFluidNodes();
- std::shared_ptr<NodeIndicesMultiGPU> nodeIndicesMultiGPU = std::make_shared<NodeIndicesMultiGPU>(gridBuilder->getGeoFluidSizes(),
- gridBuilder->getGeoFluidNodeIndices());
- para->setNodeIndicesMultiGPU(nodeIndicesMultiGPU);
SimulationFileWriter::write(gridPath + "/" + std::to_string(generatePart) + "/", gridBuilder, FILEFORMAT::BINARY);
@@ -273,7 +269,9 @@ void multipleLevel(const std::string& configPath)
gridBuilder->setVelocityBoundaryCondition(SideType::GEOMETRY, 0.0, 0.0, 0.0);
//////////////////////////////////////////////////////////////////////////
- gridBuilder->writeGridsToVtk("E:/temp/MusselOyster/" + bivalveType + "/grid/");
+ gridBuilder->findGeoFluidNodes();
+
+ //gridBuilder->writeGridsToVtk("E:/temp/MusselOyster/" + bivalveType + "/grid/");
// gridBuilder->writeArrows ("E:/temp/MusselOyster/" + bivalveType + "/arrow");
SimulationFileWriter::write(gridPath, gridBuilder, FILEFORMAT::BINARY);
diff --git a/src/gpu/GridGenerator/grid/Grid.h b/src/gpu/GridGenerator/grid/Grid.h
index f96434931..3a70c16ad 100644
--- a/src/gpu/GridGenerator/grid/Grid.h
+++ b/src/gpu/GridGenerator/grid/Grid.h
@@ -142,10 +142,9 @@ public:
virtual void repairCommunicationInices(int direction) = 0;
// needed for CUDA Streams MultiGPU
- virtual void findMatrixIDsGEO_FLUID() = 0;
- virtual uint getGeoFluidSize() const = 0;
- virtual const uint* getGeoFluidNodes() const = 0;
-
+ virtual void findFluidNodeIndices() = 0;
+ virtual uint getNumberOfFluidNodes() const = 0;
+ virtual void getFluidNodeIndices(uint *fluidNodeIndices) const = 0;
};
#endif
diff --git a/src/gpu/GridGenerator/grid/GridBuilder/GridBuilder.h b/src/gpu/GridGenerator/grid/GridBuilder/GridBuilder.h
index 8136bd38d..0d12590c1 100644
--- a/src/gpu/GridGenerator/grid/GridBuilder/GridBuilder.h
+++ b/src/gpu/GridGenerator/grid/GridBuilder/GridBuilder.h
@@ -84,6 +84,9 @@ public:
virtual void getSendIndices( int* sendIndices, int direction, int level ) = 0;
virtual void getReceiveIndices( int* sendIndices, int direction, int level ) = 0;
+ virtual uint getNumberOfFluidNodes(unsigned int level) const = 0;
+ virtual void getFluidNodeIndices(uint *fluidNodeIndices, const int level) const = 0;
+
};
#endif
diff --git a/src/gpu/GridGenerator/grid/GridBuilder/LevelGridBuilder.cpp b/src/gpu/GridGenerator/grid/GridBuilder/LevelGridBuilder.cpp
index 92bb2040f..46059f2e5 100644
--- a/src/gpu/GridGenerator/grid/GridBuilder/LevelGridBuilder.cpp
+++ b/src/gpu/GridGenerator/grid/GridBuilder/LevelGridBuilder.cpp
@@ -263,6 +263,10 @@ uint LevelGridBuilder::getNumberOfNodes(unsigned int level) const
return grids[level]->getSparseSize();
}
+uint LevelGridBuilder::getNumberOfFluidNodes(unsigned int level) const
+{
+ return grids[level]->getNumberOfFluidNodes();
+}
std::shared_ptr<Grid> LevelGridBuilder::getGrid(int level, int box)
{
@@ -294,6 +298,9 @@ void LevelGridBuilder::getNodeValues(real *xCoords, real *yCoords, real *zCoords
//TODO: add getSlipSize...
+GRIDGENERATOR_EXPORT void LevelGridBuilder::getFluidNodeIndices(uint *fluidNodeIndices, const int level) const {
+ grids[level]->getFluidNodeIndices(fluidNodeIndices);
+}
uint LevelGridBuilder::getVelocitySize(int level) const
{
diff --git a/src/gpu/GridGenerator/grid/GridBuilder/LevelGridBuilder.h b/src/gpu/GridGenerator/grid/GridBuilder/LevelGridBuilder.h
index fe7b53f52..001c54e71 100644
--- a/src/gpu/GridGenerator/grid/GridBuilder/LevelGridBuilder.h
+++ b/src/gpu/GridGenerator/grid/GridBuilder/LevelGridBuilder.h
@@ -54,9 +54,10 @@ public:
GRIDGENERATOR_EXPORT virtual std::shared_ptr<Grid> getGrid(int level, int box);
-
GRIDGENERATOR_EXPORT virtual unsigned int getNumberOfNodes(unsigned int level) const override;
+ GRIDGENERATOR_EXPORT virtual uint getNumberOfFluidNodes(unsigned int level) const override;
+ GRIDGENERATOR_EXPORT virtual void getFluidNodeIndices(uint* fluidNodeIndices, const int level) const override;
GRIDGENERATOR_EXPORT virtual void getNodeValues(real *xCoords, real *yCoords, real *zCoords,
uint *neighborX, uint *neighborY, uint *neighborZ, uint *neighborNegative,
diff --git a/src/gpu/GridGenerator/grid/GridBuilder/MultipleGridBuilder.cpp b/src/gpu/GridGenerator/grid/GridBuilder/MultipleGridBuilder.cpp
index ffa1aa776..2243ccd65 100644
--- a/src/gpu/GridGenerator/grid/GridBuilder/MultipleGridBuilder.cpp
+++ b/src/gpu/GridGenerator/grid/GridBuilder/MultipleGridBuilder.cpp
@@ -604,29 +604,10 @@ void MultipleGridBuilder::findGeoFluidNodes()
{
*logging::out << logging::Logger::INFO_HIGH << "Start findGeoFluidNodes()\n";
for (uint i = 0; i < grids.size(); i++)
- grids[i]->findMatrixIDsGEO_FLUID();
+ grids[i]->findFluidNodeIndices();
*logging::out << logging::Logger::INFO_HIGH << "Done with findGeoFluidNodes()\n";
}
-const std::vector<uint> *MultipleGridBuilder::getGeoFluidSizes() const
-{
- std::vector<uint> geoFluidSizes;
- for (uint level = 0; level < grids.size(); level++) {
- geoFluidSizes.push_back(grids[level]->getGeoFluidSize());
- }
- return &geoFluidSizes;
-}
-
-const std::vector<const uint*> *MultipleGridBuilder::getGeoFluidNodeIndices() const
-{
- std::vector<const uint*> getGeoFluidNodes;
- for (uint level = 0; level < grids.size(); level++) {
- getGeoFluidNodes.push_back(grids[level]->getGeoFluidNodes());
- }
- return &getGeoFluidNodes;
-}
-
-
void MultipleGridBuilder::writeGridsToVtk(const std::string& path) const
{
for(uint level = 0; level < grids.size(); level++)
diff --git a/src/gpu/GridGenerator/grid/GridBuilder/MultipleGridBuilder.h b/src/gpu/GridGenerator/grid/GridBuilder/MultipleGridBuilder.h
index d014613d1..8039aedd1 100644
--- a/src/gpu/GridGenerator/grid/GridBuilder/MultipleGridBuilder.h
+++ b/src/gpu/GridGenerator/grid/GridBuilder/MultipleGridBuilder.h
@@ -87,8 +87,6 @@ public:
// needed for CUDA Streams MultiGPU
void findGeoFluidNodes();
- GRIDGENERATOR_EXPORT const std::vector<uint> *getGeoFluidSizes() const;
- GRIDGENERATOR_EXPORT const std::vector<const uint*> *getGeoFluidNodeIndices() const;
};
#endif
diff --git a/src/gpu/GridGenerator/grid/GridImp.cu b/src/gpu/GridGenerator/grid/GridImp.cu
index 40d66190c..bbf0d4919 100644
--- a/src/gpu/GridGenerator/grid/GridImp.cu
+++ b/src/gpu/GridGenerator/grid/GridImp.cu
@@ -858,24 +858,18 @@ CUDA_HOST void GridImp::updateSparseIndices()
sparseSize = size - removedNodes;
}
-CUDA_HOST void GridImp::findMatrixIDsGEO_FLUID() // typeOfGridNode = para->getParD(level)->geoSP[index]
+CUDA_HOST void GridImp::findFluidNodeIndices()
{
// auf Basis von getNodeValues und updateSparseIndices
- int geoFluidSize = 0;
- int geoFluidNodesIndex = 0;
for (uint index = 0; index < size; index++) {
- if (this->sparseIndices[index] == -1) //unnötig?
+ uint sparseIndex = sparseIndices[index];
+ if (sparseIndex == -1)
continue;
- if (this->field.isFluid(index)) {
- // + 1 for numbering shift between GridGenerator and VF_GPU
- geoFluidNodes[geoFluidNodesIndex] = index + 1; //+1 notwendig?
- geoFluidNodesIndex++;
- geoFluidSize++;
- }
- /*if (typeOfGridNode[index] == GEO_FLUID)
- geoFluidNodes.push_back(index); */
+ if (this->field.isFluid(index))
+ fluidNodeIndices.push_back(sparseIndex);
}
+ numberOfFluidNodes = fluidNodeIndices.size();
}
HOSTDEVICE void GridImp::setNeighborIndices(uint index)
@@ -1767,12 +1761,8 @@ HOSTDEVICE uint GridImp::getSparseSize() const
return this->sparseSize;
}
-HOSTDEVICE uint GridImp::getGeoFluidSize() const {
- return this->geoFluidSize;
-}
-
-HOSTDEVICE const uint* GridImp::getGeoFluidNodes() const{
- return this->geoFluidNodes;
+HOSTDEVICE uint GridImp::getNumberOfFluidNodes() const {
+ return this->numberOfFluidNodes;
}
HOSTDEVICE Field GridImp::getField() const
@@ -1935,7 +1925,7 @@ CUDA_HOST void GridImp::getNodeValues(real *xCoords, real *yCoords, real *zCoord
geo[0] = GEOSOLID;
int nodeNumber = 0;
- for (uint i = 0; i < this->getSize(); i++)
+ for (uint i = 0; i < this->size; i++)
{
if (this->sparseIndices[i] == -1)
continue;
@@ -1967,6 +1957,12 @@ CUDA_HOST void GridImp::getNodeValues(real *xCoords, real *yCoords, real *zCoord
}
}
+CUDA_HOST void GridImp::getFluidNodeIndices(uint *fluidNodeIndices) const {
+ for (uint nodeNumber = 0; nodeNumber < this->numberOfFluidNodes; nodeNumber++)
+ fluidNodeIndices[nodeNumber] = this->fluidNodeIndices[nodeNumber];
+}
+
+
void GridImp::print() const
{
printf("min: (%2.4f, %2.4f, %2.4f), max: (%2.4f, %2.4f, %2.4f), size: %d, delta: %2.4f\n", startX, startY, startZ,
diff --git a/src/gpu/GridGenerator/grid/GridImp.h b/src/gpu/GridGenerator/grid/GridImp.h
index 2f4d044fb..06dc501e2 100644
--- a/src/gpu/GridGenerator/grid/GridImp.h
+++ b/src/gpu/GridGenerator/grid/GridImp.h
@@ -86,8 +86,8 @@ private:
int *neighborIndexX, *neighborIndexY, *neighborIndexZ, *neighborIndexNegative;
int *sparseIndices;
- uint *geoFluidNodes;
- uint geoFluidSize;
+ std::vector<uint> fluidNodeIndices;
+ uint numberOfFluidNodes;
uint *qIndices; //maps from matrix index to qIndex
real *qValues;
@@ -316,10 +316,10 @@ public:
void repairCommunicationInices(int direction) override;
- // needed for CUDA Streams MultiGPU
- void findMatrixIDsGEO_FLUID() override;
- uint getGeoFluidSize() const override;
- const uint* getGeoFluidNodes() const override;
+ void findFluidNodeIndices() override;
+ uint getNumberOfFluidNodes() const override;
+ CUDA_HOST void getFluidNodeIndices(uint *fluidNodeIndices) const;
+
public:
diff --git a/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridProvider.cpp b/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridProvider.cpp
index 957935756..22d3d3a38 100644
--- a/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridProvider.cpp
+++ b/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridProvider.cpp
@@ -29,6 +29,12 @@ void GridProvider::setNumberOfNodes(const int numberOfNodes, const int level) co
para->getParD(level)->mem_size_int_SP = sizeof(uint) * para->getParD(level)->size_Mat_SP;
}
+void GridProvider::setNumberOfFluidNodes(const int numberOfNodes, const int level) const
+{
+ para->getParH(level)->numberOfFluidNodes = numberOfNodes;
+ para->getParD(level)->numberOfFluidNodes = numberOfNodes;
+}
+
void GridProvider::setInitalNodeValues(const int numberOfNodes, const int level) const
{
for (int j = 1; j <= numberOfNodes; j++)
diff --git a/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridProvider.h b/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridProvider.h
index 70a2a8ec6..d8f442115 100644
--- a/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridProvider.h
+++ b/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridProvider.h
@@ -25,6 +25,8 @@ public:
virtual void allocArrays_BoundaryValues() = 0;
virtual void allocArrays_BoundaryQs() = 0;
virtual void allocArrays_OffsetScale() = 0;
+ virtual void allocArrays_fluidNodeIndices() = 0;
+
virtual void setDimensions() = 0;
virtual void setBoundingBox() = 0;
virtual void initPeriodicNeigh(std::vector<std::vector<std::vector<unsigned int> > > periodV, std::vector<std::vector<unsigned int> > periodIndex, std::string way) = 0;
@@ -39,6 +41,7 @@ public:
protected:
void setNumberOfNodes(const int numberOfNodes, const int level) const;
+ void setNumberOfFluidNodes(const int numberOfNodes, const int level) const;
virtual void setInitalNodeValues(const int numberOfNodes, const int level) const;
void setPressSizePerLevel(int level, int sizePerLevel) const;
diff --git a/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridReaderFiles/GridReader.cpp b/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridReaderFiles/GridReader.cpp
index 940dbfa61..c2e19b659 100644
--- a/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridReaderFiles/GridReader.cpp
+++ b/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridReaderFiles/GridReader.cpp
@@ -216,6 +216,10 @@ void GridReader::allocArrays_OffsetScale()
std::cout << "-----Ende OffsetScale------" << std::endl;
}
+void GridReader::allocArrays_fluidNodeIndices() {
+ // TODO
+}
+
void GridReader::setPressureValues(int channelSide) const
{
diff --git a/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridReaderFiles/GridReader.h b/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridReaderFiles/GridReader.h
index f7a4c4306..cbb0fac42 100644
--- a/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridReaderFiles/GridReader.h
+++ b/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridReaderFiles/GridReader.h
@@ -39,6 +39,7 @@ public:
void allocArrays_CoordNeighborGeo() override;
void allocArrays_BoundaryValues() override;
void allocArrays_OffsetScale() override;
+ void allocArrays_fluidNodeIndices() override;
void initalValuesDomainDecompostion(int level);
diff --git a/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridReaderGenerator/GridGenerator.cpp b/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridReaderGenerator/GridGenerator.cpp
index 3ce1093d1..ec7056460 100644
--- a/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridReaderGenerator/GridGenerator.cpp
+++ b/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridReaderGenerator/GridGenerator.cpp
@@ -87,6 +87,16 @@ void GridGenerator::allocArrays_CoordNeighborGeo()
std::cout << "-----finish Coord, Neighbor, Geo------" << std::endl;
}
+void GridGenerator::allocArrays_fluidNodeIndices() {
+ const uint numberOfLevels = builder->getNumberOfGridLevels();
+ for (uint level = 0; level < numberOfLevels; level++) {
+ setNumberOfFluidNodes(builder->getNumberOfFluidNodes(level), level);
+ cudaMemoryManager->cudaAllocFluidNodeIndices(level);
+ builder->getFluidNodeIndices(para->getParH(level)->fluidNodeIndices, level);
+ cudaMemoryManager->cudaCopyFluidNodeIndices(level);
+ }
+}
+
void GridGenerator::allocArrays_BoundaryValues()
{
std::cout << "------read BoundaryValues------" << std::endl;
diff --git a/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridReaderGenerator/GridGenerator.h b/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridReaderGenerator/GridGenerator.h
index 79f7f2b6a..096bb424d 100644
--- a/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridReaderGenerator/GridGenerator.h
+++ b/src/gpu/VirtualFluids_GPU/DataStructureInitializer/GridReaderGenerator/GridGenerator.h
@@ -29,6 +29,7 @@ public:
void allocArrays_BoundaryValues() override;
void allocArrays_BoundaryQs() override;
void allocArrays_OffsetScale() override;
+ void allocArrays_fluidNodeIndices() override;
virtual void setDimensions() override;
virtual void setBoundingBox() override;
diff --git a/src/gpu/VirtualFluids_GPU/GPU/CudaMemoryManager.cpp b/src/gpu/VirtualFluids_GPU/GPU/CudaMemoryManager.cpp
index 8e2818186..c101b9295 100644
--- a/src/gpu/VirtualFluids_GPU/GPU/CudaMemoryManager.cpp
+++ b/src/gpu/VirtualFluids_GPU/GPU/CudaMemoryManager.cpp
@@ -2606,6 +2606,7 @@ void CudaMemoryManager::cudaFreeProcessNeighborADZ(int lev, unsigned int process
checkCudaErrors( cudaFreeHost(parameter->getParH(lev)->recvProcessNeighborADZ[processNeighbor].index ));
checkCudaErrors( cudaFreeHost(parameter->getParH(lev)->recvProcessNeighborADZ[processNeighbor].f[0] ));
}
+
void CudaMemoryManager::cudaAlloc2ndOrderDerivitivesIsoTest(int lev)
{
//Host
@@ -2644,9 +2645,27 @@ void CudaMemoryManager::cudaFree2ndOrderDerivitivesIsoTest(int lev)
}
+void CudaMemoryManager::cudaAllocFluidNodeIndices(int lev) {
+ uint mem_size_geo_fluid_nodes = sizeof(uint) * parameter->getParH(lev)->numberOfFluidNodes;
+ // Host
+ checkCudaErrors(cudaMallocHost((void **)&(parameter->getParH(lev)->fluidNodeIndices), mem_size_geo_fluid_nodes));
+ // Device
+ checkCudaErrors(cudaMalloc((void **)&(parameter->getParD(lev)->fluidNodeIndices), mem_size_geo_fluid_nodes));
+ //////////////////////////////////////////////////////////////////////////
+ double tmp = (double)mem_size_geo_fluid_nodes;
+ setMemsizeGPU(tmp, false);
+}
+void CudaMemoryManager::cudaCopyFluidNodeIndices(int lev) {
+ uint mem_size_geo_fluid_nodes = sizeof(uint) * parameter->getParH(lev)->numberOfFluidNodes;
+ checkCudaErrors(cudaMemcpy(parameter->getParD(lev)->fluidNodeIndices,
+ parameter->getParH(lev)->fluidNodeIndices,
+ mem_size_geo_fluid_nodes, cudaMemcpyHostToDevice));
+}
-
+void CudaMemoryManager::cudaFreeFluidNodeIndices(int lev) {
+ checkCudaErrors(cudaFreeHost(parameter->getParH(lev)->fluidNodeIndices));
+}
diff --git a/src/gpu/VirtualFluids_GPU/GPU/CudaMemoryManager.h b/src/gpu/VirtualFluids_GPU/GPU/CudaMemoryManager.h
index 4853205f5..0588b2c32 100644
--- a/src/gpu/VirtualFluids_GPU/GPU/CudaMemoryManager.h
+++ b/src/gpu/VirtualFluids_GPU/GPU/CudaMemoryManager.h
@@ -323,7 +323,9 @@ public:
void cudaFreeProcessNeighborADZ(int lev, unsigned int processNeighbor);
-
+ void cudaAllocFluidNodeIndices(int lev);
+ void cudaCopyFluidNodeIndices(int lev);
+ void cudaFreeFluidNodeIndices(int lev);
diff --git a/src/gpu/VirtualFluids_GPU/LBM/Simulation.cpp b/src/gpu/VirtualFluids_GPU/LBM/Simulation.cpp
index b136ef7c6..cbfc3072a 100644
--- a/src/gpu/VirtualFluids_GPU/LBM/Simulation.cpp
+++ b/src/gpu/VirtualFluids_GPU/LBM/Simulation.cpp
@@ -86,6 +86,7 @@ void Simulation::init(SPtr<Parameter> para, SPtr<GridProvider> gridProvider, std
gridProvider->allocAndCopyForcing();
gridProvider->allocAndCopyQuadricLimiters();
+ gridProvider->allocArrays_fluidNodeIndices();
gridProvider->setDimensions();
gridProvider->setBoundingBox();
diff --git a/src/gpu/VirtualFluids_GPU/Parameter/NodeIndicesMultiGPU.cpp b/src/gpu/VirtualFluids_GPU/Parameter/NodeIndicesMultiGPU.cpp
deleted file mode 100644
index 43884c93a..000000000
--- a/src/gpu/VirtualFluids_GPU/Parameter/NodeIndicesMultiGPU.cpp
+++ /dev/null
@@ -1,18 +0,0 @@
-#include "NodeIndicesMultiGPU.h"
-
-NodeIndicesMultiGPU::NodeIndicesMultiGPU(const std::vector<uint> *geoFluidSizes,
- const std::vector<const uint*> *geoFluidNodesIndices)
-{
- this->geoFluidSizes = geoFluidSizes;
- this->geoFluidNodeIndices = geoFluidNodesIndices;
-}
-
-uint NodeIndicesMultiGPU::getGeoFluidSize(uint level)
-{
- return (*this->geoFluidSizes)[level];
-}
-
-const uint* NodeIndicesMultiGPU::getGeoFluidNodeIndices(uint level)
-{
- return (*this->geoFluidNodeIndices)[level];
-}
diff --git a/src/gpu/VirtualFluids_GPU/Parameter/NodeIndicesMultiGPU.h b/src/gpu/VirtualFluids_GPU/Parameter/NodeIndicesMultiGPU.h
deleted file mode 100644
index 817968f1b..000000000
--- a/src/gpu/VirtualFluids_GPU/Parameter/NodeIndicesMultiGPU.h
+++ /dev/null
@@ -1,20 +0,0 @@
-#ifndef INDICES_MULTIGPU_H
-#define INDICES_MULTIGPU_H
-
-#include <vector>
-#include <memory>
-#include "basics/Core/DataTypes.h"
-
-class NodeIndicesMultiGPU
-{
- const std::vector<uint> *geoFluidSizes;
- const std::vector<const uint*> *geoFluidNodeIndices;
-
-public:
- NodeIndicesMultiGPU(const std::vector<uint> *geoFluidSizes, const std::vector<const uint*> * geoFluidNodes);
-
- uint getGeoFluidSize(uint level);
- const uint *getGeoFluidNodeIndices(uint level);
-};
-
-#endif
\ No newline at end of file
diff --git a/src/gpu/VirtualFluids_GPU/Parameter/Parameter.cpp b/src/gpu/VirtualFluids_GPU/Parameter/Parameter.cpp
index 1ce504c09..bf03b915e 100644
--- a/src/gpu/VirtualFluids_GPU/Parameter/Parameter.cpp
+++ b/src/gpu/VirtualFluids_GPU/Parameter/Parameter.cpp
@@ -2485,15 +2485,4 @@ real Parameter::TrafoZtoMGsWorld(int CoordZ, int level)
return temp;
}
-// CUDA Streams Multi GPU
-void Parameter::setNodeIndicesMultiGPU(std::shared_ptr<NodeIndicesMultiGPU> nodeIndicesMultiGPU)
-{
- this->nodeIndicesMultiGPU = nodeIndicesMultiGPU;
-}
-
-std::shared_ptr<NodeIndicesMultiGPU> Parameter::getNodeIndicesMultiGPU()
-{
- return this->nodeIndicesMultiGPU;
-}
-
////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////
diff --git a/src/gpu/VirtualFluids_GPU/Parameter/Parameter.h b/src/gpu/VirtualFluids_GPU/Parameter/Parameter.h
index 6665190a4..80b2fac29 100644
--- a/src/gpu/VirtualFluids_GPU/Parameter/Parameter.h
+++ b/src/gpu/VirtualFluids_GPU/Parameter/Parameter.h
@@ -41,8 +41,6 @@
#include "LBM/D3Q27.h"
#include "LBM/LB.h"
-#include "VirtualFluids_GPU/Parameter/NodeIndicesMultiGPU.h"
-
#include "VirtualFluids_GPU_export.h"
@@ -309,6 +307,9 @@ struct LBMSimulationParameter
std::vector<ProcessNeighborF3> recvProcessNeighborF3Y;
std::vector<ProcessNeighborF3> recvProcessNeighborF3Z;
////////////////////////////////////////////////////////////////////////////
+
+ uint *fluidNodeIndices;
+ uint numberOfFluidNodes;
};
class VIRTUALFLUIDS_GPU_EXPORT Parameter
@@ -837,12 +838,6 @@ private:
// initial condition
std::function<void(real, real, real, real &, real &, real &, real &)> initialCondition;
-
- // CUDA Streams Multi GPU
- std::shared_ptr<NodeIndicesMultiGPU> nodeIndicesMultiGPU;
-public:
- void setNodeIndicesMultiGPU(std::shared_ptr<NodeIndicesMultiGPU> nodeIndicesMultiGPU);
- std::shared_ptr<NodeIndicesMultiGPU> getNodeIndicesMultiGPU();
};
#endif
--
GitLab