diff --git a/src/gpu/core/Kernel/Compressible/NavierStokes/B15/B15CompressibleNavierStokesBGKplus.cu b/src/gpu/core/Kernel/Compressible/NavierStokes/B15/B15CompressibleNavierStokesBGKplus.cu index 6f6befbd2cca54038ebe7a3b27747fe9b47cec1b..6f51eee72ee8b7db62cf43bb827b6d117cf26ec9 100644 --- a/src/gpu/core/Kernel/Compressible/NavierStokes/B15/B15CompressibleNavierStokesBGKplus.cu +++ b/src/gpu/core/Kernel/Compressible/NavierStokes/B15/B15CompressibleNavierStokesBGKplus.cu @@ -30,7 +30,7 @@ B15CompressibleNavierStokesBGKplus::B15CompressibleNavierStokesBGKplus(std::shar this->para = para; this->level = level; - myPreProcessorTypes.push_back(InitCompSP27); + myPreProcessorTypes.push_back(InitNavierStokesCompressible); } diff --git a/src/gpu/core/Kernel/Compressible/NavierStokes/B92/B92CompressibleNavierStokes.cu b/src/gpu/core/Kernel/Compressible/NavierStokes/B92/B92CompressibleNavierStokes.cu index 5f2d069459a3cdfc61b79771316f2d81c3039f98..1dba9343c90a213fee5193d417b135e03e855e3c 100644 --- a/src/gpu/core/Kernel/Compressible/NavierStokes/B92/B92CompressibleNavierStokes.cu +++ b/src/gpu/core/Kernel/Compressible/NavierStokes/B92/B92CompressibleNavierStokes.cu @@ -30,7 +30,7 @@ B92CompressibleNavierStokes::B92CompressibleNavierStokes(std::shared_ptr<Paramet this->para = para; this->level = level; - myPreProcessorTypes.push_back(InitCompSP27); + myPreProcessorTypes.push_back(InitNavierStokesCompressible); } diff --git a/src/gpu/core/Kernel/Compressible/NavierStokes/K15/K15CompressibleNavierStokes.cu b/src/gpu/core/Kernel/Compressible/NavierStokes/K15/K15CompressibleNavierStokes.cu index 2ba1af7a34b219d8f06e44708c354c609efc819f..4cb59734754105e50a9d9ea5d707fa493404f1d5 100644 --- a/src/gpu/core/Kernel/Compressible/NavierStokes/K15/K15CompressibleNavierStokes.cu +++ b/src/gpu/core/Kernel/Compressible/NavierStokes/K15/K15CompressibleNavierStokes.cu @@ -48,7 +48,7 @@ K15CompressibleNavierStokes::K15CompressibleNavierStokes(std::shared_ptr<Paramet this->para = para; this->level = level; - myPreProcessorTypes.push_back(InitCompSP27); + myPreProcessorTypes.push_back(InitNavierStokesCompressible); } \ No newline at end of file diff --git a/src/gpu/core/Kernel/Compressible/NavierStokes/K17/K17CompressibleNavierStokes.cu b/src/gpu/core/Kernel/Compressible/NavierStokes/K17/K17CompressibleNavierStokes.cu index 60d4a031451b73e1c4b8de1b317d6488d32e69e1..3fd0f5786e5ed9562f591bd4c264c06fb18e3f4d 100644 --- a/src/gpu/core/Kernel/Compressible/NavierStokes/K17/K17CompressibleNavierStokes.cu +++ b/src/gpu/core/Kernel/Compressible/NavierStokes/K17/K17CompressibleNavierStokes.cu @@ -116,7 +116,7 @@ void K17CompressibleNavierStokes<turbulenceModel>::runOnIndices(const unsigned i template <vf::lbm::TurbulenceModel turbulenceModel> K17CompressibleNavierStokes<turbulenceModel>::K17CompressibleNavierStokes(std::shared_ptr<Parameter> para, int level) : KernelImp(para, level) { - myPreProcessorTypes.push_back(InitCompSP27); + myPreProcessorTypes.push_back(InitNavierStokesCompressible); this->cudaGrid = vf::cuda::CudaGrid(para->getParD(level)->numberofthreads, para->getParD(level)->numberOfNodes); this->kernelUsesFluidNodeIndices = true; diff --git a/src/gpu/core/Kernel/Incompressible/NavierStokes/B15/B15IncompressibleNavierStokesBGKplus.cu b/src/gpu/core/Kernel/Incompressible/NavierStokes/B15/B15IncompressibleNavierStokesBGKplus.cu index 251cf7544eae9710d3460cacfdd4b301c9c48d79..cc0b6d43e4b7bd5fb329a0cbbf1ba7e33f466a75 100644 --- a/src/gpu/core/Kernel/Incompressible/NavierStokes/B15/B15IncompressibleNavierStokesBGKplus.cu +++ b/src/gpu/core/Kernel/Incompressible/NavierStokes/B15/B15IncompressibleNavierStokesBGKplus.cu @@ -30,7 +30,7 @@ B15IncompressibleNavierStokesBGKplus::B15IncompressibleNavierStokesBGKplus(std:: this->para = para; this->level = level; - myPreProcessorTypes.push_back(InitSP27); + myPreProcessorTypes.push_back(InitNavierStokesIncompressible); } diff --git a/src/gpu/core/Kernel/Incompressible/NavierStokes/B92/B92IncompressibleNavierStokes.cu b/src/gpu/core/Kernel/Incompressible/NavierStokes/B92/B92IncompressibleNavierStokes.cu index 4ff8d0d6ce700b0245171fe31790e320bcc9e725..49210b1fe31a0addf91401f3c86161846787b96c 100644 --- a/src/gpu/core/Kernel/Incompressible/NavierStokes/B92/B92IncompressibleNavierStokes.cu +++ b/src/gpu/core/Kernel/Incompressible/NavierStokes/B92/B92IncompressibleNavierStokes.cu @@ -30,7 +30,7 @@ B92IncompressibleNavierStokes::B92IncompressibleNavierStokes(std::shared_ptr<Par this->para = para; this->level = level; - myPreProcessorTypes.push_back(InitSP27); + myPreProcessorTypes.push_back(InitNavierStokesIncompressible); } diff --git a/src/gpu/core/Kernel/Incompressible/NavierStokes/K15/K15IncompressibleNavierStokes.cu b/src/gpu/core/Kernel/Incompressible/NavierStokes/K15/K15IncompressibleNavierStokes.cu index 40247d0f664ff842246a1a9066a69bcf5dff117f..5d64f3216ba65389a03941b55f61a7091170e0ea 100644 --- a/src/gpu/core/Kernel/Incompressible/NavierStokes/K15/K15IncompressibleNavierStokes.cu +++ b/src/gpu/core/Kernel/Incompressible/NavierStokes/K15/K15IncompressibleNavierStokes.cu @@ -30,7 +30,7 @@ K15IncompressibleNavierStokes::K15IncompressibleNavierStokes(std::shared_ptr<Par this->para = para; this->level = level; - myPreProcessorTypes.push_back(InitSP27); + myPreProcessorTypes.push_back(InitNavierStokesIncompressible); } diff --git a/src/gpu/core/KernelManager/ADKernelManager.cpp b/src/gpu/core/KernelManager/ADKernelManager.cpp index 0a531bd36eedbd0f90a1ef7f94aa29e5f1dd4ded..ac504a83944c56a746699d4c2c79944583b963f0 100644 --- a/src/gpu/core/KernelManager/ADKernelManager.cpp +++ b/src/gpu/core/KernelManager/ADKernelManager.cpp @@ -46,7 +46,7 @@ void ADKernelManager::initAD(const int level) const ////////////////////////////////////////////////////////////////////////// para->getParD(level)->isEvenTimestep = true; ////////////////////////////////////////////////////////////////////////// - InitADDev27( + initAdvectionDiffusion( para->getParD(level)->numberofthreads, para->getParD(level)->neighborX, para->getParD(level)->neighborY, @@ -62,7 +62,7 @@ void ADKernelManager::initAD(const int level) const ////////////////////////////////////////////////////////////////////////// para->getParD(level)->isEvenTimestep = false; ////////////////////////////////////////////////////////////////////////// - InitADDev27( + initAdvectionDiffusion( para->getParD(level)->numberofthreads, para->getParD(level)->neighborX, para->getParD(level)->neighborY, diff --git a/src/gpu/core/PreProcessor/PreProcessorFactory/PreProcessorFactoryImp.cpp b/src/gpu/core/PreProcessor/PreProcessorFactory/PreProcessorFactoryImp.cpp index e92be775c79b19c7bf5ca0d1840622e10cccd0fd..9f72e8515bf8c670e5391d0ad261ca45496bb5a4 100644 --- a/src/gpu/core/PreProcessor/PreProcessorFactory/PreProcessorFactoryImp.cpp +++ b/src/gpu/core/PreProcessor/PreProcessorFactory/PreProcessorFactoryImp.cpp @@ -4,11 +4,11 @@ #include "PreProcessor/PreProcessorStrategy/InitCompAD7/InitCompAD7.h" #include "PreProcessor/PreProcessorStrategy/InitCompAD27/InitCompAD27.h" -#include "PreProcessor/PreProcessorStrategy/InitCompSP27/InitCompSP27.h" -#include "PreProcessor/PreProcessorStrategy/InitF3/InitF3.h" +#include "PreProcessor/PreProcessorStrategy/InitNavierStokesCompressible/InitNavierStokesCompressible.h" +#include "PreProcessor/PreProcessorStrategy/InitK18K20NavierStokesCompressible/InitK18K20NavierStokesCompressible.h" #include "PreProcessor/PreProcessorStrategy/InitIncompAD27/InitIncompAD27.h" #include "PreProcessor/PreProcessorStrategy/InitIncompAD7/InitIncompAD7.h" -#include "PreProcessor/PreProcessorStrategy/InitSP27/InitSP27.h" +#include "PreProcessor/PreProcessorStrategy/InitNavierStokesIncompressible/InitNavierStokesIncompressible.h" std::shared_ptr<PreProcessor> PreProcessorFactoryImp::makePreProcessor(std::vector<PreProcessorType> preProcessorTypes, std::shared_ptr<Parameter> para) @@ -25,14 +25,14 @@ std::shared_ptr<PreProcessorStrategy> PreProcessorFactoryImp::makePreProcessorSt { switch (preProcessorType) { - case InitSP27: - return InitSP27::getNewInstance(para); + case InitNavierStokesIncompressible: + return InitNavierStokesIncompressible::getNewInstance(para); break; - case InitCompSP27: - return InitCompSP27::getNewInstance(para); + case InitNavierStokesCompressible: + return InitNavierStokesCompressible::getNewInstance(para); break; - case InitF3: - return InitF3::getNewInstance(para); + case InitK18K20NavierStokesCompressible: + return InitK18K20NavierStokesCompressible::getNewInstance(para); break; case InitIncompAD7: return InitIncompAD7::getNewInstance(para); diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitCompSP27/InitCompSP27.cu b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitCompSP27/InitCompSP27.cu deleted file mode 100644 index e70d5a8fe61e7c3119d83e35d81bbe742f97fa4e..0000000000000000000000000000000000000000 --- a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitCompSP27/InitCompSP27.cu +++ /dev/null @@ -1,68 +0,0 @@ -#include "InitCompSP27.h" - -#include "InitCompSP27_Device.cuh" -#include "Parameter/Parameter.h" -#include <cuda_helper/CudaGrid.h> - -std::shared_ptr<PreProcessorStrategy> InitCompSP27::getNewInstance(std::shared_ptr<Parameter> para) -{ - return std::shared_ptr<PreProcessorStrategy>(new InitCompSP27(para)); -} - -void InitCompSP27::init(int level) -{ - vf::cuda::CudaGrid grid = vf::cuda::CudaGrid(para->getParD(level)->numberofthreads, para->getParD(level)->numberOfNodes); - - if( ! para->getUseInitNeq() ) - { - LB_Init_Comp_SP_27 <<< grid.grid, grid.threads >>> ( - para->getParD(level)->neighborX, - para->getParD(level)->neighborY, - para->getParD(level)->neighborZ, - para->getParD(level)->typeOfGridNode, - para->getParD(level)->rho, - para->getParD(level)->velocityX, - para->getParD(level)->velocityY, - para->getParD(level)->velocityZ, - para->getParD(level)->numberOfNodes, - para->getParD(level)->distributions.f[0], - para->getParD(level)->isEvenTimestep); - getLastCudaError("LB_Init_Comp_SP_27 execution failed"); - } - else - { - LB_Init_Comp_Neq_SP_27 <<< grid.grid, grid.threads >>> ( - para->getParD(level)->neighborX, - para->getParD(level)->neighborY, - para->getParD(level)->neighborZ, - para->getParD(level)->neighborInverse, - para->getParD(level)->typeOfGridNode, - para->getParD(level)->rho, - para->getParD(level)->velocityX, - para->getParD(level)->velocityY, - para->getParD(level)->velocityZ, - para->getParD(level)->numberOfNodes, - para->getParD(level)->distributions.f[0], - para->getParD(level)->omega, - para->getParD(level)->isEvenTimestep); - cudaDeviceSynchronize(); - getLastCudaError("LB_Init_Comp_Neq_SP_27 execution failed"); - } - - - -} - -bool InitCompSP27::checkParameter() -{ - return false; -} - -InitCompSP27::InitCompSP27(std::shared_ptr<Parameter> para) -{ - this->para = para; -} - -InitCompSP27::InitCompSP27() -{ -} diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitCompSP27/InitCompSP27.h b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitCompSP27/InitCompSP27.h deleted file mode 100644 index 601f1a446ade70766d6b37953db77a3a92f1f67f..0000000000000000000000000000000000000000 --- a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitCompSP27/InitCompSP27.h +++ /dev/null @@ -1,23 +0,0 @@ -#ifndef INIT_COMP_SP27_H -#define INIT_COMP_SP27_H - -#include "PreProcessor/PreProcessorStrategy/PreProcessorStrategy.h" - -#include <memory> - -class Parameter; - -class InitCompSP27 : public PreProcessorStrategy -{ -public: - static std::shared_ptr<PreProcessorStrategy> getNewInstance(std::shared_ptr< Parameter> para); - void init(int level); - bool checkParameter(); - -private: - InitCompSP27(); - InitCompSP27(std::shared_ptr< Parameter> para); - std::shared_ptr< Parameter> para; -}; - -#endif \ No newline at end of file diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitCompSP27/InitCompSP27_Device.cuh b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitCompSP27/InitCompSP27_Device.cuh deleted file mode 100644 index d7f40b88bf53652de893076d529f217647c28293..0000000000000000000000000000000000000000 --- a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitCompSP27/InitCompSP27_Device.cuh +++ /dev/null @@ -1,33 +0,0 @@ -#ifndef LB_INIT_COMP_SP27_H -#define LB_INIT_COMP_SP27_H - -#include <DataTypes.h> -#include <curand.h> - -__global__ void LB_Init_Comp_SP_27(unsigned int* neighborX, - unsigned int* neighborY, - unsigned int* neighborZ, - unsigned int* geoD, - real* rho, - real* ux, - real* uy, - real* uz, - unsigned int size_Mat, - real* DD, - bool EvenOrOdd); - -__global__ void LB_Init_Comp_Neq_SP_27(unsigned int* neighborX, - unsigned int* neighborY, - unsigned int* neighborZ, - unsigned int* neighborWSB, - unsigned int* geoD, - real* rho, - real* ux, - real* uy, - real* uz, - unsigned int size_Mat, - real* DD, - real omega, - bool EvenOrOdd); - -#endif \ No newline at end of file diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitF3/InitF3.cu b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitF3/InitF3.cu deleted file mode 100644 index 82084e354ddd87c685f1913b1236bb3a01ba863b..0000000000000000000000000000000000000000 --- a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitF3/InitF3.cu +++ /dev/null @@ -1,43 +0,0 @@ -#include "InitF3.h" - -#include "InitF3_Device.cuh" -#include "Parameter/Parameter.h" -#include <cuda_helper/CudaGrid.h> - -std::shared_ptr<PreProcessorStrategy> InitF3::getNewInstance(std::shared_ptr<Parameter> para) -{ - return std::shared_ptr<PreProcessorStrategy>(new InitF3(para)); -} - -void InitF3::init(int level) -{ - vf::cuda::CudaGrid grid = vf::cuda::CudaGrid(para->getParD(level)->numberofthreads, para->getParD(level)->numberOfNodes); - - LB_Init_F3 <<< grid.grid, grid.threads >>>( - para->getParD(level)->neighborX, - para->getParD(level)->neighborY, - para->getParD(level)->neighborZ, - para->getParD(level)->typeOfGridNode, - para->getParD(level)->rho, - para->getParD(level)->velocityX, - para->getParD(level)->velocityY, - para->getParD(level)->velocityZ, - para->getParD(level)->numberOfNodes, - para->getParD(level)->g6.g[0], - para->getParD(level)->isEvenTimestep); - getLastCudaError("LB_Init_F3 execution failed"); -} - -bool InitF3::checkParameter() -{ - return false; -} - -InitF3::InitF3(std::shared_ptr<Parameter> para) -{ - this->para = para; -} - -InitF3::InitF3() -{ -} diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitF3/InitF3.h b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitF3/InitF3.h deleted file mode 100644 index c254aa906a31f61d1d82f86fa40bf048b4443fb4..0000000000000000000000000000000000000000 --- a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitF3/InitF3.h +++ /dev/null @@ -1,23 +0,0 @@ -#ifndef INITF3_H -#define INITF3_H - -#include "PreProcessor/PreProcessorStrategy/PreProcessorStrategy.h" - -#include <memory> - -class Parameter; - -class InitF3 : public PreProcessorStrategy -{ -public: - static std::shared_ptr<PreProcessorStrategy> getNewInstance(std::shared_ptr< Parameter> para); - void init(int level); - bool checkParameter(); - -private: - InitF3(); - InitF3(std::shared_ptr< Parameter> para); - std::shared_ptr< Parameter> para; -}; - -#endif \ No newline at end of file diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitF3/InitF3_Device.cuh b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitF3/InitF3_Device.cuh deleted file mode 100644 index 991a5ec75c9a05eefdfb34e0faad921cbe35d6ea..0000000000000000000000000000000000000000 --- a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitF3/InitF3_Device.cuh +++ /dev/null @@ -1,19 +0,0 @@ -#ifndef LB_INIT_F3_H -#define LB_INIT_F3_H - -#include <DataTypes.h> -#include <curand.h> - -__global__ void LB_Init_F3(unsigned int* neighborX, - unsigned int* neighborY, - unsigned int* neighborZ, - unsigned int* geoD, - real* rho, - real* ux, - real* uy, - real* uz, - unsigned int size_Mat, - real* G6, - bool EvenOrOdd); - -#endif \ No newline at end of file diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitK18K20NavierStokesCompressible/InitK18K20NavierStokesCompressible.cu b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitK18K20NavierStokesCompressible/InitK18K20NavierStokesCompressible.cu new file mode 100644 index 0000000000000000000000000000000000000000..b3f298988f64f4726743d8429f6df52da8eb8227 --- /dev/null +++ b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitK18K20NavierStokesCompressible/InitK18K20NavierStokesCompressible.cu @@ -0,0 +1,73 @@ +//======================================================================================= +// ____ ____ __ ______ __________ __ __ __ __ +// \ \ | | | | | _ \ |___ ___| | | | | / \ | | +// \ \ | | | | | |_) | | | | | | | / \ | | +// \ \ | | | | | _ / | | | | | | / /\ \ | | +// \ \ | | | | | | \ \ | | | \__/ | / ____ \ | |____ +// \ \ | | |__| |__| \__\ |__| \________/ /__/ \__\ |_______| +// \ \ | | ________________________________________________________________ +// \ \ | | | ______________________________________________________________| +// \ \| | | | __ __ __ __ ______ _______ +// \ | | |_____ | | | | | | | | | _ \ / _____) +// \ | | _____| | | | | | | | | | | \ \ \_______ +// \ | | | | |_____ | \_/ | | | | |_/ / _____ | +// \ _____| |__| |________| \_______/ |__| |______/ (_______/ +// +// This file is part of VirtualFluids. VirtualFluids is free software: you can +// redistribute it and/or modify it under the terms of the GNU General Public +// License as published by the Free Software Foundation, either version 3 of +// the License, or (at your option) any later version. +// +// VirtualFluids is distributed in the hope that it will be useful, but WITHOUT +// ANY WARRANTY; without even the implied warranty of MERCHANTABILITY or +// FITNESS FOR A PARTICULAR PURPOSE. See the GNU General Public License +// for more details. +// +// You should have received a copy of the GNU General Public License along +// with VirtualFluids (see COPYING.txt). If not, see <http://www.gnu.org/licenses/>. +// +//! \author Martin Schoenherr +//======================================================================================= +#include "InitK18K20NavierStokesCompressible.h" + +#include "InitK18K20NavierStokesCompressible_Device.cuh" +#include "Parameter/Parameter.h" +#include <cuda_helper/CudaGrid.h> + +std::shared_ptr<PreProcessorStrategy> InitK18K20NavierStokesCompressible::getNewInstance(std::shared_ptr<Parameter> para) +{ + return std::shared_ptr<PreProcessorStrategy>(new InitK18K20NavierStokesCompressible(para)); +} + +void InitK18K20NavierStokesCompressible::init(int level) +{ + vf::cuda::CudaGrid grid = vf::cuda::CudaGrid(para->getParD(level)->numberofthreads, para->getParD(level)->numberOfNodes); + + InitK18K20NavierStokesCompressible_Device <<< grid.grid, grid.threads >>>( + para->getParD(level)->neighborX, + para->getParD(level)->neighborY, + para->getParD(level)->neighborZ, + para->getParD(level)->typeOfGridNode, + para->getParD(level)->rho, + para->getParD(level)->velocityX, + para->getParD(level)->velocityY, + para->getParD(level)->velocityZ, + para->getParD(level)->numberOfNodes, + para->getParD(level)->g6.g[0], + para->getParD(level)->isEvenTimestep); + getLastCudaError("InitK18K20NavierStokesCompressible_Device execution failed"); +} + +bool InitK18K20NavierStokesCompressible::checkParameter() +{ + return false; +} + +InitK18K20NavierStokesCompressible::InitK18K20NavierStokesCompressible(std::shared_ptr<Parameter> para) +{ + this->para = para; +} + +InitK18K20NavierStokesCompressible::InitK18K20NavierStokesCompressible() +{ +} diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitK18K20NavierStokesCompressible/InitK18K20NavierStokesCompressible.h b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitK18K20NavierStokesCompressible/InitK18K20NavierStokesCompressible.h new file mode 100644 index 0000000000000000000000000000000000000000..403dc4f0cac889af75f40da0cf92f8b821692407 --- /dev/null +++ b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitK18K20NavierStokesCompressible/InitK18K20NavierStokesCompressible.h @@ -0,0 +1,53 @@ +//======================================================================================= +// ____ ____ __ ______ __________ __ __ __ __ +// \ \ | | | | | _ \ |___ ___| | | | | / \ | | +// \ \ | | | | | |_) | | | | | | | / \ | | +// \ \ | | | | | _ / | | | | | | / /\ \ | | +// \ \ | | | | | | \ \ | | | \__/ | / ____ \ | |____ +// \ \ | | |__| |__| \__\ |__| \________/ /__/ \__\ |_______| +// \ \ | | ________________________________________________________________ +// \ \ | | | ______________________________________________________________| +// \ \| | | | __ __ __ __ ______ _______ +// \ | | |_____ | | | | | | | | | _ \ / _____) +// \ | | _____| | | | | | | | | | | \ \ \_______ +// \ | | | | |_____ | \_/ | | | | |_/ / _____ | +// \ _____| |__| |________| \_______/ |__| |______/ (_______/ +// +// This file is part of VirtualFluids. VirtualFluids is free software: you can +// redistribute it and/or modify it under the terms of the GNU General Public +// License as published by the Free Software Foundation, either version 3 of +// the License, or (at your option) any later version. +// +// VirtualFluids is distributed in the hope that it will be useful, but WITHOUT +// ANY WARRANTY; without even the implied warranty of MERCHANTABILITY or +// FITNESS FOR A PARTICULAR PURPOSE. See the GNU General Public License +// for more details. +// +// You should have received a copy of the GNU General Public License along +// with VirtualFluids (see COPYING.txt). If not, see <http://www.gnu.org/licenses/>. +// +//! \author Martin Schoenherr +//======================================================================================= +#ifndef InitK18K20NavierStokesCompressible_H +#define InitK18K20NavierStokesCompressible_H + +#include "PreProcessor/PreProcessorStrategy/PreProcessorStrategy.h" + +#include <memory> + +class Parameter; + +class InitK18K20NavierStokesCompressible : public PreProcessorStrategy +{ +public: + static std::shared_ptr<PreProcessorStrategy> getNewInstance(std::shared_ptr< Parameter> para); + void init(int level); + bool checkParameter(); + +private: + InitK18K20NavierStokesCompressible(); + InitK18K20NavierStokesCompressible(std::shared_ptr< Parameter> para); + std::shared_ptr< Parameter> para; +}; + +#endif \ No newline at end of file diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitF3/InitF3_Device.cu b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitK18K20NavierStokesCompressible/InitK18K20NavierStokesCompressible_Device.cu similarity index 54% rename from src/gpu/core/PreProcessor/PreProcessorStrategy/InitF3/InitF3_Device.cu rename to src/gpu/core/PreProcessor/PreProcessorStrategy/InitK18K20NavierStokesCompressible/InitK18K20NavierStokesCompressible_Device.cu index 29050bd2807dc753051aaa78fc2d0a35fb4162ad..6beb4ad8574236857b16cebf035c2796312bdf82 100644 --- a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitF3/InitF3_Device.cu +++ b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitK18K20NavierStokesCompressible/InitK18K20NavierStokesCompressible_Device.cu @@ -1,3 +1,33 @@ +//======================================================================================= +// ____ ____ __ ______ __________ __ __ __ __ +// \ \ | | | | | _ \ |___ ___| | | | | / \ | | +// \ \ | | | | | |_) | | | | | | | / \ | | +// \ \ | | | | | _ / | | | | | | / /\ \ | | +// \ \ | | | | | | \ \ | | | \__/ | / ____ \ | |____ +// \ \ | | |__| |__| \__\ |__| \________/ /__/ \__\ |_______| +// \ \ | | ________________________________________________________________ +// \ \ | | | ______________________________________________________________| +// \ \| | | | __ __ __ __ ______ _______ +// \ | | |_____ | | | | | | | | | _ \ / _____) +// \ | | _____| | | | | | | | | | | \ \ \_______ +// \ | | | | |_____ | \_/ | | | | |_/ / _____ | +// \ _____| |__| |________| \_______/ |__| |______/ (_______/ +// +// This file is part of VirtualFluids. VirtualFluids is free software: you can +// redistribute it and/or modify it under the terms of the GNU General Public +// License as published by the Free Software Foundation, either version 3 of +// the License, or (at your option) any later version. +// +// VirtualFluids is distributed in the hope that it will be useful, but WITHOUT +// ANY WARRANTY; without even the implied warranty of MERCHANTABILITY or +// FITNESS FOR A PARTICULAR PURPOSE. See the GNU General Public License +// for more details. +// +// You should have received a copy of the GNU General Public License along +// with VirtualFluids (see COPYING.txt). If not, see <http://www.gnu.org/licenses/>. +// +//! \author Martin Schoenherr +//======================================================================================= #include "LBM/LB.h" #include "lbm/constants/D3Q27.h" #include <basics/constants/NumericConstants.h> @@ -7,7 +37,7 @@ using namespace vf::lbm::dir; #include "math.h" -__global__ void LB_Init_F3(unsigned int* neighborX, +__global__ void InitK18K20NavierStokesCompressible_Device(unsigned int* neighborX, unsigned int* neighborY, unsigned int* neighborZ, unsigned int* geoD, diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitK18K20NavierStokesCompressible/InitK18K20NavierStokesCompressible_Device.cuh b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitK18K20NavierStokesCompressible/InitK18K20NavierStokesCompressible_Device.cuh new file mode 100644 index 0000000000000000000000000000000000000000..90ecb070533a0bf40522bdf246f59921149caaa3 --- /dev/null +++ b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitK18K20NavierStokesCompressible/InitK18K20NavierStokesCompressible_Device.cuh @@ -0,0 +1,50 @@ +//======================================================================================= +// ____ ____ __ ______ __________ __ __ __ __ +// \ \ | | | | | _ \ |___ ___| | | | | / \ | | +// \ \ | | | | | |_) | | | | | | | / \ | | +// \ \ | | | | | _ / | | | | | | / /\ \ | | +// \ \ | | | | | | \ \ | | | \__/ | / ____ \ | |____ +// \ \ | | |__| |__| \__\ |__| \________/ /__/ \__\ |_______| +// \ \ | | ________________________________________________________________ +// \ \ | | | ______________________________________________________________| +// \ \| | | | __ __ __ __ ______ _______ +// \ | | |_____ | | | | | | | | | _ \ / _____) +// \ | | _____| | | | | | | | | | | \ \ \_______ +// \ | | | | |_____ | \_/ | | | | |_/ / _____ | +// \ _____| |__| |________| \_______/ |__| |______/ (_______/ +// +// This file is part of VirtualFluids. VirtualFluids is free software: you can +// redistribute it and/or modify it under the terms of the GNU General Public +// License as published by the Free Software Foundation, either version 3 of +// the License, or (at your option) any later version. +// +// VirtualFluids is distributed in the hope that it will be useful, but WITHOUT +// ANY WARRANTY; without even the implied warranty of MERCHANTABILITY or +// FITNESS FOR A PARTICULAR PURPOSE. See the GNU General Public License +// for more details. +// +// You should have received a copy of the GNU General Public License along +// with VirtualFluids (see COPYING.txt). If not, see <http://www.gnu.org/licenses/>. +// +//! \author Martin Schoenherr +//======================================================================================= +#ifndef InitK18K20NavierStokesCompressible_Device_H +#define InitK18K20NavierStokesCompressible_Device_H + +#include <DataTypes.h> +#include <curand.h> + +__global__ void InitK18K20NavierStokesCompressible_Device( + unsigned int* neighborX, + unsigned int* neighborY, + unsigned int* neighborZ, + unsigned int* geoD, + real* rho, + real* ux, + real* uy, + real* uz, + unsigned int size_Mat, + real* G6, + bool EvenOrOdd); + +#endif \ No newline at end of file diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesCompressible/InitNavierStokesCompressible.cu b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesCompressible/InitNavierStokesCompressible.cu new file mode 100644 index 0000000000000000000000000000000000000000..995607fd3278bd79a9cf36d8d5d4666699be8ae5 --- /dev/null +++ b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesCompressible/InitNavierStokesCompressible.cu @@ -0,0 +1,98 @@ +//======================================================================================= +// ____ ____ __ ______ __________ __ __ __ __ +// \ \ | | | | | _ \ |___ ___| | | | | / \ | | +// \ \ | | | | | |_) | | | | | | | / \ | | +// \ \ | | | | | _ / | | | | | | / /\ \ | | +// \ \ | | | | | | \ \ | | | \__/ | / ____ \ | |____ +// \ \ | | |__| |__| \__\ |__| \________/ /__/ \__\ |_______| +// \ \ | | ________________________________________________________________ +// \ \ | | | ______________________________________________________________| +// \ \| | | | __ __ __ __ ______ _______ +// \ | | |_____ | | | | | | | | | _ \ / _____) +// \ | | _____| | | | | | | | | | | \ \ \_______ +// \ | | | | |_____ | \_/ | | | | |_/ / _____ | +// \ _____| |__| |________| \_______/ |__| |______/ (_______/ +// +// This file is part of VirtualFluids. VirtualFluids is free software: you can +// redistribute it and/or modify it under the terms of the GNU General Public +// License as published by the Free Software Foundation, either version 3 of +// the License, or (at your option) any later version. +// +// VirtualFluids is distributed in the hope that it will be useful, but WITHOUT +// ANY WARRANTY; without even the implied warranty of MERCHANTABILITY or +// FITNESS FOR A PARTICULAR PURPOSE. See the GNU General Public License +// for more details. +// +// You should have received a copy of the GNU General Public License along +// with VirtualFluids (see COPYING.txt). If not, see <http://www.gnu.org/licenses/>. +// +//! \author Martin Schoenherr +//======================================================================================= +#include "InitNavierStokesCompressible.h" + +#include "InitNavierStokesCompressible_Device.cuh" +#include "Parameter/Parameter.h" +#include <cuda_helper/CudaGrid.h> + +std::shared_ptr<PreProcessorStrategy> InitNavierStokesCompressible::getNewInstance(std::shared_ptr<Parameter> para) +{ + return std::shared_ptr<PreProcessorStrategy>(new InitNavierStokesCompressible(para)); +} + +void InitNavierStokesCompressible::init(int level) +{ + vf::cuda::CudaGrid grid = vf::cuda::CudaGrid(para->getParD(level)->numberofthreads, para->getParD(level)->numberOfNodes); + + if( ! para->getUseInitNeq() ) + { + InitNavierStokesCompressible_Device <<< grid.grid, grid.threads >>> ( + para->getParD(level)->neighborX, + para->getParD(level)->neighborY, + para->getParD(level)->neighborZ, + para->getParD(level)->typeOfGridNode, + para->getParD(level)->rho, + para->getParD(level)->velocityX, + para->getParD(level)->velocityY, + para->getParD(level)->velocityZ, + para->getParD(level)->numberOfNodes, + para->getParD(level)->distributions.f[0], + para->getParD(level)->isEvenTimestep); + getLastCudaError("InitNavierStokesCompressible_Device execution failed"); + } + else + { + InitNavierStokesCompressibleNonEquilibrium_Device <<< grid.grid, grid.threads >>> ( + para->getParD(level)->neighborX, + para->getParD(level)->neighborY, + para->getParD(level)->neighborZ, + para->getParD(level)->neighborInverse, + para->getParD(level)->typeOfGridNode, + para->getParD(level)->rho, + para->getParD(level)->velocityX, + para->getParD(level)->velocityY, + para->getParD(level)->velocityZ, + para->getParD(level)->numberOfNodes, + para->getParD(level)->distributions.f[0], + para->getParD(level)->omega, + para->getParD(level)->isEvenTimestep); + cudaDeviceSynchronize(); + getLastCudaError("InitNavierStokesCompressibleNonEquilibrium_Device execution failed"); + } + + + +} + +bool InitNavierStokesCompressible::checkParameter() +{ + return false; +} + +InitNavierStokesCompressible::InitNavierStokesCompressible(std::shared_ptr<Parameter> para) +{ + this->para = para; +} + +InitNavierStokesCompressible::InitNavierStokesCompressible() +{ +} diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesCompressible/InitNavierStokesCompressible.h b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesCompressible/InitNavierStokesCompressible.h new file mode 100644 index 0000000000000000000000000000000000000000..804b629a5d84291c57c4ed2a0b9ef9781cac9aae --- /dev/null +++ b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesCompressible/InitNavierStokesCompressible.h @@ -0,0 +1,53 @@ +//======================================================================================= +// ____ ____ __ ______ __________ __ __ __ __ +// \ \ | | | | | _ \ |___ ___| | | | | / \ | | +// \ \ | | | | | |_) | | | | | | | / \ | | +// \ \ | | | | | _ / | | | | | | / /\ \ | | +// \ \ | | | | | | \ \ | | | \__/ | / ____ \ | |____ +// \ \ | | |__| |__| \__\ |__| \________/ /__/ \__\ |_______| +// \ \ | | ________________________________________________________________ +// \ \ | | | ______________________________________________________________| +// \ \| | | | __ __ __ __ ______ _______ +// \ | | |_____ | | | | | | | | | _ \ / _____) +// \ | | _____| | | | | | | | | | | \ \ \_______ +// \ | | | | |_____ | \_/ | | | | |_/ / _____ | +// \ _____| |__| |________| \_______/ |__| |______/ (_______/ +// +// This file is part of VirtualFluids. VirtualFluids is free software: you can +// redistribute it and/or modify it under the terms of the GNU General Public +// License as published by the Free Software Foundation, either version 3 of +// the License, or (at your option) any later version. +// +// VirtualFluids is distributed in the hope that it will be useful, but WITHOUT +// ANY WARRANTY; without even the implied warranty of MERCHANTABILITY or +// FITNESS FOR A PARTICULAR PURPOSE. See the GNU General Public License +// for more details. +// +// You should have received a copy of the GNU General Public License along +// with VirtualFluids (see COPYING.txt). If not, see <http://www.gnu.org/licenses/>. +// +//! \author Martin Schoenherr +//======================================================================================= +#ifndef InitNavierStokesCompressible_H +#define InitNavierStokesCompressible_H + +#include "PreProcessor/PreProcessorStrategy/PreProcessorStrategy.h" + +#include <memory> + +class Parameter; + +class InitNavierStokesCompressible : public PreProcessorStrategy +{ +public: + static std::shared_ptr<PreProcessorStrategy> getNewInstance(std::shared_ptr< Parameter> para); + void init(int level); + bool checkParameter(); + +private: + InitNavierStokesCompressible(); + InitNavierStokesCompressible(std::shared_ptr< Parameter> para); + std::shared_ptr< Parameter> para; +}; + +#endif \ No newline at end of file diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitCompSP27/InitCompSP27_Device.cu b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesCompressible/InitNavierStokesCompressible_Device.cu similarity index 91% rename from src/gpu/core/PreProcessor/PreProcessorStrategy/InitCompSP27/InitCompSP27_Device.cu rename to src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesCompressible/InitNavierStokesCompressible_Device.cu index f733282faf3791164bcab12354d61beea6659418..cb38eedebba41a15400c570f5520876433c90edc 100644 --- a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitCompSP27/InitCompSP27_Device.cu +++ b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesCompressible/InitNavierStokesCompressible_Device.cu @@ -1,3 +1,33 @@ +//======================================================================================= +// ____ ____ __ ______ __________ __ __ __ __ +// \ \ | | | | | _ \ |___ ___| | | | | / \ | | +// \ \ | | | | | |_) | | | | | | | / \ | | +// \ \ | | | | | _ / | | | | | | / /\ \ | | +// \ \ | | | | | | \ \ | | | \__/ | / ____ \ | |____ +// \ \ | | |__| |__| \__\ |__| \________/ /__/ \__\ |_______| +// \ \ | | ________________________________________________________________ +// \ \ | | | ______________________________________________________________| +// \ \| | | | __ __ __ __ ______ _______ +// \ | | |_____ | | | | | | | | | _ \ / _____) +// \ | | _____| | | | | | | | | | | \ \ \_______ +// \ | | | | |_____ | \_/ | | | | |_/ / _____ | +// \ _____| |__| |________| \_______/ |__| |______/ (_______/ +// +// This file is part of VirtualFluids. VirtualFluids is free software: you can +// redistribute it and/or modify it under the terms of the GNU General Public +// License as published by the Free Software Foundation, either version 3 of +// the License, or (at your option) any later version. +// +// VirtualFluids is distributed in the hope that it will be useful, but WITHOUT +// ANY WARRANTY; without even the implied warranty of MERCHANTABILITY or +// FITNESS FOR A PARTICULAR PURPOSE. See the GNU General Public License +// for more details. +// +// You should have received a copy of the GNU General Public License along +// with VirtualFluids (see COPYING.txt). If not, see <http://www.gnu.org/licenses/>. +// +//! \author Martin Schoenherr +//======================================================================================= #include "LBM/LB.h" #include "lbm/constants/D3Q27.h" #include <basics/constants/NumericConstants.h> @@ -8,7 +38,8 @@ using namespace vf::lbm::dir; #include <stdio.h> -__global__ void LB_Init_Comp_SP_27(unsigned int* neighborX, +__global__ void InitNavierStokesCompressible_Device( + unsigned int* neighborX, unsigned int* neighborY, unsigned int* neighborZ, unsigned int* geoD, @@ -179,7 +210,7 @@ __global__ void LB_Init_Comp_SP_27(unsigned int* neighborX, //////////////////////////////////////////////////////////////////////////////// -__global__ void LB_Init_Comp_Neq_SP_27( unsigned int* neighborX, +__global__ void InitNavierStokesCompressibleNonEquilibrium_Device( unsigned int* neighborX, unsigned int* neighborY, unsigned int* neighborZ, unsigned int* neighborWSB, diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesCompressible/InitNavierStokesCompressible_Device.cuh b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesCompressible/InitNavierStokesCompressible_Device.cuh new file mode 100644 index 0000000000000000000000000000000000000000..a454d7249be6948374bb614e090cc7d64b603bdd --- /dev/null +++ b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesCompressible/InitNavierStokesCompressible_Device.cuh @@ -0,0 +1,63 @@ +//======================================================================================= +// ____ ____ __ ______ __________ __ __ __ __ +// \ \ | | | | | _ \ |___ ___| | | | | / \ | | +// \ \ | | | | | |_) | | | | | | | / \ | | +// \ \ | | | | | _ / | | | | | | / /\ \ | | +// \ \ | | | | | | \ \ | | | \__/ | / ____ \ | |____ +// \ \ | | |__| |__| \__\ |__| \________/ /__/ \__\ |_______| +// \ \ | | ________________________________________________________________ +// \ \ | | | ______________________________________________________________| +// \ \| | | | __ __ __ __ ______ _______ +// \ | | |_____ | | | | | | | | | _ \ / _____) +// \ | | _____| | | | | | | | | | | \ \ \_______ +// \ | | | | |_____ | \_/ | | | | |_/ / _____ | +// \ _____| |__| |________| \_______/ |__| |______/ (_______/ +// +// This file is part of VirtualFluids. VirtualFluids is free software: you can +// redistribute it and/or modify it under the terms of the GNU General Public +// License as published by the Free Software Foundation, either version 3 of +// the License, or (at your option) any later version. +// +// VirtualFluids is distributed in the hope that it will be useful, but WITHOUT +// ANY WARRANTY; without even the implied warranty of MERCHANTABILITY or +// FITNESS FOR A PARTICULAR PURPOSE. See the GNU General Public License +// for more details. +// +// You should have received a copy of the GNU General Public License along +// with VirtualFluids (see COPYING.txt). If not, see <http://www.gnu.org/licenses/>. +// +//! \author Martin Schoenherr +//======================================================================================= +#ifndef InitNavierStokesCompressible_Device_H +#define InitNavierStokesCompressible_Device_H + +#include <DataTypes.h> +#include <curand.h> + +__global__ void InitNavierStokesCompressible_Device(unsigned int* neighborX, + unsigned int* neighborY, + unsigned int* neighborZ, + unsigned int* geoD, + real* rho, + real* ux, + real* uy, + real* uz, + unsigned int size_Mat, + real* DD, + bool EvenOrOdd); + +__global__ void InitNavierStokesCompressibleNonEquilibrium_Device(unsigned int* neighborX, + unsigned int* neighborY, + unsigned int* neighborZ, + unsigned int* neighborWSB, + unsigned int* geoD, + real* rho, + real* ux, + real* uy, + real* uz, + unsigned int size_Mat, + real* DD, + real omega, + bool EvenOrOdd); + +#endif \ No newline at end of file diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesIncompressible/InitNavierStokesIncompressible.cu b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesIncompressible/InitNavierStokesIncompressible.cu new file mode 100644 index 0000000000000000000000000000000000000000..b4c3f01cbb6ca3cca275aa28f55816e540490756 --- /dev/null +++ b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesIncompressible/InitNavierStokesIncompressible.cu @@ -0,0 +1,73 @@ +//======================================================================================= +// ____ ____ __ ______ __________ __ __ __ __ +// \ \ | | | | | _ \ |___ ___| | | | | / \ | | +// \ \ | | | | | |_) | | | | | | | / \ | | +// \ \ | | | | | _ / | | | | | | / /\ \ | | +// \ \ | | | | | | \ \ | | | \__/ | / ____ \ | |____ +// \ \ | | |__| |__| \__\ |__| \________/ /__/ \__\ |_______| +// \ \ | | ________________________________________________________________ +// \ \ | | | ______________________________________________________________| +// \ \| | | | __ __ __ __ ______ _______ +// \ | | |_____ | | | | | | | | | _ \ / _____) +// \ | | _____| | | | | | | | | | | \ \ \_______ +// \ | | | | |_____ | \_/ | | | | |_/ / _____ | +// \ _____| |__| |________| \_______/ |__| |______/ (_______/ +// +// This file is part of VirtualFluids. VirtualFluids is free software: you can +// redistribute it and/or modify it under the terms of the GNU General Public +// License as published by the Free Software Foundation, either version 3 of +// the License, or (at your option) any later version. +// +// VirtualFluids is distributed in the hope that it will be useful, but WITHOUT +// ANY WARRANTY; without even the implied warranty of MERCHANTABILITY or +// FITNESS FOR A PARTICULAR PURPOSE. See the GNU General Public License +// for more details. +// +// You should have received a copy of the GNU General Public License along +// with VirtualFluids (see COPYING.txt). If not, see <http://www.gnu.org/licenses/>. +// +//! \author Martin Schoenherr +//======================================================================================= +#include "InitNavierStokesIncompressible.h" + +#include "InitNavierStokesIncompressible_Device.cuh" +#include "Parameter/Parameter.h" +#include <cuda_helper/CudaGrid.h> + +std::shared_ptr<PreProcessorStrategy> InitNavierStokesIncompressible::getNewInstance(std::shared_ptr<Parameter> para) +{ + return std::shared_ptr<PreProcessorStrategy>(new InitNavierStokesIncompressible(para)); +} + +void InitNavierStokesIncompressible::init(int level) +{ + vf::cuda::CudaGrid grid = vf::cuda::CudaGrid(para->getParD(level)->numberofthreads, para->getParD(level)->numberOfNodes); + + InitNavierStokesIncompressible_Device<<<grid.grid, grid.threads>>>( + para->getParD(level)->neighborX, + para->getParD(level)->neighborY, + para->getParD(level)->neighborZ, + para->getParD(level)->typeOfGridNode, + para->getParD(level)->rho, + para->getParD(level)->velocityX, + para->getParD(level)->velocityY, + para->getParD(level)->velocityZ, + para->getParD(level)->numberOfNodes, + para->getParD(level)->distributions.f[0], + para->getParD(level)->isEvenTimestep); + getLastCudaError("LB_Init_SP_27 execution failed"); +} + +bool InitNavierStokesIncompressible::checkParameter() +{ + return false; +} + +InitNavierStokesIncompressible::InitNavierStokesIncompressible(std::shared_ptr<Parameter> para) +{ + this->para = para; +} + +InitNavierStokesIncompressible::InitNavierStokesIncompressible() +{ +} diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesIncompressible/InitNavierStokesIncompressible.h b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesIncompressible/InitNavierStokesIncompressible.h new file mode 100644 index 0000000000000000000000000000000000000000..1bf9f945e450945d660c2f996ada29b175c77886 --- /dev/null +++ b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesIncompressible/InitNavierStokesIncompressible.h @@ -0,0 +1,53 @@ +//======================================================================================= +// ____ ____ __ ______ __________ __ __ __ __ +// \ \ | | | | | _ \ |___ ___| | | | | / \ | | +// \ \ | | | | | |_) | | | | | | | / \ | | +// \ \ | | | | | _ / | | | | | | / /\ \ | | +// \ \ | | | | | | \ \ | | | \__/ | / ____ \ | |____ +// \ \ | | |__| |__| \__\ |__| \________/ /__/ \__\ |_______| +// \ \ | | ________________________________________________________________ +// \ \ | | | ______________________________________________________________| +// \ \| | | | __ __ __ __ ______ _______ +// \ | | |_____ | | | | | | | | | _ \ / _____) +// \ | | _____| | | | | | | | | | | \ \ \_______ +// \ | | | | |_____ | \_/ | | | | |_/ / _____ | +// \ _____| |__| |________| \_______/ |__| |______/ (_______/ +// +// This file is part of VirtualFluids. VirtualFluids is free software: you can +// redistribute it and/or modify it under the terms of the GNU General Public +// License as published by the Free Software Foundation, either version 3 of +// the License, or (at your option) any later version. +// +// VirtualFluids is distributed in the hope that it will be useful, but WITHOUT +// ANY WARRANTY; without even the implied warranty of MERCHANTABILITY or +// FITNESS FOR A PARTICULAR PURPOSE. See the GNU General Public License +// for more details. +// +// You should have received a copy of the GNU General Public License along +// with VirtualFluids (see COPYING.txt). If not, see <http://www.gnu.org/licenses/>. +// +//! \author Martin Schoenherr +//======================================================================================= +#ifndef INIT_SP27_H +#define INIT_SP27_H + +#include "PreProcessor/PreProcessorStrategy/PreProcessorStrategy.h" + +#include <memory> + +class Parameter; + +class InitNavierStokesIncompressible : public PreProcessorStrategy +{ +public: + static std::shared_ptr<PreProcessorStrategy> getNewInstance(std::shared_ptr< Parameter> para); + void init(int level); + bool checkParameter(); + +private: + InitNavierStokesIncompressible(); + InitNavierStokesIncompressible(std::shared_ptr< Parameter> para); + std::shared_ptr< Parameter> para; +}; + +#endif \ No newline at end of file diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitSP27/InitSP27_Device.cu b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesIncompressible/InitNavierStokesIncompressible_Device.cu similarity index 79% rename from src/gpu/core/PreProcessor/PreProcessorStrategy/InitSP27/InitSP27_Device.cu rename to src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesIncompressible/InitNavierStokesIncompressible_Device.cu index 8677e0f2eb3b1c5917d6550ef478c94c8400098b..3b49f44274717c41d10a9e724639f99395b6265b 100644 --- a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitSP27/InitSP27_Device.cu +++ b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesIncompressible/InitNavierStokesIncompressible_Device.cu @@ -1,3 +1,33 @@ +//======================================================================================= +// ____ ____ __ ______ __________ __ __ __ __ +// \ \ | | | | | _ \ |___ ___| | | | | / \ | | +// \ \ | | | | | |_) | | | | | | | / \ | | +// \ \ | | | | | _ / | | | | | | / /\ \ | | +// \ \ | | | | | | \ \ | | | \__/ | / ____ \ | |____ +// \ \ | | |__| |__| \__\ |__| \________/ /__/ \__\ |_______| +// \ \ | | ________________________________________________________________ +// \ \ | | | ______________________________________________________________| +// \ \| | | | __ __ __ __ ______ _______ +// \ | | |_____ | | | | | | | | | _ \ / _____) +// \ | | _____| | | | | | | | | | | \ \ \_______ +// \ | | | | |_____ | \_/ | | | | |_/ / _____ | +// \ _____| |__| |________| \_______/ |__| |______/ (_______/ +// +// This file is part of VirtualFluids. VirtualFluids is free software: you can +// redistribute it and/or modify it under the terms of the GNU General Public +// License as published by the Free Software Foundation, either version 3 of +// the License, or (at your option) any later version. +// +// VirtualFluids is distributed in the hope that it will be useful, but WITHOUT +// ANY WARRANTY; without even the implied warranty of MERCHANTABILITY or +// FITNESS FOR A PARTICULAR PURPOSE. See the GNU General Public License +// for more details. +// +// You should have received a copy of the GNU General Public License along +// with VirtualFluids (see COPYING.txt). If not, see <http://www.gnu.org/licenses/>. +// +//! \author Martin Schoenherr +//======================================================================================= #include "LBM/LB.h" #include "lbm/constants/D3Q27.h" #include <basics/constants/NumericConstants.h> @@ -6,7 +36,7 @@ using namespace vf::basics::constant; using namespace vf::lbm::dir; #include "math.h" -__global__ void LB_Init_SP_27(unsigned int* neighborX, +__global__ void InitNavierStokesIncompressible_Device(unsigned int* neighborX, unsigned int* neighborY, unsigned int* neighborZ, unsigned int* geoD, diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesIncompressible/InitNavierStokesIncompressible_Device.cuh b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesIncompressible/InitNavierStokesIncompressible_Device.cuh new file mode 100644 index 0000000000000000000000000000000000000000..44015e0100e53ba70bee8b1d720869b93c2da813 --- /dev/null +++ b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitNavierStokesIncompressible/InitNavierStokesIncompressible_Device.cuh @@ -0,0 +1,50 @@ +//======================================================================================= +// ____ ____ __ ______ __________ __ __ __ __ +// \ \ | | | | | _ \ |___ ___| | | | | / \ | | +// \ \ | | | | | |_) | | | | | | | / \ | | +// \ \ | | | | | _ / | | | | | | / /\ \ | | +// \ \ | | | | | | \ \ | | | \__/ | / ____ \ | |____ +// \ \ | | |__| |__| \__\ |__| \________/ /__/ \__\ |_______| +// \ \ | | ________________________________________________________________ +// \ \ | | | ______________________________________________________________| +// \ \| | | | __ __ __ __ ______ _______ +// \ | | |_____ | | | | | | | | | _ \ / _____) +// \ | | _____| | | | | | | | | | | \ \ \_______ +// \ | | | | |_____ | \_/ | | | | |_/ / _____ | +// \ _____| |__| |________| \_______/ |__| |______/ (_______/ +// +// This file is part of VirtualFluids. VirtualFluids is free software: you can +// redistribute it and/or modify it under the terms of the GNU General Public +// License as published by the Free Software Foundation, either version 3 of +// the License, or (at your option) any later version. +// +// VirtualFluids is distributed in the hope that it will be useful, but WITHOUT +// ANY WARRANTY; without even the implied warranty of MERCHANTABILITY or +// FITNESS FOR A PARTICULAR PURPOSE. See the GNU General Public License +// for more details. +// +// You should have received a copy of the GNU General Public License along +// with VirtualFluids (see COPYING.txt). If not, see <http://www.gnu.org/licenses/>. +// +//! \author Martin Schoenherr +//======================================================================================= +#ifndef InitNavierStokesIncompressible_Device_H +#define InitNavierStokesIncompressible_Device_H + +#include <DataTypes.h> +#include <curand.h> + +__global__ void InitNavierStokesIncompressible_Device( + unsigned int* neighborX, + unsigned int* neighborY, + unsigned int* neighborZ, + unsigned int* geoD, + real* rho, + real* ux, + real* uy, + real* uz, + unsigned int size_Mat, + real* DD, + bool EvenOrOdd); + +#endif \ No newline at end of file diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitSP27/InitSP27.cu b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitSP27/InitSP27.cu deleted file mode 100644 index 78994e997c34d6b47cf1b8b49dd0d70442f91588..0000000000000000000000000000000000000000 --- a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitSP27/InitSP27.cu +++ /dev/null @@ -1,43 +0,0 @@ -#include "InitSP27.h" - -#include "InitSP27_Device.cuh" -#include "Parameter/Parameter.h" -#include <cuda_helper/CudaGrid.h> - -std::shared_ptr<PreProcessorStrategy> InitSP27::getNewInstance(std::shared_ptr<Parameter> para) -{ - return std::shared_ptr<PreProcessorStrategy>(new InitSP27(para)); -} - -void InitSP27::init(int level) -{ - vf::cuda::CudaGrid grid = vf::cuda::CudaGrid(para->getParD(level)->numberofthreads, para->getParD(level)->numberOfNodes); - - LB_Init_SP_27 <<< grid.grid, grid.threads >>>( - para->getParD(level)->neighborX, - para->getParD(level)->neighborY, - para->getParD(level)->neighborZ, - para->getParD(level)->typeOfGridNode, - para->getParD(level)->rho, - para->getParD(level)->velocityX, - para->getParD(level)->velocityY, - para->getParD(level)->velocityZ, - para->getParD(level)->numberOfNodes, - para->getParD(level)->distributions.f[0], - para->getParD(level)->isEvenTimestep); - getLastCudaError("LB_Init_SP_27 execution failed"); -} - -bool InitSP27::checkParameter() -{ - return false; -} - -InitSP27::InitSP27(std::shared_ptr<Parameter> para) -{ - this->para = para; -} - -InitSP27::InitSP27() -{ -} diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitSP27/InitSP27.h b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitSP27/InitSP27.h deleted file mode 100644 index def7ccd175168ba6f463e6bc0b755db26e233212..0000000000000000000000000000000000000000 --- a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitSP27/InitSP27.h +++ /dev/null @@ -1,23 +0,0 @@ -#ifndef INIT_SP27_H -#define INIT_SP27_H - -#include "PreProcessor/PreProcessorStrategy/PreProcessorStrategy.h" - -#include <memory> - -class Parameter; - -class InitSP27 : public PreProcessorStrategy -{ -public: - static std::shared_ptr<PreProcessorStrategy> getNewInstance(std::shared_ptr< Parameter> para); - void init(int level); - bool checkParameter(); - -private: - InitSP27(); - InitSP27(std::shared_ptr< Parameter> para); - std::shared_ptr< Parameter> para; -}; - -#endif \ No newline at end of file diff --git a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitSP27/InitSP27_Device.cuh b/src/gpu/core/PreProcessor/PreProcessorStrategy/InitSP27/InitSP27_Device.cuh deleted file mode 100644 index c00a191b4afd5fd9168485932cb162dfc5a05694..0000000000000000000000000000000000000000 --- a/src/gpu/core/PreProcessor/PreProcessorStrategy/InitSP27/InitSP27_Device.cuh +++ /dev/null @@ -1,19 +0,0 @@ -#ifndef LB_INIT_SP27_H -#define LB_INIT_SP27_H - -#include <DataTypes.h> -#include <curand.h> - -__global__ void LB_Init_SP_27(unsigned int* neighborX, - unsigned int* neighborY, - unsigned int* neighborZ, - unsigned int* geoD, - real* rho, - real* ux, - real* uy, - real* uz, - unsigned int size_Mat, - real* DD, - bool EvenOrOdd); - -#endif \ No newline at end of file diff --git a/src/gpu/core/PreProcessor/PreProcessorType.h b/src/gpu/core/PreProcessor/PreProcessorType.h index 6bf54afc08fa99d0e77e3e57897cfa379392f585..1fb9b78fc0c20f6bc8cefaa55d7e8ab539e75129 100644 --- a/src/gpu/core/PreProcessor/PreProcessorType.h +++ b/src/gpu/core/PreProcessor/PreProcessorType.h @@ -1,11 +1,41 @@ +//======================================================================================= +// ____ ____ __ ______ __________ __ __ __ __ +// \ \ | | | | | _ \ |___ ___| | | | | / \ | | +// \ \ | | | | | |_) | | | | | | | / \ | | +// \ \ | | | | | _ / | | | | | | / /\ \ | | +// \ \ | | | | | | \ \ | | | \__/ | / ____ \ | |____ +// \ \ | | |__| |__| \__\ |__| \________/ /__/ \__\ |_______| +// \ \ | | ________________________________________________________________ +// \ \ | | | ______________________________________________________________| +// \ \| | | | __ __ __ __ ______ _______ +// \ | | |_____ | | | | | | | | | _ \ / _____) +// \ | | _____| | | | | | | | | | | \ \ \_______ +// \ | | | | |_____ | \_/ | | | | |_/ / _____ | +// \ _____| |__| |________| \_______/ |__| |______/ (_______/ +// +// This file is part of VirtualFluids. VirtualFluids is free software: you can +// redistribute it and/or modify it under the terms of the GNU General Public +// License as published by the Free Software Foundation, either version 3 of +// the License, or (at your option) any later version. +// +// VirtualFluids is distributed in the hope that it will be useful, but WITHOUT +// ANY WARRANTY; without even the implied warranty of MERCHANTABILITY or +// FITNESS FOR A PARTICULAR PURPOSE. See the GNU General Public License +// for more details. +// +// You should have received a copy of the GNU General Public License along +// with VirtualFluids (see COPYING.txt). If not, see <http://www.gnu.org/licenses/>. +// +//! \author Martin Schoenherr +//======================================================================================= #ifndef PRE_PROCESSOR_TYPE_H #define PRE_PROCESSOR_TYPE_H enum PreProcessorType { - InitSP27, - InitCompSP27, - InitF3, + InitNavierStokesIncompressible, + InitNavierStokesCompressible, + InitK18K20NavierStokesCompressible, InitIncompAD7, InitIncompAD27, InitCompAD7,