diff --git a/ci/build-rocm.sh b/ci/build-rocm.sh index 59744d018..c5cda141e 100755 --- a/ci/build-rocm.sh +++ b/ci/build-rocm.sh @@ -2,8 +2,6 @@ set -xeuo pipefail -export SPACK_SYSTEM_CONFIG_PATH=/user-environment/config - # make sure we keep the stage direcorty spack config --scope=spack add config:build_stage:/dev/shm/spack-stage # we might need to install dependencies too, e.g. nlcglib in case of API changes @@ -11,11 +9,15 @@ spack config --scope=spack add config:install_tree:root:/dev/shm/spack-stage spack env create -d ./spack-env # add local repository with current sirius recipe -spack -e ./spack-env repo add $REPO +# spack -e ./spack-env repo add $REPO +spack -e ./spack-env config add 'include:[/user-environment/config]' spack -e ./spack-env config add "packages:all:variants:[amdgpu_target=${ROCM_ARCH},amdgpu_target_sram_ecc=${ROCM_ARCH},+rocm]" +# TODO: update once this is has changed upstream +# spack -e ./spack-env config add "packages:all:providers:mpi:[cray-mpich@9.1.0+rocm]" spack -e ./spack-env add $SPEC +spack -e ./spack-env add cray-mpich@9.1.0+rocm # build sirius from source spack -e ./spack-env develop -p $PWD sirius@develop diff --git a/ci/cscs-beverin.yml b/ci/cscs-beverin.yml index dd13072d8..d3007630b 100644 --- a/ci/cscs-beverin.yml +++ b/ci/cscs-beverin.yml @@ -32,7 +32,7 @@ build sirius mi300: variables: SPEC: sirius@develop +rocm +apps +tests +scalapack ~python +dftd4 +dftd3 +vdwxc build_type=RelWithDebInfo ROCM_ARCH: gfx942 - REPO: ./ci/sirius-uenv-recipe-mi300/spack_repo/sirius_ci + REPO: ./ci/spack_repo/sirius_ci script: - | # @@ -100,15 +100,14 @@ beverin test gpu band parallel: USE_MPI: 'YES' TEST_COMMAND: ctest -L gpu_band_parallel --timeout 3600 -# skip this, it wants to use cusolver -# beverin test gpu k-point parallel: -# extends: [.run_tests_beverin] -# variables: -# SLURM_NTASKS: 12 -# SLURM_GPUS_PER_TASK: 1 -# SLURM_CPU_BIND: 'verbose' -# USE_MPI: 'YES' -# TEST_COMMAND: ctest -L gpu_k_point_parallel +beverin test gpu k-point parallel: + extends: [.run_tests_beverin] + variables: + SLURM_NTASKS: 12 + SLURM_GPUS_PER_TASK: 1 + SLURM_CPU_BIND: 'verbose' + USE_MPI: 'YES' + TEST_COMMAND: ctest -L gpu_k_point_parallel # eiger test cpu serial: # extends: [.run_tests_eiger] diff --git a/ci/sirius-uenv-recipe-mi300/config.yaml b/ci/sirius-uenv-recipe-mi300/config.yaml index 599cd787d..809c2c76b 100644 --- a/ci/sirius-uenv-recipe-mi300/config.yaml +++ b/ci/sirius-uenv-recipe-mi300/config.yaml @@ -5,6 +5,6 @@ spack: repo: https://github.com/spack/spack.git packages: repo: https://github.com/spack/spack-packages.git - commit: 5f20b9190596e0b875141e8cee03f0d3847ad65c + commit: 876d656ad17366f92e2233d843ba285d655150dd # March 9 store: /user-environment description: SIRIUS dependencies for SIRIUS CI diff --git a/ci/sirius-uenv-recipe-mi300/environments.yaml b/ci/sirius-uenv-recipe-mi300/environments.yaml index 09ffe5891..8f48c4eb7 100644 --- a/ci/sirius-uenv-recipe-mi300/environments.yaml +++ b/ci/sirius-uenv-recipe-mi300/environments.yaml @@ -1,16 +1,15 @@ sirius-env: compiler: [gcc] network: - mpi: cray-mpich@8.1.32+rocm + mpi: cray-mpich@9.1.0+rocm specs: [ 'libfabric@2.3 +rocm' ] - unify: when_possible + unify: true specs: - - sirius@develop ~python +vdwxc +apps +tests +dlaf +scalapack +rocm +dftd4 +dftd3 + - sirius@develop ~python +vdwxc +apps +tests +dlaf +scalapack +rocm +dftd4 +dftd3 ~elpa - kokkos@4.7.02+serial - - sirius@develop +elpa - nlcglib@1.3.0 - eigen@3 - - hip@6.3.3 ^mesa@23.3.6 + - hip@7.1.0 ^mesa variants: - amdgpu_target=gfx942 - amdgpu_target_sram_ecc=gfx942 diff --git a/ci/sirius-uenv-recipe-mi300/repo/packages b/ci/sirius-uenv-recipe-mi300/repo/packages new file mode 120000 index 000000000..7e3a36444 --- /dev/null +++ b/ci/sirius-uenv-recipe-mi300/repo/packages @@ -0,0 +1 @@ +../../spack_repo/sirius_ci/packages \ No newline at end of file diff --git a/ci/sirius-uenv-recipe-mi300/spack_repo/sirius_ci/packages b/ci/sirius-uenv-recipe-mi300/spack_repo/sirius_ci/packages deleted file mode 120000 index f2abe005a..000000000 --- a/ci/sirius-uenv-recipe-mi300/spack_repo/sirius_ci/packages +++ /dev/null @@ -1 +0,0 @@ -../../../../spack_repo/sirius_repo/packages \ No newline at end of file diff --git a/ci/spack_repo/sirius_ci/packages/nlcglib/package.py b/ci/spack_repo/sirius_ci/packages/nlcglib/package.py new file mode 100644 index 000000000..8997c85a0 --- /dev/null +++ b/ci/spack_repo/sirius_ci/packages/nlcglib/package.py @@ -0,0 +1,155 @@ +# Copyright Spack Project Developers. See COPYRIGHT file for details. +# +# SPDX-License-Identifier: (Apache-2.0 OR MIT) + +from spack_repo.builtin.build_systems.cmake import CMakePackage +from spack_repo.builtin.build_systems.cuda import CudaPackage +from spack_repo.builtin.build_systems.rocm import ROCmPackage + +from spack.package import * + + +class Nlcglib(CMakePackage, CudaPackage, ROCmPackage): + """Nonlinear CG methods for wave-function optimization in DFT.""" + + homepage = "https://github.com/simonpintarelli/nlcglib" + git = "https://github.com/simonpintarelli/nlcglib.git" + url = "https://github.com/simonpintarelli/nlcglib/archive/v0.9.tar.gz" + + maintainers("simonpintarelli") + + license("BSD-3-Clause") + + version("develop", branch="develop") + version("1.3.0", sha256="d6adfe97407be2ecf41eda6f530e9724052f5298c5f884113969a9204530a745") + version("1.2.0", sha256="bb3676472cf7cc9effe06e416ecf4ef38e6c58c3e423dc70b8b6b32890bc89f2") + version("1.1.0", sha256="9e7c2eea84a5ce191bd9af08f6c890717f7b6e88be7bd15cfe774eb0e0dabd8a") + version("1.0b", sha256="086c46f06a117f267cbdf1df4ad42a8512689a9610885763f463469fb15e82dc") + version("0.9", sha256="8d5bc6b85ee714fb3d6480f767e7f43e5e7d569116cf60e48f533a7f50a37a08") + + variant("openmp", default=True, description="Use OpenMP") + variant("tests", default=False, description="Build tests") + variant( + "build_type", + default="Release", + description="CMake build type", + values=("Debug", "Release", "RelWithDebInfo"), + ) + + variant( + "gpu_direct", + default=False, + description="Enable GPU direct. Required to support distributed wave-functions.", + ) + + + depends_on("cxx", type="build") # generated + + depends_on("cmake@3.23:", type="build") + depends_on("mpi") + depends_on("lapack") + depends_on("fmt", when="@1.3:") + + requires( + "^[virtuals=lapack] openblas", + "^[virtuals=lapack] intel-oneapi-mkl", + policy="one_of", + msg="Only mkl or openblas are supported for blas/lapack with ldak", + ) + + depends_on("kokkos~cuda~rocm", when="~cuda~rocm") + depends_on("kokkos+openmp", when="+openmp") + + depends_on("googletest", type="build", when="+tests") + depends_on("nlohmann-json") + depends_on("kokkos+serial@4:", when="@1.1:") + + # MKLConfig.cmake introduced in 2021.3 + conflicts("intel-oneapi-mkl@:2021.2", when="^intel-oneapi-mkl") + conflicts("+gpu_direct", when="~rocm~cuda") + + with when("@:0.9"): + conflicts("+rocm") + conflicts("^kokkos@4:") + + with when("+rocm"): + variant("magma", default=True, description="Use magma eigenvalue solver (AMDGPU)") + depends_on("magma+rocm", when="+magma") + depends_on("kokkos+rocm") + depends_on("rocblas") + depends_on("rocsolver") + + with when("+cuda"): + depends_on("kokkos+cuda_lambda+wrapper", when="%gcc") + depends_on("kokkos+cuda") + for arch in CudaPackage.cuda_arch_values: + depends_on(f"kokkos cuda_arch={arch}", when=f"cuda_arch={arch}") + + def cmake_args(self): + options = [ + self.define_from_variant("USE_OPENMP", "openmp"), + self.define_from_variant("BUILD_TESTS", "tests"), + self.define_from_variant("USE_ROCM", "rocm"), + self.define_from_variant("USE_GPU_DIRECT", "gpu_direct"), + self.define_from_variant("USE_MAGMA", "magma"), + self.define_from_variant("USE_CUDA", "cuda"), + ] + + if self.spec.satisfies("^[virtuals=lapack] intel-oneapi-mkl"): + mkl_mapper = { + "threading": {"none": "sequential", "openmp": "gnu_thread", "tbb": "tbb_thread"}, + "mpi": {"intel-oneapi-mpi": "intelmpi", "mpich": "mpich", "openmpi": "openmpi"}, + } + + mkl_threads = mkl_mapper["threading"][ + self.spec["intel-oneapi-mkl"].variants["threads"].value + ] + + mpi_provider = self.spec["mpi"].name + if mpi_provider in ["mpich", "cray-mpich", "mvapich", "mvapich2"]: + mkl_mpi = mkl_mapper["mpi"]["mpich"] + else: + mkl_mpi = mkl_mapper["mpi"][mpi_provider] + + options.extend( + [ + self.define("LAPACK_VENDOR", "MKLONEAPI"), + self.define("MKL_INTERFACE", "lp64"), + self.define("MKL_THREADING", mkl_threads), + self.define("MKL_MPI", mkl_mpi), + ] + ) + else: + options.append(self.define("LAPACK_VENDOR", "OpenBLAS")) + + if "+cuda%gcc" in self.spec: + options.append( + self.define("CMAKE_CXX_COMPILER", self["kokkos-nvcc-wrapper"].kokkos_cxx) + ) + + if "+cuda" in self.spec: + cuda_archs = self.spec.variants["cuda_arch"].value + if "@:0.9" in self.spec: + cuda_flags = " ".join( + ["-gencode arch=compute_{0},code=sm_{0}".format(x) for x in cuda_archs] + ) + options.append(self.define("CMAKE_CUDA_FLAGS", cuda_flags)) + else: + options.append(self.define("CMAKE_CUDA_ARCHITECTURES", cuda_archs)) + + if "^cuda+allow-unsupported-compilers" in self.spec: + options.append(self.define("CMAKE_CUDA_FLAGS", "--allow-unsupported-compiler")) + + if "+rocm" in self.spec: + archs = ",".join(self.spec.variants["amdgpu_target"].value) + options.extend( + [ + self.define("CMAKE_CXX_COMPILER", self.spec["hip"].hipcc), + self.define("HIP_HCC_FLAGS", f"--amdgpu-target={archs}"), + self.define( + "CMAKE_CXX_FLAGS", f"--amdgpu-target={archs} --offload-arch={archs}" + ), + ] + ) + + return options diff --git a/ci/spack_repo/sirius_ci/packages/q_e_sirius/package.py b/ci/spack_repo/sirius_ci/packages/q_e_sirius/package.py new file mode 100644 index 000000000..c5cf93a2c --- /dev/null +++ b/ci/spack_repo/sirius_ci/packages/q_e_sirius/package.py @@ -0,0 +1,100 @@ +# Copyright Spack Project Developers. See COPYRIGHT file for details. +# +# SPDX-License-Identifier: (Apache-2.0 OR MIT) +# adapted from official quantum espresso package + + +from spack_repo.builtin.build_systems.cmake import CMakePackage + +from spack.package import * + + +class QESirius(CMakePackage): + """SIRIUS enabled fork of QuantumESPRESSO.""" + + homepage = "https://github.com/electronic-structure/q-e-sirius/" + url = "https://github.com/electronic-structure/q-e-sirius/archive/refs/tags/q-e-sirius/1.0.0.tar.gz" + git = "https://github.com/electronic-structure/q-e-sirius.git" + + maintainers("simonpintarelli") + + license("GPL-2.0-or-later") + + version("develop-ristretto", branch="ristretto", submodules=True) + version("1.0.2", sha256="6885979d6c23b81b49d4a96c7f73f2eac618adccb0181bfb96ae2318664d9828") + version("1.0.1", sha256="512f982aa60fe9fd1cc588fa270e74427c66b62cb2d02ac1cb6cd07dcbe72204") + version("1.0.0", sha256="d85485db8e9252a0bcd67a6a348b2a74626030183199b0edeb97f14c33bca15b") + + variant("openmp", default=True, description="Enables OpenMP support") + variant("libxc", default=False, description="Support functionals through libxc") + variant("sirius_apps", default=False, description="Build SIRIUS standalone binaries") + # Support for HDF5 has been added starting in QE 6.1.0 and is + # still experimental + variant( + "hdf5", + default="none", + description="Orbital and density data I/O with HDF5", + values=("parallel", "serial", "none"), + multi=False, + ) + + depends_on("fortran", type="build") + depends_on("c", type="build") + depends_on("cxx", type="build") + + depends_on("sirius +fortran") + depends_on("sirius +apps", when="+sirius_apps") + depends_on("sirius ~apps", when="~sirius_apps") + depends_on("sirius +openmp", when="+openmp") + depends_on("sirius@develop", when="@develop-ristretto") + + depends_on("mpi") + depends_on("elpa", when="+elpa") + depends_on("libxc", when="+libxc") + depends_on("fftw-api@3") + depends_on("blas") + depends_on("lapack") + depends_on("git", type="build") + depends_on("pkgconfig", type="build") + + variant("scalapack", default=True, description="Enables scalapack support") + + with when("+scalapack"): + depends_on("scalapack") + variant("elpa", default=False, description="Uses elpa as an eigenvalue solver") + + # Versions of HDF5 prior to 1.8.16 lead to QE runtime errors + depends_on("hdf5@1.8.16:+fortran+hl+mpi", when="hdf5=parallel") + depends_on("hdf5@1.8.16:+fortran+hl~mpi", when="hdf5=serial") + + with when("+openmp"): + requires("^fftw+openmp", when="^[virtuals=fftw-api] fftw") + requires("^openblas threads=openmp", when="^[virtuals=blas] openblas") + requires("^intel-oneapi-mkl threads=openmp", when="^[virtuals=blas] intel-oneapi-mkl") + + def cmake_args(self): + args = [ + "-DQE_ENABLE_SIRIUS=ON", + "-DQE_ENABLE_CUDA=OFF", + "-DQE_LAPACK_INTERNAL=OFF", + "-DQE_ENABLE_DOC=OFF", + "-DQE_ENABLE_MPI=ON", + self.define_from_variant("QE_ENABLE_OPENMP", "openmp"), + self.define_from_variant("QE_ENABLE_ELPA", "elpa"), + self.define_from_variant("QE_ENABLE_LIBXC", "libxc"), + self.define_from_variant("QE_ENABLE_SCALAPACK", "scalapack"), + ] + + if not self.spec.satisfies("hdf5=none"): + args.append(self.define("QE_ENABLE_HDF5", True)) + + # Work around spack issue #19970 where spack sets + # rpaths for MKL just during make, but cmake removes + # them during make install. + if self.spec.satisfies("^[virtuals=lapack] intel-oneapi-mkl"): + args.append("-DCMAKE_INSTALL_RPATH_USE_LINK_PATH=ON") + spec = self.spec + args.append(self.define("BLAS_LIBRARIES", spec["blas"].libs.joined(";"))) + args.append(self.define("LAPACK_LIBRARIES", spec["lapack"].libs.joined(";"))) + + return args diff --git a/ci/spack_repo/sirius_ci/packages/rocblas/0001-remove-blas-override-7.2.patch b/ci/spack_repo/sirius_ci/packages/rocblas/0001-remove-blas-override-7.2.patch new file mode 100644 index 000000000..1470bd924 --- /dev/null +++ b/ci/spack_repo/sirius_ci/packages/rocblas/0001-remove-blas-override-7.2.patch @@ -0,0 +1,19 @@ +diff --git a/projects/rocblas/clients/CMakeLists.txt b/projects/rocblas/clients/CMakeLists.txt +index 0f32675..65e458d 100644 +--- a/projects/rocblas/clients/CMakeLists.txt ++++ b/projects/rocblas/clients/CMakeLists.txt +@@ -163,14 +163,6 @@ if( BUILD_CLIENTS_BENCHMARKS OR BUILD_CLIENTS_TESTS) + find_package(PkgConfig) + if(NOT PKG_CONFIG_FOUND) + message( FATAL_ERROR "Could not find libblis and pkgconfig is not available" ) +- else() +- pkg_search_module(PKGBLAS cblas) +- if(PKGBLAS_FOUND) +- set( BLAS_LIBRARY ${PKGBLAS_LIBRARIES} ) +- set( BLAS_INCLUDE_DIR ${PKGBLAS_INCLUDE_DIRS} ) +- else() +- message( FATAL_ERROR "Could not find libblis and pkgconfig can not find any other implementation of cblas" ) +- endif() + endif() + endif() + else() diff --git a/ci/spack_repo/sirius_ci/packages/rocblas/0001-remove-blas-override.patch b/ci/spack_repo/sirius_ci/packages/rocblas/0001-remove-blas-override.patch new file mode 100644 index 000000000..982f21143 --- /dev/null +++ b/ci/spack_repo/sirius_ci/packages/rocblas/0001-remove-blas-override.patch @@ -0,0 +1,19 @@ +diff --git a/clients/CMakeLists.txt b/clients/CMakeLists.txt +index 0f32675..65e458d 100644 +--- a/clients/CMakeLists.txt ++++ b/clients/CMakeLists.txt +@@ -163,14 +163,6 @@ if( BUILD_CLIENTS_BENCHMARKS OR BUILD_CLIENTS_TESTS) + find_package(PkgConfig) + if(NOT PKG_CONFIG_FOUND) + message( FATAL_ERROR "Could not find libblis and pkgconfig is not available" ) +- else() +- pkg_search_module(PKGBLAS cblas) +- if(PKGBLAS_FOUND) +- set( BLAS_LIBRARY ${PKGBLAS_LIBRARIES} ) +- set( BLAS_INCLUDE_DIR ${PKGBLAS_INCLUDE_DIRS} ) +- else() +- message( FATAL_ERROR "Could not find libblis and pkgconfig can not find any other implementation of cblas" ) +- endif() + endif() + endif() + else() diff --git a/ci/spack_repo/sirius_ci/packages/rocblas/0007-add-rocm-openmp-extras-include-dir.patch b/ci/spack_repo/sirius_ci/packages/rocblas/0007-add-rocm-openmp-extras-include-dir.patch new file mode 100644 index 000000000..4e265011b --- /dev/null +++ b/ci/spack_repo/sirius_ci/packages/rocblas/0007-add-rocm-openmp-extras-include-dir.patch @@ -0,0 +1,20 @@ +diff --git a/clients/gtest/CMakeLists.txt b/clients/gtest/CMakeLists.txt +index 55e9bbe..32b350c 100644 +--- a/clients/gtest/CMakeLists.txt ++++ b/clients/gtest/CMakeLists.txt +@@ -164,6 +164,7 @@ target_include_directories( rocblas-test + $ + $ # may be blank if not used + $ ++ $ + ) + target_include_directories( rocblas_v3-test + SYSTEM PRIVATE +@@ -171,6 +172,7 @@ target_include_directories( rocblas_v3-test + $ + $ # may be blank if not used + $ ++ $ + ) + + if( BUILD_FORTRAN_CLIENTS ) diff --git a/ci/spack_repo/sirius_ci/packages/rocblas/0008-link-roctracer.patch b/ci/spack_repo/sirius_ci/packages/rocblas/0008-link-roctracer.patch new file mode 100644 index 000000000..9e55ee44a --- /dev/null +++ b/ci/spack_repo/sirius_ci/packages/rocblas/0008-link-roctracer.patch @@ -0,0 +1,16 @@ +diff --git a/library/CMakeLists.txt b/library/CMakeLists.txt +index 762580d..1692578 100644 +--- a/library/CMakeLists.txt ++++ b/library/CMakeLists.txt +@@ -81,7 +81,10 @@ function( rocblas_library_settings lib_target_ ) + else() + target_link_libraries( ${lib_target_} PRIVATE hip::device -lstdc++fs --rtlib=compiler-rt --unwindlib=libgcc) + if (BUILD_SHARED_LIBS) +- target_link_libraries(${lib_target_} PRIVATE -lroctx64) ++ find_path(ROCTRACER_INCLUDE_DIR "roctracer/roctx.h") ++ target_include_directories(${lib_target_} PUBLIC ${ROCTRACER_INCLUDE_DIR}) ++ find_library(ROCTX_LIB NAMES roctx64) ++ target_link_libraries(${lib_target_} PRIVATE ${ROCTX_LIB}) + endif() + endif() + target_link_libraries( ${lib_target_} PRIVATE Threads::Threads ) diff --git a/ci/spack_repo/sirius_ci/packages/rocblas/0009-use-rocm-smi-config.patch b/ci/spack_repo/sirius_ci/packages/rocblas/0009-use-rocm-smi-config.patch new file mode 100644 index 000000000..0b474bc02 --- /dev/null +++ b/ci/spack_repo/sirius_ci/packages/rocblas/0009-use-rocm-smi-config.patch @@ -0,0 +1,15 @@ +diff --git a/clients/CMakeLists.txt b/clients/CMakeLists.txt +index 6194286..292bd8d 100644 +--- a/clients/CMakeLists.txt ++++ b/clients/CMakeLists.txt +@@ -178,8 +178,8 @@ if( BUILD_CLIENTS_BENCHMARKS OR BUILD_CLIENTS_TESTS) + + # Find the package ROCmSMI + if(NOT WIN32) +- find_package(ROCmSMI REQUIRED) +- list( APPEND COMMON_LINK_LIBS rocm_smi ) ++ find_package(rocm_smi CONFIG REQUIRED) ++ list( APPEND COMMON_LINK_LIBS rocm_smi64 ) + endif() + + find_package( GTest REQUIRED ) diff --git a/ci/spack_repo/sirius_ci/packages/rocblas/package.py b/ci/spack_repo/sirius_ci/packages/rocblas/package.py new file mode 100644 index 000000000..514887c7d --- /dev/null +++ b/ci/spack_repo/sirius_ci/packages/rocblas/package.py @@ -0,0 +1,326 @@ +# Copyright Spack Project Developers. See COPYRIGHT file for details. +# +# SPDX-License-Identifier: (Apache-2.0 OR MIT) + +import re + +from spack_repo.builtin.build_systems.cmake import CMakePackage +from spack_repo.builtin.build_systems.rocm import ROCmPackage + +from spack.package import * + + +class Rocblas(CMakePackage): + """Radeon Open Compute BLAS library""" + + homepage = "https://github.com/ROCm/rocBLAS/" + git = "https://github.com/ROCm/rocm-libraries.git" + + tags = ["rocm"] + maintainers("cgmb", "srekolam", "renjithravindrankannath", "haampie", "afzpatel") + libraries = ["librocblas"] + license("MIT") + + def url_for_version(self, version): + if version <= Version("7.1.1"): + url = "https://github.com/ROCm/rocBLAS/archive/refs/tags/rocm-{0}.tar.gz" + else: + url = "https://github.com/ROCm/rocm-libraries/archive/rocm-{0}.tar.gz" + return url.format(version) + + version("7.2.0", sha256="8ad5f4a11f1ed8a7b927f2e65f24083ca6ce902a42021a66a815190a91ccb654") + version("7.1.1", sha256="29d43270ccf5d4818d261993f964d4fce4bd0a55c2b6dde60d1529b6c227a873") + version("7.1.0", sha256="54f38222d0e58344cf5c86f151d418c071b59145297fd2ed953bb561df1e12c3") + version("7.0.2", sha256="8398cda68242db2386abc9eaf00c3588bb27e2b382e29be2bc5624c2d4ac8a99") + version("7.0.0", sha256="337a77cec31927e484672002d245d3aebf7a67e95658a8477fc593c95cf281fb") + version("6.4.3", sha256="754dcc88b30468a2293d2406d7fe40f78dc92dd77c193758f937532217ecdad3") + version("6.4.2", sha256="703226c458bb3dd1155aad8bdc02cdae2ff789c6b44e41e4a49ae28e40baff98") + version("6.4.1", sha256="517950ff6b3715dee8b2bcfbdd3968c65e1910e4b8e353e148574ae08aa6dc73") + version("6.4.0", sha256="ab8e75c9f98d17817a650aa4f06ff1e6c6af92cd143079e361cb6a0c96676aaa") + version("6.3.3", sha256="73e91bd50c920b818742fa5bf9990c0676be5bfbafe321d5781607dc2ce27060") + version("6.3.2", sha256="455cad760d926c21101594197c4456f617e5873a8f17bb3e14bd762018545a9e") + version("6.3.1", sha256="88d2de6ce6b23a157eea8be63408350848935e4dfc3e27e5f2add78834c6d6ba") + version("6.3.0", sha256="051f53bb69a9aba55a0c66c32688bf6af80e29e4a6b56b380b3c427e7a6aff9d") + version("6.2.4", sha256="8bacf74e3499c445f1bb0a8048df1ef3ce6f72388739b1823b5784fd1e8aa22a") + version("6.2.1", sha256="cf3bd7b47694f95f387803191615e2ff5c1106175473be7a5b2e8eb6fb99179f") + version("6.2.0", sha256="184e9b39dcbed57c25f351b047d44c613f8a2bbab3314a20c335f024a12ad4e5") + version("6.1.2", sha256="1e83918bd7b28ec9ee292c6fb7eb0fc5f4db2d5d831a9a3db541f14a90c20a1a") + version("6.1.1", sha256="c920742fb8f45512c360cdb40e37d0ac767f042e52f1981264853dab5ec2c876") + version("6.1.0", sha256="af00357909da60d82618038aa9a3cc1f9d4ce1bdfb54db20ec746b592d478edf") + version("6.0.2", sha256="d1bf31063a2d349797b88c994c91d05f94e681bafb5550ad9b53529703d89dbb") + version("6.0.0", sha256="befa4a75f1de0ea37f2358d4c2de5406d7bce671ca9936e2294b64d3b3bafb60") + version("5.7.1", sha256="2984a5ed0ea5a05d40996ee3fddecb24399cbe8ea3e4921fc254e54d8f52fe4f") + version("5.7.0", sha256="024edd98de9687ee5394badc4dd4c543eef4eb3f71c96ff64100705d851e1744") + + amdgpu_targets = ROCmPackage.amdgpu_targets + + variant( + "amdgpu_target", + description="AMD GPU architecture", + values=auto_or_any_combination_of(*amdgpu_targets), + sticky=True, + ) + variant("tensile", default=True, description="Use Tensile as a backend") + variant("asan", default=False, description="Build with address-sanitizer enabled or disabled") + variant("hipblaslt", default=True, when="@6.3:", description="Build with hipblaslt") + + conflicts("+asan", when="os=rhel9") + conflicts("+asan", when="os=centos7") + conflicts("+asan", when="os=centos8") + + depends_on("c", type="build") # generated + depends_on("cxx", type="build") # generated + depends_on("fortran", type="build") # generated + + depends_on("cmake@3.16.8:", type="build") + + depends_on("googletest@1.10.0:", type="test") + depends_on("amdblis", type="test") + + for ver in [ + "6.2.0", + "6.2.1", + "6.2.4", + "6.3.0", + "6.3.1", + "6.3.2", + "6.3.3", + "6.4.0", + "6.4.1", + "6.4.2", + "6.4.3", + "7.0.0", + "7.0.2", + "7.1.0", + "7.1.1", + "7.2.0", + ]: + depends_on(f"rocm-smi-lib@{ver}", type="test", when=f"@{ver}") + + for ver in [ + "5.7.0", + "5.7.1", + "6.0.0", + "6.0.2", + "6.1.0", + "6.1.1", + "6.1.2", + "6.2.0", + "6.2.1", + "6.2.4", + "6.3.0", + "6.3.1", + "6.3.2", + "6.3.3", + "6.4.0", + "6.4.1", + "6.4.2", + "6.4.3", + "7.0.0", + "7.0.2", + "7.1.0", + "7.1.1", + ]: + depends_on(f"rocm-openmp-extras@{ver}", type="test", when=f"@{ver}") + + for ver in [ + "5.7.0", + "5.7.1", + "6.0.0", + "6.0.2", + "6.1.0", + "6.1.1", + "6.1.2", + "6.2.0", + "6.2.1", + "6.2.4", + "6.3.0", + "6.3.1", + "6.3.2", + "6.3.3", + "6.4.0", + "6.4.1", + "6.4.2", + "6.4.3", + "7.0.0", + "7.0.2", + "7.1.0", + "7.1.1", + "7.2.0", + ]: + depends_on(f"hip@{ver}", when=f"@{ver}") + depends_on(f"llvm-amdgpu@{ver}", type="build", when=f"@{ver}") + depends_on(f"rocminfo@{ver}", type="build", when=f"@{ver}") + depends_on(f"rocm-cmake@{ver}", type="build", when=f"@{ver}") + + for ver in [ + "6.3.0", + "6.3.1", + "6.3.2", + "6.3.3", + "6.4.0", + "6.4.1", + "6.4.2", + "6.4.3", + "7.0.0", + "7.0.2", + "7.1.0", + "7.1.1", + "7.2.0", + ]: + depends_on(f"hipblaslt@{ver}", when=f"@{ver} +hipblaslt") + + for ver in ["6.4.0", "6.4.1", "6.4.2", "6.4.3", "7.0.0", "7.0.2", "7.1.0", "7.1.1", "7.2.0"]: + depends_on(f"roctracer-dev@{ver}", when=f"@{ver}") + + depends_on("python@3.6:", type="build") + + with when("+tensile"): + depends_on("msgpack-c@3:") + + depends_on("py-virtualenv", type="build") + depends_on("perl-file-which", type="build") + depends_on("py-pyyaml", type="build") + depends_on("py-wheel", type="build") + depends_on("py-msgpack", type="build") + depends_on("py-pip", type="build") + depends_on("py-joblib", type="build") + depends_on("procps", type="build") + + for t_version, t_commit in [ + ("@5.7.0", "97e0cfc2c8cb87a1e38901d99c39090dc4181652"), + ("@5.7.1", "97e0cfc2c8cb87a1e38901d99c39090dc4181652"), + ("@6.0.0", "17df881bde80fc20f997dfb290f4bb4b0e05a7e9"), + ("@6.0.2", "17df881bde80fc20f997dfb290f4bb4b0e05a7e9"), + ("@6.1.0", "2b55ccf58712f67b3df0ca53b0445f094fcb96b2"), + ("@6.1.1", "2b55ccf58712f67b3df0ca53b0445f094fcb96b2"), + ("@6.1.2", "2b55ccf58712f67b3df0ca53b0445f094fcb96b2"), + ("@6.2.0", "dbc2062dced66e4cbee8e0591d76e0a1588a4c70"), + ("@6.2.1", "dbc2062dced66e4cbee8e0591d76e0a1588a4c70"), + ("@6.2.4", "81ae9537671627fe541332c0a5d953bfd6af71d6"), + ("@6.3.0", "aca95d1743c243dd0dd0c8b924608bc915ce1ae7"), + ("@6.3.1", "aca95d1743c243dd0dd0c8b924608bc915ce1ae7"), + ("@6.3.2", "aca95d1743c243dd0dd0c8b924608bc915ce1ae7"), + ("@6.3.3", "aca95d1743c243dd0dd0c8b924608bc915ce1ae7"), + ("@6.4.0", "be49885fce2a61b600ae4593f1c2d00c8b4fa11e"), + ("@6.4.1", "be49885fce2a61b600ae4593f1c2d00c8b4fa11e"), + ("@6.4.2", "be49885fce2a61b600ae4593f1c2d00c8b4fa11e"), + ("@6.4.3", "be49885fce2a61b600ae4593f1c2d00c8b4fa11e"), + ("@7.0.0", "cca3c8136aa812109629e6291ce9f0ca846b68d3"), + ("@7.0.2", "63c27e505cb532ff8e568d737bfdbd9e1d024665"), + ("@7.1.0", "0c8314da90fee8cf3b16dcb1bbc75bc1266e123f"), + ("@7.1.1", "0c8314da90fee8cf3b16dcb1bbc75bc1266e123f"), + ]: + resource( + name="Tensile", + git="https://github.com/ROCm/Tensile.git", + commit=t_commit, + placement="Tensile", + when=f"{t_version} +tensile", + ) + + patch("0007-add-rocm-openmp-extras-include-dir.patch", when="@5.7") + patch("0008-link-roctracer.patch", when="@6.4") + patch("0009-use-rocm-smi-config.patch", when="@6.4:7.1") + patch("0001-remove-blas-override.patch", when="@7.1") + patch("0001-remove-blas-override-7.2.patch", when="@7.2:") + patch( + "https://github.com/ROCm/rocm-libraries/commit/b3b20a3ea53051a14f30c28e577620c0beeea57c.patch?full_index=1", + sha256="1f436c5ad03c8fdc021f309a1ad84d4356f30c39c4cc940bb8267841561bf5f1", + when="@7.2", + ) + + @property + def root_cmakelists_dir(self): + if self.spec.satisfies("@7.2:"): + return "projects/rocblas" + else: + return "." + + def setup_build_environment(self, env: EnvironmentModifications) -> None: + env.set("CXX", self.spec["hip"].hipcc) + if self.spec.satisfies("+asan"): + env.set("CC", f"{self.spec['llvm-amdgpu'].prefix}/bin/clang") + env.set("CXX", f"{self.spec['llvm-amdgpu'].prefix}/bin/clang++") + env.set("ASAN_OPTIONS", "detect_leaks=0") + env.set("CFLAGS", "-fsanitize=address -shared-libasan") + env.set("CXXFLAGS", "-fsanitize=address -shared-libasan") + env.set("LDFLAGS", "-fuse-ld=lld") + + @classmethod + def determine_version(cls, lib): + match = re.search(r"lib\S*\.so\.\d+\.\d+\.(\d)(\d\d)(\d\d)", lib) + if match: + ver = "{0}.{1}.{2}".format( + int(match.group(1)), int(match.group(2)), int(match.group(3)) + ) + else: + ver = None + return ver + + def cmake_args(self): + args = [ + self.define("BUILD_CLIENTS_TESTS", self.run_tests), + self.define("BUILD_CLIENTS_BENCHMARKS", "OFF"), + self.define("BUILD_CLIENTS_SAMPLES", "OFF"), + self.define("RUN_HEADER_TESTING", "OFF"), + self.define_from_variant("BUILD_WITH_TENSILE", "tensile"), + self.define("CMAKE_INSTALL_LIBDIR", "lib"), + self.define("Tensile_CODE_OBJECT_VERSION", "default"), + ] + if self.run_tests: + args.append(self.define("LINK_BLIS", "ON")) + args.append( + self.define("BLIS_INCLUDE_DIR", self.spec["amdblis"].prefix + "/include/blis/") + ) + args.append( + self.define("BLAS_LIBRARY", self.spec["amdblis"].prefix + "/lib/libblis.a") + ) + if self.spec.satisfies("@:7.1"): + args.append( + self.define("ROCM_OPENMP_EXTRAS_DIR", self.spec["rocm-openmp-extras"].prefix) + ) + + if "+tensile" in self.spec: + if self.spec.satisfies("@:6.2"): + tensile_compiler = "hipcc" + else: + tensile_compiler = f"{self.spec['llvm-amdgpu'].prefix}/bin/amdclang++" + args += [ + self.define("Tensile_COMPILER", tensile_compiler), + self.define("Tensile_LOGIC", "asm_full"), + self.define("BUILD_WITH_TENSILE_HOST", "@3.7.0:" in self.spec), + self.define("Tensile_LIBRARY_FORMAT", "msgpack"), + ] + if self.spec.satisfies("@:7.1"): + tensile_path = join_path(self.stage.source_path, "Tensile") + args.append(self.define("Tensile_TEST_LOCAL_PATH", tensile_path)) + # Restrict the number of jobs Tensile can spawn. + # If we don't specify otherwise, Tensile creates a job per available core, + # and that consumes a lot of system memory. + # https://github.com/ROCm/Tensile/blob/93e10678a0ced7843d9332b80bc17ebf9a166e8e/Tensile/Parallel.py#L38 + args.append(self.define("Tensile_CPU_THREADS", min(16, make_jobs))) + + if "auto" not in self.spec.variants["amdgpu_target"]: + if self.spec.satisfies("@7.1:"): + args.append(self.define_from_variant("GPU_TARGETS", "amdgpu_target")) + else: + args.append(self.define_from_variant("AMDGPU_TARGETS", "amdgpu_target")) + + # See https://github.com/ROCm/rocBLAS/issues/1196 + if self.spec.satisfies("^cmake@3.21.0:3.21.2"): + args.append(self.define("__skip_rocmclang", "ON")) + + if self.spec.satisfies("@:6.3.1"): + args.append(self.define("BUILD_FILE_REORG_BACKWARD_COMPATIBILITY", True)) + if self.spec.satisfies("@6.3:"): + args.append(self.define_from_variant("BUILD_WITH_HIPBLASLT", "hipblaslt")) + if self.spec.satisfies("@7.1:"): + args.append(self.define("ROCTX_PATH", self.spec["roctracer-dev"].prefix)) + return args + + @run_after("build") + @on_package_attributes(run_tests=True) + def check_build(self): + exe = Executable(join_path(self.build_directory, "clients", "staging", "rocblas-test")) + exe("--gtest_filter=*quick*-*known_bug*") diff --git a/ci/spack_repo/sirius_ci/packages/rocprofiler_systems/package.py b/ci/spack_repo/sirius_ci/packages/rocprofiler_systems/package.py new file mode 100644 index 000000000..f6f996cab --- /dev/null +++ b/ci/spack_repo/sirius_ci/packages/rocprofiler_systems/package.py @@ -0,0 +1,360 @@ +# Copyright Spack Project Developers. See COPYRIGHT file for details. +# +# SPDX-License-Identifier: (Apache-2.0 OR MIT) + +import re + +from spack_repo.builtin.build_systems.cmake import CMakePackage + +from spack.package import * + + +def submodules(package): + submodules = [ + "projects/rocprofiler-systems/external/timemory", + "projects/rocprofiler-systems/external/perfetto", + "projects/rocprofiler-systems/external/elfio", + "projects/rocprofiler-systems/external/dyninst", + "projects/rocprofiler-systems/external/PTL", + "projects/rocprofiler-systems/external/papi", + "projects/rocprofiler-systems/external/pybind11", + "projects/rocprofiler-systems/examples/openmp/external/ompvv", + ] + return submodules + + +class RocprofilerSystems(CMakePackage): + """Application Profiling, Tracing, and Analysis""" + + homepage = "https://github.com/ROCm/rocprofiler-systems" + git = "https://github.com/ROCm/rocm-systems.git" + url = "https://github.com/ROCm/rocprofiler-systems/archive/refs/tags/rocm-6.4.3.tar.gz" + executables = ["rocprof-sys-sample"] + tags = ["rocm"] + + maintainers("dgaliffiAMD", "afzpatel", "srekolam", "renjithravindrankannath") + + license("MIT") + + version( + "7.2.0", + git="https://github.com/ROCm/rocm-systems.git", + tag="rocm-7.2.0", + commit="fc0010cf6a5a972d42b276df946510f30343d493", + submodules=submodules, + ) + version( + "7.1.1", + git="https://github.com/ROCm/rocprofiler-systems", + branch="release/rocm-rel-7.1.1", + submodules=True, + ) + version( + "7.1.0", + git="https://github.com/ROCm/rocprofiler-systems", + tag="rocm-7.1.0", + commit="427f656162559f21fc3d6cb0e3688f3d31ae374c", + submodules=True, + ) + version( + "7.0.2", + git="https://github.com/ROCm/rocprofiler-systems", + tag="rocm-7.0.2", + commit="8bad624afb06ea2b567e985484d1b5d604865743", + submodules=True, + ) + version( + "7.0.0", + git="https://github.com/ROCm/rocprofiler-systems", + tag="rocm-7.0.0", + commit="1030d99db9934a07f1d276f6aadd0eb810b5f5f9", + submodules=True, + ) + version( + "6.4.3", + git="https://github.com/ROCm/rocprofiler-systems", + tag="rocm-6.4.3", + commit="ba0bfe8cf344294347cbb854084ab5b5df1b1a43", + submodules=True, + ) + version( + "6.4.2", + git="https://github.com/ROCm/rocprofiler-systems", + tag="rocm-6.4.2", + commit="ba0bfe8cf344294347cbb854084ab5b5df1b1a43", + submodules=True, + ) + version( + "6.4.1", + git="https://github.com/ROCm/rocprofiler-systems", + tag="rocm-6.4.1", + commit="2e945e4a08781e13a822f568814e2c434fd8858f", + submodules=True, + ) + version( + "6.4.0", + git="https://github.com/ROCm/rocprofiler-systems", + tag="rocm-6.4.0", + commit="c4cec593b6021f2b84294838f2ffe388ed8e911a", + submodules=True, + ) + version( + "6.3.3", + git="https://github.com/ROCm/rocprofiler-systems", + tag="rocm-6.3.3", + commit="f03ef1dd9a4e984e3e72056352532e6149e742fc", + submodules=True, + ) + version( + "6.3.2", + git="https://github.com/ROCm/rocprofiler-systems", + tag="rocm-6.3.2", + commit="2fd5fbbef941ff219a1ecef702f8cfaae6e8e5ba", + submodules=True, + ) + version( + "6.3.1", + git="https://github.com/ROCm/rocprofiler-systems", + tag="rocm-6.3.1", + commit="04a84dd0b0df3dfd61f7765696e0e474ec29f10b", + submodules=True, + ) + + version( + "6.3.0", + git="https://github.com/ROCm/rocprofiler-systems", + tag="rocm-6.3.0", + commit="71a5e271b5e07efd2948fb6e7b451db5e8e40cb8", + submodules=True, + ) + + variant( + "rocm", + default=True, + description="Enable ROCm API, kernel tracing, and GPU HW counters support", + ) + variant("strip", default=False, description="Faster binary instrumentation, worse debugging") + variant( + "python", default=False, description="Enable support for Python function profiling and API" + ) + variant("papi", default=True, description="Enable HW counters support via PAPI") + variant("ompt", default=True, description="Enable OpenMP Tools support") + variant( + "tau", + default=False, + description="Enable support for using TAU markers in omnitrace instrumentation", + ) + variant( + "caliper", + default=False, + description="Enable support for using Caliper markers in omnitrace instrumentation", + ) + variant( + "perfetto_tools", + default=False, + description="Install perfetto tools (e.g. traced, perfetto)", + ) + variant( + "mpi", + default=False, + description=( + "Enable intercepting MPI functions and aggregating output during finalization " + "(requires target application to use same MPI installation)" + ), + ) + variant( + "mpi_headers", + default=True, + description=( + "Enable intercepting MPI functions but w/o support for aggregating output " + "(target application can use any MPI installation)" + ), + ) + variant("internal-dyninst", default=False, description="build internal dyninst") + variant("internal-boost", default=False, description="build internal boost") + variant("internal-tbb", default=False, description="build internal tbb") + + conflicts("%rocmcc", when="+internal-tbb") + conflicts("%clang", when="+internal-tbb") + extends("python", when="+python") + + depends_on("c", type="build") # generated + depends_on("cxx", type="build") # generated + depends_on("fortran", type="build") # generated + + # hard dependencies + depends_on("cmake@3.16:", type="build") + depends_on("dyninst@:12", when="@6 ~internal-dyninst") + depends_on("dyninst@13", when="@7 ~internal-dyninst") + depends_on( + "boost@:1.88" + "+atomic+chrono+date_time+filesystem+system+thread+timer+container+random+exception", + when="@:7.1.0 +internal-dyninst", + ) + depends_on( + "boost@:1.88" + "+atomic+chrono+date_time+filesystem+system+thread+timer+container+random+exception", + when="@7.1.1:", + ) + depends_on("libiberty+pic", when="+internal-dyninst") + depends_on("intel-tbb@2019:2020.3", when="~internal-tbb") + depends_on("sqlite", when="@7.1:") + depends_on("elfutils") + depends_on("m4") + depends_on("texinfo") + depends_on("libunwind", type=("build", "run")) + depends_on("papi+shared", when="+papi") + depends_on("mpi", when="+mpi") + depends_on("tau", when="+tau") + depends_on("caliper", when="+caliper") + depends_on("python@3:", when="+python", type=("build", "run")) + depends_on("libunwind", when="+rocm") + depends_on("autoconf", when="+rocm") + depends_on("automake", when="+rocm") + depends_on("libtool", when="+rocm") + depends_on("sqlite", when="@7.1:") + with when("+rocm"): + for ver in ["6.3.0", "6.3.1", "6.3.2", "6.3.3"]: + depends_on(f"roctracer-dev@{ver}", when=f"@{ver}") + depends_on(f"rocprofiler-dev@{ver}", when=f"@{ver}") + + for ver in ["6.3.0", "6.3.1", "6.3.2", "6.3.3", "6.4.0", "6.4.1", "6.4.2", "6.4.3"]: + depends_on(f"rocm-smi-lib@{ver}", when=f"@{ver}") + + for ver in [ + "6.3.0", + "6.3.1", + "6.3.2", + "6.3.3", + "6.4.0", + "6.4.1", + "6.4.2", + "6.4.3", + "7.0.0", + "7.0.2", + "7.1.0", + "7.1.1", + "7.2.0", + ]: + depends_on(f"hip@{ver}", when=f"@{ver}") + + for ver in [ + "6.4.0", + "6.4.1", + "6.4.2", + "6.4.3", + "7.0.0", + "7.0.2", + "7.1.0", + "7.1.1", + "7.2.0", + ]: + depends_on(f"rocprofiler-sdk@{ver}", when=f"@{ver}") + + for ver in ["7.0.0", "7.0.2", "7.1.0", "7.1.1", "7.2.0"]: + depends_on(f"amdsmi@{ver}", when=f"@{ver}") + + # Fix GCC 13 build failure caused by a missing include of in dyninst + patch( + "https://github.com/ROCm/dyninst/commit/09e781d414c83b4ad587083d449a3e976546937d.patch?full_index=1", + sha256="e64c6b75393e7fbd711c0bd0233628c176a352cd10b4057f00eec283426eaf0a", + when="@:6.4.0 +internal-dyninst", + working_dir="external/dyninst", + ) + patch( + "https://github.com/ROCm/timemory/commit/b5e41aa9e4b83ab0868211d81924ac4f639bd998.patch?full_index=1", + sha256="2696f59dd9b6e74bf44bfcc56a0536c3f1f3845c29fac18f0224dee72bd9225f", + when="%rocmcc", + working_dir="external/timemory", + ) + + @property + def root_cmakelists_dir(self): + if self.spec.satisfies("@:7.1"): + return "." + else: + return "projects/rocprofiler-systems" + + def cmake_args(self): + spec = self.spec + + args = [ + self.define("SPACK_BUILD", True), + self.define("ROCPROFSYS_BUILD_PAPI", False), + self.define("ROCPROFSYS_BUILD_PYTHON", True), + self.define("ROCPROFSYS_BUILD_LIBUNWIND", False), + self.define("ROCPROFSYS_BUILD_STATIC_LIBGCC", False), + self.define("ROCPROFSYS_BUILD_STATIC_LIBSTDCXX", False), + self.define_from_variant("ROCPROFSYS_BUILD_DYNINST", "internal-dyninst"), + self.define_from_variant("ROCPROFSYS_BUILD_LTO", "ipo"), + self.define_from_variant("ROCPROFSYS_USE_MPI", "mpi"), + self.define_from_variant("ROCPROFSYS_USE_OMPT", "ompt"), + self.define_from_variant("ROCPROFSYS_USE_PAPI", "papi"), + self.define_from_variant("ROCPROFSYS_USE_RCCL", "rocm"), + self.define_from_variant("ROCPROFSYS_USE_PYTHON", "python"), + self.define_from_variant("ROCPROFSYS_USE_MPI_HEADERS", "mpi_headers"), + self.define_from_variant("ROCPROFSYS_STRIP_LIBRARIES", "strip"), + self.define_from_variant("ROCPROFSYS_INSTALL_PERFETTO_TOOLS", "perfetto_tools"), + self.define("ElfUtils_ROOT_DIR", spec["elfutils"].prefix), + # timemory arguments + self.define("TIMEMORY_BUILD_CALIPER", False), + self.define_from_variant("TIMEMORY_USE_TAU", "tau"), + self.define_from_variant("TIMEMORY_USE_CALIPER", "caliper"), + ] + if spec.satisfies("@:6.3"): + args.append(self.define_from_variant("ROCPROFSYS_USE_ROCM_SMI", "rocm")) + args.append(self.define_from_variant("ROCPROFSYS_USE_HIP", "rocm")) + args.append(self.define_from_variant("ROCPROFSYS_USE_ROCTRACER", "rocm")) + args.append(self.define_from_variant("ROCPROFSYS_USE_ROCPROFILER", "rocm")) + else: + args.append(self.define_from_variant("ROCPROFSYS_USE_ROCM", "rocm")) + + if "+tau" in spec: + tau_root = spec["tau"].prefix + args.append(self.define("TAU_ROOT_DIR", tau_root)) + + if "+mpi" in spec: + args.append(self.define("MPI_C_COMPILER", spec["mpi"].mpicc)) + args.append(self.define("MPI_CXX_COMPILER", spec["mpi"].mpicxx)) + + if spec.satisfies("@6.3:"): + args.append(self.define("dl_LIBRARY", "dl")) + args.append( + self.define("libunwind_INCLUDE_DIR", self.spec["libunwind"].prefix.include) + ) + if spec.satisfies("@7.0:"): + args.append(self.define_from_variant("ROCPROFSYS_BUILD_TBB", "internal-tbb")) + if spec.satisfies("+internal-dyninst"): + args.append(self.define_from_variant("DYNINST_BUILD_TBB", "internal-tbb")) + if spec.satisfies("+internal-boost"): + args.append(self.define_from_variant("ROCPROFSYS_BUILD_BOOST", "internal-boost")) + if spec.satisfies("@7.2:"): + args.append(self.define("libunwind_ROOT", self.spec["libunwind"].prefix)) + return args + + ## newer glibc have lintl inlcuded + # def flag_handler(self, name, flags): + # if self.spec.satisfies("@6.3:"): + # if name == "ldflags": + # flags.append("-lintl") + # return (flags, None, None) + + @classmethod + def determine_version(cls, exe): + output = Executable(exe)("--version", output=str, error=str) + match = re.search(r"rocm: v(\d+)\.(\d+)", output) + if match: + ver = "{0}.{1}".format(int(match.group(1)), int(match.group(2))) + else: + ver = None + return ver + + def setup_build_environment(self, env: EnvironmentModifications) -> None: + if "+tau" in self.spec: + import glob + + # below is how TAU_MAKEFILE is set in packages/tau/package.py + pattern = join_path(self.spec["tau"].prefix.lib, "Makefile.*") + files = glob.glob(pattern) + if files: + env.set("TAU_MAKEFILE", files[0]) diff --git a/ci/spack_repo/sirius_ci/packages/sirius/fj.patch b/ci/spack_repo/sirius_ci/packages/sirius/fj.patch new file mode 100644 index 000000000..567ced347 --- /dev/null +++ b/ci/spack_repo/sirius_ci/packages/sirius/fj.patch @@ -0,0 +1,13 @@ +diff --git a/src/hamiltonian/hamiltonian.cpp b/src/hamiltonian/hamiltonian.cpp +index 54a91df..ea66ecf 100644 +--- a/src/hamiltonian/hamiltonian.cpp ++++ b/src/hamiltonian/hamiltonian.cpp +@@ -74,7 +74,7 @@ Hamiltonian0::Hamiltonian0(Potential& potential__, bool precompute_lapw__) + for (int j1 = 0; j1 <= j2; j1++) { + int lm1 = type.indexb(j1).lm; + int idxrf1 = type.indexb(j1).idxrf; +- hmt_[ia](j1, j2) = atom.radial_integrals_sum_L3(idxrf1, idxrf2, ++ hmt_[ia](j1, j2) = atom.template radial_integrals_sum_L3(idxrf1, idxrf2, + type.gaunt_coefs().gaunt_vector(lm1, lm2)); + hmt_[ia](j2, j1) = std::conj(hmt_[ia](j1, j2)); + } diff --git a/ci/spack_repo/sirius_ci/packages/sirius/libxc7.patch b/ci/spack_repo/sirius_ci/packages/sirius/libxc7.patch new file mode 100644 index 000000000..d6d388e13 --- /dev/null +++ b/ci/spack_repo/sirius_ci/packages/sirius/libxc7.patch @@ -0,0 +1,16 @@ +--- a/src/potential/xc_functional_base.hpp ++++ b/src/potential/xc_functional_base.hpp +@@ -15,6 +15,13 @@ + #define __XC_FUNCTIONAL_BASE_HPP__ + + #include ++ ++/* libxc >= 7 split the functional definition in a different file from xc.h */ ++ ++#if (XC_MAJOR_VERSION >= 7) ++#include ++#endif ++ + #include + #include + #include diff --git a/ci/spack_repo/sirius_ci/packages/sirius/package.py b/ci/spack_repo/sirius_ci/packages/sirius/package.py new file mode 100644 index 000000000..81da54d4c --- /dev/null +++ b/ci/spack_repo/sirius_ci/packages/sirius/package.py @@ -0,0 +1,325 @@ +# Copyright Spack Project Developers. See COPYRIGHT file for details. +# +# SPDX-License-Identifier: (Apache-2.0 OR MIT) + +import os + +from spack_repo.builtin.build_systems.cmake import CMakePackage +from spack_repo.builtin.build_systems.cuda import CudaPackage +from spack_repo.builtin.build_systems.rocm import ROCmPackage + +from spack.package import * + + +class Sirius(CMakePackage, CudaPackage, ROCmPackage): + """Domain specific library for electronic structure calculations""" + + homepage = "https://github.com/electronic-structure/SIRIUS" + url = "https://github.com/electronic-structure/SIRIUS/archive/v6.1.5.tar.gz" + list_url = "https://github.com/electronic-structure/SIRIUS/releases" + git = "https://github.com/electronic-structure/SIRIUS.git" + + maintainers( + "simonpintarelli", "haampie", "dev-zero", "AdhocMan", "toxa81", "RMeli", "mtaillefumier" + ) + + license("BSD-2-Clause") + + version("develop", branch="develop") + version("master", branch="master") + + version("7.10.0", sha256="ef02f2a899ccd72803bed3c4b7460457e67d99a68e0d6e5ab157566a1e6abf4f") + version("7.9.0", sha256="c36e9a00637b9626f83c0db740751440bfe06ef4c8d7cadb9822dd3cf03c1046") + version("7.8.0", sha256="2cd2f98d35fb9e0a8f6d68714c6f8d682895781d564e91ef6685d92569ffd413") + version("7.7.1", sha256="6039c84197d9e719e826f98b840cff19bc513887b443f97c0099d3c8b908efed") + version("7.7.0", sha256="be0bdc76db9eb8afdcb950f0ccaf7535b8e85d72a4232dc92246f54fa68d9d7b") + version("7.6.2", sha256="1ba92942aa39b49771677cc8bf1c94a0b4350eb45bf3009318a6c2350b46a276") + version("7.6.1", sha256="16a114dc17e28697750585820e69718a96e6929f88406d266c75cf9a7cdbdaaa") + version("7.6.0", sha256="e424206fecb35bb2082b5c87f0865a9536040e984b88b041e6f7d531f8a65b20") + version("7.5.2", sha256="9ae01935578532c84f1d0d673dbbcdd490e26be22efa6c4acf7129f9dc1a0c60") + version("7.5.1", sha256="aadfa7976e90a109aeb1677042454388a8d1a50d75834d59c86c8aef06bc12e4") + version("7.5.0", sha256="c583f88ffc02e9acac24e786bc35c7c32066882d2f70a1e0c14b5780b510365d") + version("7.4.3", sha256="015679a60a39fa750c5d1bd8fb1ce73945524bef561270d8a171ea2fd4687fec") + version("7.4.0", sha256="f9360a695a1e786d8cb9d6702c82dd95144a530c4fa7e8115791c7d1e92b020b") + version("7.3.2", sha256="a256508de6b344345c295ad8642dbb260c4753cd87cc3dd192605c33542955d7") + version("7.3.1", sha256="8bf9848b8ebf0b43797fd359adf8c84f00822de4eb677e3049f22baa72735e98") + version("7.3.0", sha256="69b5cf356adbe181be6c919032859c4e0160901ff42a885d7e7ea0f38cc772e2") + + variant("shared", default=True, description="Build shared libraries") + variant("openmp", default=True, description="Build with OpenMP support") + variant("fortran", default=False, description="Build Fortran bindings") + variant("python", default=False, description="Build Python bindings") + variant("memory_pool", default=True, description="Build with memory pool") + variant("elpa", default=False, description="Use ELPA") + variant("dlaf", default=False, when="@7.5.0:", description="Use DLA-Future") + variant("vdwxc", default=False, description="Enable libvdwxc support") + variant("scalapack", default=False, description="Enable scalapack support") + variant("magma", default=False, description="Enable MAGMA support") + variant("nlcglib", default=False, description="Enable robust wave function optimization") + variant("vcsqnm", default=False, description="Enable lattice relaxation") + variant("wannier90", default=False, description="Enable Wannier90 library") + variant( + "build_type", + default="Release", + description="CMake build type", + values=("Debug", "Release", "RelWithDebInfo"), + ) + variant("apps", default=True, description="Build applications") + variant("tests", default=False, description="Build tests") + variant("single_precision", default=False, description="Use single precision arithmetics") + variant( + "profiler", default=True, description="Use internal profiler to measure execution time" + ) + variant("nvtx", default=False, description="Use NVTX profiler") + variant("dftd3", default=False, description="Enable dft-d3 corrections", when="@7.9.0:") + variant("dftd4", default=False, description="Enable dft-d4 corrections", when="@7.9.0:") + + with when("@7.6:"): + variant( + "pugixml", + default=False, + description="Enable direct reading of UPF v2 pseudopotentials", + ) + conflicts("+tests~pugixml") + depends_on("pugixml", when="+pugixml") + + depends_on("cxx", type="build") + depends_on("c", type="build") + depends_on("fortran", type="build") + + depends_on("cmake@3.25:", type="build") + depends_on("mpi") + depends_on("gsl") + depends_on("blas") + depends_on("lapack") + depends_on("fftw-api@3") + depends_on("libxc@3.0.0:") + depends_on("libxc@4.0.0:", when="@7.2.0:") + depends_on("libxc@:7", when="@:7.5") + depends_on("spglib") + depends_on("hdf5+hl") + depends_on("pkgconfig", type="build") + depends_on("fmt", when="@7.8:") + depends_on("simple-dftd3 build_system=cmake", when="+dftd3") + depends_on("dftd4 build_system=cmake", when="+dftd4") + + # Python module + depends_on("python", when="+python", type=("build", "run")) + depends_on("python", when="@:6", type=("build", "run")) + depends_on("py-numpy", when="+python", type=("build", "run")) + depends_on("py-scipy", when="+python", type=("build", "run")) + depends_on("py-h5py", when="+python", type=("build", "run")) + depends_on("py-mpi4py", when="+python", type=("build", "run")) + depends_on("py-pyyaml", when="+python", type=("build", "run")) + depends_on("py-mpi4py", when="+python", type=("build", "run")) + depends_on("py-voluptuous", when="+python", type=("build", "run")) + depends_on("py-pybind11", when="+python", type=("build", "run")) + extends("python", when="+python") + + depends_on("magma", when="+magma") + + with when("@7.0.1:"): + depends_on("spfft@0.9.13:") + depends_on("spfft+single_precision", when="+single_precision") + depends_on("spfft+cuda", when="+cuda") + depends_on("spfft+rocm", when="+rocm") + depends_on("spfft+openmp", when="+openmp") + + with when("@7.0.2:"): + depends_on("spla@1.1.0:") + depends_on("spla+cuda", when="+cuda") + depends_on("spla+rocm", when="+rocm") + # spla removed the openmp option in 1.6.0 + conflicts("^spla@:1.5~openmp", when="+openmp") + + with when("@7.8:"): + conflicts("nlcglib@:1.2") + + patch("libxc7.patch", when="@7.6.0:7.6.1") + patch( + "https://github.com/electronic-structure/SIRIUS/commit/dd07010f7b49f31b7e3bb1b4e47f3d9ac3a0c0b4.patch?full_index=1", + sha256="dd680f8c47a0fc29097cae5cd1e72dfdbcf95f93089f73fb3f2fe9e750125d6f", + when="@7.6.0:7.6.1 +pugixml", + ) + + depends_on("nlcglib", when="+nlcglib") + depends_on("nlcglib+rocm", when="+nlcglib+rocm") + depends_on("nlcglib+cuda", when="+nlcglib+cuda") + + depends_on("libvdwxc@0.3.0:+mpi", when="+vdwxc") + + depends_on("scalapack", when="+scalapack") + + with when("+dlaf"): + depends_on("dla-future@0.3.0:") + depends_on("dla-future +scalapack", when="+scalapack") + depends_on("dla-future +cuda", when="+cuda") + depends_on("dla-future +rocm", when="+rocm") + + conflicts("^pika@:0.22.1", when="+cuda") + conflicts("^pika@:0.22.1", when="+rocm") + + depends_on("rocblas", when="+rocm") + depends_on("rocsolver", when="@7.5.0: +rocm") + + # FindHIP cmake script only works for < 4.1, but HIP 4.1 is not provided by spack anymore + conflicts("+rocm", when="@:7.2.0") + + conflicts("^libxc@5.0.0") # known to produce incorrect results + conflicts("+single_precision", when="@:7.2.4") + conflicts("+scalapack", when="^cray-libsci") + + # Propagate openmp to blas + depends_on("openblas threads=openmp", when="+openmp ^[virtuals=blas,lapack] openblas") + depends_on("amdblis threads=openmp", when="+openmp ^[virtuals=blas] amdblis") + depends_on("blis threads=openmp", when="+openmp ^[virtuals=blas] blis") + depends_on( + "intel-oneapi-mkl threads=openmp", + when="+openmp ^[virtuals=blas,lapack,fftw-api] intel-oneapi-mkl", + ) + depends_on( + "intel-oneapi-mkl+cluster", + when="+scalapack ^[virtuals=blas,lapack,fftw-api] intel-oneapi-mkl", + ) + + # MKLConfig.cmake introduced in 2021.3 + conflicts("intel-oneapi-mkl@:2021.2", when="^intel-oneapi-mkl") + + depends_on("wannier90", when="@7.5.0: +wannier90") + depends_on("wannier90+shared", when="@7.5.0: +wannier90+shared") + + depends_on("elpa+openmp", when="+elpa+openmp") + depends_on("elpa~openmp", when="+elpa~openmp") + + depends_on("eigen@3.4.0:", when="@7.3.2: +tests") + depends_on("eigen@3.4.0:", when="@7.7: +vcsqnm") + + depends_on("costa+shared", when="@7.3.2:") + + with when("@7.5: +memory_pool"): + depends_on("umpire~cuda~rocm", when="~cuda~rocm") + depends_on("umpire+cuda~device_alloc", when="+cuda") + depends_on("umpire+rocm~device_alloc", when="+rocm") + + patch("fj.patch", when="@7.3.2: %fj") + + def cmake_args(self): + spec = self.spec + + cm_label = "" + if "@7.5:" in spec: + cm_label = "SIRIUS_" + + args = [ + self.define_from_variant(cm_label + "USE_OPENMP", "openmp"), + self.define_from_variant(cm_label + "USE_ELPA", "elpa"), + self.define_from_variant(cm_label + "USE_MAGMA", "magma"), + self.define_from_variant(cm_label + "USE_NLCGLIB", "nlcglib"), + self.define_from_variant(cm_label + "USE_VDWXC", "vdwxc"), + self.define_from_variant(cm_label + "USE_MEMORY_POOL", "memory_pool"), + self.define_from_variant(cm_label + "USE_SCALAPACK", "scalapack"), + self.define_from_variant(cm_label + "USE_DLAF", "dlaf"), + self.define_from_variant(cm_label + "CREATE_FORTRAN_BINDINGS", "fortran"), + self.define_from_variant(cm_label + "CREATE_PYTHON_MODULE", "python"), + self.define_from_variant(cm_label + "USE_CUDA", "cuda"), + self.define_from_variant(cm_label + "USE_ROCM", "rocm"), + self.define_from_variant(cm_label + "BUILD_APPS", "apps"), + self.define_from_variant(cm_label + "USE_FP32", "single_precision"), + self.define_from_variant(cm_label + "USE_PROFILER", "profiler"), + self.define_from_variant(cm_label + "USE_NVTX", "nvtx"), + self.define_from_variant(cm_label + "USE_WANNIER90", "wannier90"), + self.define_from_variant(cm_label + "USE_PUGIXML", "pugixml"), + self.define_from_variant(cm_label + "USE_DFTD3", "dftd3"), + self.define_from_variant(cm_label + "USE_DFTD4", "dftd4"), + self.define_from_variant(cm_label + "USE_VCSQNM", "vcsqnm"), + self.define_from_variant("BUILD_SHARED_LIBS", "shared"), + self.define_from_variant("BUILD_TESTING", "tests"), + ] + + lapack = spec["lapack"] + blas = spec["blas"] + + args.extend( + [ + self.define("LAPACK_FOUND", "true"), + self.define("LAPACK_LIBRARIES", lapack.libs.joined(";")), + self.define("BLAS_FOUND", "true"), + self.define("BLAS_LIBRARIES", blas.libs.joined(";")), + ] + ) + + if "+scalapack" in spec and "^cray-libsci" not in spec: + args.extend( + [ + self.define(cm_label + "SCALAPACK_FOUND", "true"), + self.define( + cm_label + "SCALAPACK_INCLUDE_DIRS", spec["scalapack"].prefix.include + ), + self.define( + cm_label + "SCALAPACK_LIBRARIES", spec["scalapack"].libs.joined(";") + ), + ] + ) + + if "^cray-libsci" in spec: + args.append(self.define(cm_label + "USE_CRAY_LIBSCI", "ON")) + + if spec.satisfies("^[virtuals=blas] intel-oneapi-mkl"): + args.append(self.define(cm_label + "USE_MKL", "ON")) + + if spec.satisfies("@7.6.0:"): + mkl_mapper = { + "threading": { + "none": "sequential", + "openmp": "gnu_thread", + "tbb": "tbb_thread", + }, + "mpi": { + "intel-oneapi-mpi": "intelmpi", + "mpich": "mpich", + "openmpi": "openmpi", + }, + } + + mkl_threads = mkl_mapper["threading"][ + spec["intel-oneapi-mkl"].variants["threads"].value + ] + + mpi_provider = spec["mpi"].name + if mpi_provider in ["mpich", "cray-mpich", "mvapich", "mvapich2"]: + mkl_mpi = mkl_mapper["mpi"]["mpich"] + else: + mkl_mpi = mkl_mapper["mpi"][mpi_provider] + + args.extend( + [ + self.define("MKL_INTERFACE", "lp64"), + self.define("MKL_THREADING", mkl_threads), + self.define("MKL_MPI", mkl_mpi), + ] + ) + + if "+scalapack" in self.spec: + # options provided by `MKLConfig.cmake` + args.extend( + [self.define("ENABLE_BLACS", "On"), self.define("ENABLE_SCALAPACK", "On")] + ) + + if "+elpa" in spec: + elpa_incdir = os.path.join(spec["elpa"].headers.directories[0], "elpa") + args.append(self.define(cm_label + "ELPA_INCLUDE_DIR", elpa_incdir)) + + if "+cuda" in spec: + cuda_arch = spec.variants["cuda_arch"].value + if cuda_arch[0] != "none": + # Make SIRIUS handle it + if "@:7.4.3" in spec: + args.append(self.define("CMAKE_CUDA_ARCH", ";".join(cuda_arch))) + else: + args.append(self.define("CMAKE_CUDA_ARCHITECTURES", ";".join(cuda_arch))) + + if "+rocm" in spec: + archs = ",".join(self.spec.variants["amdgpu_target"].value) + args.extend([self.define("CMAKE_HIP_ARCHITECTURES", archs)]) + + return args diff --git a/ci/sirius-uenv-recipe-mi300/spack_repo/sirius_ci/repo.yaml b/ci/spack_repo/sirius_ci/repo.yaml similarity index 100% rename from ci/sirius-uenv-recipe-mi300/spack_repo/sirius_ci/repo.yaml rename to ci/spack_repo/sirius_ci/repo.yaml diff --git a/src/core/acc/acc_blas.hpp b/src/core/acc/acc_blas.hpp index d2b7404ed..241d0378f 100644 --- a/src/core/acc/acc_blas.hpp +++ b/src/core/acc/acc_blas.hpp @@ -144,7 +144,7 @@ get_gpublasDiagonal_t(char c) { \ acc::blas_api::status_t status; \ if ((status = func__ args__) != acc::blas_api::status::Success) { \ - error_message(status); \ + acc::blas::error_message(status); \ char nm[1024]; \ gethostname(nm, 1024); \ std::printf("hostname: %s\n", nm); \ diff --git a/src/core/acc/acc_blas_api.hpp b/src/core/acc/acc_blas_api.hpp index ad681390f..64fb912a3 100644 --- a/src/core/acc/acc_blas_api.hpp +++ b/src/core/acc/acc_blas_api.hpp @@ -22,8 +22,8 @@ #elif defined(SIRIUS_ROCM) #include -#else -#error Either SIRIUS_CUDA or SIRIUS_ROCM must be defined! +// #else +// #error Either SIRIUS_CUDA or SIRIUS_ROCM must be defined! #endif namespace sirius { diff --git a/src/core/acc/rocsolver.hpp b/src/core/acc/rocsolver.hpp index f8399fd8a..5f6e64440 100644 --- a/src/core/acc/rocsolver.hpp +++ b/src/core/acc/rocsolver.hpp @@ -20,6 +20,7 @@ #include "acc.hpp" #include "acc_blas_api.hpp" #include "core/rte/rte.hpp" +#include "core/memory.hpp" namespace sirius { @@ -68,178 +69,355 @@ get_rocblas_operation(char trans) return op; } -/// _sy_mmetric or _he_rmitian STANDARD eigenvalue problem | double +/// _sy_mmetric or _he_rmitian STANDARD eigenvalue problem | double,float template -std::enable_if_t::value> -syheevd(rocblas_handle handle, const rocblas_evect evect, const rocblas_fill uplo, int n, T* A, int lda, T* D, T* E, - int* info) -{ - CALL_ROCSOLVER(rocsolver_dsyevd, (handle, evect, uplo, n, A, lda, D, E, info)); -} - -/// _sy_mmetric or _he_rmitian STANDARD eigenvalue problem | float -template -std::enable_if_t::value> +void syheevd(rocblas_handle handle, const rocblas_evect evect, const rocblas_fill uplo, int n, T* A, int lda, T* D, T* E, int* info) { - CALL_ROCSOLVER(rocsolver_ssyevd, (handle, evect, uplo, n, A, lda, D, E, info)); + if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_dsyevd, (handle, evect, uplo, n, A, lda, D, E, info)); + } else if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_ssyevd, (handle, evect, uplo, n, A, lda, D, E, info)); + } } -/// _sy_mmetric or _he_rmitian STANDARD eigenvalue problem | complex double +/// _sy_mmetric or _he_rmitian STANDARD eigenvalue problem | complex double, float template -std::enable_if_t::value> +void syheevd(rocblas_handle handle, const rocblas_evect evect, const rocblas_fill uplo, int n, std::complex* A, int lda, T* D, T* E, int* info) { - CALL_ROCSOLVER(rocsolver_zheevd, - (handle, evect, uplo, n, reinterpret_cast(A), lda, D, E, info)); -} + if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_zheevd, + (handle, evect, uplo, n, reinterpret_cast(A), lda, D, E, info)); -/// _sy_mmetric or _he_rmitian STANDARD eigenvalue problem | complex float -template -std::enable_if_t::value> -syheevd(rocblas_handle handle, const rocblas_evect evect, const rocblas_fill uplo, int n, std::complex* A, int lda, - T* D, T* E, int* info) -{ - CALL_ROCSOLVER(rocsolver_cheevd, - (handle, evect, uplo, n, reinterpret_cast(A), lda, D, E, info)); + } else if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_cheevd, + (handle, evect, uplo, n, reinterpret_cast(A), lda, D, E, info)); + } } -/// _sy_mmetric or _he_rmitian GENERALIZED eigenvalue problem | double +/// _sy_mmetric or _he_rmitian GENERALIZED eigenvalue problem | double, float template -std::enable_if_t::value> +int syhegvd(rocblas_handle handle, const rocblas_eform itype, const rocblas_evect evect, const rocblas_fill uplo, int n, - T* A, int lda, T* B, int ldb, T* D, T* E, int* info) + T* A, int lda, T* B, int ldb, T* D, T* E) { - CALL_ROCSOLVER(rocsolver_dsygvd, (handle, itype, evect, uplo, n, A, lda, B, ldb, D, E, info)); + auto& mpd = get_memory_pool(memory_t::device); + auto d_info = mpd.get_unique_ptr(1); + if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_dsygvd, (handle, itype, evect, uplo, n, A, lda, B, ldb, D, E, d_info.get())); + } else if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_ssygvd, (handle, itype, evect, uplo, n, A, lda, B, ldb, D, E, d_info.get())); + } + int info; + acc::copyout(&info, d_info.get(), 1); + return info; } -/// _sy_mmetric or _he_rmitian GENERALIZED eigenvalue problem | float +/// _sy_mmetric or _he_rmitian GENERALIZED eigenvalue problem | complex double, float template -std::enable_if_t::value> +int syhegvd(rocblas_handle handle, const rocblas_eform itype, const rocblas_evect evect, const rocblas_fill uplo, int n, - T* A, int lda, T* B, int ldb, T* D, T* E, int* info) + std::complex* A, int lda, std::complex* B, int ldb, T* D, T* E) { - CALL_ROCSOLVER(rocsolver_ssygvd, (handle, itype, evect, uplo, n, A, lda, B, ldb, D, E, info)); + auto& mpd = get_memory_pool(memory_t::device); + auto d_info = mpd.get_unique_ptr(1); + if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_zhegvd, (handle, itype, evect, uplo, n, reinterpret_cast(A), + lda, reinterpret_cast(B), ldb, D, E, d_info.get())); + } else if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_chegvd, (handle, itype, evect, uplo, n, reinterpret_cast(A), + lda, reinterpret_cast(B), ldb, D, E, d_info.get())); + } + int info; + acc::copyout(&info, d_info.get(), 1); + return info; } -/// _sy_mmetric or _he_rmitian GENERALIZED eigenvalue problem | complex float +#if (ROCSOLVER_VERSION_MAJOR > 3) || ((ROCSOLVER_VERSION_MAJOR == 3) && (ROCSOLVER_VERSION_MINOR >= 19)) +/// x versions +/// ----------------------------------------------------------------------------------------------------------------- +struct rocsolver_evx_return_type +{ + /// If info = 0, the first nev elements of ifail are zero. Otherwise, + /// contains the indices of those eigenvectors that failed to converge. Not + /// referenced if evect is rocblas_evect_none. + std::vector ifail; + /// If info = 0, successful exit. If info = i > 0, the algorithm did not + /// converge. i columns of Z did not converge. + int info; + /// The total number of eigenvalues found. If erange is rocblas_erange_all, + /// nev = n. If erange is rocblas_erange_index, nev = iu - il + 1. + /// Otherwise, 0 <= nev <= n. + int nev; +}; template -std::enable_if_t::value> -syhegvd(rocblas_handle handle, const rocblas_eform itype, const rocblas_evect evect, const rocblas_fill uplo, int n, - std::complex* A, int lda, std::complex* B, int ldb, T* D, T* E, int* info) +rocsolver_evx_return_type +syheevx(rocblas_handle handle, const rocblas_evect evect, rocblas_erange erange, const rocblas_fill uplo, int n, T* A, + int lda, double vl, double vu, int il, int iu, double abstol, T* D, T* Z, int ldz) { - CALL_ROCSOLVER(rocsolver_chegvd, (handle, itype, evect, uplo, n, reinterpret_cast(A), lda, - reinterpret_cast(B), ldb, D, E, info)); + auto& mpd = get_memory_pool(memory_t::device); + auto d_info = mpd.get_unique_ptr(1); + auto d_nev = mpd.get_unique_ptr(1); + auto d_ifail = mpd.get_unique_ptr(n); + + if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_ssyevx, (handle, evect, rocblas_erange::rocblas_erange_index, uplo, n, A, lda, vl, vu, + il, iu, abstol, d_nev.get(), D, Z, ldz, d_ifail.get(), d_info.get())); + } else if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_dsyevx, (handle, evect, rocblas_erange::rocblas_erange_index, uplo, n, A, lda, vl, vu, + il, iu, abstol, d_nev.get(), D, Z, ldz, d_ifail.get(), d_info.get())); + } + + rocsolver_evx_return_type ret; + + acc::copyout(&ret.info, d_info.get(), 1); + acc::copyout(&ret.nev, d_nev.get(), 1); + + if (evect != rocblas_evect_none) { + ret.ifail = std::vector(n); + acc::copyout(ret.ifail.data(), d_ifail.get(), n); + } + + return ret; } -/// _sy_mmetric or _he_rmitian GENERALIZED eigenvalue problem | complex double +/// Hermitian | complex double template -std::enable_if_t::value> -syhegvd(rocblas_handle handle, const rocblas_eform itype, const rocblas_evect evect, const rocblas_fill uplo, int n, - std::complex* A, int lda, std::complex* B, int ldb, T* D, T* E, int* info) +rocsolver_evx_return_type +syheevx(rocblas_handle handle, const rocblas_evect evect, const rocblas_fill uplo, int n, std::complex* A, int lda, + double vu, double vl, int il, int iu, double abstol, T* D, std::complex* Z, int ldz) { - CALL_ROCSOLVER(rocsolver_zhegvd, (handle, itype, evect, uplo, n, reinterpret_cast(A), lda, - reinterpret_cast(B), ldb, D, E, info)); + auto& mpd = get_memory_pool(memory_t::device); + auto d_info = mpd.get_unique_ptr(1); + auto d_nev = mpd.get_unique_ptr(1); + auto d_ifail = mpd.get_unique_ptr(n); + + if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_cheevx, + (handle, evect, rocblas_erange::rocblas_erange_index, uplo, n, + reinterpret_cast(A), lda, vl, vu, il, iu, abstol, d_nev.get(), D, + reinterpret_cast(Z), ldz, d_ifail.get(), d_info.get())); + + } else if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_zheevx, + (handle, evect, rocblas_erange::rocblas_erange_index, uplo, n, + reinterpret_cast(A), lda, vl, vu, il, iu, abstol, d_nev.get(), D, + reinterpret_cast(Z), ldz, d_ifail.get(), d_info.get())); + } + rocsolver_evx_return_type ret; + + acc::copyout(&ret.info, d_info.get(), 1); + acc::copyout(&ret.nev, d_nev.get(), 1); + + if (evect != rocblas_evect_none) { + ret.ifail = std::vector(n); + acc::copyout(ret.ifail.data(), d_ifail.get(), n); + } + + return ret; } -#if (ROCSOLVER_VERSION_MAJOR > 3) || ((ROCSOLVER_VERSION_MAJOR == 3) && (ROCSOLVER_VERSION_MINOR >= 19)) /// x versions /// ----------------------------------------------------------------------------------------------------------------- template -std::enable_if_t::value> -syheevx(rocblas_handle handle, const rocblas_evect evect, const rocblas_fill uplo, int n, T* A, int lda, int il, int iu, - double abstol, int* nev, T* D, T* Z, int ldz, int* ifail, int* info) +rocsolver_evx_return_type +syhegvx(rocblas_handle handle, const rocblas_eform itype, const rocblas_evect evect, const rocblas_fill uplo, int n, + T* A, int lda, T* B, int ldb, double vu, double vl, int il, int iu, double abstol, T* D, T* Z, int ldz) { - double vl, vu{0}; // ingored if erange = erange_index - rocsolver_dsyevx(handle, evect, rocblas_erange::rocblas_erange_index, uplo, n, A, lda, vl, vu, il, iu, abstol, nev, - D, Z, ldz, ifail, info); -} + auto& mpd = get_memory_pool(memory_t::device); + auto d_info = mpd.get_unique_ptr(1); + auto d_nev = mpd.get_unique_ptr(1); + auto d_ifail = mpd.get_unique_ptr(n); -template -std::enable_if_t::value> -syheevx(rocblas_handle handle, const rocblas_evect evect, const rocblas_fill uplo, int n, T* A, int lda, int il, int iu, - double abstol, int* nev, T* D, T* Z, int ldz, int* ifail, int* info) -{ - double vl, vu{0}; // ingored if erange = erange_index - rocsolver_ssyevx(handle, evect, rocblas_erange::rocblas_erange_index, uplo, n, A, lda, vl, vu, il, iu, abstol, nev, - D, Z, ldz, ifail, info); + if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_ssygvx, + (handle, itype, evect, rocblas_erange::rocblas_erange_index, uplo, n, A, lda, B, ldb, vl, vu, il, + iu, abstol, d_nev.get(), D, Z, ldz, d_ifail.get(), d_info.get())); + } else if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_dsygvx, + (handle, itype, evect, rocblas_erange::rocblas_erange_index, uplo, n, A, lda, B, ldb, vl, vu, il, + iu, abstol, d_nev.get(), D, Z, ldz, d_ifail.get(), d_info.get())); + } + + rocsolver_evx_return_type ret; + + acc::copyout(&ret.info, d_info.get(), 1); + acc::copyout(&ret.nev, d_nev.get(), 1); + + if (evect != rocblas_evect_none) { + ret.ifail = std::vector(n); + acc::copyout(ret.ifail.data(), d_ifail.get(), n); + } + + return ret; } /// Hermitian | complex double template -std::enable_if_t::value> -syheevx(rocblas_handle handle, const rocblas_evect evect, const rocblas_fill uplo, int n, std::complex* A, - int lda, int il, int iu, double abstol, int* nev, T* D, std::complex* Z, int ldz, int* ifail, int* info) +rocsolver_evx_return_type +syhegvx(rocblas_handle handle, const rocblas_eform itype, const rocblas_evect evect, const rocblas_fill uplo, int n, + std::complex* A, int lda, std::complex* B, int ldb, double vl, double vu, int il, int iu, + double abstol, T* D, std::complex* Z, int ldz) { - double vl, vu{0}; // ingored if erange = erange_index - rocsolver_zheevx(handle, evect, rocblas_erange::rocblas_erange_index, uplo, n, - reinterpret_cast(A), lda, vl, vu, il, iu, abstol, nev, D, - reinterpret_cast(Z), ldz, ifail, info); -} + auto& mpd = get_memory_pool(memory_t::device); + auto d_info = mpd.get_unique_ptr(1); + auto d_nev = mpd.get_unique_ptr(1); + auto d_ifail = mpd.get_unique_ptr(n); -template -std::enable_if_t::value> -syheevx(rocblas_handle handle, const rocblas_evect evect, const rocblas_fill uplo, int n, std::complex* A, - int lda, int il, int iu, double abstol, int* nev, T* D, std::complex* Z, int ldz, int* ifail, int* info) -{ - double vl, vu{0}; // ingored if erange = erange_index - rocsolver_cheevx(handle, evect, rocblas_erange::rocblas_erange_index, uplo, n, - reinterpret_cast(A), lda, vl, vu, il, iu, abstol, nev, D, - reinterpret_cast(Z), ldz, ifail, info); + if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_chegvx, + (handle, itype, evect, rocblas_erange::rocblas_erange_index, uplo, n, + reinterpret_cast(A), lda, reinterpret_cast(B), + ldb, vl, vu, il, iu, abstol, d_nev.get(), D, reinterpret_cast(Z), ldz, + d_ifail.get(), d_info.get())); + } else if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_zhegvx, + (handle, itype, evect, rocblas_erange::rocblas_erange_index, uplo, n, + reinterpret_cast(A), lda, reinterpret_cast(B), + ldb, vl, vu, il, iu, abstol, d_nev.get(), D, reinterpret_cast(Z), ldz, + d_ifail.get(), d_info.get())); + } + + rocsolver_evx_return_type ret; + + acc::copyout(&ret.info, d_info.get(), 1); + acc::copyout(&ret.nev, d_nev.get(), 1); + + if (evect != rocblas_evect_none) { + ret.ifail = std::vector(n); + acc::copyout(ret.ifail.data(), d_ifail.get(), n); + } + + return ret; } -/// x versions +#endif // rocsolver >=3.19.0 + +// rocm 7.2.0 +#if (ROCSOLVER_VERSION_MAJOR > 3) || ((ROCSOLVER_VERSION_MAJOR == 3) && (ROCSOLVER_VERSION_MINOR >= 30)) +struct rocblas_evdx_return_type +{ + int info; + int nev; +}; +/// dx versions (divide-and-conquer + subset selection) /// ----------------------------------------------------------------------------------------------------------------- template -std::enable_if_t::value> -syhegvx(rocblas_handle handle, const rocblas_eform itype, const rocblas_evect evect, const rocblas_fill uplo, int n, - T* A, int lda, T* B, int ldb, int il, int iu, double abstol, int* nev, T* D, T* Z, int ldz, int* ifail, - int* info) +rocblas_evdx_return_type +syheevdx(rocblas_handle handle, const rocblas_evect evect, const rocblas_erange erange, const rocblas_fill uplo, int n, + T* A, int lda, T vl, T vu, int il, int iu, T* w, T* E, int ldz) { - double vl, vu{0}; // ingored if erange = erange_index - rocsolver_dsygvx(handle, itype, evect, rocblas_erange::rocblas_erange_index, uplo, n, A, lda, B, ldb, vl, vu, il, - iu, abstol, nev, D, Z, ldz, ifail, info); + int info; + int nev; + + auto& mpd = get_memory_pool(memory_t::device); + auto d_info = mpd.get_unique_ptr(1); + auto d_nev = mpd.get_unique_ptr(1); + + if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_ssyevdx, + (handle, evect, erange, uplo, n, A, lda, vl, vu, il, iu, d_nev.get(), w, E, ldz, d_info.get())); + } else if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_dsyevdx, + (handle, evect, erange, uplo, n, A, lda, vl, vu, il, iu, d_nev.get(), w, E, ldz, d_info.get())); + } + acc::copyout(&info, d_info.get(), 1); + acc::copyout(&nev, d_nev.get(), 1); + + return {.info = info, .nev = nev}; } +/// Hermitian | complex double template -std::enable_if_t::value> -syhegvx(rocblas_handle handle, const rocblas_eform itype, const rocblas_evect evect, const rocblas_fill uplo, int n, - T* A, int lda, T* B, int ldb, int il, int iu, double abstol, int* nev, T* D, T* Z, int ldz, int* ifail, - int* info) +rocblas_evdx_return_type +syheevdx(rocblas_handle handle, const rocblas_evect evect, const rocblas_erange erange, const rocblas_fill uplo, int n, + std::complex* A, int lda, T vl, T vu, int il, int iu, T* w, std::complex* E, int ldz) { - double vl, vu{0}; // ingored if erange = erange_index - rocsolver_ssygvx(handle, itype, evect, rocblas_erange::rocblas_erange_index, uplo, n, A, lda, B, ldb, vl, vu, il, - iu, abstol, nev, D, Z, ldz, ifail, info); + int info; + int nev; + + auto& mpd = get_memory_pool(memory_t::device); + auto d_info = mpd.get_unique_ptr(1); + auto d_nev = mpd.get_unique_ptr(1); + + if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_cheevdx, + (handle, evect, erange, uplo, n, reinterpret_cast(A), lda, vl, vu, il, + iu, d_nev.get(), w, reinterpret_cast(E), ldz, d_info.get())); + } else if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_zheevdx, + (handle, evect, erange, uplo, n, reinterpret_cast(A), lda, vl, vu, il, + iu, d_nev.get(), w, reinterpret_cast(E), ldz, d_info.get())); + } + acc::copyout(&info, d_info.get(), 1); + acc::copyout(&nev, d_nev.get(), 1); + + return {.info = info, .nev = nev}; } -/// Hermitian | complex double +/// General eigenproblem template -std::enable_if_t::value> -syhegvx(rocblas_handle handle, const rocblas_eform itype, const rocblas_evect evect, const rocblas_fill uplo, int n, - std::complex* A, int lda, std::complex* B, int ldb, int il, int iu, double abstol, int* nev, - T* D, std::complex* Z, int ldz, int* ifail, int* info) +rocblas_evdx_return_type +syhegvdx(rocblas_handle handle, const rocblas_eform itype, const rocblas_evect evect, const rocblas_erange erange, + const rocblas_fill uplo, int n, T* A, int lda, T* B, int ldb, T vl, T vu, int il, int iu, T* w, T* Z, int ldz) { - double vl, vu{0}; // ingored if erange = erange_index - rocsolver_zhegvx(handle, itype, evect, rocblas_erange::rocblas_erange_index, uplo, n, - reinterpret_cast(A), lda, reinterpret_cast(B), - ldb, vl, vu, il, iu, abstol, nev, D, reinterpret_cast(Z), ldz, ifail, - info); + int info; + int nev; + + auto& mpd = get_memory_pool(memory_t::device); + auto d_info = mpd.get_unique_ptr(1); + auto d_nev = mpd.get_unique_ptr(1); + + if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_ssygvdx, (handle, itype, evect, erange, uplo, n, A, lda, B, ldb, vl, vu, il, iu, + d_nev.get(), w, Z, ldz, d_info.get())); + } else if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_dsygvdx, (handle, itype, evect, erange, uplo, n, A, lda, B, ldb, vl, vu, il, iu, + d_nev.get(), w, Z, ldz, d_info.get())); + } + + acc::copyout(&info, d_info.get(), 1); + acc::copyout(&nev, d_nev.get(), 1); + + return {.info = info, .nev = nev}; } +// complex template -std::enable_if_t::value> -syhegvx(rocblas_handle handle, const rocblas_eform itype, const rocblas_evect evect, const rocblas_fill uplo, int n, - std::complex* A, int lda, std::complex* B, int ldb, int il, int iu, double abstol, int* nev, T* D, - std::complex* Z, int ldz, int* ifail, int* info) +rocblas_evdx_return_type +syhegvdx(rocblas_handle handle, const rocblas_eform itype, const rocblas_evect evect, const rocblas_erange erange, + const rocblas_fill uplo, int n, std::complex* A, int lda, std::complex* B, int ldb, T vl, T vu, int il, + int iu, T* w, std::complex* Z, int ldz) { - double vl, vu{0}; // ingored if erange = erange_index - rocsolver_chegvx(handle, itype, evect, rocblas_erange::rocblas_erange_index, uplo, n, - reinterpret_cast(A), lda, reinterpret_cast(B), ldb, - vl, vu, il, iu, abstol, nev, D, reinterpret_cast(Z), ldz, ifail, info); + int info; + int nev; + + auto& mpd = get_memory_pool(memory_t::device); + auto d_info = mpd.get_unique_ptr(1); + auto d_nev = mpd.get_unique_ptr(1); + + if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_chegvdx, + (handle, itype, evect, erange, uplo, n, reinterpret_cast(A), lda, + reinterpret_cast(B), ldb, vl, vu, il, iu, d_nev.get(), w, + reinterpret_cast(Z), ldz, d_info.get())); + } else if constexpr (std::is_same_v) { + CALL_ROCSOLVER(rocsolver_zhegvdx, + (handle, itype, evect, erange, uplo, n, reinterpret_cast(A), lda, + reinterpret_cast(B), ldb, vl, vu, il, iu, d_nev.get(), w, + reinterpret_cast(Z), ldz, d_info.get())); + } + + acc::copyout(&info, d_info.get(), 1); + acc::copyout(&nev, d_nev.get(), 1); + + return {.info = info, .nev = nev}; } -#endif // rocsolver >=5.3.0 + +#endif // rocsolver >= 3.30.0 (aka rocm 7.0.1) /// Linear Solvers void @@ -255,5 +433,5 @@ zgetrf(rocblas_handle handle, int m, int n, acc_complex_double_t* A, int* devIpi } // namespace sirius -#endif -#endif +#endif // __ROCSOLVER_HPP__ +#endif // SIRIUS_ROCM diff --git a/src/core/la/eigenproblem.hpp b/src/core/la/eigenproblem.hpp index bb9ba71f8..b1bb39cd1 100644 --- a/src/core/la/eigenproblem.hpp +++ b/src/core/la/eigenproblem.hpp @@ -2040,24 +2040,22 @@ class Eigensolver_cuda : public Eigensolver auto uplo = rocblas_fill::rocblas_fill_lower; auto& handle = acc::rocsolver::rocsolver_handle(); - auto& mpd = get_memory_pool(memory_t::device); - auto w = mpd.get_unique_ptr>(matrix_size__); - auto work = mpd.get_unique_ptr>(matrix_size__); + auto& mpd = get_memory_pool(memory_t::device); + auto evals = mpd.get_unique_ptr>(matrix_size__); acc::copyin(A__.at(memory_t::device), A__.ld(), A__.at(memory_t::host), A__.ld(), matrix_size__, matrix_size__); acc::copyin(B__.at(memory_t::device), B__.ld(), B__.at(memory_t::host), B__.ld(), matrix_size__, matrix_size__); - int info; - auto dinfo = mpd.get_unique_ptr(1); + auto ret = acc::rocsolver::syhegvdx(handle, itype, jobz, rocblas_erange::rocblas_erange_index, uplo, + matrix_size__, A__.at(memory_t::device), A__.ld(), B__.at(memory_t::device), + B__.ld(), real_type{0}, real_type(0), 1, nev__, evals.get(), + Z__.at(memory_t::device), Z__.ld()); - acc::rocsolver::syhegvd(handle, itype, jobz, uplo, matrix_size__, A__.at(memory_t::device), A__.ld(), - B__.at(memory_t::device), B__.ld(), w.get(), work.get(), dinfo.get()); - - acc::copyout(&info, dinfo.get(), 1); - if (!info) { - acc::copyout(eval__, w.get(), nev__); - acc::copyout(Z__.at(memory_t::host), Z__.ld(), A__.at(memory_t::device), A__.ld(), matrix_size__, nev__); + if (ret.info == rocblas_status::rocblas_status_success) { + acc::copyout(eval__, evals.get(), nev__); + acc::copyout(Z__.at(memory_t::host), Z__.ld(), Z__.at(memory_t::device), Z__.ld(), matrix_size__, nev__); } - return info; + + return ret.info; } template @@ -2068,26 +2066,20 @@ class Eigensolver_cuda : public Eigensolver auto uplo = rocblas_fill::rocblas_fill_lower; auto& handle = acc::rocsolver::rocsolver_handle(); - auto& mpd = get_memory_pool(memory_t::device); - auto w = mpd.get_unique_ptr>(matrix_size__); + auto& mpd = get_memory_pool(memory_t::device); + auto evals = mpd.get_unique_ptr>(matrix_size__); acc::copyin(A__.at(memory_t::device), A__.ld(), A__.at(memory_t::host), A__.ld(), matrix_size__, matrix_size__); - int lwork; - - auto work = mpd.get_unique_ptr>(lwork); - - int info; - auto dinfo = mpd.get_unique_ptr(1); - - acc::rocsolver::syheevd(handle, jobz, uplo, matrix_size__, A__.at(memory_t::device), A__.ld(), w.get(), - work.get(), dinfo.get()); + auto ret = acc::rocsolver::syheevdx(handle, jobz, rocblas_erange::rocblas_erange_index, uplo, matrix_size__, + A__.at(memory_t::device), A__.ld(), real_type{0}, real_type{0}, 1, + nev__, evals.get(), Z__.at(memory_t::device), Z__.ld()); - acc::copyout(&info, dinfo.get(), 1); - if (!info) { - acc::copyout(eval__, w.get(), nev__); - acc::copyout(Z__.at(memory_t::host), Z__.ld(), A__.at(memory_t::device), A__.ld(), matrix_size__, nev__); + if (ret.info == rocblas_status::rocblas_status_success) { + acc::copyout(eval__, evals.get(), nev__); + acc::copyout(Z__.at(memory_t::host), Z__.ld(), Z__.at(memory_t::device), Z__.ld(), matrix_size__, nev__); } - return info; + + return ret.info; } public: diff --git a/src/core/la/eigensolver.cpp b/src/core/la/eigensolver.cpp index 94a6d5600..c21e4590b 100644 --- a/src/core/la/eigensolver.cpp +++ b/src/core/la/eigensolver.cpp @@ -61,7 +61,7 @@ Eigensolver_factory(std::string name__) break; } #endif -#if defined(SIRIUS_CUDA) +#if defined(SIRIUS_CUDA) || defined(SIRIUS_ROCM) case ev_solver_t::cusolver: { ptr = new Eigensolver_cuda(); break; diff --git a/src/core/wf/wave_functions.hpp b/src/core/wf/wave_functions.hpp index 26429b402..91b8abbd4 100644 --- a/src/core/wf/wave_functions.hpp +++ b/src/core/wf/wave_functions.hpp @@ -19,6 +19,7 @@ #include #include #include +#include #include "core/la/linalg.hpp" #include "core/strong_type.hpp" #include "core/hdf5_tree.hpp" @@ -26,6 +27,7 @@ #include "core/env/env.hpp" #include "core/rte/rte.hpp" #include "core/time_tools.hpp" +#include "wf_inner_local.hpp" namespace sirius { @@ -1309,29 +1311,15 @@ inner_diag_local(memory_t mem__, wf::Wave_functions const& lhs__, wf::Wave_fu if (std::is_same>::value) { reduced = lhs__.comm().rank() + 1; } - mdarray result_gpu({num_wf__.get()}); - result_gpu.allocate(mem__).zero(mem__); for (auto s = spins__.begin(); s != spins__.end(); s++) { - auto s1 = lhs__.actual_spin_index(s); - auto s2 = rhs__.actual_spin_index(s); - auto ptr1 = lhs__.at(mem__, 0, s1, wf::band_index(0)); - auto ptr2 = rhs__.at(mem__, 0, s2, wf::band_index(0)); - if (std::is_same::value) { + auto s1 = lhs__.actual_spin_index(s); + auto s2 = rhs__.actual_spin_index(s); + auto ptr1 = lhs__.at(mem__, 0, s1, wf::band_index(0)); + auto ptr2 = rhs__.at(mem__, 0, s2, wf::band_index(0)); + int ngv_loc = lhs__.ld(); - if (std::is_same::value) { - inner_diag_local_gpu_double_double(ptr1, lhs__.ld(), ptr2, rhs__.ld(), lhs__.ld(), num_wf__.get(), - reduced, result_gpu.at(mem__)); - } - if (std::is_same>::value) { - inner_diag_local_gpu_double_complex_double(ptr1, lhs__.ld(), ptr2, rhs__.ld(), lhs__.ld(), - num_wf__.get(), result_gpu.at(mem__)); - } - } - } - result_gpu.copy_to(memory_t::host); - for (int i = 0; i < num_wf__.get(); i++) { - result[i] = result_gpu[i]; + inner_product_local_gpu(ptr1, rhs__.ld(), ptr2, lhs__.ld(), ngv_loc, num_wf__.get(), reduced, result); } #endif } diff --git a/src/core/wf/wf_inner_local.hpp b/src/core/wf/wf_inner_local.hpp new file mode 100644 index 000000000..e115d201d --- /dev/null +++ b/src/core/wf/wf_inner_local.hpp @@ -0,0 +1,97 @@ +#ifndef __WF_INNER_LOCAL_HPP__ +#define __WF_INNER_LOCAL_HPP__ + +#if defined(SIRIUS_ROCM) || defined(SIRIUS_CUDA) + +#include +#include "core/acc/acc.hpp" +#include "core/memory.hpp" +#include "core/acc/acc_blas.hpp" +#include "core/typedefs.hpp" + +#if defined(SIRIUS_ROCM) +#include +#endif + +#if defined(SIRIUS_CUDA) +#include +#endif + +namespace sirius { + +template +void +inner_product_local_gpu(T const* wf1, int ld1, T const* wf2, int ld2, int n, int num_bands, int reduced, + std::vector& result) +{ + auto mem_pool = get_memory_pool(memory_t::device); + auto d_work = mem_pool.get_unique_ptr(num_bands); + auto stream = acc::blas::stream_handle(0); + std::vector result_local(num_bands, 0); + +#if defined(SIRIUS_ROCM) + CALL_GPU_BLAS(rocblas_zdotc_strided_batched, (stream, n, reinterpret_cast(wf1), 1, + ld1, reinterpret_cast(wf2), 1, ld2, + num_bands, reinterpret_cast(d_work.get()))); +#elif defined(SIRIUS_CUDA) + T beta{0}; + T alpha{1}; + // https://docs.nvidia.com/cuda/cublas/#cublas-t-gemmstridedbatched + CALL_GPU_BLAS(cublasZgemmStridedBatched, + (stream, CUBLAS_OP_C, CUBLAS_OP_N, 1, 1, n, reinterpret_cast(&alpha), + reinterpret_cast(wf1), n, ld1, reinterpret_cast(wf2), + n, ld2, reinterpret_cast(&beta), + reinterpret_cast(d_work.get()), 1, 1, num_bands)); +#endif + + acc::copyout(result_local.data(), d_work.get(), num_bands); + if constexpr (std::is_same_v>) { + /* non Γ-point case */ + } else if constexpr (std::is_same_v) { + /* Γ-point case */ + // #if defined(SIRIUS_ROCM) + // rocblas_ddot_strided_batched(stream, n, reinterpret_cast(wf1), 1, ld1, + // reinterpret_cast(wf2), 1, ld2, num_bands, + // reinterpret_cast(d_work.get())); + // #elif defined(SIRIUS_CUDA) + // T beta{0}; + // T alpha{1}; + // // https://docs.nvidia.com/cuda/cublas/#cublas-t-gemmstridedbatched + // cublasDgemmStridedBatched(stream, CUBLAS_OP_T, CUBLAS_OP_N, 1, 1, n, &alpha, wf1, n, ld1, wf2, n, ld2, &beta, + // d_work.get(), 1, 1, num_bands); + // #endif + // acc::copyout(result_local.data(), d_work.get(), num_bands); + for (int i = 0; i < num_bands; ++i) { + result_local[i] = F{2} * result_local[i]; + } + + if (reduced == 1) { + /* rank owns G=0 */ + std::vector> wf1_h_g0(num_bands); + std::vector> wf2_h_g0(num_bands); + // copy G=0 to host + acc::copyout(wf1_h_g0.data(), 1, wf1, ld1, 1, num_bands); + acc::copyout(wf2_h_g0.data(), 1, wf2, ld2, 1, num_bands); + for (int i = 0; i < num_bands; ++i) { + result_local[i] -= std::conj(wf1_h_g0[i]) * wf2_h_g0[i]; + } + } else if (reduced == 2) { + /* rank doesn't own G=0 */ + // nothing todo + } + } + + // add to `result` (because of spin-components) + for (int i = 0; i < num_bands; ++i) { + if constexpr (is_real_v) { + result[i] += std::real(result_local[i]); + } else { + result[i] += result_local[i]; + } + } +} + +} // namespace sirius +#endif // defined(SIRIUS_ROCM) || defined(SIRIUS_CUDA) + +#endif /* __WF_INNER_LOCAL_HPP__ */ diff --git a/src/potential/xc.cpp b/src/potential/xc.cpp index 317d51073..8aa197648 100644 --- a/src/potential/xc.cpp +++ b/src/potential/xc.cpp @@ -346,7 +346,10 @@ Potential::xc_rg_magnetic(Density const& density__, bool use_lapl__, const bool mdarray exc({num_points}, mdarray_label("exc_tmp")); mdarray vxc_up({num_points}, mdarray_label("vxc_up_tmp")); mdarray vxc_dn({num_points}, mdarray_label("vxc_dn_dmp")); + +#if defined(SIRIUS_USE_VDWXC) std::array stress_kernel; +#endif /* loop over XC functionals */ for (auto& ixc : xc_func_) {