Merge branch 'feature/gks' into 'develop'

Checking GKS development version

See merge request irmb/VirtualFluids_dev!79

CMake/3rd/vtk.cmake

@@ -3,9 +3,9 @@
 # VTK_DIR needs to bet set to the VTK build directory in the config file.
 #########################################################################
 find_package(VTK REQUIRED)
-    vf_get_library_name(library_name)
+vf_get_library_name(library_name)
 
-    include(${VTK_USE_FILE})
-    target_include_directories(${library_name} PRIVATE ${VTK_INCLUDE_DIRS})
+include(${VTK_USE_FILE})
+target_include_directories(${library_name} PRIVATE ${VTK_INCLUDE_DIRS})
 
-    target_link_libraries(${library_name} PRIVATE ${VTK_LIBRARIES})
+target_link_libraries(${library_name} PRIVATE ${VTK_LIBRARIES})

CMakeLists.txt

@@ -49,7 +49,6 @@ option(BUILD_USE_OPENMP "Build VirtualFluids with openmp" ON)
 option(BUILD_USE_BOOST "Build VirtualFluids with boost" OFF)
 
 # vf gpu
-option(BUILD_VF_GPU          "Build VirtualFluids GPU"     ON )
 option(BUILD_VF_GKS          "Build VirtualFluids GKS"     OFF )
 option(BUILD_VF_TRAFFIC      "Build VirtualFluids Traffic" OFF)
 option(BUILD_JSONCPP         "Builds json cpp "            OFF)
@@ -97,7 +96,7 @@ IF( BUILD_VF_DOUBLE_ACCURACY )
     list(APPEND VF_COMPILER_DEFINITION VF_DOUBLE_ACCURACY)
 ENDIF()
 
-if(BUILD_VF_GPU)
+if(BUILD_VF_GPU OR BUILD_VF_GKS)
     include(CheckLanguage)
     check_language(CUDA)
 
@@ -192,6 +191,6 @@ add_subdirectory(src/lbm)
 if (BUILD_VF_CPU)
     include (cpu.cmake)
 endif()
-if(BUILD_VF_GPU)
+if(BUILD_VF_GPU OR BUILD_VF_GKS)
     include (gpu.cmake)
 endif()

apps/gpu/GKS/Flame7cm/CMakeLists.txt

-PROJECT(Flame7cm)
+PROJECT(Flame7cm LANGUAGES CUDA CXX)
 
-vf_add_library(BUILDTYPE binary PRIVATE_LINK basics GridGenerator GksMeshAdapter GksVtkAdapter GksGpu FILES Flame7cm.cpp )
+vf_add_library(BUILDTYPE binary PRIVATE_LINK basics GridGenerator GksMeshAdapter GksVtkAdapter GksGpu MPI::MPI_CXX FILES Flame7cm.cpp )
 
-include (${VF_CMAKE_DIR}/3rd/cuda.cmake)
-include (${VF_CMAKE_DIR}/3rd/mpi.cmake)
+set_source_files_properties(Flame7cm.cpp PROPERTIES LANGUAGE CUDA)

apps/gpu/GKS/Flame7cm/Flame7cm.cpp

@@ -152,11 +152,10 @@ void thermalCavity( std::string path, std::string simulationName, uint _gpuIndex
 
     ////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////
 
-    auto gridFactory = GridFactory::make();
-    gridFactory->setGridStrategy(Device::CPU);
-    gridFactory->setTriangularMeshDiscretizationMethod(TriangularMeshDiscretizationMethod::POINT_IN_OBJECT);
+    // auto gridFactory = GridFactory::make();
+    // gridFactory->setTriangularMeshDiscretizationMethod(TriangularMeshDiscretizationMethod::POINT_IN_OBJECT);
 
-    auto gridBuilder = MultipleGridBuilder::makeShared(gridFactory);
+    auto gridBuilder = MultipleGridBuilder::makeShared();
 
     ////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////
 
@@ -476,14 +475,14 @@ int main( int argc, char* argv[])
     {
         thermalCavity( path, simulationName, gpuIndex, nx, useTempLimiter, restartIter );
     }
-    catch (const std::exception& e)
-    {     
-        *logging::out << logging::Logger::LOGGER_ERROR << e.what() << "\n";
-    }
     catch (const std::bad_alloc& e)
     {  
         *logging::out << logging::Logger::LOGGER_ERROR << "Bad Alloc:" << e.what() << "\n";
     }
+    catch (const std::exception& e)
+    {     
+        *logging::out << logging::Logger::LOGGER_ERROR << e.what() << "\n";
+    }
     catch (...)
     {
         *logging::out << logging::Logger::LOGGER_ERROR << "Unknown exception!\n";

src/gpu/GksGpu/Analyzer/EnstrophyAnalyzer.cu

@@ -59,6 +59,8 @@ bool EnstrophyAnalyzer::run(uint iter)
     this->enstrophyTimeSeries.push_back( EnstrophyTmp );
 
     //*logging::out << logging::Logger::INFO_HIGH << "EKin = " << EKin << "\n";
+
+    return true;
 }
 
 ////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////

src/gpu/GksGpu/Analyzer/HeatFluxAnalyzer.cu

@@ -63,6 +63,8 @@ bool HeatFluxAnalyzer::run(uint iter, Parameters parameters)
     this->heatFluxTimeSeries.push_back( q / qIdeal );
 
     if( iter % this->outputIter == 0 ) *logging::out << logging::Logger::INFO_HIGH << "q = " << q / qIdeal << "\n";
+
+    return true;
 }
 
 //////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////

src/gpu/GksGpu/Analyzer/TurbulenceAnalyzer.cu

@@ -51,6 +51,8 @@ bool TurbulenceAnalyzer::run(uint iter, Parameters parameters)
     getLastCudaError("TurbulenceAnalyzer::run(uint iter, Parameters parameters)");
 
     this->counter++;
+
+    return true;
 }
 
 ////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////

src/gpu/GksGpu/BoundaryConditions/CreepingMassFlux.cu

@@ -91,7 +91,7 @@ __host__ __device__ inline void boundaryConditionFunction(const DataBaseStruct&
                                                           const uint startIndex,
                                                           const uint index)
 {
-    uint ghostCellIdx  = boundaryCondition.ghostCells [ startIndex + index ];
+    // uint ghostCellIdx  = boundaryCondition.ghostCells [ startIndex + index ];
     uint domainCellIdx = boundaryCondition.domainCells[ startIndex + index ];
 
     ////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////

src/gpu/GksGpu/BoundaryConditions/HeatFlux.cu

@@ -91,7 +91,7 @@ __host__ __device__ inline void boundaryConditionFunction(const DataBaseStruct&
                                                           const uint startIndex,
                                                           const uint index)
 {
-    uint ghostCellIdx  = boundaryCondition.ghostCells [ startIndex + index ];
+    // uint ghostCellIdx  = boundaryCondition.ghostCells [ startIndex + index ];
     uint domainCellIdx = boundaryCondition.domainCells[ startIndex + index ];
 
     ////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////

src/gpu/GksGpu/BoundaryConditions/Inflow.cu

@@ -86,7 +86,7 @@ __host__ __device__ inline void boundaryConditionFunction(const DataBaseStruct&
 {
     uint ghostCellIdx  = boundaryCondition.ghostCells [ startIndex + index ];
     uint domainCellIdx = boundaryCondition.domainCells[ startIndex + index ];
-    uint secondCellIdx = boundaryCondition.secondCells[ startIndex + index ];
+    // uint secondCellIdx = boundaryCondition.secondCells[ startIndex + index ];
 
     PrimitiveVariables ghostCellPrim;
     {

src/gpu/GksGpu/BoundaryConditions/MassCompensation.cu

@@ -91,7 +91,7 @@ __host__ __device__ inline void boundaryConditionFunction(const DataBaseStruct&
                                                           const uint startIndex,
                                                           const uint index)
 {
-    uint ghostCellIdx  = boundaryCondition.ghostCells [ startIndex + index ];
+    // uint ghostCellIdx  = boundaryCondition.ghostCells [ startIndex + index ];
     uint domainCellIdx = boundaryCondition.domainCells[ startIndex + index ];
 
     ////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////

src/gpu/GksGpu/CMakeLists.txt

 project(GksGpu LANGUAGES CUDA CXX)
 
-vf_add_library(PRIVATE_LINK basics GksMeshAdapter OpenMP::OpenMP_CXX MPI::MPI_CXX)
+vf_add_library(PRIVATE_LINK basics lbmCuda GksMeshAdapter OpenMP::OpenMP_CXX MPI::MPI_CXX)
+
+target_include_directories(GksGpu PRIVATE "${VF_THIRD_DIR}/cuda_samples/")
+target_compile_options(GksGpu PRIVATE "-fPIC")
\ No newline at end of file

src/gpu/GksGpu/CellProperties/CellProperties.cuh

@@ -4,9 +4,13 @@
 #ifdef __CUDACC__
 #include <cuda_runtime.h>
 #else
+#ifndef __host__
 #define __host__
+#endif
+#ifndef __device__
 #define __device__
 #endif
+#endif
 
 //////////////////////////////////////////////////////////////////////////
 

src/gpu/GksGpu/CellUpdate/Reaction.cuh

@@ -36,14 +36,14 @@ inline __host__ __device__ real getTurbulentViscosityDeardorff(const DataBaseStr
         real uHead = c1o2 * prim.U;
 
         {
-            uint neighborCellIndex = dataBase.cellToCell[CELL_TO_CELL(cellIndex, 0, dataBase.numberOfCells)];
+            // uint neighborCellIndex = dataBase.cellToCell[CELL_TO_CELL(cellIndex, 0, dataBase.numberOfCells)];
             readCellData(cellIndex, dataBase, neighborCons);
             neighborPrim = toPrimitiveVariables(neighborCons, parameters.K);
 
             uHead += c1o4 * neighborPrim.U;
         }
         {
-            uint neighborCellIndex = dataBase.cellToCell[CELL_TO_CELL(cellIndex, 1, dataBase.numberOfCells)];
+            // uint neighborCellIndex = dataBase.cellToCell[CELL_TO_CELL(cellIndex, 1, dataBase.numberOfCells)];
             readCellData(cellIndex, dataBase, neighborCons);
             neighborPrim = toPrimitiveVariables(neighborCons, parameters.K);
 
@@ -57,14 +57,14 @@ inline __host__ __device__ real getTurbulentViscosityDeardorff(const DataBaseStr
         real vHead = c1o2 * prim.V;
 
         {
-            uint neighborCellIndex = dataBase.cellToCell[CELL_TO_CELL(cellIndex, 2, dataBase.numberOfCells)];
+            // uint neighborCellIndex = dataBase.cellToCell[CELL_TO_CELL(cellIndex, 2, dataBase.numberOfCells)];
             readCellData(cellIndex, dataBase, neighborCons);
             neighborPrim = toPrimitiveVariables(neighborCons, parameters.K);
 
             vHead += c1o4 * neighborPrim.V;
         }
         {
-            uint neighborCellIndex = dataBase.cellToCell[CELL_TO_CELL(cellIndex, 3, dataBase.numberOfCells)];
+            // uint neighborCellIndex = dataBase.cellToCell[CELL_TO_CELL(cellIndex, 3, dataBase.numberOfCells)];
             readCellData(cellIndex, dataBase, neighborCons);
             neighborPrim = toPrimitiveVariables(neighborCons, parameters.K);
 
@@ -78,14 +78,14 @@ inline __host__ __device__ real getTurbulentViscosityDeardorff(const DataBaseStr
         real wHead = c1o2 * prim.W;
 
         {
-            uint neighborCellIndex = dataBase.cellToCell[CELL_TO_CELL(cellIndex, 4, dataBase.numberOfCells)];
+            // uint neighborCellIndex = dataBase.cellToCell[CELL_TO_CELL(cellIndex, 4, dataBase.numberOfCells)];
             readCellData(cellIndex, dataBase, neighborCons);
             neighborPrim = toPrimitiveVariables(neighborCons, parameters.K);
 
             wHead += c1o4 * neighborPrim.W;
         }
         {
-            uint neighborCellIndex = dataBase.cellToCell[CELL_TO_CELL(cellIndex, 5, dataBase.numberOfCells)];
+            // uint neighborCellIndex = dataBase.cellToCell[CELL_TO_CELL(cellIndex, 5, dataBase.numberOfCells)];
             readCellData(cellIndex, dataBase, neighborCons);
             neighborPrim = toPrimitiveVariables(neighborCons, parameters.K);
 

src/gpu/GksGpu/DataBase/DataBase.cpp

@@ -14,6 +14,10 @@
 #include "GksMeshAdapter/GksMeshAdapter.h"
 #include "Communication/Communicator.h"
 
+#include <lbm/constants/NumericConstants.h>
+
+using namespace vf::lbm::constant;
+
 namespace GksGpu {
 
 DataBase::DataBase( std::string type ) 

src/gpu/GksGpu/FlowStateData/AccessDeviceData.cuh

@@ -4,9 +4,13 @@
 #ifdef __CUDACC__
 #include <cuda_runtime.h>
 #else
+#ifndef __host__
 #define __host__
+#endif
+#ifndef __device__
 #define __device__
 #endif
+#endif
 
 #include "Core/DataTypes.h"
 #include "Core/RealConstants.h"

src/gpu/GksGpu/FlowStateData/FlowStateData.cuh

@@ -4,15 +4,22 @@
 #ifdef __CUDACC__
 #include <cuda_runtime.h>
 #else
+#ifndef __host__
 #define __host__
+#endif
+#ifndef __device__
 #define __device__
 #endif
+#endif
 
 #include "Core/DataTypes.h"
-#include "Core/RealConstants.h"
 
 #include "Definitions/PassiveScalar.h"
 
+#include <lbm/constants/NumericConstants.h>
+
+using namespace vf::lbm::constant;
+
 namespace GksGpu {
 
 //////////////////////////////////////////////////////////////////////////

src/gpu/GksGpu/FlowStateData/FlowStateDataConversion.cuh

@@ -4,9 +4,13 @@
 #ifdef __CUDACC__
 #include <cuda_runtime.h>
 #else
+#ifndef __host__
 #define __host__
+#endif
+#ifndef __device__
 #define __device__
 #endif
+#endif
 
 #include "Core/DataTypes.h"
 #include "Core/RealConstants.h"

src/gpu/GksGpu/FlowStateData/HeatCapacities.cuh

 //#ifndef HeatCapacities_H
 //#define HeatCapacities_H
 //
-//#ifdef __CUDACC__
-//#include <cuda_runtime.h>
-//#else
-//#define __host__
-//#define __device__
-//#endif
+// #ifdef __CUDACC__
+// #include <cuda_runtime.h>
+// #else
+// #ifndef __host__
+// #define __host__
+// #endif
+// #ifndef __device__
+// #define __device__
+// #endif
+// #endif
 //
 //#include "Core/DataTypes.h"
 //#include "Core/RealConstants.h"

src/gpu/GksGpu/FlowStateData/ThermalDependencies.cuh

@@ -4,9 +4,13 @@
 #ifdef __CUDACC__
 #include <cuda_runtime.h>
 #else
+#ifndef __host__
 #define __host__
+#endif
+#ifndef __device__
 #define __device__
 #endif
+#endif
 
 #include <math.h>