From 75c58dc453564db4b41b3c610e7987fdf3fdcc85 Mon Sep 17 00:00:00 2001 From: Suyash Tandon Date: Mon, 25 Mar 2024 15:12:20 -0500 Subject: [PATCH 1/7] ENH: Implementing `AtomicAccumulator` class to perform atomic operations on scalars or vector A templated class, which defines the operators: `operator+=` and `operator-=`. The class declaration looks like: ``` template class AtomicAccumulator { // Private Data T* val; public: // Constructors //- Construct from components AtomicAccumulator(T& ref); // Member Operators inline void operator+=(const T& rhs); inline void operator-=(const T& rhs); }; ``` and the constructor and member operators are defined in AtomicAccumulatorI.H --- .../AtomicAccumulator/AtomicAccumulator.H | 61 +++++++++++++ .../AtomicAccumulatorScalar.H | 90 +++++++++++++++++++ .../AtomicAccumulatorScalarI.H | 53 +++++++++++ .../AtomicAccumulatorVector.H | 90 +++++++++++++++++++ .../AtomicAccumulatorVectorI.H | 52 +++++++++++ 5 files changed, 346 insertions(+) create mode 100644 src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulator.H create mode 100644 src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulatorScalar.H create mode 100644 src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulatorScalarI.H create mode 100644 src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulatorVector.H create mode 100644 src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulatorVectorI.H diff --git a/src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulator.H b/src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulator.H new file mode 100644 index 0000000000..dfa8f6a159 --- /dev/null +++ b/src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulator.H @@ -0,0 +1,61 @@ +/*---------------------------------------------------------------------------*\ + ========= | + \\ / F ield | OpenFOAM: The Open Source CFD Toolbox + \\ / O peration | + \\ / A nd | www.openfoam.com + \\/ M anipulation | +------------------------------------------------------------------------------- + Copyright (C) 2023 Advanced Micro Devices, Inc. All rights reserved. +------------------------------------------------------------------------------- +License + This file is part of OpenFOAM. + + OpenFOAM 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. + + OpenFOAM 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 OpenFOAM. If not, see . + +Class + Foam::AtomicAccumulator + +Description + Atomic operations of scalars and vectors + +SourceFiles + AtomicAccumulatorI.H + +\*---------------------------------------------------------------------------*/ + +#ifndef AtomicAccumulator_H +#define AtomicAccumulator_H + +#include "VectorSpace.H" +#include "VectorSpaceOps.H" +#include "macros.H" +#include "ops.H" +#include + +#ifdef USE_OMP +#include + #ifndef OMP_UNIFIED_MEMORY_REQUIRED + #define OMP_UNIFIED_MEMORY_REQUIRED + #pragma omp requires unified_shared_memory + #endif +#endif + +#include "AtomicAccumulatorScalar.H" +#include "AtomicAccumulatorVector.H" + +// * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * // + +#endif + +// ************************************************************************* // diff --git a/src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulatorScalar.H b/src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulatorScalar.H new file mode 100644 index 0000000000..e594dffd21 --- /dev/null +++ b/src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulatorScalar.H @@ -0,0 +1,90 @@ +/*---------------------------------------------------------------------------*\ + ========= | + \\ / F ield | OpenFOAM: The Open Source CFD Toolbox + \\ / O peration | + \\ / A nd | www.openfoam.com + \\/ M anipulation | +------------------------------------------------------------------------------- + Copyright (C) 2023 Advanced Micro Devices, Inc. All rights reserved. +------------------------------------------------------------------------------- +License + This file is part of OpenFOAM. + + OpenFOAM 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. + + OpenFOAM 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 OpenFOAM. If not, see . + +Class + Foam::AtomicAccumulator + +Description + Atomic operations of scalars + +SourceFiles + AtomicAccumulatorScalarI.H + +\*---------------------------------------------------------------------------*/ + +#ifndef AtomicAccumulatorScalar_H +#define AtomicAccumulatorScalar_H + +// * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * // + +namespace Foam +{ + +/*---------------------------------------------------------------------------*\ + Class AtomicAccumulator Declaration +\*---------------------------------------------------------------------------*/ + +template +class AtomicAccumulator +{ + // Private Data + + T* val; + +public: + + // Constructors + + //- Construct from components + AtomicAccumulator(T& ref); + + // Member Operators + + inline void operator+=(const T& rhs); + + inline void operator-=(const T& rhs); +}; + +// * * * * * * * * * * * * * * * * * Traits * * * * * * * * * * * * * * * * // + +template +inline typename std::enable_if::value, AtomicAccumulator>::type atomicAccumulator(T& ref) +{ + return AtomicAccumulator(ref); +} + +// * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * // + +#include "AtomicAccumulatorScalarI.H" + +// * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * // + +} // End namespace Foam + +// * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * // + +#endif + +// ************************************************************************* // diff --git a/src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulatorScalarI.H b/src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulatorScalarI.H new file mode 100644 index 0000000000..556e387ec9 --- /dev/null +++ b/src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulatorScalarI.H @@ -0,0 +1,53 @@ +/*---------------------------------------------------------------------------*\ + ========= | + \\ / F ield | OpenFOAM: The Open Source CFD Toolbox + \\ / O peration | + \\ / A nd | www.openfoam.com + \\/ M anipulation | +------------------------------------------------------------------------------- + Copyright (C) 2023 Advanced Micro Devices, Inc. All rights reserved. +------------------------------------------------------------------------------- +License + This file is part of OpenFOAM. + + OpenFOAM 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. + + OpenFOAM 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 OpenFOAM. If not, see . + +\*---------------------------------------------------------------------------*/ + +// * * * * * * * * * * * * * * * * Constructors * * * * * * * * * * * * * * // +template +AtomicAccumulator::AtomicAccumulator (T& ref) +: + val(&ref) +{} + +// * * * * * * * * * * * * * * * Member Operators * * * * * * * * * * * * * // +template +inline void AtomicAccumulator::operator+=(const T& rhs) +{ + #pragma omp atomic update + *val += rhs; +} + +template +inline void AtomicAccumulator::operator-=(const T& rhs) +{ + #pragma omp atomic update + *val -= rhs; +} + +// * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * // + + +// ************************************************************************* // diff --git a/src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulatorVector.H b/src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulatorVector.H new file mode 100644 index 0000000000..9d55b4f8c8 --- /dev/null +++ b/src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulatorVector.H @@ -0,0 +1,90 @@ +/*---------------------------------------------------------------------------*\ + ========= | + \\ / F ield | OpenFOAM: The Open Source CFD Toolbox + \\ / O peration | + \\ / A nd | www.openfoam.com + \\/ M anipulation | +------------------------------------------------------------------------------- + Copyright (C) 2023 Advanced Micro Devices, Inc. All rights reserved. +------------------------------------------------------------------------------- +License + This file is part of OpenFOAM. + + OpenFOAM 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. + + OpenFOAM 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 OpenFOAM. If not, see . + +Class + Foam::AtomicAccumulator + +Description + Atomic operations of vectors + +SourceFiles + AtomicAccumulatorVectorI.H + +\*---------------------------------------------------------------------------*/ + +#ifndef AtomicAccumulatorVector_H +#define AtomicAccumulatorVector_H + +// * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * // + +namespace Foam +{ + +/*---------------------------------------------------------------------------*\ + Class AtomicAccumulator> Declaration +\*---------------------------------------------------------------------------*/ + +template +class AtomicAccumulator> +{ + // Private Data + VectorSpace* val; + +public: + + // Constructors + + //- Construct from components + AtomicAccumulator(VectorSpace &ref); + + // Member Operators + + inline void operator+=(const VectorSpace& rhs); + + inline void operator-=(const VectorSpace& rhs); +}; + + +// * * * * * * * * * * * * * * * * * Traits * * * * * * * * * * * * * * * * // + +template +inline AtomicAccumulator> atomicAccumulator(VectorSpace& ref) +{ + return AtomicAccumulator>(ref); +} + +// * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * // + +#include "AtomicAccumulatorVectorI.H" + +// * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * // + +} // End namespace Foam + +// * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * // + +#endif + +// ************************************************************************* // diff --git a/src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulatorVectorI.H b/src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulatorVectorI.H new file mode 100644 index 0000000000..3cf2b1fcd7 --- /dev/null +++ b/src/OpenFOAM/primitives/AtomicAccumulator/AtomicAccumulatorVectorI.H @@ -0,0 +1,52 @@ +/*---------------------------------------------------------------------------*\ + ========= | + \\ / F ield | OpenFOAM: The Open Source CFD Toolbox + \\ / O peration | + \\ / A nd | www.openfoam.com + \\/ M anipulation | +------------------------------------------------------------------------------- + Copyright (C) 2023 Advanced Micro Devices, Inc. All rights reserved. +------------------------------------------------------------------------------- +License + This file is part of OpenFOAM. + + OpenFOAM 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. + + OpenFOAM 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 OpenFOAM. If not, see . + +\*---------------------------------------------------------------------------*/ + +// * * * * * * * * * * * * * * * * Constructors * * * * * * * * * * * * * * // +template +AtomicAccumulator>::AtomicAccumulator (VectorSpace &ref) +: + val(&ref) +{} + +// * * * * * * * * * * * * * * * Member Operators * * * * * * * * * * * * * // +template +inline void AtomicAccumulator>::operator+=(const VectorSpace& rhs) +{ + VectorSpaceOps::eqOp(*val, rhs, atomicPlusEqOp()); +} + +template +inline void AtomicAccumulator>::operator-=(const VectorSpace& rhs) +{ + VectorSpaceOps::eqOp(*val, rhs, atomicMinusEqOp()); +} + +// * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * // + + + +// ************************************************************************* // From 03b73ddd0529190f962058b154bcd0fd5eb2c46b Mon Sep 17 00:00:00 2001 From: Suyash Tandon Date: Mon, 25 Mar 2024 20:20:00 -0500 Subject: [PATCH 2/7] ENH: Adding more target regions in lduMatrix using atomics Using the newly added AtomicAccumulator class to add more target regions in lduMatrix assemby and operations. For example, the `sumDiag()` operation has following offloaded region: ``` #pragma omp target teams distribute parallel for if (target:l.size()>10000) for (label face=0; face10000) + for (label face=0; face10000) + for (label face=0; face10000) + for (label face=0; face10000) + for (label face=0; face Foam::lduMatrix::H1() const const label nFaces = upper().size(); + #ifdef USE_OMP + #pragma omp target teams distribute parallel for thread_limit(64) if (target:nFaces>10000) + for (label face=0; face10000) + for (label face=0; face10000) + for (label face=0; face10000) + for (label face = 0; face < l.size(); face++) + { + atomicAccumulator(sumOff[u[face]]) += mag(Lower[face]); + atomicAccumulator(sumOff[l[face]]) += mag(Upper[face]); + } +#else for (label face = 0; face < l.size(); face++) { sumOff[u[face]] += mag(Lower[face]); sumOff[l[face]] += mag(Upper[face]); } +#endif } diff --git a/src/OpenFOAM/matrices/lduMatrix/lduMatrix/lduMatrixTemplates.C b/src/OpenFOAM/matrices/lduMatrix/lduMatrix/lduMatrixTemplates.C index 251e167a4c..62f4a79d19 100644 --- a/src/OpenFOAM/matrices/lduMatrix/lduMatrix/lduMatrixTemplates.C +++ b/src/OpenFOAM/matrices/lduMatrix/lduMatrix/lduMatrixTemplates.C @@ -37,6 +37,8 @@ Description #define OMP_UNIFIED_MEMORY_REQUIRED #pragma omp requires unified_shared_memory #endif + +#include "AtomicAccumulator.H" #endif // * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * // @@ -60,11 +62,23 @@ Foam::tmp> Foam::lduMatrix::H(const Field& psi) const const label nFaces = upper().size(); + #ifdef USE_OMP + #pragma omp target teams distribute parallel for thread_limit(64) if (target:nFaces>10000) + for (label face=0; face& psi) const Upper[face]*psi[u[face]] - Lower[face]*psi[l[face]]; } - + return tfaceHpsi; } From 47ce7aa9b19e05bcdf4cf703b0053d2037d1aa7c Mon Sep 17 00:00:00 2001 From: Suyash Tandon Date: Tue, 26 Mar 2024 13:05:15 -0500 Subject: [PATCH 3/7] ENH: Adding OpenMP target offloading with unified shared memory in `lduAddressing` Adding target offload regions in `lduMatrix/lduAddressing` with unified shared memory and atomics. --- .../lduCalculatedProcessorField.C | 19 +++++++++++++++++++ .../lduInterfaceField/lduInterfaceField.H | 11 +++++++++++ .../lduInterfaceFieldTemplates.C | 19 +++++++++++++++++++ 3 files changed, 49 insertions(+) diff --git a/src/OpenFOAM/matrices/lduMatrix/lduAddressing/lduInterface/lduCalculatedProcessorField/lduCalculatedProcessorField.C b/src/OpenFOAM/matrices/lduMatrix/lduAddressing/lduInterface/lduCalculatedProcessorField/lduCalculatedProcessorField.C index 0fb082115b..950044296b 100644 --- a/src/OpenFOAM/matrices/lduMatrix/lduAddressing/lduInterface/lduCalculatedProcessorField/lduCalculatedProcessorField.C +++ b/src/OpenFOAM/matrices/lduMatrix/lduAddressing/lduInterface/lduCalculatedProcessorField/lduCalculatedProcessorField.C @@ -6,6 +6,7 @@ \\/ M anipulation | ------------------------------------------------------------------------------- Copyright (C) 2022-2023 OpenCFD Ltd. + Copyright (C) 2023 Advanced Micro Devices, Inc. All rights reserved. ------------------------------------------------------------------------------- License This file is part of OpenFOAM. @@ -148,17 +149,35 @@ void Foam::lduCalculatedProcessorField::addToInternalField if (add) { + #ifdef USE_OMP + const label fCells = faceCells.size(); + #pragma omp target teams distribute parallel for if (target:fCells>10000) + for (label elemI = 0; elemI < fCells; elemI++) + { + atomicAccumulator(result[faceCells[elemI]]) += (coeffs[elemI]*vals[elemI]); + } + #else forAll(faceCells, elemI) { result[faceCells[elemI]] += coeffs[elemI]*vals[elemI]; } + #endif } else { + #ifdef USE_OMP + const label fCells = faceCells.size(); + #pragma omp target teams distribute parallel for if (target:fCells>10000) + for (label elemI = 0; elemI < fCells; elemI++) + { + atomicAccumulator(result[faceCells[elemI]]) -= (coeffs[elemI]*vals[elemI]); + } + #else forAll(faceCells, elemI) { result[faceCells[elemI]] -= coeffs[elemI]*vals[elemI]; } + #endif } } diff --git a/src/OpenFOAM/matrices/lduMatrix/lduAddressing/lduInterfaceFields/lduInterfaceField/lduInterfaceField.H b/src/OpenFOAM/matrices/lduMatrix/lduAddressing/lduInterfaceFields/lduInterfaceField/lduInterfaceField.H index 3eac6cb2d6..87ad6ddda8 100644 --- a/src/OpenFOAM/matrices/lduMatrix/lduAddressing/lduInterfaceFields/lduInterfaceField/lduInterfaceField.H +++ b/src/OpenFOAM/matrices/lduMatrix/lduAddressing/lduInterfaceFields/lduInterfaceField/lduInterfaceField.H @@ -7,6 +7,7 @@ ------------------------------------------------------------------------------- Copyright (C) 2011-2013 OpenFOAM Foundation Copyright (C) 2019-2023 OpenCFD Ltd. + Copyright (C) 2023 Advanced Micro Devices, Inc. All rights reserved. ------------------------------------------------------------------------------- License This file is part of OpenFOAM. @@ -44,6 +45,16 @@ SourceFiles #include "primitiveFieldsFwd.H" #include "Pstream.H" +#ifdef USE_OMP +#include + #ifndef OMP_UNIFIED_MEMORY_REQUIRED + #define OMP_UNIFIED_MEMORY_REQUIRED + #pragma omp requires unified_shared_memory + #endif + +#include "AtomicAccumulator.H" +#endif + // * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * * // namespace Foam diff --git a/src/OpenFOAM/matrices/lduMatrix/lduAddressing/lduInterfaceFields/lduInterfaceField/lduInterfaceFieldTemplates.C b/src/OpenFOAM/matrices/lduMatrix/lduAddressing/lduInterfaceFields/lduInterfaceField/lduInterfaceFieldTemplates.C index 9190412c22..81272f7773 100644 --- a/src/OpenFOAM/matrices/lduMatrix/lduAddressing/lduInterfaceFields/lduInterfaceField/lduInterfaceFieldTemplates.C +++ b/src/OpenFOAM/matrices/lduMatrix/lduAddressing/lduInterfaceFields/lduInterfaceField/lduInterfaceFieldTemplates.C @@ -6,6 +6,7 @@ \\/ M anipulation | ------------------------------------------------------------------------------- Copyright (C) 2017-2019 OpenCFD Ltd. + Copyright (C) 2023 Advanced Micro Devices, Inc. All rights reserved. ------------------------------------------------------------------------------- License This file is part of OpenFOAM. @@ -39,17 +40,35 @@ void Foam::lduInterfaceField::addToInternalField { if (add) { + #ifdef USE_OMP + const label fCells = faceCells.size(); + #pragma omp target teams distribute parallel for if (target:fCells>10000) + for (label elemi = 0; elemi < fCells; elemi++) + { + atomicAccumulator(result[faceCells[elemi]]) += (coeffs[elemi]*vals[elemi]); + } + #else forAll(faceCells, elemi) { result[faceCells[elemi]] += coeffs[elemi]*vals[elemi]; } + #endif } else { + #ifdef USE_OMP + const label fCells = faceCells.size(); + #pragma omp target teams distribute parallel for if (target:fCells>10000) + for (label elemi = 0; elemi < fCells; elemi++) + { + atomicAccumulator(result[faceCells[elemi]]) -= (coeffs[elemi]*vals[elemi]); + } + #else forAll(faceCells, elemi) { result[faceCells[elemi]] -= coeffs[elemi]*vals[elemi]; } + #endif } } From dd8d3b7e00a68b8bc3109a63ca6b70caec8281b0 Mon Sep 17 00:00:00 2001 From: Giuseppe Giaquinto Date: Fri, 31 May 2024 13:59:12 +0200 Subject: [PATCH 4/7] update Readme --- README.md | 12 ++++++++++++ wmake/rules/General/general | 1 + 2 files changed, 13 insertions(+) diff --git a/README.md b/README.md index ad37c448d1..dffcea0134 100644 --- a/README.md +++ b/README.md @@ -1,3 +1,15 @@ +## notes on installation +Given that some roctx profiling calls are in openfoam macros it is necessary to +add the relevant compiler flags to succesfully compile the code, this is done +via FOAM_EXTRA environment variables + + export ROCM4FOAM="path to your rocm installation" + export FOAM_EXTRA_CFLAGS="-DUSE_ROCTX -I${ROCM4FOAM}/roctracer/include/" + export FOAM_EXTRA_CXXFLAGS="-DUSE_ROCTX -I${ROCM4FOAM}/roctracer/include/" + export FOAM_EXTRA_LDFLAGS="${ROCM4FOAM}/lib/libroctx64.so -L${ROCM4FOAM}/lib + +these must be set before compiling openfoam. + ## About OpenFOAM OpenFOAM is a free, open source CFD software [released and developed by OpenCFD Ltd since 2004](http://www.openfoam.com/history/). It has a large user base across most areas of engineering and science, from both commercial and academic organisations. diff --git a/wmake/rules/General/general b/wmake/rules/General/general index ad982b7a1c..ee37f862a7 100644 --- a/wmake/rules/General/general +++ b/wmake/rules/General/general @@ -58,6 +58,7 @@ endif ifeq (,$(findstring ~openmp,$(WM_COMPILE_CONTROL))) ifneq (,$(findstring +openmp,$(WM_COMPILE_CONTROL))) c++FLAGS += $(COMP_OPENMP) + cFLAGS += $(COMP_OPENMP) endif endif From f1721f447a25fd5db165792e32336972e54364ea Mon Sep 17 00:00:00 2001 From: Giuseppe Giaquinto Date: Fri, 31 May 2024 12:15:31 +0200 Subject: [PATCH 5/7] update bashrc and openmp flags --- etc/bashrc | 2 ++ wmake/rules/General/Clang/openmp | 19 ++++++++++++++++++- wmake/rules/General/Clang/openmp-gomp | 9 --------- wmake/rules/General/Clang/openmp-omp | 6 ------ 4 files changed, 20 insertions(+), 16 deletions(-) mode change 120000 => 100644 wmake/rules/General/Clang/openmp delete mode 100644 wmake/rules/General/Clang/openmp-gomp delete mode 100644 wmake/rules/General/Clang/openmp-omp diff --git a/etc/bashrc b/etc/bashrc index 548f726cd2..8464945803 100644 --- a/etc/bashrc +++ b/etc/bashrc @@ -90,6 +90,7 @@ export WM_COMPILE_OPTION=Opt # ~libz : without libz compression # ~rpath : without rpath handling [MacOS] # +openmp : with openmp +# for clang optional add offload arch, i.e for amd MI300A +openmp:gfx942 # ~openmp : without openmp # +ccache : use ccache # +xcrun : use xcrun and native compilers [MacOS] @@ -97,6 +98,7 @@ export WM_COMPILE_OPTION=Opt # ccache=... : ccache command (unquoted, single/double or <> quoted) # version=... : compiler suffix (eg, version=11 -> gcc-11) #export WM_COMPILE_CONTROL="+strict" +#export WM_COMPILE_CONTROL="+openmp:gfx942 +link-ld" # [WM_MPLIB] - MPI implementation: # = SYSTEMOPENMPI | OPENMPI | SYSTEMMPI | MPI | MPICH | MPICH-GM | diff --git a/wmake/rules/General/Clang/openmp b/wmake/rules/General/Clang/openmp deleted file mode 120000 index 1320b58055..0000000000 --- a/wmake/rules/General/Clang/openmp +++ /dev/null @@ -1 +0,0 @@ -openmp-gomp \ No newline at end of file diff --git a/wmake/rules/General/Clang/openmp b/wmake/rules/General/Clang/openmp new file mode 100644 index 0000000000..5cef4a941a --- /dev/null +++ b/wmake/rules/General/Clang/openmp @@ -0,0 +1,18 @@ +# Flags for compiling/linking openmp +# - + +COMP_OPENMP = -fopenmp -fopenmp-version=51 + +LINK_OPENMP = + +WM_OMP_OFFLOAD_ARCH=$(patsubst +openmp:%,%,$(filter +openmp:%,$(WM_COMPILE_CONTROL))) + +ifneq ($(WM_OMP_OFFLOAD_ARCH),) + + COMP_OPENMP += -fno-fast-math --offload-arch=${WM_OMP_OFFLOAD_ARCH} + +endif + +#specific to openfoam_hmm +COMP_OPENMP += -DUSE_OMP +# ---------------------------------------------------------------------------- diff --git a/wmake/rules/General/Clang/openmp-gomp b/wmake/rules/General/Clang/openmp-gomp deleted file mode 100644 index b49ba9482f..0000000000 --- a/wmake/rules/General/Clang/openmp-gomp +++ /dev/null @@ -1,9 +0,0 @@ -# Flags for compiling/linking openmp -# - -# Clang provides 'omp' and a link for 'gomp'. -# With 'gomp' we can also use system libs. - -COMP_OPENMP = -DUSE_OMP -fopenmp --offload-arch=gfx90a -fopenmp-target-fast -fopenmp-version=51 -fno-fast-math -Wno-conditional-type-mismatch -LINK_OPENMP = -lgomp - -# ----------------------------------------------------------------------------- diff --git a/wmake/rules/General/Clang/openmp-omp b/wmake/rules/General/Clang/openmp-omp deleted file mode 100644 index b5088598e3..0000000000 --- a/wmake/rules/General/Clang/openmp-omp +++ /dev/null @@ -1,6 +0,0 @@ -# Flags for compiling/linking openmp - -COMP_OPENMP = -fopenmp -LINK_OPENMP = -lomp - -# ----------------------------------------------------------------------------- From af6bfd41741776f26502ee37ccba7f3201ab8d72 Mon Sep 17 00:00:00 2001 From: Giuseppe Giaquinto Date: Thu, 28 Mar 2024 17:45:22 +0100 Subject: [PATCH 6/7] Add Clang rules for openmp integration --- wmake/rules/General/Clang/c++ | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/wmake/rules/General/Clang/c++ b/wmake/rules/General/Clang/c++ index a476d44f90..dd0c34533d 100644 --- a/wmake/rules/General/Clang/c++ +++ b/wmake/rules/General/Clang/c++ @@ -3,7 +3,7 @@ #------------------------------------------------------------------------------ SUFFIXES += .C .cc .cpp .cxx -CC := clang++$(COMPILER_VERSION) -std=c++14 +CC := clang++$(COMPILER_VERSION) -std=c++17 c++ARCH := c++DBUG := From dda0ade130919662279a849800411948677edffa Mon Sep 17 00:00:00 2001 From: GiuseppeGq20 <71777513+GiuseppeGq20@users.noreply.github.com> Date: Mon, 3 Jun 2024 17:10:17 +0200 Subject: [PATCH 7/7] remove openmp CFLAGS --- wmake/rules/General/general | 1 - 1 file changed, 1 deletion(-) diff --git a/wmake/rules/General/general b/wmake/rules/General/general index ee37f862a7..ad982b7a1c 100644 --- a/wmake/rules/General/general +++ b/wmake/rules/General/general @@ -58,7 +58,6 @@ endif ifeq (,$(findstring ~openmp,$(WM_COMPILE_CONTROL))) ifneq (,$(findstring +openmp,$(WM_COMPILE_CONTROL))) c++FLAGS += $(COMP_OPENMP) - cFLAGS += $(COMP_OPENMP) endif endif