diff --git a/Libraries/rocSPARSE/level_2/CMakeLists.txt b/Libraries/rocSPARSE/level_2/CMakeLists.txt index 4919ca406..e44a93bf7 100644 --- a/Libraries/rocSPARSE/level_2/CMakeLists.txt +++ b/Libraries/rocSPARSE/level_2/CMakeLists.txt @@ -1,6 +1,6 @@ # MIT License # -# Copyright (c) 2023 Advanced Micro Devices, Inc. All rights reserved. +# Copyright (c) 2023-2024 Advanced Micro Devices, Inc. All rights reserved. # # Permission is hereby granted, free of charge, to any person obtaining a copy # of this software and associated documentation files (the "Software"), to deal @@ -27,10 +27,12 @@ add_subdirectory(bsrmv) add_subdirectory(bsrsv) add_subdirectory(bsrxmv) add_subdirectory(coomv) +add_subdirectory(csritsv) add_subdirectory(csrmv) add_subdirectory(csrsv) add_subdirectory(ellmv) add_subdirectory(gebsrmv) add_subdirectory(gemvi) +add_subdirectory(spitsv) add_subdirectory(spmv) add_subdirectory(spsv) diff --git a/Libraries/rocSPARSE/level_2/Makefile b/Libraries/rocSPARSE/level_2/Makefile index 671824c9f..52981cd85 100644 --- a/Libraries/rocSPARSE/level_2/Makefile +++ b/Libraries/rocSPARSE/level_2/Makefile @@ -1,6 +1,6 @@ # MIT License # -# Copyright (c) 2023 Advanced Micro Devices, Inc. All rights reserved. +# Copyright (c) 2023-2024 Advanced Micro Devices, Inc. All rights reserved. # # Permission is hereby granted, free of charge, to any person obtaining a copy # of this software and associated documentation files (the "Software"), to deal @@ -25,11 +25,13 @@ EXAMPLES := \ bsrsv \ bsrxmv \ coomv \ + csritsv \ csrmv \ csrsv \ ellmv \ gebsrmv \ gemvi \ + spitsv \ spmv \ spsv diff --git a/Libraries/rocSPARSE/level_2/csritsv/.gitignore b/Libraries/rocSPARSE/level_2/csritsv/.gitignore new file mode 100644 index 000000000..9352a595a --- /dev/null +++ b/Libraries/rocSPARSE/level_2/csritsv/.gitignore @@ -0,0 +1 @@ +rocsparse_csritsv diff --git a/Libraries/rocSPARSE/level_2/csritsv/CMakeLists.txt b/Libraries/rocSPARSE/level_2/csritsv/CMakeLists.txt new file mode 100644 index 000000000..adc3eddc3 --- /dev/null +++ b/Libraries/rocSPARSE/level_2/csritsv/CMakeLists.txt @@ -0,0 +1,62 @@ +# MIT License +# +# Copyright (c) 2023-2024 Advanced Micro Devices, Inc. All rights reserved. +# +# Permission is hereby granted, free of charge, to any person obtaining a copy +# of this software and associated documentation files (the "Software"), to deal +# in the Software without restriction, including without limitation the rights +# to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +# copies of the Software, and to permit persons to whom the Software is +# furnished to do so, subject to the following conditions: +# +# The above copyright notice and this permission notice shall be included in all +# copies or substantial portions of the Software. +# +# THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +# IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +# FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +# AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +# LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +# OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE +# SOFTWARE. + +set(example_name rocsparse_csritsv) + +cmake_minimum_required(VERSION 3.21 FATAL_ERROR) +project(${example_name} LANGUAGES CXX) + +if(GPU_RUNTIME STREQUAL "CUDA") + message(STATUS "rocSPARSE examples do not support the CUDA runtime") + return() +endif() + +# This example does not contain device code, thereby it can be compiled with any conforming C++ compiler. + +set(CMAKE_CXX_STANDARD 17) +set(CMAKE_CXX_EXTENSIONS OFF) +set(CMAKE_CXX_STANDARD_REQUIRED ON) + +if(WIN32) + set(ROCM_ROOT "$ENV{HIP_PATH}" CACHE PATH "Root directory of the ROCm installation") +else() + set(ROCM_ROOT "/opt/rocm" CACHE PATH "Root directory of the ROCm installation") +endif() + +list(APPEND CMAKE_PREFIX_PATH "${ROCM_ROOT}") + +find_package(rocsparse REQUIRED) + +add_executable(${example_name} main.cpp) +# Make example runnable using ctest +add_test(${example_name} ${example_name}) + +# Link to example library +target_link_libraries(${example_name} PRIVATE roc::rocsparse hip::host) + +target_include_directories(${example_name} PRIVATE "../../../../Common") + +install(TARGETS ${example_name}) + +if(CMAKE_SYSTEM_NAME MATCHES Windows) + install(IMPORTED_RUNTIME_ARTIFACTS roc::rocsparse) +endif() diff --git a/Libraries/rocSPARSE/level_2/csritsv/Makefile b/Libraries/rocSPARSE/level_2/csritsv/Makefile new file mode 100644 index 000000000..c41fe5f7a --- /dev/null +++ b/Libraries/rocSPARSE/level_2/csritsv/Makefile @@ -0,0 +1,58 @@ +# MIT License +# +# Copyright (c) 2023-2024 Advanced Micro Devices, Inc. All rights reserved. +# +# Permission is hereby granted, free of charge, to any person obtaining a copy +# of this software and associated documentation files (the "Software"), to deal +# in the Software without restriction, including without limitation the rights +# to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +# copies of the Software, and to permit persons to whom the Software is +# furnished to do so, subject to the following conditions: +# +# The above copyright notice and this permission notice shall be included in all +# copies or substantial portions of the Software. +# +# THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +# IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +# FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +# AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +# LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +# OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE +# SOFTWARE. + +EXAMPLE := rocsparse_csritsv +COMMON_INCLUDE_DIR := ../../../../Common +GPU_RUNTIME := HIP + +ifneq ($(GPU_RUNTIME), HIP) + $(error GPU_RUNTIME is set to "$(GPU_RUNTIME)". GPU_RUNTIME must be HIP.) +endif + +ROCM_INSTALL_DIR := /opt/rocm + +HIP_INCLUDE_DIR := $(ROCM_INSTALL_DIR)/include +ROCSPARSE_INCLUDE_DIR := $(HIP_INCLUDE_DIR) + +CXX ?= g++ + +# Common variables and flags +CXX_STD := c++17 +ICXXFLAGS := -std=$(CXX_STD) +ICPPFLAGS := -isystem $(ROCSPARSE_INCLUDE_DIR) -isystem $(HIP_INCLUDE_DIR) -I $(COMMON_INCLUDE_DIR) -D__HIP_PLATFORM_AMD__ +ILDFLAGS := -L $(ROCM_INSTALL_DIR)/lib +ILDLIBS := -lrocsparse -lamdhip64 + +CXXFLAGS ?= -Wall -Wextra + +ICXXFLAGS += $(CXXFLAGS) +ICPPFLAGS += $(CPPFLAGS) +ILDFLAGS += $(LDFLAGS) +ILDLIBS += $(LDLIBS) + +$(EXAMPLE): main.cpp $(COMMON_INCLUDE_DIR)/example_utils.hpp $(COMMON_INCLUDE_DIR)/rocsparse_utils.hpp + $(CXX) $(ICXXFLAGS) $(ICPPFLAGS) $(ILDFLAGS) -o $@ $< $(ILDLIBS) + +clean: + $(RM) $(EXAMPLE) + +.PHONY: clean diff --git a/Libraries/rocSPARSE/level_2/csritsv/README.md b/Libraries/rocSPARSE/level_2/csritsv/README.md new file mode 100644 index 000000000..97a2d0121 --- /dev/null +++ b/Libraries/rocSPARSE/level_2/csritsv/README.md @@ -0,0 +1,156 @@ +# rocSPARSE Level 2 CSR Iterative Triangular Matrix-Vector Multiplication + +## Description + +This example illustrates the use of the `rocSPARSE` level 2 iterative triangular solver using the CSR storage format. + +This triangular solver is used to find an iterative solution with Jacobi method for a linear system of the form + +$$ +A' y \approx \alpha x, +$$ + +with a `tolerance` and a `max_iter` maximal number of iterations where + +- $A$ is a sparse triangular matrix of order $n$ whose elements are the coefficients of the equations, +- $A'$ is one of the following: + - $A' = A$ (identity) + - $A' = A^T$ (transpose $A$: $A_{ij}^T = A_{ji}$) + - $A' = A^H$ (conjugate transpose/Hermitian $A$: $A_{ij}^H = \bar A_{ji}$), +- $\alpha$ is a scalar, +- $x$ is a dense vector of size $n$ containing the constant terms of the equations, and +- $y$ is a dense vector of size $n$ which contains the unknowns of the system. + +Obtaining solution for such a system consists on finding concrete values of all the unknowns such that the above equality holds. + +### Application flow + +1. Setup input data. +2. Allocate device memory and offload input data to device. +3. Initialize rocSPARSE by creating a handle. +4. Prepare utility variables for rocSPARSE csritsv invocation. +5. Perform analysis step. +6. Perform triangular solve $A' y = \alpha x$. +7. Check results obtained. +8. Copy solution vector $y$ from device to host and compare with expected result. +9. Free rocSPARSE resources and device memory. +10. Print validation result. + +## Key APIs and Concepts + +### CSR Matrix Storage Format + +The [Compressed Sparse Row (CSR) storage format](https://rocsparse.readthedocs.io/en/latest/usermanual.html#csr-storage-format) describes an $m \times n$ sparse matrix with three arrays. + +Defining + +- `m`: number of rows +- `n`: number of columns +- `nnz`: number of non-zero elements + +we can describe a sparse matrix using the following arrays: + +- `csr_val`: array storing the non-zero elements of the matrix. +- `csr_row_ptr`: given $i \in [0, m]$ + - if $` 0 \leq i < m `$, `csr_row_ptr[i]` stores the index of the first non-zero element in row $i$ of the matrix + - if $i = m$, `csr_row_ptr[i]` stores `nnz`. + + This way, row $j \in [0, m)$ contains the non-zero elements of indices from `csr_row_ptr[j]` to `csr_row_ptr[j+1]-1`. Therefore, the corresponding values in `csr_val` can be accessed from `csr_row_ptr[j]` to `csr_row_ptr[j+1]-1`. +- `csr_col_ind`: given $i \in [0, nnz-1]$, `csr_col_ind[i]` stores the column of the $i^{th}$ non-zero element in the matrix. + +The CSR matrix is sorted by column indices in the same row, and each pair of indices appear only once. + +For instance, consider a sparse matrix as + +$$ +A= +\left( +\begin{array}{ccccc} +1 & 2 & 0 & 3 & 0 \\ +0 & 4 & 5 & 0 & 0 \\ +6 & 0 & 0 & 7 & 8 +\end{array} +\right) +$$ + +Therefore, the CSR representation of $A$ is: + +```c++ +m = 3 + +n = 5 + +nnz = 8 + +csr_val = { 1, 2, 3, 4, 5, 6, 7, 8 } + +csr_row_ptr = { 0, 3, 5, 8 } + +csr_col_ind = { 0, 1, 3, 1, 2, 0, 3, 4 } +``` + +### rocSPARSE + +- rocSPARSE is initialized by calling `rocsparse_create_handle(rocsparse_handle*)` and is terminated by calling `rocsparse_destroy_handle(rocsparse_handle)`. +- `rocsparse_pointer_mode` controls whether scalar parameters must be allocated on the host (`rocsparse_pointer_mode_host`) or on the device (`rocsparse_pointer_mode_device`). It is controlled by `rocsparse_set_pointer_mode`. +- `rocsparse_operation trans`: matrix operation applied to the given input matrix. The following values are accepted: + - `rocsparse_operation_none`: identity operation $A' = A$. + - `rocsparse_operation_transpose`: transpose operation $A' = A^\mathrm{T}$. + - `rocsparse_operation_conjugate_transpose`: conjugate transpose operation (Hermitian matrix) $A' = A^\mathrm{H}$. This operation is not yet supported. +- `rocsparse_mat_descr descr`: holds all properties of a matrix. The properties set in this example are the following: + - `rocsparse_diag_type`: indicates whether the diagonal entries of a matrix are unit elements (`rocsparse_diag_type_unit`) or not (`rocsparse_diag_type_non_unit`). + - `rocsparse_fill_mode`: indicates whether a (triangular) matrix is lower (`rocsparse_fill_mode_lower`) or upper (`rocsparse_fill_mode_upper`) triangular. +- `rocsparse_[sdcz]csritsv_buffer_size` allows to obtain the size (in bytes) of the temporary storage buffer required for the `rocsparse_[sdcz]csritsv_analysis` and `rocsparse_[sdcz]csritsv_solve` functions. The character matched in `[sdcz]` coincides with the one matched in any of the mentioned functions. +- `rocsparse_solve_policy policy`: specifies the policy to follow for triangular solvers and factorizations. The only value accepted is `rocsparse_solve_policy_auto`. +- `rocsparse_[sdcz]csritsv_solve` solves a sparse triangular linear system $A' y = \alpha x$. The correct function signature should be chosen based on the datatype of the input matrix: + - `s` single-precision real (`float`) + - `d` double-precision real (`double`) + - `c` single-precision complex (`rocsparse_float_complex`) + - `z` double-precision complex (`rocsparse_double_complex`) +- `rocsparse_analysis_policy analysis`: specifies the policy to follow for analysis data. The following values are accepted: + - `rocsparse_analysis_policy_reuse`: the analysis data gathered is re-used. + - `rocsparse_analysis_policy_force`: the analysis data will be re-built. +- `rocsparse_[sdcz]csritsv_analysis` performs the analysis step for `rocsparse_[sdcz]csritsv_solve`. The character matched in `[sdcz]` coincides with the one matched in `rocsparse_[sdcz]csritsv_solve`. +- `rocsparse_csritsv_zero_pivot(rocsparse_handle, rocsparse_mat_info, rocsparse_int *position)` returns `rocsparse_status_zero_pivot` if either a structural or numerical zero has been found during the execution of `rocsparse_[sbcz]csritsv_solve(....)` and stores in `position` the index $i$ of the first zero pivot $A_{ii}$ found. If no zero pivot is found it returns `rocsparse_status_success`. + +## Demonstrated API Calls + +### rocSPARSE + +- `rocsparse_analysis_policy` +- `rocsparse_analysis_policy_reuse` +- `rocsparse_create_handle` +- `rocsparse_create_mat_descr` +- `rocsparse_create_mat_info` +- `rocsparse_csritsv_zero_pivot` +- `rocsparse_dcsritsv_analysis` +- `rocsparse_dcsritsv_buffer_size` +- `rocsparse_dcsritsv_solve` +- `rocsparse_destroy_handle` +- `rocsparse_destroy_mat_descr` +- `rocsparse_destroy_mat_info` +- `rocsparse_diag_type_non_unit` +- `rocsparse_fill_mode_lower` +- `rocsparse_handle` +- `rocsparse_int` +- `rocsparse_mat_descr` +- `rocsparse_mat_info` +- `rocsparse_operation` +- `rocsparse_operation_none` +- `rocsparse_pointer_mode_host` +- `rocsparse_set_mat_diag_type` +- `rocsparse_set_mat_fill_mode` +- `rocsparse_set_pointer_mode` +- `rocsparse_solve_policy` +- `rocsparse_solve_policy_auto` +- `rocsparse_status` +- `rocsparse_status_zero_pivot` + +### HIP runtime + +- `hipDeviceSynchronize` +- `hipFree` +- `hipMalloc` +- `hipMemcpy` +- `hipMemcpyDeviceToHost` +- `hipMemcpyHostToDevice` diff --git a/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2017.sln b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2017.sln new file mode 100644 index 000000000..5ca5a8b3b --- /dev/null +++ b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2017.sln @@ -0,0 +1,24 @@ +Microsoft Visual Studio Solution File, Format Version 12.00 +# Visual Studio 15 +VisualStudioVersion = 15.0.33026.149 +MinimumVisualStudioVersion = 10.0.40219.1 +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "csritsv_vs2017", "csritsv_vs2017.vcxproj", "{F0AF1DEB-4B07-4FDC-8566-FB53F60D10B7}" +EndProject +Global + GlobalSection(SolutionConfigurationPlatforms) = preSolution + Debug|x64 = Debug|x64 + Release|x64 = Release|x64 + EndGlobalSection + GlobalSection(ProjectConfigurationPlatforms) = postSolution + {F0AF1DEB-4B07-4FDC-8566-FB53F60D10B7}.Debug|x64.ActiveCfg = Debug|x64 + {F0AF1DEB-4B07-4FDC-8566-FB53F60D10B7}.Debug|x64.Build.0 = Debug|x64 + {F0AF1DEB-4B07-4FDC-8566-FB53F60D10B7}.Release|x64.ActiveCfg = Release|x64 + {F0AF1DEB-4B07-4FDC-8566-FB53F60D10B7}.Release|x64.Build.0 = Release|x64 + EndGlobalSection + GlobalSection(SolutionProperties) = preSolution + HideSolutionNode = FALSE + EndGlobalSection + GlobalSection(ExtensibilityGlobals) = postSolution + SolutionGuid = {E11DC4C1-CA8A-46CA-93BB-3CB480169DA5} + EndGlobalSection +EndGlobal diff --git a/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2017.vcxproj b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2017.vcxproj new file mode 100644 index 000000000..861c69af1 --- /dev/null +++ b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2017.vcxproj @@ -0,0 +1,108 @@ + + + + + Debug + x64 + + + Release + x64 + + + + 5.7 + 15.0 + {F0AF1DEB-4B07-4FDC-8566-FB53F60D10B7} + Win32Proj + csritsv_vs2017 + $(LatestTargetPlatformVersion) + + + + + + + + + + + PreserveNewest + + + + + Application + true + HIP clang 5.7 + Unicode + + + Application + false + HIP clang 5.7 + Unicode + + + + + + + + + + + + + + + + true + + + + true + rocsparse_$(ProjectName) + + + false + rocsparse_$(ProjectName) + + + + true + + + + + Level2 + __clang__;__HIP__;_DEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + true + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + Level2 + __clang__;__HIP__;NDEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2017.vcxproj.filters b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2017.vcxproj.filters new file mode 100644 index 000000000..a8f092a01 --- /dev/null +++ b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2017.vcxproj.filters @@ -0,0 +1,30 @@ + + + + + {f53b0509-d71c-4fbf-81c4-ffdb65b0c6d1} + cpp;c;cc;cxx;c++;def;odl;idl;hpj;bat;asm;asmx;hip;cu + + + {708cbb35-729f-487a-aed6-c0b64ac29a1f} + h;hh;hpp;hxx;h++;hm;inl;inc;ipp;xsd;cuh + + + {ff6fe45d-d5b5-4c0f-8b71-e1a0cfedc652} + rc;ico;cur;bmp;dlg;rc2;rct;bin;rgs;gif;jpg;jpeg;jpe;resx;tiff;tif;png;wav;mfcribbon-ms + + + + + Source Files + + + + + Header Files + + + Header Files + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2019.sln b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2019.sln new file mode 100644 index 000000000..00307617e --- /dev/null +++ b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2019.sln @@ -0,0 +1,24 @@ +Microsoft Visual Studio Solution File, Format Version 12.00 +# Visual Studio Version 16 +VisualStudioVersion = 16.0.32630.194 +MinimumVisualStudioVersion = 10.0.40219.1 +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "csritsv_vs2019", "csritsv_vs2019.vcxproj", "{99A25D0A-93FE-47F2-8223-7313E53E7951}" +EndProject +Global + GlobalSection(SolutionConfigurationPlatforms) = preSolution + Debug|x64 = Debug|x64 + Release|x64 = Release|x64 + EndGlobalSection + GlobalSection(ProjectConfigurationPlatforms) = postSolution + {99A25D0A-93FE-47F2-8223-7313E53E7951}.Debug|x64.ActiveCfg = Debug|x64 + {99A25D0A-93FE-47F2-8223-7313E53E7951}.Debug|x64.Build.0 = Debug|x64 + {99A25D0A-93FE-47F2-8223-7313E53E7951}.Release|x64.ActiveCfg = Release|x64 + {99A25D0A-93FE-47F2-8223-7313E53E7951}.Release|x64.Build.0 = Release|x64 + EndGlobalSection + GlobalSection(SolutionProperties) = preSolution + HideSolutionNode = FALSE + EndGlobalSection + GlobalSection(ExtensibilityGlobals) = postSolution + SolutionGuid = {0162D119-0F9F-4EAD-8799-A82601DA4643} + EndGlobalSection +EndGlobal diff --git a/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2019.vcxproj b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2019.vcxproj new file mode 100644 index 000000000..439e568d1 --- /dev/null +++ b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2019.vcxproj @@ -0,0 +1,108 @@ + + + + + Debug + x64 + + + Release + x64 + + + + 5.7 + 16.0 + {99A25D0A-93FE-47F2-8223-7313E53E7951} + Win32Proj + csritsv_vs2019 + $(LatestTargetPlatformVersion) + + + + + + + + + + + PreserveNewest + + + + + Application + true + HIP clang 5.7 + Unicode + + + Application + false + HIP clang 5.7 + Unicode + + + + + + + + + + + + + + + + true + + + + true + rocsparse_$(ProjectName) + + + false + rocsparse_$(ProjectName) + + + + true + + + + + Level2 + __clang__;__HIP__;_DEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + true + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + Level2 + __clang__;__HIP__;NDEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2019.vcxproj.filters b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2019.vcxproj.filters new file mode 100644 index 000000000..fa4e0bfed --- /dev/null +++ b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2019.vcxproj.filters @@ -0,0 +1,30 @@ + + + + + {d3fab5a1-9b45-4f95-9c26-09cd7bb42c89} + cpp;c;cc;cxx;c++;def;odl;idl;hpj;bat;asm;asmx;hip;cu + + + {b0eb2520-65f5-4cd7-bcdb-cd42ee6f743a} + h;hh;hpp;hxx;h++;hm;inl;inc;ipp;xsd;cuh + + + {27f7e884-059f-464d-a9de-897cd6a157ed} + rc;ico;cur;bmp;dlg;rc2;rct;bin;rgs;gif;jpg;jpeg;jpe;resx;tiff;tif;png;wav;mfcribbon-ms + + + + + Source Files + + + + + Header Files + + + Header Files + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2022.sln b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2022.sln new file mode 100644 index 000000000..7adcd069f --- /dev/null +++ b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2022.sln @@ -0,0 +1,24 @@ +Microsoft Visual Studio Solution File, Format Version 12.00 +# Visual Studio Version 17 +VisualStudioVersion = 17.4.33213.308 +MinimumVisualStudioVersion = 10.0.40219.1 +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "csritsv_vs2022", "csritsv_vs2022.vcxproj", "{DC1DF216-BC97-4797-8EA7-8DDCC38DFDCF}" +EndProject +Global + GlobalSection(SolutionConfigurationPlatforms) = preSolution + Debug|x64 = Debug|x64 + Release|x64 = Release|x64 + EndGlobalSection + GlobalSection(ProjectConfigurationPlatforms) = postSolution + {DC1DF216-BC97-4797-8EA7-8DDCC38DFDCF}.Debug|x64.ActiveCfg = Debug|x64 + {DC1DF216-BC97-4797-8EA7-8DDCC38DFDCF}.Debug|x64.Build.0 = Debug|x64 + {DC1DF216-BC97-4797-8EA7-8DDCC38DFDCF}.Release|x64.ActiveCfg = Release|x64 + {DC1DF216-BC97-4797-8EA7-8DDCC38DFDCF}.Release|x64.Build.0 = Release|x64 + EndGlobalSection + GlobalSection(SolutionProperties) = preSolution + HideSolutionNode = FALSE + EndGlobalSection + GlobalSection(ExtensibilityGlobals) = postSolution + SolutionGuid = {88A781BF-32EA-4A62-826D-4FD05019CD70} + EndGlobalSection +EndGlobal diff --git a/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2022.vcxproj b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2022.vcxproj new file mode 100644 index 000000000..f91c4bab7 --- /dev/null +++ b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2022.vcxproj @@ -0,0 +1,108 @@ + + + + + Debug + x64 + + + Release + x64 + + + + 5.7 + 17.0 + {DC1DF216-BC97-4797-8EA7-8DDCC38DFDCF} + Win32Proj + csritsv_vs2022 + $(LatestTargetPlatformVersion) + + + + + + + + + + + PreserveNewest + + + + + Application + true + HIP clang 5.7 + Unicode + + + Application + false + HIP clang 5.7 + Unicode + + + + + + + + + + + + + + + + true + + + + true + rocsparse_$(ProjectName) + + + false + rocsparse_$(ProjectName) + + + + true + + + + + Level2 + __clang__;__HIP__;_DEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + true + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + Level2 + __clang__;__HIP__;NDEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2022.vcxproj.filters b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2022.vcxproj.filters new file mode 100644 index 000000000..3af781b92 --- /dev/null +++ b/Libraries/rocSPARSE/level_2/csritsv/csritsv_vs2022.vcxproj.filters @@ -0,0 +1,30 @@ + + + + + {3197089a-f577-477c-b7f6-fd9093a8d9f6} + cpp;c;cc;cxx;c++;def;odl;idl;hpj;bat;asm;asmx;hip;cu + + + {a63c63f6-e0b8-4ad7-8bf0-2ae0717b0a3a} + h;hh;hpp;hxx;h++;hm;inl;inc;ipp;xsd;cuh + + + {92377426-1df4-4bb1-858b-1d0072c8e639} + rc;ico;cur;bmp;dlg;rc2;rct;bin;rgs;gif;jpg;jpeg;jpe;resx;tiff;tif;png;wav;mfcribbon-ms + + + + + Source Files + + + + + Header Files + + + Header Files + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/level_2/csritsv/main.cpp b/Libraries/rocSPARSE/level_2/csritsv/main.cpp new file mode 100644 index 000000000..01e3e1529 --- /dev/null +++ b/Libraries/rocSPARSE/level_2/csritsv/main.cpp @@ -0,0 +1,239 @@ +// MIT License +// +// Copyright (c) 2023-2024 Advanced Micro Devices, Inc. All rights reserved. +// +// Permission is hereby granted, free of charge, to any person obtaining a copy +// of this software and associated documentation files (the "Software"), to deal +// in the Software without restriction, including without limitation the rights +// to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +// copies of the Software, and to permit persons to whom the Software is +// furnished to do so, subject to the following conditions: +// +// The above copyright notice and this permission notice shall be included in all +// copies or substantial portions of the Software. +// +// THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +// IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +// FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +// AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +// LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +// OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE +// SOFTWARE. + +#include "example_utils.hpp" +#include "rocsparse_utils.hpp" + +#include +#include + +#include +#include +#include +#include + +int main() +{ + // 1. Setup input data. + + // alpha * A' * y = x + // 1.0 * ( 1.0 0.0 0.0 0.0 ) * ( 1 ) = ( 1.0 ) + // ( 2.0 3.0 0.0 0.0 ) * ( 0 ) ( 2.0 ) + // ( 4.0 5.0 6.0 0.0 ) * ( -1/6 ) ( 3.0 ) + // ( 7.0 0.0 8.0 9.0 ) * ( -5/27 ) ( 4.0 ) + + // Number of rows and columns of the input matrix. + constexpr rocsparse_int m = 4; + constexpr rocsparse_int n = 4; + + // Number of non-zero elements. + constexpr rocsparse_int nnz = 9; + + // CSR values. + constexpr std::array h_csr_val = {1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0, 9.0}; + + // CSR row indices. + constexpr std::array h_csr_row_ptr = {0, 1, 3, 6, 9}; + + // CSR column indices. + constexpr std::array h_csr_col_ind = {0, 0, 1, 0, 1, 2, 0, 2, 3}; + + // Operation applied to the matrix. + constexpr rocsparse_operation trans = rocsparse_operation_none; + + // Analysis and solve policies. + constexpr rocsparse_analysis_policy analysis_policy = rocsparse_analysis_policy_reuse; + constexpr rocsparse_solve_policy solve_policy = rocsparse_solve_policy_auto; + + // Scalar alpha. + constexpr double alpha = 1.0; + + // Host vectors for the right hand side and solution of the linear system. + std::array h_x = {1.0, 2.0, 3.0, 4.0}; + std::array h_y; + + // 2. Allocate device memory and offload input data to device. + rocsparse_int* d_csr_row_ptr; + rocsparse_int* d_csr_col_ind; + double* d_csr_val; + double* d_x; + double* d_y; + + constexpr size_t x_size = sizeof(*d_x) * n; + constexpr size_t y_size = sizeof(*d_y) * m; + constexpr size_t val_size = sizeof(*d_csr_val) * nnz; + constexpr size_t row_ptr_size = sizeof(*d_csr_row_ptr) * (m + 1); + constexpr size_t col_ind_size = sizeof(*d_csr_col_ind) * nnz; + + HIP_CHECK(hipMalloc(&d_csr_row_ptr, row_ptr_size)); + HIP_CHECK(hipMalloc(&d_csr_col_ind, col_ind_size)); + HIP_CHECK(hipMalloc(&d_csr_val, val_size)); + HIP_CHECK(hipMalloc(&d_x, x_size)); + HIP_CHECK(hipMalloc(&d_y, y_size)); + + HIP_CHECK(hipMemcpy(d_csr_row_ptr, h_csr_row_ptr.data(), row_ptr_size, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(d_csr_col_ind, h_csr_col_ind.data(), col_ind_size, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(d_csr_val, h_csr_val.data(), val_size, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(d_x, h_x.data(), x_size, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(d_y, h_y.data(), y_size, hipMemcpyHostToDevice)); + + // 3. Initialize rocSPARSE by creating a handle. + rocsparse_handle handle; + ROCSPARSE_CHECK(rocsparse_create_handle(&handle)); + ROCSPARSE_CHECK(rocsparse_set_pointer_mode(handle, rocsparse_pointer_mode_host)); + + // 4. Prepare utility variables for rocSPARSE csrmv invocation. + // Matrix descriptor. + rocsparse_mat_descr descr; + ROCSPARSE_CHECK(rocsparse_create_mat_descr(&descr)); + + // Matrix fill mode. + ROCSPARSE_CHECK(rocsparse_set_mat_fill_mode(descr, rocsparse_fill_mode_lower)); + + // Matrix diagonal type. + ROCSPARSE_CHECK(rocsparse_set_mat_diag_type(descr, rocsparse_diag_type_non_unit)); + + // Matrix info structure. + rocsparse_mat_info info; + ROCSPARSE_CHECK(rocsparse_create_mat_info(&info)); + + // Set maximum iteration number. + constexpr rocsparse_int max_iter = 200; + rocsparse_int iter_counter = max_iter; + + // Set tolerance. + constexpr double tolerance = 1e-4; + + // Set up convergence history. + std::array history; + + // Obtain required buffer size in bytes for analysis and solve stages. + size_t buffer_size; + ROCSPARSE_CHECK(rocsparse_dcsritsv_buffer_size(handle, + trans, + m, + nnz, + descr, + d_csr_val, + d_csr_row_ptr, + d_csr_col_ind, + info, + &buffer_size)); + + // Synchronize threads. + HIP_CHECK(hipDeviceSynchronize()); + + // Allocate temporary buffer. + void* temp_buffer{}; + HIP_CHECK(hipMalloc(&temp_buffer, buffer_size)); + + // 5. Perform analysis step. + ROCSPARSE_CHECK(rocsparse_dcsritsv_analysis(handle, + trans, + m, + nnz, + descr, + d_csr_val, + d_csr_row_ptr, + d_csr_col_ind, + info, + analysis_policy, + solve_policy, + temp_buffer)); + + // 6. Perform triangular solve A' y = alpha * x. + ROCSPARSE_CHECK(rocsparse_dcsritsv_solve(handle, + &iter_counter, + &tolerance, + history.data(), + trans, + m, + nnz, + &alpha, + descr, + d_csr_val, + d_csr_row_ptr, + d_csr_col_ind, + info, + d_x, + d_y, + solve_policy, + temp_buffer)); + + // Synchronize threads. + HIP_CHECK(hipDeviceSynchronize()); + + // 7. Check results obtained. + int errors = 0; + + if(iter_counter >= max_iter) + { + std::cout << "The iteration did not converged in " << iter_counter << " steps." + << std::endl; + ++errors; + } + + rocsparse_int position; + rocsparse_status pivot_status = rocsparse_csritsv_zero_pivot(handle, descr, info, &position); + + if(pivot_status == rocsparse_status_zero_pivot) + { + std::cout << "Found zero pivot in matrix row " << position << std::endl; + ++errors; + } + else + { + ROCSPARSE_CHECK(pivot_status); + } + + // 8. Compare solution vector with reference. + // Copy solution vector to host. + HIP_CHECK(hipMemcpy(h_y.data(), d_y, sizeof(*d_y) * n, hipMemcpyDeviceToHost)); + + std::cout << "Solution successfully computed: "; + std::cout << "y = " << format_range(h_y.begin(), h_y.end()) << std::endl; + + // Define expected result. + constexpr std::array expected_y = {1.0, 0.0, -1.0 / 6.0, -5.0 / 27.0}; + + // Compare solution with the expected result. + const double eps = 1.0e5 * std::numeric_limits::epsilon(); + for(size_t i = 0; i < h_y.size(); ++i) + { + errors += std::fabs(h_y[i] - expected_y[i]) > eps; + } + + // 9. Free rocSPARSE resources and device memory. + ROCSPARSE_CHECK(rocsparse_destroy_handle(handle)); + ROCSPARSE_CHECK(rocsparse_destroy_mat_descr(descr)); + ROCSPARSE_CHECK(rocsparse_destroy_mat_info(info)); + + HIP_CHECK(hipFree(d_csr_row_ptr)); + HIP_CHECK(hipFree(d_csr_col_ind)); + HIP_CHECK(hipFree(d_csr_val)); + HIP_CHECK(hipFree(d_x)); + HIP_CHECK(hipFree(d_y)); + HIP_CHECK(hipFree(temp_buffer)); + + // 10. Print validation result. + return report_validation_result(errors); +} diff --git a/Libraries/rocSPARSE/level_2/csrsv/README.md b/Libraries/rocSPARSE/level_2/csrsv/README.md index e38c87aeb..8fd2b047f 100644 --- a/Libraries/rocSPARSE/level_2/csrsv/README.md +++ b/Libraries/rocSPARSE/level_2/csrsv/README.md @@ -28,7 +28,7 @@ Obtaining solution for such a system consists on finding concrete values of all 1. Setup input data. 2. Allocate device memory and offload input data to device. 3. Initialize rocSPARSE by creating a handle. -4. Prepare utility variables for rocSPARSE csrmv invocation. +4. Prepare utility variables for rocSPARSE csrsv invocation. 5. Perform analysis step. 6. Perform triangular solve $Ay = \alpha x$. 7. Check results obtained. If no zero-pivots, copy solution vector $y$ from device to host and compare with expected result. diff --git a/Libraries/rocSPARSE/level_2/spitsv/.gitignore b/Libraries/rocSPARSE/level_2/spitsv/.gitignore new file mode 100644 index 000000000..8cc73bc08 --- /dev/null +++ b/Libraries/rocSPARSE/level_2/spitsv/.gitignore @@ -0,0 +1 @@ +rocsparse_spitsv diff --git a/Libraries/rocSPARSE/level_2/spitsv/CMakeLists.txt b/Libraries/rocSPARSE/level_2/spitsv/CMakeLists.txt new file mode 100644 index 000000000..777cae125 --- /dev/null +++ b/Libraries/rocSPARSE/level_2/spitsv/CMakeLists.txt @@ -0,0 +1,62 @@ +# MIT License +# +# Copyright (c) 2023-2024 Advanced Micro Devices, Inc. All rights reserved. +# +# Permission is hereby granted, free of charge, to any person obtaining a copy +# of this software and associated documentation files (the "Software"), to deal +# in the Software without restriction, including without limitation the rights +# to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +# copies of the Software, and to permit persons to whom the Software is +# furnished to do so, subject to the following conditions: +# +# The above copyright notice and this permission notice shall be included in all +# copies or substantial portions of the Software. +# +# THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +# IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +# FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +# AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +# LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +# OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE +# SOFTWARE. + +set(example_name rocsparse_spitsv) + +cmake_minimum_required(VERSION 3.21 FATAL_ERROR) +project(${example_name} LANGUAGES CXX) + +if(GPU_RUNTIME STREQUAL "CUDA") + message(STATUS "rocSPARSE examples do not support the CUDA runtime") + return() +endif() + +# This example does not contain device code, thereby it can be compiled with any conforming C++ compiler. + +set(CMAKE_CXX_STANDARD 17) +set(CMAKE_CXX_EXTENSIONS OFF) +set(CMAKE_CXX_STANDARD_REQUIRED ON) + +if(WIN32) + set(ROCM_ROOT "$ENV{HIP_PATH}" CACHE PATH "Root directory of the ROCm installation") +else() + set(ROCM_ROOT "/opt/rocm" CACHE PATH "Root directory of the ROCm installation") +endif() + +list(APPEND CMAKE_PREFIX_PATH "${ROCM_ROOT}") + +find_package(rocsparse REQUIRED) + +add_executable(${example_name} main.cpp) +# Make example runnable using ctest +add_test(${example_name} ${example_name}) + +# Link to example library +target_link_libraries(${example_name} PRIVATE roc::rocsparse hip::host) + +target_include_directories(${example_name} PRIVATE "../../../../Common") + +install(TARGETS ${example_name}) + +if(CMAKE_SYSTEM_NAME MATCHES Windows) + install(IMPORTED_RUNTIME_ARTIFACTS roc::rocsparse) +endif() diff --git a/Libraries/rocSPARSE/level_2/spitsv/Makefile b/Libraries/rocSPARSE/level_2/spitsv/Makefile new file mode 100644 index 000000000..596626694 --- /dev/null +++ b/Libraries/rocSPARSE/level_2/spitsv/Makefile @@ -0,0 +1,58 @@ +# MIT License +# +# Copyright (c) 2023-2024 Advanced Micro Devices, Inc. All rights reserved. +# +# Permission is hereby granted, free of charge, to any person obtaining a copy +# of this software and associated documentation files (the "Software"), to deal +# in the Software without restriction, including without limitation the rights +# to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +# copies of the Software, and to permit persons to whom the Software is +# furnished to do so, subject to the following conditions: +# +# The above copyright notice and this permission notice shall be included in all +# copies or substantial portions of the Software. +# +# THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +# IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +# FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +# AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +# LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +# OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE +# SOFTWARE. + +EXAMPLE := rocsparse_spitsv +COMMON_INCLUDE_DIR := ../../../../Common +GPU_RUNTIME := HIP + +ifneq ($(GPU_RUNTIME), HIP) + $(error GPU_RUNTIME is set to "$(GPU_RUNTIME)". GPU_RUNTIME must be HIP.) +endif + +ROCM_INSTALL_DIR := /opt/rocm + +HIP_INCLUDE_DIR := $(ROCM_INSTALL_DIR)/include +ROCSPARSE_INCLUDE_DIR := $(HIP_INCLUDE_DIR) + +CXX ?= g++ + +# Common variables and flags +CXX_STD := c++17 +ICXXFLAGS := -std=$(CXX_STD) +ICPPFLAGS := -isystem $(ROCSPARSE_INCLUDE_DIR) -isystem $(HIP_INCLUDE_DIR) -I $(COMMON_INCLUDE_DIR) -D__HIP_PLATFORM_AMD__ +ILDFLAGS := -L $(ROCM_INSTALL_DIR)/lib +ILDLIBS := -lrocsparse -lamdhip64 + +CXXFLAGS ?= -Wall -Wextra + +ICXXFLAGS += $(CXXFLAGS) +ICPPFLAGS += $(CPPFLAGS) +ILDFLAGS += $(LDFLAGS) +ILDLIBS += $(LDLIBS) + +$(EXAMPLE): main.cpp $(COMMON_INCLUDE_DIR)/example_utils.hpp $(COMMON_INCLUDE_DIR)/rocsparse_utils.hpp + $(CXX) $(ICXXFLAGS) $(ICPPFLAGS) $(ILDFLAGS) -o $@ $< $(ILDLIBS) + +clean: + $(RM) $(EXAMPLE) + +.PHONY: clean diff --git a/Libraries/rocSPARSE/level_2/spitsv/README.md b/Libraries/rocSPARSE/level_2/spitsv/README.md new file mode 100644 index 000000000..cedbfde9e --- /dev/null +++ b/Libraries/rocSPARSE/level_2/spitsv/README.md @@ -0,0 +1,132 @@ +# rocSPARSE Level 2 Triangular Solver Example + +## Description + +This example illustrates the use of the `rocSPARSE` level 2 iterative triangular solver with a CSR sparse format. + +This triangular solver is used to solve a linear system of the form + +$$ +A'y = \alpha x, +$$ + +where + +- $A$ is a sparse triangular matrix of order $n$ whose elements are the coefficients of the equations, +- $A'$ is one of the following: + - $A' = A$ (identity) + - $A' = A^T$ (transpose $A$: $A_{ij}^T = A_{ji}$) + - $A' = A^H$ (conjugate transpose/Hermitian $A$: $A_{ij}^H = \bar A_{ji}$), +- $\alpha$ is a scalar, +- $x$ is a dense vector of size $n$ containing the constant terms of the equations, and +- $y$ is a dense vector of size $n$ which contains the unknowns of the system. + +### Application flow + +1. Set up input data. +2. Allocate device memory and offload input data to device. +3. Initialize rocSPARSE by creating a handle. +4. Prepare device for rocSPARSE iterative SpSV invocation. +5. Perform analysis step. +6. Perform triangular solve $Ay = \alpha x$. +7. Copy solution vector $y$ from device to host. +8. Free rocSPARSE resources and device memory. +9. Fetch convergence data. +10. Print solution vector $y$ to the standard output. +11. Print validation result. + +## Key APIs and Concepts + +### rocSPARSE + +- rocSPARSE is initialized by calling `rocsparse_create_handle(rocsparse_handle*)` and is terminated by calling `rocsparse_destroy_handle(rocsparse_handle)`. +- `rocsparse_pointer_mode` controls whether scalar parameters must be allocated on the host (`rocsparse_pointer_mode_host`) or on the device (`rocsparse_pointer_mode_device`). It is controlled by `rocsparse_set_pointer_mode`. +- `rocsparse_operation`: matrix operation applied to the given input matrix. The following values are accepted: + - `rocsparse_operation_none`: identity operation $A' = A$. + - `rocsparse_operation_transpose`: transpose operation $A' = A^\mathrm{T}$. + - `rocsparse_operation_conjugate_transpose`: conjugate transpose operation (Hermitian matrix) $A' = A^\mathrm{H}$. This operation is not yet supported. + +- `rocsparse_spitsv()` performs a stage of the triangular linear system solver of a sparse matrix in CSR format, iteratively. + +- `rocsparse_datatype`: data type of rocSPARSE vector and matrix elements. + - `rocsparse_datatype_f32_r`: real 32-bit floating point type + - `rocsparse_datatype_f64_r`: real 64-bit floating point type + - `rocsparse_datatype_f32_c`: complex 32-bit floating point type + - `rocsparse_datatype_f64_c`: complex 64-bit floating point type + - `rocsparse_datatype_i8_r`: real 8-bit signed integer + - `rocsparse_datatype_u8_r`: real 8-bit unsigned integer + - `rocsparse_datatype_i32_r`: real 32-bit signed integer + - `rocsparse_datatype_u32_r` real 32-bit unsigned integer + +- `rocsparse_indextype` indicates the index type of a rocSPARSE index vector. + - `rocsparse_indextype_u16`: 16-bit unsigned integer + - `rocsparse_indextype_i32`: 32-bit signed integer + - `rocsparse_indextype_i64`: 64-bit signed integer + +- `rocsparse_index_base` indicates the index base of indices. + - `rocsparse_index_base_zero`: zero based indexing. + - `rocsparse_index_base_one`: one based indexing. + +- `rocsparse_spitsv_alg`: list of iterative SpSV algorithms. + - `rocsparse_spitsv_alg_default`: default iterative SpSV algorithm for the given format (the only available option) + +- `rocsparse_spmat_descr`: sparse matrix descriptor. +- `rocsparse_create_csr_descr` creates a sparse matrix descriptor in CSR format. + + The descriptor should be destroyed at the end by `rocsparse_destroy_spmat_descr`. +- `rocsparse_destroy_spmat_descr`: Destroy a sparse matrix descriptor and release used resources allocated by the descriptor. + +- `rocsparse_dnvec_descr` is a dense vector descriptor. +- `rocsparse_create_dnvec_descr` creates a dense vector descriptor. + + The descriptor should be destroyed at the end by `rocsparse_destroy_dnvec_descr`. +- `rocsparse_destroy_dnvec_descr` destroys a dense vector descriptor. + +- `rocsparse_spitsv_stage`: list of possible stages during iterative SpSV computation. Typical order is `rocsparse_spitsv_buffer_size`, `rocsparse_spitsv_preprocess`, `rocsparse_spitsv_compute`. + - `rocsparse_spitsv_stage_buffer_size` returns the required buffer size. + - `rocsparse_spitsv_stage_preprocess` preprocesses data. + - `rocsparse_spitsv_stage_compute` performs the actual iterative SpSV computation. + - `rocsparse_spitsv_stage_auto`: automatic stage detection. + - If `temp_buffer` is equal to `nullptr`, the required buffer size will be returned. + - If `buffer_size` is equal to `nullptr`, analysis will be performed. + - Otherwise, the iterative SpSV preprocess and the iterative SpSV algorithm will be executed. + +## Demonstrated API Calls + +### rocSPARSE + +- `rocsparse_create_csr_descr` +- `rocsparse_create_dnvec_descr` +- `rocsparse_create_handle` +- `rocsparse_datatype` +- `rocsparse_datatype_f64_r` +- `rocsparse_destroy_dnvec_descr` +- `rocsparse_destroy_handle` +- `rocsparse_destroy_spmat_descr` +- `rocsparse_dnvec_descr` +- `rocsparse_handle` +- `rocsparse_index_base` +- `rocsparse_index_base_zero` +- `rocsparse_indextype` +- `rocsparse_indextype_i32` +- `rocsparse_int` +- `rocsparse_operation` +- `rocsparse_operation_none` +- `rocsparse_pointer_mode_host` +- `rocsparse_set_pointer_mode` +- `rocsparse_spitsv` +- `rocsparse_spitsv_alg` +- `rocsparse_spitsv_alg_default` +- `rocsparse_spitsv_stage_buffer_size` +- `rocsparse_spitsv_stage_compute` +- `rocsparse_spitsv_stage_preprocess` +- `rocsparse_spmat_descr` + +### HIP runtime + +- `hipDeviceSynchronize` +- `hipFree` +- `hipMalloc` +- `hipMemcpy` +- `hipMemcpyDeviceToHost` +- `hipMemcpyHostToDevice` diff --git a/Libraries/rocSPARSE/level_2/spitsv/main.cpp b/Libraries/rocSPARSE/level_2/spitsv/main.cpp new file mode 100644 index 000000000..fb0a45da4 --- /dev/null +++ b/Libraries/rocSPARSE/level_2/spitsv/main.cpp @@ -0,0 +1,236 @@ +// MIT License +// +// Copyright (c) 2023-2024 Advanced Micro Devices, Inc. All rights reserved. +// +// Permission is hereby granted, free of charge, to any person obtaining a copy +// of this software and associated documentation files (the "Software"), to deal +// in the Software without restriction, including without limitation the rights +// to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +// copies of the Software, and to permit persons to whom the Software is +// furnished to do so, subject to the following conditions: +// +// The above copyright notice and this permission notice shall be included in all +// copies or substantial portions of the Software. +// +// THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +// IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +// FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +// AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +// LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +// OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE +// SOFTWARE. + +#include "example_utils.hpp" +#include "rocsparse_utils.hpp" + +#include +#include + +#include +#include + +int main() +{ + // 1. Set up input data. + // A' * y = alpha * x + // + // ( 1.0 0.0 0.0 0.0 ) ( 1.0 ) ( 0.4 ) + // ( 2.0 3.0 0.0 0.0 ) * ( 2.0 ) = 2.5 * ( 3.2 ) + // ( 4.0 5.0 6.0 0.0 ) ( 3.0 ) ( 12.8 ) + // ( 7.0 0.0 8.0 9.0 ) ( 4.0 ) ( 26.8 ) + + // Number of rows and columns of the input matrix. + constexpr rocsparse_int m = 4; + constexpr rocsparse_int n = 4; + + // Number of non-zero elements. + constexpr rocsparse_int nnz = 9; + + // CSR values. + constexpr std::array h_csr_val = {1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0, 9.0}; + + // CSR row indices. + constexpr std::array h_csr_row_ptr = {0, 1, 3, 6, 9}; + + // CSR column indices. + constexpr std::array h_csr_col_ind = {0, 0, 1, 0, 1, 2, 0, 2, 3}; + + // Operation applied to the matrix. + constexpr rocsparse_operation trans = rocsparse_operation_none; + + // Scalar alpha. + constexpr double alpha = 2.5; + + // Host vectors for the right hand side and solution of the linear system. + constexpr std::array h_x = {0.4, 3.2, 12.8, 26.8}; + std::array h_y; + + // Index and data type. + constexpr rocsparse_indextype index_type = rocsparse_indextype_i32; + constexpr rocsparse_datatype data_type = rocsparse_datatype_f64_r; + + // Index base. + constexpr rocsparse_index_base index_base = rocsparse_index_base_zero; + + // Use default algorithm. + constexpr rocsparse_spitsv_alg alg = rocsparse_spitsv_alg_default; + + // 2. Allocate device memory and offload input data to device. + double* d_x{}; + double* d_y{}; + double* d_csr_val{}; + rocsparse_int* d_csr_row_ptr{}; + rocsparse_int* d_csr_col_ind{}; + + constexpr size_t val_size = sizeof(*d_csr_val) * nnz; + constexpr size_t row_ptr_size = sizeof(*d_csr_row_ptr) * (m + 1); + constexpr size_t col_ind_size = sizeof(*d_csr_col_ind) * nnz; + constexpr size_t size_vector = sizeof(*d_x) * n; + + HIP_CHECK(hipMalloc(&d_x, size_vector)); + HIP_CHECK(hipMalloc(&d_y, size_vector)); + HIP_CHECK(hipMalloc(&d_csr_row_ptr, row_ptr_size)); + HIP_CHECK(hipMalloc(&d_csr_col_ind, col_ind_size)); + HIP_CHECK(hipMalloc(&d_csr_val, val_size)); + + HIP_CHECK(hipMemcpy(d_x, h_x.data(), size_vector, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(d_csr_row_ptr, h_csr_row_ptr.data(), row_ptr_size, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(d_csr_col_ind, h_csr_col_ind.data(), col_ind_size, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(d_csr_val, h_csr_val.data(), val_size, hipMemcpyHostToDevice)); + + // 3. Initialize rocSPARSE by creating a handle. + rocsparse_handle handle; + ROCSPARSE_CHECK(rocsparse_create_handle(&handle)); + ROCSPARSE_CHECK(rocsparse_set_pointer_mode(handle, rocsparse_pointer_mode_host)); + + // Create sparse matrix and dense vector descriptors. + rocsparse_spmat_descr descr_A{}; + rocsparse_dnvec_descr descr_x{}; + rocsparse_dnvec_descr descr_y{}; + + rocsparse_create_csr_descr(&descr_A, + m, + n, + nnz, + d_csr_row_ptr, + d_csr_col_ind, + d_csr_val, + index_type, + index_type, + index_base, + data_type); + + rocsparse_create_dnvec_descr(&descr_x, n, d_x, data_type); + rocsparse_create_dnvec_descr(&descr_y, n, d_y, data_type); + + // Set maximum iteration number. + constexpr rocsparse_int max_iter = 200; + rocsparse_int iter_counter = max_iter; + + // Set tolerance. + constexpr double tolerance = 1e-4; + + // Set up convergence history. + std::array residual_history; + + // 4. Prepare device for rocSPARSE iterative SpSV invocation. + // Obtain required buffer size in bytes for analysis and solve stages. + size_t buffer_size; + ROCSPARSE_CHECK(rocsparse_spitsv(handle, + &iter_counter, + &tolerance, + residual_history.data(), + trans, + &alpha, + descr_A, + descr_x, + descr_y, + data_type, + alg, + rocsparse_spitsv_stage_buffer_size, + &buffer_size, + nullptr)); + // Synchronize threads. + HIP_CHECK(hipDeviceSynchronize()); + + // 5. Analysis. + // Allocate temporary buffer. + void* temp_buffer{}; + HIP_CHECK(hipMalloc(&temp_buffer, buffer_size)); + + // Perform analysis. + ROCSPARSE_CHECK(rocsparse_spitsv(handle, + &iter_counter, + &tolerance, + residual_history.data(), + trans, + &alpha, + descr_A, + descr_x, + descr_y, + data_type, + alg, + rocsparse_spitsv_stage_preprocess, + nullptr, + temp_buffer)); + + // 6. Perform triangular solve Ay = alpha * x. + ROCSPARSE_CHECK(rocsparse_spitsv(handle, + &iter_counter, + &tolerance, + residual_history.data(), + trans, + &alpha, + descr_A, + descr_x, + descr_y, + data_type, + alg, + rocsparse_spitsv_stage_compute, + &buffer_size, + temp_buffer)); + // Synchronize threads. + HIP_CHECK(hipDeviceSynchronize()); + + // 7. Copy result from device to host. + HIP_CHECK(hipMemcpy(h_y.data(), d_y, sizeof(*d_y) * n, hipMemcpyDeviceToHost)); + + // 8. Free rocSPARSE resources and device memory. + ROCSPARSE_CHECK(rocsparse_destroy_handle(handle)); + ROCSPARSE_CHECK(rocsparse_destroy_spmat_descr(descr_A)); + ROCSPARSE_CHECK(rocsparse_destroy_dnvec_descr(descr_x)); + ROCSPARSE_CHECK(rocsparse_destroy_dnvec_descr(descr_y)); + + HIP_CHECK(hipFree(d_x)); + HIP_CHECK(hipFree(d_y)); + HIP_CHECK(hipFree(d_csr_val)); + HIP_CHECK(hipFree(d_csr_row_ptr)); + HIP_CHECK(hipFree(d_csr_col_ind)); + HIP_CHECK(hipFree(temp_buffer)); + + // 9. Fetch the convergence data. + std::cout << "Iterations performed: " << iter_counter << std::endl; + + // Check last residual computed to confirm that convergence was successful. + const double last_residual = residual_history.back(); + const bool is_converged = last_residual < tolerance; + + int errors{}; + + if(!is_converged) + { + std::cout << "Residual in last iteration " << last_residual + << " is greater or equal than tolerance " << tolerance + << ". Iterative incomplete LU factorization was unsuccessful." << std::endl; + ++errors; + } + else + { + + // 10. Print result: (1, 2, 3, 4). + std::cout << "y = " << format_range(h_y.begin(), h_y.end()) << std::endl; + } + + // 11. Print validation result. + return report_validation_result(errors); +} diff --git a/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2017.sln b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2017.sln new file mode 100644 index 000000000..1718456bd --- /dev/null +++ b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2017.sln @@ -0,0 +1,24 @@ +Microsoft Visual Studio Solution File, Format Version 12.00 +# Visual Studio 15 +VisualStudioVersion = 15.0.33026.149 +MinimumVisualStudioVersion = 10.0.40219.1 +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "spitsv_vs2017", "spitsv_vs2017.vcxproj", "{EFD1A0EC-2699-443C-BC18-8A3ACFEFB807}" +EndProject +Global + GlobalSection(SolutionConfigurationPlatforms) = preSolution + Debug|x64 = Debug|x64 + Release|x64 = Release|x64 + EndGlobalSection + GlobalSection(ProjectConfigurationPlatforms) = postSolution + {EFD1A0EC-2699-443C-BC18-8A3ACFEFB807}.Debug|x64.ActiveCfg = Debug|x64 + {EFD1A0EC-2699-443C-BC18-8A3ACFEFB807}.Debug|x64.Build.0 = Debug|x64 + {EFD1A0EC-2699-443C-BC18-8A3ACFEFB807}.Release|x64.ActiveCfg = Release|x64 + {EFD1A0EC-2699-443C-BC18-8A3ACFEFB807}.Release|x64.Build.0 = Release|x64 + EndGlobalSection + GlobalSection(SolutionProperties) = preSolution + HideSolutionNode = FALSE + EndGlobalSection + GlobalSection(ExtensibilityGlobals) = postSolution + SolutionGuid = {15DF38FA-FD41-4D39-AD2E-1F94776F13F5} + EndGlobalSection +EndGlobal diff --git a/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2017.vcxproj b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2017.vcxproj new file mode 100644 index 000000000..b6b7baaca --- /dev/null +++ b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2017.vcxproj @@ -0,0 +1,108 @@ + + + + + Debug + x64 + + + Release + x64 + + + + 5.7 + 15.0 + {EFD1A0EC-2699-443C-BC18-8A3ACFEFB807} + Win32Proj + spitsv_vs2017 + $(LatestTargetPlatformVersion) + + + + + + + + + + + PreserveNewest + + + + + Application + true + HIP clang 5.7 + Unicode + + + Application + false + HIP clang 5.7 + Unicode + + + + + + + + + + + + + + + + true + + + + true + rocsparse_$(ProjectName) + + + false + rocsparse_$(ProjectName) + + + + true + + + + + Level2 + __clang__;__HIP__;_DEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + true + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + Level2 + __clang__;__HIP__;NDEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2017.vcxproj.filters b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2017.vcxproj.filters new file mode 100644 index 000000000..3552c37fb --- /dev/null +++ b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2017.vcxproj.filters @@ -0,0 +1,30 @@ + + + + + {848d870c-ed0e-4709-aa7e-52ebf3ae41d4} + cpp;c;cc;cxx;c++;def;odl;idl;hpj;bat;asm;asmx;hip;cu + + + {d5a3f85d-9d34-471b-944c-15c810f3a846} + h;hh;hpp;hxx;h++;hm;inl;inc;ipp;xsd;cuh + + + {28d2603a-1cfd-4934-9120-e90d33ac4d0a} + rc;ico;cur;bmp;dlg;rc2;rct;bin;rgs;gif;jpg;jpeg;jpe;resx;tiff;tif;png;wav;mfcribbon-ms + + + + + Source Files + + + + + Header Files + + + Header Files + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2019.sln b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2019.sln new file mode 100644 index 000000000..900a81017 --- /dev/null +++ b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2019.sln @@ -0,0 +1,24 @@ +Microsoft Visual Studio Solution File, Format Version 12.00 +# Visual Studio Version 16 +VisualStudioVersion = 16.0.32630.194 +MinimumVisualStudioVersion = 10.0.40219.1 +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "spitsv_vs2019", "spitsv_vs2019.vcxproj", "{E92723FC-411A-4656-9C0F-88D5D9F01EBD}" +EndProject +Global + GlobalSection(SolutionConfigurationPlatforms) = preSolution + Debug|x64 = Debug|x64 + Release|x64 = Release|x64 + EndGlobalSection + GlobalSection(ProjectConfigurationPlatforms) = postSolution + {E92723FC-411A-4656-9C0F-88D5D9F01EBD}.Debug|x64.ActiveCfg = Debug|x64 + {E92723FC-411A-4656-9C0F-88D5D9F01EBD}.Debug|x64.Build.0 = Debug|x64 + {E92723FC-411A-4656-9C0F-88D5D9F01EBD}.Release|x64.ActiveCfg = Release|x64 + {E92723FC-411A-4656-9C0F-88D5D9F01EBD}.Release|x64.Build.0 = Release|x64 + EndGlobalSection + GlobalSection(SolutionProperties) = preSolution + HideSolutionNode = FALSE + EndGlobalSection + GlobalSection(ExtensibilityGlobals) = postSolution + SolutionGuid = {83525873-59D2-45B6-8E6D-389396D22C29} + EndGlobalSection +EndGlobal diff --git a/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2019.vcxproj b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2019.vcxproj new file mode 100644 index 000000000..6dcd7585c --- /dev/null +++ b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2019.vcxproj @@ -0,0 +1,108 @@ + + + + + Debug + x64 + + + Release + x64 + + + + 5.7 + 16.0 + {E92723FC-411A-4656-9C0F-88D5D9F01EBD} + Win32Proj + spitsv_vs2019 + $(LatestTargetPlatformVersion) + + + + + + + + + + + PreserveNewest + + + + + Application + true + HIP clang 5.7 + Unicode + + + Application + false + HIP clang 5.7 + Unicode + + + + + + + + + + + + + + + + true + + + + true + rocsparse_$(ProjectName) + + + false + rocsparse_$(ProjectName) + + + + true + + + + + Level2 + __clang__;__HIP__;_DEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + true + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + Level2 + __clang__;__HIP__;NDEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2019.vcxproj.filters b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2019.vcxproj.filters new file mode 100644 index 000000000..2dc16b583 --- /dev/null +++ b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2019.vcxproj.filters @@ -0,0 +1,30 @@ + + + + + {1a923f87-f0c1-4a02-b14b-ed2250bf8216} + cpp;c;cc;cxx;c++;def;odl;idl;hpj;bat;asm;asmx;hip;cu + + + {a0f37083-694c-454c-923c-45bee1885f64} + h;hh;hpp;hxx;h++;hm;inl;inc;ipp;xsd;cuh + + + {c7fde6df-93b9-4e40-b218-b8842a2e18a4} + rc;ico;cur;bmp;dlg;rc2;rct;bin;rgs;gif;jpg;jpeg;jpe;resx;tiff;tif;png;wav;mfcribbon-ms + + + + + Source Files + + + + + Header Files + + + Header Files + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2022.sln b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2022.sln new file mode 100644 index 000000000..1eef71cc0 --- /dev/null +++ b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2022.sln @@ -0,0 +1,24 @@ +Microsoft Visual Studio Solution File, Format Version 12.00 +# Visual Studio Version 17 +VisualStudioVersion = 17.4.33213.308 +MinimumVisualStudioVersion = 10.0.40219.1 +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "spitsv_vs2022", "spitsv_vs2022.vcxproj", "{A987BF4A-988D-410A-B3EF-1140AEA10960}" +EndProject +Global + GlobalSection(SolutionConfigurationPlatforms) = preSolution + Debug|x64 = Debug|x64 + Release|x64 = Release|x64 + EndGlobalSection + GlobalSection(ProjectConfigurationPlatforms) = postSolution + {A987BF4A-988D-410A-B3EF-1140AEA10960}.Debug|x64.ActiveCfg = Debug|x64 + {A987BF4A-988D-410A-B3EF-1140AEA10960}.Debug|x64.Build.0 = Debug|x64 + {A987BF4A-988D-410A-B3EF-1140AEA10960}.Release|x64.ActiveCfg = Release|x64 + {A987BF4A-988D-410A-B3EF-1140AEA10960}.Release|x64.Build.0 = Release|x64 + EndGlobalSection + GlobalSection(SolutionProperties) = preSolution + HideSolutionNode = FALSE + EndGlobalSection + GlobalSection(ExtensibilityGlobals) = postSolution + SolutionGuid = {4849E864-88B5-4CF5-8A35-FA0294FAFBE7} + EndGlobalSection +EndGlobal diff --git a/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2022.vcxproj b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2022.vcxproj new file mode 100644 index 000000000..fc37fb541 --- /dev/null +++ b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2022.vcxproj @@ -0,0 +1,108 @@ + + + + + Debug + x64 + + + Release + x64 + + + + 5.7 + 17.0 + {A987BF4A-988D-410A-B3EF-1140AEA10960} + Win32Proj + spitsv_vs2022 + $(LatestTargetPlatformVersion) + + + + + + + + + + + PreserveNewest + + + + + Application + true + HIP clang 5.7 + Unicode + + + Application + false + HIP clang 5.7 + Unicode + + + + + + + + + + + + + + + + true + + + + true + rocsparse_$(ProjectName) + + + false + rocsparse_$(ProjectName) + + + + true + + + + + Level2 + __clang__;__HIP__;_DEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + true + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + Level2 + __clang__;__HIP__;NDEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2022.vcxproj.filters b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2022.vcxproj.filters new file mode 100644 index 000000000..11da7956f --- /dev/null +++ b/Libraries/rocSPARSE/level_2/spitsv/spitsv_vs2022.vcxproj.filters @@ -0,0 +1,30 @@ + + + + + {f2a53d78-a1fb-4958-b8bd-0cf1f9c37abf} + cpp;c;cc;cxx;c++;def;odl;idl;hpj;bat;asm;asmx;hip;cu + + + {3f3c948b-7fc2-4f3e-a026-d1c430b78f6a} + h;hh;hpp;hxx;h++;hm;inl;inc;ipp;xsd;cuh + + + {ea59d625-79a8-46b7-a9d3-45887c7435be} + rc;ico;cur;bmp;dlg;rc2;rct;bin;rgs;gif;jpg;jpeg;jpe;resx;tiff;tif;png;wav;mfcribbon-ms + + + + + Source Files + + + + + Header Files + + + Header Files + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/preconditioner/CMakeLists.txt b/Libraries/rocSPARSE/preconditioner/CMakeLists.txt index 4afd56699..409eae76e 100644 --- a/Libraries/rocSPARSE/preconditioner/CMakeLists.txt +++ b/Libraries/rocSPARSE/preconditioner/CMakeLists.txt @@ -1,6 +1,6 @@ # MIT License # -# Copyright (c) 2023 Advanced Micro Devices, Inc. All rights reserved. +# Copyright (c) 2023-2024 Advanced Micro Devices, Inc. All rights reserved. # # Permission is hereby granted, free of charge, to any person obtaining a copy # of this software and associated documentation files (the "Software"), to deal @@ -28,3 +28,5 @@ add_subdirectory(bsrilu0) add_subdirectory(csric0) add_subdirectory(csrilu0) add_subdirectory(csritilu0) +add_subdirectory(gpsv) +add_subdirectory(gtsv) diff --git a/Libraries/rocSPARSE/preconditioner/Makefile b/Libraries/rocSPARSE/preconditioner/Makefile index 4511ae4e9..acfbad627 100644 --- a/Libraries/rocSPARSE/preconditioner/Makefile +++ b/Libraries/rocSPARSE/preconditioner/Makefile @@ -1,6 +1,6 @@ # MIT License # -# Copyright (c) 2023 Advanced Micro Devices, Inc. All rights reserved. +# Copyright (c) 2023-2024 Advanced Micro Devices, Inc. All rights reserved. # # Permission is hereby granted, free of charge, to any person obtaining a copy # of this software and associated documentation files (the "Software"), to deal @@ -25,7 +25,9 @@ EXAMPLES := \ bsrilu0 \ csric0 \ csrilu0 \ - csritilu0 + csritilu0 \ + gpsv \ + gtsv all: $(EXAMPLES) diff --git a/Libraries/rocSPARSE/preconditioner/csritilu0/main.cpp b/Libraries/rocSPARSE/preconditioner/csritilu0/main.cpp index 124fcdd79..23d7fc1b7 100644 --- a/Libraries/rocSPARSE/preconditioner/csritilu0/main.cpp +++ b/Libraries/rocSPARSE/preconditioner/csritilu0/main.cpp @@ -230,15 +230,15 @@ int main() // Check last residual computed to confirm that convergence was successful. HIP_CHECK(hipMemcpy(data.data(), d_data, size_data, hipMemcpyDeviceToHost)); - const double last_residual = data.back(); - const rocsparse_int csritilu0_converges = last_residual < tol; + const double last_residual = data.back(); + const bool csritilu0_converges = last_residual < tol; int errors{}; if(!csritilu0_converges) { std::cout << "Residual in last iteration " << last_residual - << "is greater or equal than tolerance " << tol + << " is greater or equal than tolerance " << tol << ". Iterative incomplete LU factorization was unsuccessful." << std::endl; errors++; } diff --git a/Libraries/rocSPARSE/preconditioner/gpsv/.gitignore b/Libraries/rocSPARSE/preconditioner/gpsv/.gitignore new file mode 100644 index 000000000..962b2ca0b --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gpsv/.gitignore @@ -0,0 +1 @@ +rocsparse_gpsv diff --git a/Libraries/rocSPARSE/preconditioner/gpsv/CMakeLists.txt b/Libraries/rocSPARSE/preconditioner/gpsv/CMakeLists.txt new file mode 100644 index 000000000..e697c95e2 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gpsv/CMakeLists.txt @@ -0,0 +1,62 @@ +# MIT License +# +# Copyright (c) 2023-2024 Advanced Micro Devices, Inc. All rights reserved. +# +# Permission is hereby granted, free of charge, to any person obtaining a copy +# of this software and associated documentation files (the "Software"), to deal +# in the Software without restriction, including without limitation the rights +# to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +# copies of the Software, and to permit persons to whom the Software is +# furnished to do so, subject to the following conditions: +# +# The above copyright notice and this permission notice shall be included in all +# copies or substantial portions of the Software. +# +# THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +# IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +# FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +# AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +# LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +# OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE +# SOFTWARE. + +set(example_name rocsparse_gpsv) + +cmake_minimum_required(VERSION 3.21 FATAL_ERROR) +project(${example_name} LANGUAGES CXX) + +if(GPU_RUNTIME STREQUAL "CUDA") + message(STATUS "rocSPARSE examples do not support the CUDA runtime") + return() +endif() + +# This example does not contain device code, thereby it can be compiled with any conforming C++ compiler. + +set(CMAKE_CXX_STANDARD 17) +set(CMAKE_CXX_EXTENSIONS OFF) +set(CMAKE_CXX_STANDARD_REQUIRED ON) + +if(WIN32) + set(ROCM_ROOT "$ENV{HIP_PATH}" CACHE PATH "Root directory of the ROCm installation") +else() + set(ROCM_ROOT "/opt/rocm" CACHE PATH "Root directory of the ROCm installation") +endif() + +list(APPEND CMAKE_PREFIX_PATH "${ROCM_ROOT}") + +find_package(rocsparse REQUIRED) + +add_executable(${example_name} main.cpp) +# Make example runnable using ctest +add_test(${example_name} ${example_name}) + +# Link to example library +target_link_libraries(${example_name} PRIVATE roc::rocsparse hip::host) + +target_include_directories(${example_name} PRIVATE "../../../../Common") + +install(TARGETS ${example_name}) + +if(CMAKE_SYSTEM_NAME MATCHES Windows) + install(IMPORTED_RUNTIME_ARTIFACTS roc::rocsparse) +endif() diff --git a/Libraries/rocSPARSE/preconditioner/gpsv/Makefile b/Libraries/rocSPARSE/preconditioner/gpsv/Makefile new file mode 100644 index 000000000..c40f6ca91 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gpsv/Makefile @@ -0,0 +1,58 @@ +# MIT License +# +# Copyright (c) 2023-2024 Advanced Micro Devices, Inc. All rights reserved. +# +# Permission is hereby granted, free of charge, to any person obtaining a copy +# of this software and associated documentation files (the "Software"), to deal +# in the Software without restriction, including without limitation the rights +# to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +# copies of the Software, and to permit persons to whom the Software is +# furnished to do so, subject to the following conditions: +# +# The above copyright notice and this permission notice shall be included in all +# copies or substantial portions of the Software. +# +# THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +# IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +# FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +# AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +# LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +# OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE +# SOFTWARE. + +EXAMPLE := rocsparse_gpsv +COMMON_INCLUDE_DIR := ../../../../Common +GPU_RUNTIME := HIP + +ifneq ($(GPU_RUNTIME), HIP) + $(error GPU_RUNTIME is set to "$(GPU_RUNTIME)". GPU_RUNTIME must be HIP.) +endif + +ROCM_INSTALL_DIR := /opt/rocm + +HIP_INCLUDE_DIR := $(ROCM_INSTALL_DIR)/include +ROCSPARSE_INCLUDE_DIR := $(HIP_INCLUDE_DIR) + +CXX ?= g++ + +# Common variables and flags +CXX_STD := c++17 +ICXXFLAGS := -std=$(CXX_STD) +ICPPFLAGS := -isystem $(ROCSPARSE_INCLUDE_DIR) -isystem $(HIP_INCLUDE_DIR) -I $(COMMON_INCLUDE_DIR) -D__HIP_PLATFORM_AMD__ +ILDFLAGS := -L $(ROCM_INSTALL_DIR)/lib +ILDLIBS := -lrocsparse -lamdhip64 + +CXXFLAGS ?= -Wall -Wextra + +ICXXFLAGS += $(CXXFLAGS) +ICPPFLAGS += $(CPPFLAGS) +ILDFLAGS += $(LDFLAGS) +ILDLIBS += $(LDLIBS) + +$(EXAMPLE): main.cpp $(COMMON_INCLUDE_DIR)/example_utils.hpp $(COMMON_INCLUDE_DIR)/rocsparse_utils.hpp + $(CXX) $(ICXXFLAGS) $(ICPPFLAGS) $(ILDFLAGS) -o $@ $< $(ILDLIBS) + +clean: + $(RM) $(EXAMPLE) + +.PHONY: clean diff --git a/Libraries/rocSPARSE/preconditioner/gpsv/README.md b/Libraries/rocSPARSE/preconditioner/gpsv/README.md new file mode 100644 index 000000000..22232895b --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gpsv/README.md @@ -0,0 +1,78 @@ +# rocSPARSE Preconditioner Pentadiagonal Solver + +## Description + + This example illustrates the use of the `rocSPARSE` pentadiagonal solver with multiple batches: + + $$ A_i \cdot x_i = b_i $$ + + where + +- $s$ is the batch ID. +- $A_i$ is a $m \times n$ pentadiagonal matrix for each batch: + + $$\begin{pmatrix} + d_0 & u_0 & w_0 & 0 & 0 & 0 & 0 & \cdots & 0 \\ + l_1 & d_1 & u_1 & w_1 & 0 & 0 & 0 & \cdots & 0 \\ + s_2 & l_2 & d_2 & u_2 & w_2 & 0 & 0 & \cdots & 0 \\ + 0 & s_3 & l_3 & d_3 & u_3 & w_3 & 0 & \cdots & 0 \\ + \vdots & \ddots & \ddots & \ddots & \ddots & \ddots & \ddots & \cdots & \vdots \\ + 0 & \cdots & 0 & s_{m-2} & l_{m-2} & d_{m-2} & u_{m-2} & w_{m-2} & 0 \\ + 0 & \cdots & 0 & 0 & s_{m-1} & l_{m-1} & d_{m-1} & u_{m-1} & w_{m-1} \\ + 0 & \cdots & 0 & 0 & 0 & s_m & l_m & d_m & u_m + \end{pmatrix}$$ + + - $b_i$ is a dense right hand side vector for each batch. + - $x_i$ is the dense solution vector for each batch. + +### Application flow + + 1. Set up input data. + 2. Allocate device memory and offload input data to the device. + 3. Initialize rocSPARSE by creating a handle. + 4. Obtain the required buffer size. + 5. Call interleaved batched pentadiagonal solver. + 6. Copy result matrix to host. + 7. Free rocSPARSE resources and device memory. + 8. Check convergence. + 9. Print result matrix, check errors. + +## Key APIs and Concepts + + The components of the pentadiagonal system are stored in length $m$ vectors for each batch: + +- second upper diagonal with last two elements 0: $\mathbf{w} = (w_0, w_1, \dots, w_{m-2}, 0, 0)$ +- upper diagonal with last elements 0: $\mathbf{u} = (u_0, u_1, \dots, u_{m-1}, 0)$ +- main diagonal: $\mathbf{d} = (d_0, d_1, \dots, d_{m-1}, d_m)$ +- lower diagonal with first elements 0: $\mathbf{l} = (0, l_1, \dots, l_{m-1}, l_m)$ +- lower diagonal with first two elements 0: $\mathbf{s} = (0, 0, s_2, \dots, s_{m-1}, s_m)$ + +### rocSPARSE + +- rocSPARSE is initialized by calling `rocsparse_create_handle(rocsparse_handle*)` and is terminated by calling `rocsparse_destroy_handle(rocsparse_handle)`. +- `rocsparse_[sdcz]gpsv_interleaved_batch` computes the solution for a pentadiagonal matrix with multiple batches. The correct function signature should be chosen based on the datatype of the input matrix: + - `s` single-precision real (`float`) + - `d` double-precision real (`double`) + - `c` single-precision complex (`rocsparse_float_complex`) + - `z` double-precision complex (`rocsparse_double_complex`) +- `rocsparse_[sdcz]gpsv_interleaved_batch_buffer_size` allows to obtain the size (in bytes) of the temporary storage buffer required for the calculation. The character matched in `[sdcz]` coincides with the one matched in any of the mentioned functions. + +## Demonstrated API Calls + +### rocSPARSE + +- `rocsparse_create_handle` +- `rocsparse_destroy_handle` +- `rocsparse_dgpsv_interleaved_batch` +- `rocsparse_dgpsv_interleaved_batch_buffer_size` +- `rocsparse_handle` +- `rocsparse_int` + +### HIP runtime + +- `hipDeviceSynchronize` +- `hipFree` +- `hipMalloc` +- `hipMemcpy` +- `hipMemcpyDeviceToHost` +- `hipMemcpyHostToDevice` diff --git a/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2017.sln b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2017.sln new file mode 100644 index 000000000..998d43fd7 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2017.sln @@ -0,0 +1,24 @@ +Microsoft Visual Studio Solution File, Format Version 12.00 +# Visual Studio 15 +VisualStudioVersion = 15.0.33026.149 +MinimumVisualStudioVersion = 10.0.40219.1 +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "gpsv_vs2017", "gpsv_vs2017.vcxproj", "{FBD46E48-5689-44EA-817A-BBAA6EB006BD}" +EndProject +Global + GlobalSection(SolutionConfigurationPlatforms) = preSolution + Debug|x64 = Debug|x64 + Release|x64 = Release|x64 + EndGlobalSection + GlobalSection(ProjectConfigurationPlatforms) = postSolution + {FBD46E48-5689-44EA-817A-BBAA6EB006BD}.Debug|x64.ActiveCfg = Debug|x64 + {FBD46E48-5689-44EA-817A-BBAA6EB006BD}.Debug|x64.Build.0 = Debug|x64 + {FBD46E48-5689-44EA-817A-BBAA6EB006BD}.Release|x64.ActiveCfg = Release|x64 + {FBD46E48-5689-44EA-817A-BBAA6EB006BD}.Release|x64.Build.0 = Release|x64 + EndGlobalSection + GlobalSection(SolutionProperties) = preSolution + HideSolutionNode = FALSE + EndGlobalSection + GlobalSection(ExtensibilityGlobals) = postSolution + SolutionGuid = {D98D48D7-82F4-420B-816A-95913EFD3B4E} + EndGlobalSection +EndGlobal diff --git a/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2017.vcxproj b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2017.vcxproj new file mode 100644 index 000000000..0af70380c --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2017.vcxproj @@ -0,0 +1,108 @@ + + + + + Debug + x64 + + + Release + x64 + + + + 5.7 + 15.0 + {FBD46E48-5689-44EA-817A-BBAA6EB006BD} + Win32Proj + gpsv_vs2017 + $(LatestTargetPlatformVersion) + + + + + + + + + + + PreserveNewest + + + + + Application + true + HIP clang 5.7 + Unicode + + + Application + false + HIP clang 5.7 + Unicode + + + + + + + + + + + + + + + + true + + + + true + rocsparse_$(ProjectName) + + + false + rocsparse_$(ProjectName) + + + + true + + + + + Level2 + __clang__;__HIP__;_DEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + true + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + Level2 + __clang__;__HIP__;NDEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2017.vcxproj.filters b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2017.vcxproj.filters new file mode 100644 index 000000000..e017db9c6 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2017.vcxproj.filters @@ -0,0 +1,30 @@ + + + + + {bc04d50c-2d64-4512-a821-09388a647b11} + cpp;c;cc;cxx;c++;def;odl;idl;hpj;bat;asm;asmx;hip;cu + + + {70f1781a-73e8-4686-9aff-1f74dac38810} + h;hh;hpp;hxx;h++;hm;inl;inc;ipp;xsd;cuh + + + {df975475-76ce-4cb8-9c27-2a73c46f6a4f} + rc;ico;cur;bmp;dlg;rc2;rct;bin;rgs;gif;jpg;jpeg;jpe;resx;tiff;tif;png;wav;mfcribbon-ms + + + + + Source Files + + + + + Header Files + + + Header Files + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2019.sln b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2019.sln new file mode 100644 index 000000000..13e37ceb5 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2019.sln @@ -0,0 +1,24 @@ +Microsoft Visual Studio Solution File, Format Version 12.00 +# Visual Studio Version 16 +VisualStudioVersion = 16.0.32630.194 +MinimumVisualStudioVersion = 10.0.40219.1 +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "gpsv_vs2019", "gpsv_vs2019.vcxproj", "{17E97A94-213D-413B-A2EB-0164CEEFDEFC}" +EndProject +Global + GlobalSection(SolutionConfigurationPlatforms) = preSolution + Debug|x64 = Debug|x64 + Release|x64 = Release|x64 + EndGlobalSection + GlobalSection(ProjectConfigurationPlatforms) = postSolution + {17E97A94-213D-413B-A2EB-0164CEEFDEFC}.Debug|x64.ActiveCfg = Debug|x64 + {17E97A94-213D-413B-A2EB-0164CEEFDEFC}.Debug|x64.Build.0 = Debug|x64 + {17E97A94-213D-413B-A2EB-0164CEEFDEFC}.Release|x64.ActiveCfg = Release|x64 + {17E97A94-213D-413B-A2EB-0164CEEFDEFC}.Release|x64.Build.0 = Release|x64 + EndGlobalSection + GlobalSection(SolutionProperties) = preSolution + HideSolutionNode = FALSE + EndGlobalSection + GlobalSection(ExtensibilityGlobals) = postSolution + SolutionGuid = {2A677BE7-CB8D-4920-B0A2-A3188737E380} + EndGlobalSection +EndGlobal diff --git a/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2019.vcxproj b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2019.vcxproj new file mode 100644 index 000000000..47cc24c57 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2019.vcxproj @@ -0,0 +1,108 @@ + + + + + Debug + x64 + + + Release + x64 + + + + 5.7 + 16.0 + {17E97A94-213D-413B-A2EB-0164CEEFDEFC} + Win32Proj + gpsv_vs2019 + $(LatestTargetPlatformVersion) + + + + + + + + + + + PreserveNewest + + + + + Application + true + HIP clang 5.7 + Unicode + + + Application + false + HIP clang 5.7 + Unicode + + + + + + + + + + + + + + + + true + + + + true + rocsparse_$(ProjectName) + + + false + rocsparse_$(ProjectName) + + + + true + + + + + Level2 + __clang__;__HIP__;_DEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + true + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + Level2 + __clang__;__HIP__;NDEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2019.vcxproj.filters b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2019.vcxproj.filters new file mode 100644 index 000000000..d78edafd9 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2019.vcxproj.filters @@ -0,0 +1,30 @@ + + + + + {fbc83606-f80c-4774-8a04-5398d7957434} + cpp;c;cc;cxx;c++;def;odl;idl;hpj;bat;asm;asmx;hip;cu + + + {206c474e-905a-4270-bb8d-bf220f12227a} + h;hh;hpp;hxx;h++;hm;inl;inc;ipp;xsd;cuh + + + {74fd3cb6-86f9-4978-b812-e82ea26029ac} + rc;ico;cur;bmp;dlg;rc2;rct;bin;rgs;gif;jpg;jpeg;jpe;resx;tiff;tif;png;wav;mfcribbon-ms + + + + + Source Files + + + + + Header Files + + + Header Files + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2022.sln b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2022.sln new file mode 100644 index 000000000..45ff0b9eb --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2022.sln @@ -0,0 +1,24 @@ +Microsoft Visual Studio Solution File, Format Version 12.00 +# Visual Studio Version 17 +VisualStudioVersion = 17.4.33213.308 +MinimumVisualStudioVersion = 10.0.40219.1 +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "gpsv_vs2022", "gpsv_vs2022.vcxproj", "{65DD89E3-AB8C-4EAE-B0AB-65FD1B120DC6}" +EndProject +Global + GlobalSection(SolutionConfigurationPlatforms) = preSolution + Debug|x64 = Debug|x64 + Release|x64 = Release|x64 + EndGlobalSection + GlobalSection(ProjectConfigurationPlatforms) = postSolution + {65DD89E3-AB8C-4EAE-B0AB-65FD1B120DC6}.Debug|x64.ActiveCfg = Debug|x64 + {65DD89E3-AB8C-4EAE-B0AB-65FD1B120DC6}.Debug|x64.Build.0 = Debug|x64 + {65DD89E3-AB8C-4EAE-B0AB-65FD1B120DC6}.Release|x64.ActiveCfg = Release|x64 + {65DD89E3-AB8C-4EAE-B0AB-65FD1B120DC6}.Release|x64.Build.0 = Release|x64 + EndGlobalSection + GlobalSection(SolutionProperties) = preSolution + HideSolutionNode = FALSE + EndGlobalSection + GlobalSection(ExtensibilityGlobals) = postSolution + SolutionGuid = {02E7AE71-203B-4108-B087-EAA2B109D976} + EndGlobalSection +EndGlobal diff --git a/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2022.vcxproj b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2022.vcxproj new file mode 100644 index 000000000..8dcf6db86 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2022.vcxproj @@ -0,0 +1,108 @@ + + + + + Debug + x64 + + + Release + x64 + + + + 5.7 + 17.0 + {65DD89E3-AB8C-4EAE-B0AB-65FD1B120DC6} + Win32Proj + gpsv_vs2022 + $(LatestTargetPlatformVersion) + + + + + + + + + + + PreserveNewest + + + + + Application + true + HIP clang 5.7 + Unicode + + + Application + false + HIP clang 5.7 + Unicode + + + + + + + + + + + + + + + + true + + + + true + rocsparse_$(ProjectName) + + + false + rocsparse_$(ProjectName) + + + + true + + + + + Level2 + __clang__;__HIP__;_DEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + true + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + Level2 + __clang__;__HIP__;NDEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2022.vcxproj.filters b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2022.vcxproj.filters new file mode 100644 index 000000000..132585d64 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gpsv/gpsv_vs2022.vcxproj.filters @@ -0,0 +1,30 @@ + + + + + {74c67968-508a-444b-9319-c67fc07646a0} + cpp;c;cc;cxx;c++;def;odl;idl;hpj;bat;asm;asmx;hip;cu + + + {b4aecf27-1e0f-4176-bd9a-1a42b83a4a9a} + h;hh;hpp;hxx;h++;hm;inl;inc;ipp;xsd;cuh + + + {df588610-3f4c-4333-98a0-900f73cedd33} + rc;ico;cur;bmp;dlg;rc2;rct;bin;rgs;gif;jpg;jpeg;jpe;resx;tiff;tif;png;wav;mfcribbon-ms + + + + + Source Files + + + + + Header Files + + + Header Files + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/preconditioner/gpsv/main.cpp b/Libraries/rocSPARSE/preconditioner/gpsv/main.cpp new file mode 100644 index 000000000..2d7e74631 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gpsv/main.cpp @@ -0,0 +1,265 @@ +// MIT License +// +// Copyright (c) 2023-2024 Advanced Micro Devices, Inc. All rights reserved. +// +// Permission is hereby granted, free of charge, to any person obtaining a copy +// of this software and associated documentation files (the "Software"), to deal +// in the Software without restriction, including without limitation the rights +// to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +// copies of the Software, and to permit persons to whom the Software is +// furnished to do so, subject to the following conditions: +// +// The above copyright notice and this permission notice shall be included in all +// copies or substantial portions of the Software. +// +// THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +// IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +// FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +// AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +// LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +// OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE +// SOFTWARE. + +#include "example_utils.hpp" +#include "rocsparse_utils.hpp" + +#include +#include + +#include +#include +#include +#include + +int main() +{ + // 1. Set up input data. + + // Number of unknowns + constexpr rocsparse_int m = 5; + + // Number of batches and stride + constexpr rocsparse_int batch_count = 2; + constexpr rocsparse_int batch_stride = 543; + + constexpr rocsparse_int N_b = m * batch_stride * batch_count; + + std::array h_s; + std::array h_l; + std::array h_d; + std::array h_u; + std::array h_w; + std::array h_b; + + // A_1 * x_1 = b_1 + // ( 6 -2 1 0 0 ) ( 1 ) ( 5 ) + // ( 8 3 4 -3 0 ) ( 2 ) ( 14 ) + // ( 2 0 -2 3 2 ) * ( 3 ) = ( 18 ) + // ( 0 10 7 1 4 ) ( 4 ) ( 65 ) + // ( 0 0 3 -1 5 ) ( 5 ) ( 30 ) + + // Pentadiagonal matrix A_1 + rocsparse_int batch_id = 0; + rocsparse_int batch_index = batch_id; + + h_s[batch_index] = 0; // 0 by definition + h_l[batch_index] = 0; // 0 by definition + h_d[batch_index] = 6; + h_u[batch_index] = -2; + h_w[batch_index] = 1; + h_b[batch_index] = 5; // solution vector + + batch_index += batch_stride; + h_s[batch_index] = 0; // 0 by definition + h_l[batch_index] = 8; + h_d[batch_index] = 3; + h_u[batch_index] = 4; + h_w[batch_index] = -3; + h_b[batch_index] = 14; // solution vector + + batch_index += batch_stride; + h_s[batch_index] = 2; + h_l[batch_index] = 0; + h_d[batch_index] = -2; + h_u[batch_index] = 3; + h_w[batch_index] = 2; + h_b[batch_index] = 18; // solution vector + + batch_index += batch_stride; + h_s[batch_index] = 10; + h_l[batch_index] = 7; + h_d[batch_index] = 1; + h_u[batch_index] = 4; + h_w[batch_index] = 0; // 0 by definition + h_b[batch_index] = 65; // solution vector + + batch_index += batch_stride; + h_s[batch_index] = 3; + h_l[batch_index] = -1; + h_d[batch_index] = 5; + h_u[batch_index] = 0; // 0 by definition + h_w[batch_index] = 0; // 0 by definition + h_b[batch_index] = 30; // solution vector + + // A_2 * x_2 = b_2 + // ( 1 -2 3 0 0 ) ( 3 ) ( 20 ) + // ( -4 5 -6 7 0 ) ( 14 ) ( 612 ) + // ( -8 9 10 -9 8 ) * ( 15 ) = ( -56 ) + // ( 0 -7 6 -5 4 ) ( 92 ) ( -208 ) + // ( 0 0 -3 2 -1 ) ( 65 ) ( 74 ) + + // Pentadiagonal matrix A_2 + + batch_id = 1; + batch_index = batch_id; + + h_s[batch_index] = 0; // 0 by definition + h_l[batch_index] = 0; // 0 by definition + h_d[batch_index] = 1; + h_u[batch_index] = -2; + h_w[batch_index] = 3; + h_b[batch_index] = 20; // solution vector + + batch_index += batch_stride; + h_s[batch_index] = 0; // 0 by definition + h_l[batch_index] = -4; + h_d[batch_index] = 5; + h_u[batch_index] = -6; + h_w[batch_index] = 7; + h_b[batch_index] = 612; // solution vector + + batch_index += batch_stride; + h_s[batch_index] = -8; + h_l[batch_index] = 9; + h_d[batch_index] = 10; + h_u[batch_index] = -9; + h_w[batch_index] = 8; + h_b[batch_index] = -56; // solution vector + + batch_index += batch_stride; + h_s[batch_index] = -7; + h_l[batch_index] = 6; + h_d[batch_index] = -5; + h_u[batch_index] = 4; + h_w[batch_index] = 0; // 0 by definition + h_b[batch_index] = -208; // solution vector + + batch_index += batch_stride; + h_s[batch_index] = -3; + h_l[batch_index] = 2; + h_d[batch_index] = -1; + h_u[batch_index] = 0; // 0 by definition + h_w[batch_index] = 0; // 0 by definition + h_b[batch_index] = 74; // solution vector + + // 2. Allocate device memory and offload input data to the device. + double* d_w{}; + double* d_u{}; + double* d_d{}; + double* d_l{}; + double* d_s{}; + double* d_b{}; + + constexpr size_t d_size = sizeof(*d_d) * N_b; + + HIP_CHECK(hipMalloc(&d_w, d_size)); + HIP_CHECK(hipMalloc(&d_u, d_size)); + HIP_CHECK(hipMalloc(&d_d, d_size)); + HIP_CHECK(hipMalloc(&d_l, d_size)); + HIP_CHECK(hipMalloc(&d_s, d_size)); + HIP_CHECK(hipMalloc(&d_b, d_size)); + + HIP_CHECK(hipMemcpy(d_w, h_w.data(), d_size, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(d_u, h_u.data(), d_size, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(d_d, h_d.data(), d_size, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(d_l, h_l.data(), d_size, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(d_s, h_s.data(), d_size, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(d_b, h_b.data(), d_size, hipMemcpyHostToDevice)); + + // 3. Initialize rocSPARSE by creating a handle. + rocsparse_handle handle; + ROCSPARSE_CHECK(rocsparse_create_handle(&handle)); + + // 4. Obtain the required buffer size. + size_t buffer_size; + ROCSPARSE_CHECK(rocsparse_dgpsv_interleaved_batch_buffer_size(handle, + rocsparse_gpsv_interleaved_alg_qr, + m, + d_s, + d_l, + d_d, + d_u, + d_w, + d_b, + batch_count, + batch_stride, + &buffer_size)); + + // Synchronize threads. + HIP_CHECK(hipDeviceSynchronize()); + + // Allocate temporary buffer. + void* temp_buffer{}; + HIP_CHECK(hipMalloc(&temp_buffer, buffer_size)); + + // 5. Call interleaved batched pentadiagonal solver. + ROCSPARSE_CHECK(rocsparse_dgpsv_interleaved_batch(handle, + rocsparse_gpsv_interleaved_alg_qr, + m, + d_s, + d_l, + d_d, + d_u, + d_w, + d_b, + batch_count, + batch_stride, + temp_buffer)); + + // Synchronize threads. + HIP_CHECK(hipDeviceSynchronize()); + + // 6. Copy result matrix to host. + HIP_CHECK(hipMemcpy(h_b.data(), d_b, d_size, hipMemcpyDeviceToHost)); + + // 7. Free rocSPARSE resources and device memory. + ROCSPARSE_CHECK(rocsparse_destroy_handle(handle)); + HIP_CHECK(hipFree(temp_buffer)); + HIP_CHECK(hipFree(d_w)); + HIP_CHECK(hipFree(d_u)); + HIP_CHECK(hipFree(d_d)); + HIP_CHECK(hipFree(d_l)); + HIP_CHECK(hipFree(d_s)); + HIP_CHECK(hipFree(d_b)); + + // 8. Check convergence. + if(std::isnan(h_b[0])) + { + std::cout << "gpsv does not converge." << std::endl; + } + + // 9. Print result matrix, check errors. + // clang-format off + constexpr std::array expected = { + 1, 2, 3, 4, 5, + 3, 14, 15, 92, 65 + }; + // clang-format on + + constexpr double eps = 1.0e5 * std::numeric_limits::epsilon(); + rocsparse_int errors = 0; + for(rocsparse_int b = 0; b < batch_count; ++b) + { + std::cout << "batch #" << b << ":\t"; + for(rocsparse_int i = 0; i < m; ++i) + { + rocsparse_int j = batch_stride * i + b; + errors += std::fabs(h_b[j] - expected[b * m + i]) > eps; + std::cout << h_b[j] << " "; + } + std::cout << std::endl; + } + + // Print validation result. + return report_validation_result(errors); +} diff --git a/Libraries/rocSPARSE/preconditioner/gtsv/.gitignore b/Libraries/rocSPARSE/preconditioner/gtsv/.gitignore new file mode 100644 index 000000000..91bb8a450 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gtsv/.gitignore @@ -0,0 +1 @@ +rocsparse_gtsv diff --git a/Libraries/rocSPARSE/preconditioner/gtsv/CMakeLists.txt b/Libraries/rocSPARSE/preconditioner/gtsv/CMakeLists.txt new file mode 100644 index 000000000..9cf908e68 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gtsv/CMakeLists.txt @@ -0,0 +1,62 @@ +# MIT License +# +# Copyright (c) 2023-2024 Advanced Micro Devices, Inc. All rights reserved. +# +# Permission is hereby granted, free of charge, to any person obtaining a copy +# of this software and associated documentation files (the "Software"), to deal +# in the Software without restriction, including without limitation the rights +# to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +# copies of the Software, and to permit persons to whom the Software is +# furnished to do so, subject to the following conditions: +# +# The above copyright notice and this permission notice shall be included in all +# copies or substantial portions of the Software. +# +# THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +# IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +# FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +# AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +# LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +# OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE +# SOFTWARE. + +set(example_name rocsparse_gtsv) + +cmake_minimum_required(VERSION 3.21 FATAL_ERROR) +project(${example_name} LANGUAGES CXX) + +if(GPU_RUNTIME STREQUAL "CUDA") + message(STATUS "rocSPARSE examples do not support the CUDA runtime") + return() +endif() + +# This example does not contain device code, thereby it can be compiled with any conforming C++ compiler. + +set(CMAKE_CXX_STANDARD 17) +set(CMAKE_CXX_EXTENSIONS OFF) +set(CMAKE_CXX_STANDARD_REQUIRED ON) + +if(WIN32) + set(ROCM_ROOT "$ENV{HIP_PATH}" CACHE PATH "Root directory of the ROCm installation") +else() + set(ROCM_ROOT "/opt/rocm" CACHE PATH "Root directory of the ROCm installation") +endif() + +list(APPEND CMAKE_PREFIX_PATH "${ROCM_ROOT}") + +find_package(rocsparse REQUIRED) + +add_executable(${example_name} main.cpp) +# Make example runnable using ctest +add_test(${example_name} ${example_name}) + +# Link to example library +target_link_libraries(${example_name} PRIVATE roc::rocsparse hip::host) + +target_include_directories(${example_name} PRIVATE "../../../../Common") + +install(TARGETS ${example_name}) + +if(CMAKE_SYSTEM_NAME MATCHES Windows) + install(IMPORTED_RUNTIME_ARTIFACTS roc::rocsparse) +endif() diff --git a/Libraries/rocSPARSE/preconditioner/gtsv/Makefile b/Libraries/rocSPARSE/preconditioner/gtsv/Makefile new file mode 100644 index 000000000..e3b434649 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gtsv/Makefile @@ -0,0 +1,58 @@ +# MIT License +# +# Copyright (c) 2023-2024 Advanced Micro Devices, Inc. All rights reserved. +# +# Permission is hereby granted, free of charge, to any person obtaining a copy +# of this software and associated documentation files (the "Software"), to deal +# in the Software without restriction, including without limitation the rights +# to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +# copies of the Software, and to permit persons to whom the Software is +# furnished to do so, subject to the following conditions: +# +# The above copyright notice and this permission notice shall be included in all +# copies or substantial portions of the Software. +# +# THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +# IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +# FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +# AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +# LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +# OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE +# SOFTWARE. + +EXAMPLE := rocsparse_gtsv +COMMON_INCLUDE_DIR := ../../../../Common +GPU_RUNTIME := HIP + +ifneq ($(GPU_RUNTIME), HIP) + $(error GPU_RUNTIME is set to "$(GPU_RUNTIME)". GPU_RUNTIME must be HIP.) +endif + +ROCM_INSTALL_DIR := /opt/rocm + +HIP_INCLUDE_DIR := $(ROCM_INSTALL_DIR)/include +ROCSPARSE_INCLUDE_DIR := $(HIP_INCLUDE_DIR) + +CXX ?= g++ + +# Common variables and flags +CXX_STD := c++17 +ICXXFLAGS := -std=$(CXX_STD) +ICPPFLAGS := -isystem $(ROCSPARSE_INCLUDE_DIR) -isystem $(HIP_INCLUDE_DIR) -I $(COMMON_INCLUDE_DIR) -D__HIP_PLATFORM_AMD__ +ILDFLAGS := -L $(ROCM_INSTALL_DIR)/lib +ILDLIBS := -lrocsparse -lamdhip64 + +CXXFLAGS ?= -Wall -Wextra + +ICXXFLAGS += $(CXXFLAGS) +ICPPFLAGS += $(CPPFLAGS) +ILDFLAGS += $(LDFLAGS) +ILDLIBS += $(LDLIBS) + +$(EXAMPLE): main.cpp $(COMMON_INCLUDE_DIR)/example_utils.hpp $(COMMON_INCLUDE_DIR)/rocsparse_utils.hpp + $(CXX) $(ICXXFLAGS) $(ICPPFLAGS) $(ILDFLAGS) -o $@ $< $(ILDLIBS) + +clean: + $(RM) $(EXAMPLE) + +.PHONY: clean diff --git a/Libraries/rocSPARSE/preconditioner/gtsv/README.md b/Libraries/rocSPARSE/preconditioner/gtsv/README.md new file mode 100644 index 000000000..d7e4e7d12 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gtsv/README.md @@ -0,0 +1,74 @@ +# rocSPARSE Preconditioner Tridiagonal Solver + +## Description + +This example illustrates the use of the `rocSPARSE` tridiagonal solver with multiple right hand sides: + +$$ A \cdot X = B $$ + +where + +- $A$ is a $m \times n$ tridiagonal matrix: + +$$\begin{pmatrix} + d_0 & u_0 & 0 & 0 & 0 & \cdots & 0 \\ + l_1 & d_1 & u_1 & 0 & 0 & \cdots & 0 \\ + 0 & l_2 & d_2 & u_2 & 0 & \cdots & 0 \\ + \vdots & \ddots & \ddots & \ddots & \ddots & \ddots & \vdots \\ + 0 & \cdots & 0 & l_{m-2} & d_{m-2} & u_{m-2} & 0 \\ + 0 & \cdots & 0 & 0 & l_{m-1} & d_{m-1} & u_{m-1} \\ + 0 & \cdots & 0 & 0 & 0 & l_m & d_m +\end{pmatrix}$$ + +- $B$ is dense right hand side matrix. +- $X$ is the dense solution matrix. + +### Application flow + +1. Set up input data. +2. Allocate device memory and offload input data to the device. +3. Initialize rocSPARSE by creating a handle. +4. Obtain the required buffer size. +5. Call dgtsv tridiagonal solver. +6. Copy result matrix to host. +7. Free rocSPARSE resources and device memory. +8. Check convergence. +9. Print result matrix, check errors. + +## Key APIs and Concepts + +The components of the tridiagonal system are stored in length $m$ vectors: + +- upper diagonal with last elements 0: $\mathbf{u} = (u_0, u_1, \dots, u_{m-1}, 0)$ +- main diagonal: $\mathbf{d} = (d_0, d_1, \dots, d_{m-1}, d_m)$ +- lower diagonal with first elements 0: $\mathbf{l} = (0, l_1, \dots, l_{m-1}, l_m)$ + +### rocSPARSE + +- rocSPARSE is initialized by calling `rocsparse_create_handle(rocsparse_handle*)` and is terminated by calling `rocsparse_destroy_handle(rocsparse_handle)`. +- `rocsparse_[sdcz]gtsv` computes the solution for a tridiagonal matrix with multiple right hand side. The correct function signature should be chosen based on the datatype of the input matrix: + - `s` single-precision real (`float`) + - `d` double-precision real (`double`) + - `c` single-precision complex (`rocsparse_float_complex`) + - `z` double-precision complex (`rocsparse_double_complex`) +- `rocsparse_[sdcz]gtsv_buffer_size` allows to obtain the size (in bytes) of the temporary storage buffer required for the calculation. The character matched in `[sdcz]` coincides with the one matched in any of the mentioned functions. + +## Demonstrated API Calls + +### rocSPARSE + +- `rocsparse_create_handle` +- `rocsparse_destroy_handle` +- `rocsparse_dgtsv` +- `rocsparse_dgtsv_buffer_size` +- `rocsparse_handle` +- `rocsparse_int` + +### HIP runtime + +- `hipDeviceSynchronize` +- `hipFree` +- `hipMalloc` +- `hipMemcpy` +- `hipMemcpyDeviceToHost` +- `hipMemcpyHostToDevice` diff --git a/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2017.sln b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2017.sln new file mode 100644 index 000000000..e216e0e55 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2017.sln @@ -0,0 +1,24 @@ +Microsoft Visual Studio Solution File, Format Version 12.00 +# Visual Studio 15 +VisualStudioVersion = 15.0.33026.149 +MinimumVisualStudioVersion = 10.0.40219.1 +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "gtsv_vs2017", "gtsv_vs2017.vcxproj", "{93C05FE6-788A-424B-9713-2B55AFA0C361}" +EndProject +Global + GlobalSection(SolutionConfigurationPlatforms) = preSolution + Debug|x64 = Debug|x64 + Release|x64 = Release|x64 + EndGlobalSection + GlobalSection(ProjectConfigurationPlatforms) = postSolution + {93C05FE6-788A-424B-9713-2B55AFA0C361}.Debug|x64.ActiveCfg = Debug|x64 + {93C05FE6-788A-424B-9713-2B55AFA0C361}.Debug|x64.Build.0 = Debug|x64 + {93C05FE6-788A-424B-9713-2B55AFA0C361}.Release|x64.ActiveCfg = Release|x64 + {93C05FE6-788A-424B-9713-2B55AFA0C361}.Release|x64.Build.0 = Release|x64 + EndGlobalSection + GlobalSection(SolutionProperties) = preSolution + HideSolutionNode = FALSE + EndGlobalSection + GlobalSection(ExtensibilityGlobals) = postSolution + SolutionGuid = {F272D9E9-9A95-4410-A685-C44B817810CD} + EndGlobalSection +EndGlobal diff --git a/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2017.vcxproj b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2017.vcxproj new file mode 100644 index 000000000..329dc38d8 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2017.vcxproj @@ -0,0 +1,108 @@ + + + + + Debug + x64 + + + Release + x64 + + + + 5.7 + 15.0 + {93C05FE6-788A-424B-9713-2B55AFA0C361} + Win32Proj + gtsv_vs2017 + $(LatestTargetPlatformVersion) + + + + + + + + + + + PreserveNewest + + + + + Application + true + HIP clang 5.7 + Unicode + + + Application + false + HIP clang 5.7 + Unicode + + + + + + + + + + + + + + + + true + + + + true + rocsparse_$(ProjectName) + + + false + rocsparse_$(ProjectName) + + + + true + + + + + Level2 + __clang__;__HIP__;_DEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + true + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + Level2 + __clang__;__HIP__;NDEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2017.vcxproj.filters b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2017.vcxproj.filters new file mode 100644 index 000000000..f4eec9d88 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2017.vcxproj.filters @@ -0,0 +1,30 @@ + + + + + {dba58508-be21-48cb-83b1-cb92a573d8ba} + cpp;c;cc;cxx;c++;def;odl;idl;hpj;bat;asm;asmx;hip;cu + + + {666a5706-41b3-414a-90dd-4a987f7a1654} + h;hh;hpp;hxx;h++;hm;inl;inc;ipp;xsd;cuh + + + {bece2999-8b52-4e33-b94b-0ac40742c8fb} + rc;ico;cur;bmp;dlg;rc2;rct;bin;rgs;gif;jpg;jpeg;jpe;resx;tiff;tif;png;wav;mfcribbon-ms + + + + + Source Files + + + + + Header Files + + + Header Files + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2019.sln b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2019.sln new file mode 100644 index 000000000..3a77b5ba8 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2019.sln @@ -0,0 +1,24 @@ +Microsoft Visual Studio Solution File, Format Version 12.00 +# Visual Studio Version 16 +VisualStudioVersion = 16.0.32630.194 +MinimumVisualStudioVersion = 10.0.40219.1 +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "gtsv_vs2019", "gtsv_vs2019.vcxproj", "{65A27F2B-C070-4FDD-93EC-0DFEA35C2E53}" +EndProject +Global + GlobalSection(SolutionConfigurationPlatforms) = preSolution + Debug|x64 = Debug|x64 + Release|x64 = Release|x64 + EndGlobalSection + GlobalSection(ProjectConfigurationPlatforms) = postSolution + {65A27F2B-C070-4FDD-93EC-0DFEA35C2E53}.Debug|x64.ActiveCfg = Debug|x64 + {65A27F2B-C070-4FDD-93EC-0DFEA35C2E53}.Debug|x64.Build.0 = Debug|x64 + {65A27F2B-C070-4FDD-93EC-0DFEA35C2E53}.Release|x64.ActiveCfg = Release|x64 + {65A27F2B-C070-4FDD-93EC-0DFEA35C2E53}.Release|x64.Build.0 = Release|x64 + EndGlobalSection + GlobalSection(SolutionProperties) = preSolution + HideSolutionNode = FALSE + EndGlobalSection + GlobalSection(ExtensibilityGlobals) = postSolution + SolutionGuid = {262D2511-B1E0-49DD-BD31-63D31ABC65AD} + EndGlobalSection +EndGlobal diff --git a/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2019.vcxproj b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2019.vcxproj new file mode 100644 index 000000000..fe8f09384 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2019.vcxproj @@ -0,0 +1,108 @@ + + + + + Debug + x64 + + + Release + x64 + + + + 5.7 + 16.0 + {65A27F2B-C070-4FDD-93EC-0DFEA35C2E53} + Win32Proj + gtsv_vs2019 + $(LatestTargetPlatformVersion) + + + + + + + + + + + PreserveNewest + + + + + Application + true + HIP clang 5.7 + Unicode + + + Application + false + HIP clang 5.7 + Unicode + + + + + + + + + + + + + + + + true + + + + true + rocsparse_$(ProjectName) + + + false + rocsparse_$(ProjectName) + + + + true + + + + + Level2 + __clang__;__HIP__;_DEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + true + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + Level2 + __clang__;__HIP__;NDEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2019.vcxproj.filters b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2019.vcxproj.filters new file mode 100644 index 000000000..504a69fcb --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2019.vcxproj.filters @@ -0,0 +1,30 @@ + + + + + {8d92b371-c985-41e4-8d49-dd61eb5a0941} + cpp;c;cc;cxx;c++;def;odl;idl;hpj;bat;asm;asmx;hip;cu + + + {7750b0ee-7e0b-4d44-a844-a2be8d79acec} + h;hh;hpp;hxx;h++;hm;inl;inc;ipp;xsd;cuh + + + {54d81fd6-b668-46d6-b41f-7f36a21957fa} + rc;ico;cur;bmp;dlg;rc2;rct;bin;rgs;gif;jpg;jpeg;jpe;resx;tiff;tif;png;wav;mfcribbon-ms + + + + + Source Files + + + + + Header Files + + + Header Files + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2022.sln b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2022.sln new file mode 100644 index 000000000..cde888472 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2022.sln @@ -0,0 +1,24 @@ +Microsoft Visual Studio Solution File, Format Version 12.00 +# Visual Studio Version 17 +VisualStudioVersion = 17.4.33213.308 +MinimumVisualStudioVersion = 10.0.40219.1 +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "gtsv_vs2022", "gtsv_vs2022.vcxproj", "{3F19C0A7-CD2E-438C-88B1-9AF2F8B63B8E}" +EndProject +Global + GlobalSection(SolutionConfigurationPlatforms) = preSolution + Debug|x64 = Debug|x64 + Release|x64 = Release|x64 + EndGlobalSection + GlobalSection(ProjectConfigurationPlatforms) = postSolution + {3F19C0A7-CD2E-438C-88B1-9AF2F8B63B8E}.Debug|x64.ActiveCfg = Debug|x64 + {3F19C0A7-CD2E-438C-88B1-9AF2F8B63B8E}.Debug|x64.Build.0 = Debug|x64 + {3F19C0A7-CD2E-438C-88B1-9AF2F8B63B8E}.Release|x64.ActiveCfg = Release|x64 + {3F19C0A7-CD2E-438C-88B1-9AF2F8B63B8E}.Release|x64.Build.0 = Release|x64 + EndGlobalSection + GlobalSection(SolutionProperties) = preSolution + HideSolutionNode = FALSE + EndGlobalSection + GlobalSection(ExtensibilityGlobals) = postSolution + SolutionGuid = {4A1F593F-4BC3-4C81-B8AA-466ED1B83168} + EndGlobalSection +EndGlobal diff --git a/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2022.vcxproj b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2022.vcxproj new file mode 100644 index 000000000..f51a72499 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2022.vcxproj @@ -0,0 +1,108 @@ + + + + + Debug + x64 + + + Release + x64 + + + + 5.7 + 17.0 + {3F19C0A7-CD2E-438C-88B1-9AF2F8B63B8E} + Win32Proj + gtsv_vs2022 + $(LatestTargetPlatformVersion) + + + + + + + + + + + PreserveNewest + + + + + Application + true + HIP clang 5.7 + Unicode + + + Application + false + HIP clang 5.7 + Unicode + + + + + + + + + + + + + + + + true + + + + true + rocsparse_$(ProjectName) + + + false + rocsparse_$(ProjectName) + + + + true + + + + + Level2 + __clang__;__HIP__;_DEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + true + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + Level2 + __clang__;__HIP__;NDEBUG;_CONSOLE;%(PreprocessorDefinitions) + $(MSBuildProjectDirectory)\..\..\..\..\Common;%(AdditionalIncludeDirectories) + stdcpp17 + true + + + Console + rocsparse.lib;kernel32.lib;user32.lib;gdi32.lib;winspool.lib;comdlg32.lib;advapi32.lib;shell32.lib;ole32.lib;oleaut32.lib;uuid.lib;odbc32.lib;odbccp32.lib;%(AdditionalDependencies) + + + + + + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2022.vcxproj.filters b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2022.vcxproj.filters new file mode 100644 index 000000000..f9cbe1631 --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gtsv/gtsv_vs2022.vcxproj.filters @@ -0,0 +1,30 @@ + + + + + {27fd04f5-fe17-4a0c-bde4-60b5acb7a1d4} + cpp;c;cc;cxx;c++;def;odl;idl;hpj;bat;asm;asmx;hip;cu + + + {6c0f158b-a877-4613-a5dc-d0d0d7c224f7} + h;hh;hpp;hxx;h++;hm;inl;inc;ipp;xsd;cuh + + + {cce389de-98f3-488d-ae85-62dc8835a244} + rc;ico;cur;bmp;dlg;rc2;rct;bin;rgs;gif;jpg;jpeg;jpe;resx;tiff;tif;png;wav;mfcribbon-ms + + + + + Source Files + + + + + Header Files + + + Header Files + + + \ No newline at end of file diff --git a/Libraries/rocSPARSE/preconditioner/gtsv/main.cpp b/Libraries/rocSPARSE/preconditioner/gtsv/main.cpp new file mode 100644 index 000000000..17fcd692e --- /dev/null +++ b/Libraries/rocSPARSE/preconditioner/gtsv/main.cpp @@ -0,0 +1,150 @@ +// MIT License +// +// Copyright (c) 2023-2024 Advanced Micro Devices, Inc. All rights reserved. +// +// Permission is hereby granted, free of charge, to any person obtaining a copy +// of this software and associated documentation files (the "Software"), to deal +// in the Software without restriction, including without limitation the rights +// to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +// copies of the Software, and to permit persons to whom the Software is +// furnished to do so, subject to the following conditions: +// +// The above copyright notice and this permission notice shall be included in all +// copies or substantial portions of the Software. +// +// THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +// IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +// FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +// AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +// LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +// OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE +// SOFTWARE. + +#include "example_utils.hpp" +#include "rocsparse_utils.hpp" + +#include +#include + +#include +#include +#include +#include + +int main() +{ + // 1. Set up input data. + // A * X = B + // ( 1 -5 0 0 ) ( 1 2 0 ) ( 1 7 -2.5 ) + // ( 5 -2 -3 0 ) ( 0 -1 0.5 ) ( -7 3 8 ) + // ( 0 3 1 4 ) * ( 4 3 -3 ) = ( -4 0 6.5 ) + // ( 0 0 1 2 ) ( -2 0 2 ) ( 0 3 1 ) + + // Number of rows and columns of the input matrix. + constexpr rocsparse_int m = 4; + constexpr rocsparse_int n = 3; + constexpr rocsparse_int ldb = m; + + // A matrix + std::array h_u = {-5, -3, 4, 0}; + std::array h_d = {1, -2, 1, 2}; + std::array h_l = {0, 5, 3, 1}; + + // B matrix + constexpr size_t N_B = ldb * n; + + // clang-format off + std::array h_B = { + 1, -7, -4, 0, + 7, 3, 0, 3, + -2.5, 8, 6.5, 1 + }; + // clang-format on + + // 2. Allocate device memory and offload input data to the device. + double* d_u{}; + double* d_d{}; + double* d_l{}; + double* d_B{}; + + constexpr size_t d_size = sizeof(*d_d) * m; + constexpr size_t B_size = sizeof(*d_B) * N_B; + + HIP_CHECK(hipMalloc(&d_u, d_size)); + HIP_CHECK(hipMalloc(&d_d, d_size)); + HIP_CHECK(hipMalloc(&d_l, d_size)); + HIP_CHECK(hipMalloc(&d_B, B_size)); + + HIP_CHECK(hipMemcpy(d_u, h_u.data(), d_size, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(d_d, h_d.data(), d_size, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(d_l, h_l.data(), d_size, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy(d_B, h_B.data(), B_size, hipMemcpyHostToDevice)); + + // 3. Initialize rocSPARSE by creating a handle. + rocsparse_handle handle; + ROCSPARSE_CHECK(rocsparse_create_handle(&handle)); + + // 4. Obtain the required buffer size. + size_t buffer_size; + ROCSPARSE_CHECK( + rocsparse_dgtsv_buffer_size(handle, m, n, d_l, d_d, d_u, d_B, ldb, &buffer_size)); + + // Synchronize threads. + HIP_CHECK(hipDeviceSynchronize()); + + // Allocate temporary buffer. + void* temp_buffer{}; + HIP_CHECK(hipMalloc(&temp_buffer, buffer_size)); + + // 5. Call GTSV tridiagonal solver. + ROCSPARSE_CHECK(rocsparse_dgtsv(handle, m, n, d_l, d_d, d_u, d_B, ldb, temp_buffer)); + + // Synchronize threads. + HIP_CHECK(hipDeviceSynchronize()); + + // 6. Copy result matrix to host. + HIP_CHECK(hipMemcpy(h_B.data(), d_B, B_size, hipMemcpyDeviceToHost)); + + // 7. Free rocSPARSE resources and device memory. + ROCSPARSE_CHECK(rocsparse_destroy_handle(handle)); + HIP_CHECK(hipFree(temp_buffer)); + HIP_CHECK(hipFree(d_u)); + HIP_CHECK(hipFree(d_d)); + HIP_CHECK(hipFree(d_l)); + HIP_CHECK(hipFree(d_B)); + + // 8. Check convergence. + if(std::isnan(h_B[0])) + { + std::cout << "GTSV does not converge." << std::endl; + } + + // 9. Print result matrix, check errors. + // clang-format off + constexpr std::array expected = { + 1, 0, 4, -2, + 2, -1, 3, 0, + 0, 0.5, -3, 2 + }; + // clang-format on + + constexpr double eps = 1.0e5 * std::numeric_limits::epsilon(); + int errors = 0; + + std::cout << "X = " << std::endl; + for(int i = 0; i < m; ++i) + { + std::cout << " ("; + for(int j = 0; j < n; ++j) + { + int idx = i + j * m; + std::printf("%7.2lf", h_B[idx]); + + errors += std::fabs(h_B[idx] - expected[idx]) > eps; + } + std::cout << ")" << std::endl; + } + + // Print validation result. + return report_validation_result(errors); +} diff --git a/README.md b/README.md index d1e61e1b6..56a30d83a 100644 --- a/README.md +++ b/README.md @@ -95,11 +95,13 @@ A collection of examples to enable new users to start using ROCm. Advanced users - [bsrxmv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/bsrxmv/): Showcases a masked sparse matrix-vector multiplication using BSR storage format. - [bsrsv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/bsrsv/): Showcases how to solve a linear system of equations whose coefficients are stored in a BSR sparse triangular matrix. - [coomv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/coomv/): Showcases a sparse matrix-vector multiplication using COO storage format. + - [csritsv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/csritsv/): Showcases how find an iterative solution with the Jacobi method for a linear system of equations whose coefficients are stored in a CSR sparse triangular matrix. - [csrmv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/csrmv/): Showcases a sparse matrix-vector multiplication using CSR storage format. - [csrsv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/csrsv/): Showcases how to solve a linear system of equations whose coefficients are stored in a CSR sparse triangular matrix. - [ellmv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/ellmv/): Showcases a sparse matrix-vector multiplication using ELL storage format. - [gebsrmv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/gebsrmv/): Showcases a sparse matrix-dense vector multiplication using GEBSR storage format. - [gemvi](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/gemvi/): Showcases a dense matrix-sparse vector multiplication. + - [spitsv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/spitsv/): Showcases how to solve iteratively a linear system of equations whose coefficients are stored in a CSR sparse triangular matrix. - [spmv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/spmv/): Showcases a general sparse matrix-dense vector multiplication. - [spsv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/spsv/): Showcases how to solve a linear system of equations whose coefficients are stored in a sparse triangular matrix. - [level_3](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_3/): Operations between sparse and dense matrices. @@ -118,6 +120,8 @@ A collection of examples to enable new users to start using ROCm. Advanced users - [csric0](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/preconditioner/csric0/): Shows how to compute the incomplete Cholesky decomposition of a Hermitian positive-definite sparse CSR matrix. - [csrilu0](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/preconditioner/csrilu0/): Showcases how to obtain the incomplete LU decomposition of a sparse CSR square matrix. - [csritilu0](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/preconditioner/csritilu0/): Showcases how to obtain iteratively the incomplete LU decomposition of a sparse CSR square matrix. + - [gpsv](https://github.com/amd/rocm-examples/tree/develop/Libraries/rocSPARSE/preconditioner/gpsv/): Shows how to compute the solution of pentadiagonal linear system. + - [gtsv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/preconditioner/gtsv/): Shows how to compute the solution of a tridiagonal linear system. - [rocThrust](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocThrust/) - [device_ptr](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocThrust/device_ptr/): Simple program that showcases the usage of the `thrust::device_ptr` template. - [norm](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocThrust/norm/): An example that computes the Euclidean norm of a `thrust::device_vector`. @@ -125,6 +129,89 @@ A collection of examples to enable new users to start using ROCm. Advanced users - [remove_points](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocThrust/remove_points/): Simple program that demonstrates the usage of the `thrust` random number generation, host vector, generation, tuple, zip iterator, and conditional removal templates. It generates a number of random points in a unit square and then removes all of them outside the unit circle. - [saxpy](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocThrust/saxpy/): Simple program that implements the SAXPY operation (`y[i] = a * x[i] + y[i]`) using rocThrust and showcases the usage of the vector and functor templates and of `thrust::fill` and `thrust::transform` operations. - [vectors](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocThrust/vectors/): Simple program that showcases the `host_vector` and the `device_vector` of rocThrust. + - [hipBLAS](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipBLAS/) + - [gemm_strided_batched](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipBLAS/gemm_strided_batched/): Showcases the general matrix product operation with strided and batched matrices. + - [her](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipBLAS/her/): Showcases a rank-2 update of a Hermitian matrix with complex values. + - [scal](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipBLAS/scal/): Simple program that showcases vector scaling (SCAL) operation. + - [hipCUB](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipCUB/) + - [device_radix_sort](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipCUB/device_radix_sort/): Simple program that showcases `hipcub::DeviceRadixSort::SortPairs`. + - [device_sum](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipCUB/device_sum/): Simple program that showcases `hipcub::DeviceReduce::Sum`. + - [hipSOLVER](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipSOLVER/) + - [gels](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipSOLVER/gels/): Solve a linear system of the form $A\times X=B$. + - [geqrf](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipSOLVER/geqrf/): Program that showcases how to obtain a QR decomposition with the hipSOLVER API. + - [gesvd](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipSOLVER/gesvd/): Program that showcases how to obtain a singular value decomposition with the hipSOLVER API. + - [getrf](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipSOLVER/getrf): Program that showcases how to perform a LU factorization with hipSOLVER. + - [potrf](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipSOLVER/potrf/): Perform Cholesky factorization and solve linear system with result. + - [syevd](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipSOLVER/syevd/): Program that showcases how to calculate the eigenvalues of a matrix using a divide-and-conquer algorithm in hipSOLVER. + - [syevdx](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipSOLVER/syevdx/): Shows how to compute a subset of the eigenvalues and the corresponding eigenvectors of a real symmetric matrix A using the Compatibility API of hipSOLVER. + - [sygvd](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipSOLVER/sygvd/): Showcases how to obtain a solution $(X, \Lambda)$ for a generalized symmetric-definite eigenvalue problem of the form $A \cdot X = B\cdot X \cdot \Lambda$. + - [syevj](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipSOLVER/syevj): Calculates the eigenvalues and eigenvectors from a real symmetric matrix using the Jacobi method. + - [syevj_batched](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipSOLVER/syevj_batched): Showcases how to compute the eigenvalues and eigenvectors (via Jacobi method) of each matrix in a batch of real symmetric matrices. + - [sygvj](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/hipSOLVER/sygvj): Calculates the generalized eigenvalues and eigenvectors from a pair of real symmetric matrices using the Jacobi method. + - [rocBLAS](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocBLAS/) + - [level_1](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocBLAS/level_1/): Operations between vectors and vectors. + - [axpy](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocBLAS/level_1/axpy/): Simple program that showcases the AXPY operation. + - [dot](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocBLAS/level_1/dot/): Simple program that showcases dot product. + - [nrm2](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocBLAS/level_1/nrm2/): Simple program that showcases Euclidean norm of a vector. + - [scal](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocBLAS/level_1/scal/): Simple program that showcases vector scaling (SCAL) operation. + - [swap](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocBLAS/level_1/swap/): Showcases exchanging elements between two vectors. + - [level_2](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocBLAS/level_2/): Operations between vectors and matrices. + - [her](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocBLAS/level_2/her/): Showcases a rank-1 update of a Hermitian matrix with complex values. + - [gemv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocBLAS/level_2/gemv/): Showcases the general matrix-vector product operation. + - [level_3](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocBLAS/level_3/): Operations between matrices and matrices. + - [gemm](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocBLAS/level_3/gemm/): Showcases the general matrix product operation. + - [gemm_strided_batched](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocBLAS/level_3/gemm_strided_batched/): Showcases the general matrix product operation with strided and batched matrices. + - [rocPRIM](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocPRIM/) + - [block_sum](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocPRIM/block_sum/): Simple program that showcases `rocprim::block_reduce` with an addition operator. + - [device_sum](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocPRIM/device_sum/): Simple program that showcases `rocprim::reduce` with an addition operator. + - [rocRAND](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocRAND/) + - [simple_distributions_cpp](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocRAND/simple_distributions_cpp/): A command-line app to compare random number generation on the CPU and on the GPU with rocRAND. + - [rocSOLVER](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSOLVER/) + - [getf2](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSOLVER/getf2): Program that showcases how to perform a LU factorization with rocSOLVER. + - [getri](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSOLVER/getri): Program that showcases matrix inversion by LU-decomposition using rocSOLVER. + - [syev](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSOLVER/syev): Shows how to compute the eigenvalues and eigenvectors from a symmetrical real matrix. + - [syev_batched](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSOLVER/syev_batched): Shows how to compute the eigenvalues and eigenvectors for each matrix in a batch of real symmetric matrices. + - [syev_strided_batched](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSOLVER/syev_strided_batched): Shows how to compute the eigenvalues and eigenvectors for multiple symmetrical real matrices, that are stored with an arbitrary stride. + - [rocSPARSE](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/) + - [level_2](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/): Operations between sparse matrices and dense vectors. + - [bsrmv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/bsrmv/): Showcases a sparse matrix-vector multiplication using BSR storage format. + - [bsrxmv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/bsrxmv/): Showcases a masked sparse matrix-vector multiplication using BSR storage format. + - [bsrsv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/bsrsv/): Showcases how to solve a linear system of equations whose coefficients are stored in a BSR sparse triangular matrix. + - [coomv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/coomv/): Showcases a sparse matrix-vector multiplication using COO storage format. + - [csritsv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/csritsv/): Showcases how find an iterative solution with the Jacobi method for a linear system of equations whose coefficients are stored in a CSR sparse triangular matrix. + - [csrmv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/csrmv/): Showcases a sparse matrix-vector multiplication using CSR storage format. + - [csrsv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/csrsv/): Showcases how to solve a linear system of equations whose coefficients are stored in a CSR sparse triangular matrix. + - [ellmv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/ellmv/): Showcases a sparse matrix-vector multiplication using ELL storage format. + - [gebsrmv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/gebsrmv/): Showcases a sparse matrix-dense vector multiplication using GEBSR storage format. + - [gemvi](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/gemvi/): Showcases a dense matrix-sparse vector multiplication. + - [spitsv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/spitsv/): Showcases how to solve iteratively a linear system of equations whose coefficients are stored in a CSR sparse triangular matrix. + - [spmv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/spmv/): Showcases a general sparse matrix-dense vector multiplication. + - [spsv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_2/spsv/): Showcases how to solve a linear system of equations whose coefficients are stored in a sparse triangular matrix. + - [level_3](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_3/): Operations between sparse and dense matrices. + - [bsrmm](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_3/bsrmm/): Showcases a sparse matrix-matrix multiplication using BSR storage format. + - [bsrsm](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_3/bsrsm): Showcases how to solve a linear system of equations whose coefficients are stored in a BSR sparse triangular matrix, with solution and right-hand side stored in dense matrices. + - [csrmm](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_3/csrmm/): Showcases a sparse matrix-matrix multiplication using CSR storage format. + - [csrsm](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_3/csrsm): Showcases how to solve a linear system of equations whose coefficients are stored in a CSR sparse triangular matrix, with solution and right-hand side stored in dense matrices. + - [gebsrmm](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_3/gebsrmm/): Showcases a sparse matrix-matrix multiplication using GEBSR storage format. + - [gemmi](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_3/gemmi/): Showcases a dense matrix sparse matrix multiplication using CSR storage format. + - [sddmm](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_3/sddmm/): Showcases a sampled dense-dense matrix multiplication using CSR storage format. + - [spmm](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_3/spmm/): Showcases a sparse matrix-dense matrix multiplication. + - [spsm](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/level_3/spsm/): Showcases a sparse triangular linear system solver using CSR storage format. + - [preconditioner](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/preconditioner/): Manipulations on sparse matrices to obtain sparse preconditioner matrices. + - [bsric0](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/preconditioner/bsric0/): Shows how to compute the incomplete Cholesky decomposition of a Hermitian positive-definite sparse BSR matrix. + - [bsrilu0](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/preconditioner/bsrilu0/): Showcases how to obtain the incomplete LU decomposition of a sparse BSR square matrix. + - [csric0](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/preconditioner/csric0/): Shows how to compute the incomplete Cholesky decomposition of a Hermitian positive-definite sparse CSR matrix. + - [csrilu0](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/preconditioner/csrilu0/): Showcases how to obtain the incomplete LU decomposition of a sparse CSR square matrix. + - [csritilu0](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/preconditioner/csritilu0/): Showcases how to obtain iteratively the incomplete LU decomposition of a sparse CSR square matrix. + - [gpsv](https://github.com/amd/rocm-examples/tree/develop/Libraries/rocSPARSE/preconditioner/gpsv/): Shows how to compute the solution of pentadiagonal linear system. + - [gtsv](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocSPARSE/preconditioner/gtsv/): Shows how to compute the solution of a tridiagonal linear system. + - [rocThrust](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocThrust/) + - [device_ptr](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocThrust/device_ptr/): Simple program that showcases the usage of the `thrust::device_ptr` template. + - [norm](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocThrust/norm/): An example that computes the Euclidean norm of a `thrust::device_vector`. + - [reduce_sum](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocThrust/reduce_sum/): An example that computes the sum of a `thrust::device_vector` integer vector using the `thrust::reduce()` generalized summation and the `thrust::plus` operator. + - [remove_points](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocThrust/remove_points/): Simple program that demonstrates the usage of the `thrust` random number generation, host vector, generation, tuple, zip iterator, and conditional removal templates. It generates a number of random points in a unit square and then removes all of them outside the unit circle. + - [saxpy](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocThrust/saxpy/): Simple program that implements the SAXPY operation (`y[i] = a * x[i] + y[i]`) using rocThrust and showcases the usage of the vector and functor templates and of `thrust::fill` and `thrust::transform` operations. + - [vectors](https://github.com/ROCm/rocm-examples/tree/develop/Libraries/rocThrust/vectors/): Simple program that showcases the `host_vector` and the `device_vector` of rocThrust. ## Prerequisites diff --git a/ROCm-Examples-VS2017.sln b/ROCm-Examples-VS2017.sln index b4e253f31..d35251809 100644 --- a/ROCm-Examples-VS2017.sln +++ b/ROCm-Examples-VS2017.sln @@ -230,6 +230,14 @@ Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "spmm_vs2017", "Libraries\ro EndProject Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "monte_carlo_pi_vs2017", "Applications\monte_carlo_pi\monte_carlo_pi_vs2017.vcxproj", "{2ACD5660-DAA6-490B-8C5D-F0B178A80D16}" EndProject +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "gtsv_vs2017", "Libraries\rocSPARSE\preconditioner\gtsv\gtsv_vs2017.vcxproj", "{93C05FE6-788A-424B-9713-2B55AFA0C361}" +EndProject +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "csritsv_vs2017", "Libraries\rocSPARSE\level_2\csritsv\csritsv_vs2017.vcxproj", "{F0AF1DEB-4B07-4FDC-8566-FB53F60D10B7}" +EndProject +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "spitsv_vs2017", "Libraries\rocSPARSE\level_2\spitsv\spitsv_vs2017.vcxproj", "{EFD1A0EC-2699-443C-BC18-8A3ACFEFB807}" +EndProject +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "gpsv_vs2017", "Libraries\rocSPARSE\preconditioner\gpsv\gpsv_vs2017.vcxproj", "{FBD46E48-5689-44EA-817A-BBAA6EB006BD}" +EndProject Global GlobalSection(SolutionConfigurationPlatforms) = preSolution Debug|x64 = Debug|x64 @@ -620,6 +628,22 @@ Global {2ACD5660-DAA6-490B-8C5D-F0B178A80D16}.Debug|x64.Build.0 = Debug|x64 {2ACD5660-DAA6-490B-8C5D-F0B178A80D16}.Release|x64.ActiveCfg = Release|x64 {2ACD5660-DAA6-490B-8C5D-F0B178A80D16}.Release|x64.Build.0 = Release|x64 + {93C05FE6-788A-424B-9713-2B55AFA0C361}.Debug|x64.ActiveCfg = Debug|x64 + {93C05FE6-788A-424B-9713-2B55AFA0C361}.Debug|x64.Build.0 = Debug|x64 + {93C05FE6-788A-424B-9713-2B55AFA0C361}.Release|x64.ActiveCfg = Release|x64 + {93C05FE6-788A-424B-9713-2B55AFA0C361}.Release|x64.Build.0 = Release|x64 + {F0AF1DEB-4B07-4FDC-8566-FB53F60D10B7}.Debug|x64.ActiveCfg = Debug|x64 + {F0AF1DEB-4B07-4FDC-8566-FB53F60D10B7}.Debug|x64.Build.0 = Debug|x64 + {F0AF1DEB-4B07-4FDC-8566-FB53F60D10B7}.Release|x64.ActiveCfg = Release|x64 + {F0AF1DEB-4B07-4FDC-8566-FB53F60D10B7}.Release|x64.Build.0 = Release|x64 + {EFD1A0EC-2699-443C-BC18-8A3ACFEFB807}.Debug|x64.ActiveCfg = Debug|x64 + {EFD1A0EC-2699-443C-BC18-8A3ACFEFB807}.Debug|x64.Build.0 = Debug|x64 + {EFD1A0EC-2699-443C-BC18-8A3ACFEFB807}.Release|x64.ActiveCfg = Release|x64 + {EFD1A0EC-2699-443C-BC18-8A3ACFEFB807}.Release|x64.Build.0 = Release|x64 + {FBD46E48-5689-44EA-817A-BBAA6EB006BD}.Debug|x64.ActiveCfg = Debug|x64 + {FBD46E48-5689-44EA-817A-BBAA6EB006BD}.Debug|x64.Build.0 = Debug|x64 + {FBD46E48-5689-44EA-817A-BBAA6EB006BD}.Release|x64.ActiveCfg = Release|x64 + {FBD46E48-5689-44EA-817A-BBAA6EB006BD}.Release|x64.Build.0 = Release|x64 EndGlobalSection GlobalSection(SolutionProperties) = preSolution HideSolutionNode = FALSE @@ -736,6 +760,10 @@ Global {7830AAFE-B001-40B5-BBF4-99EE8AAC519A} = {4581A6EF-211D-4B00-A65E-C29F55CEE886} {DA4B2E3F-E114-49B2-91F6-02061F6AEF1A} = {79082CA5-3D7F-41AC-862B-E16EE6EB25A0} {2ACD5660-DAA6-490B-8C5D-F0B178A80D16} = {0328C27A-BB25-46F6-89F7-4EEF7AC225D8} + {93C05FE6-788A-424B-9713-2B55AFA0C361} = {2586BC68-9BEF-4AC4-9096-353D503EABA6} + {F0AF1DEB-4B07-4FDC-8566-FB53F60D10B7} = {4581A6EF-211D-4B00-A65E-C29F55CEE886} + {EFD1A0EC-2699-443C-BC18-8A3ACFEFB807} = {4581A6EF-211D-4B00-A65E-C29F55CEE886} + {FBD46E48-5689-44EA-817A-BBAA6EB006BD} = {2586BC68-9BEF-4AC4-9096-353D503EABA6} EndGlobalSection GlobalSection(ExtensibilityGlobals) = postSolution SolutionGuid = {5C96FD63-6F26-4E6F-B6D0-7FB9E1833081} diff --git a/ROCm-Examples-VS2019.sln b/ROCm-Examples-VS2019.sln index 21d7f9f43..2550a744c 100644 --- a/ROCm-Examples-VS2019.sln +++ b/ROCm-Examples-VS2019.sln @@ -230,6 +230,14 @@ Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "spmm_vs2019", "Libraries\ro EndProject Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "monte_carlo_pi_vs2019", "Applications\monte_carlo_pi\monte_carlo_pi_vs2019.vcxproj", "{6E278B7D-E928-4151-8613-08E91FC6D4D5}" EndProject +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "gtsv_vs2019", "Libraries\rocSPARSE\preconditioner\gtsv\gtsv_vs2019.vcxproj", "{65A27F2B-C070-4FDD-93EC-0DFEA35C2E53}" +EndProject +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "csritsv_vs2019", "Libraries\rocSPARSE\level_2\csritsv\csritsv_vs2019.vcxproj", "{99A25D0A-93FE-47F2-8223-7313E53E7951}" +EndProject +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "spitsv_vs2019", "Libraries\rocSPARSE\level_2\spitsv\spitsv_vs2019.vcxproj", "{E92723FC-411A-4656-9C0F-88D5D9F01EBD}" +EndProject +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "gpsv_vs2019", "Libraries\rocSPARSE\preconditioner\gpsv\gpsv_vs2019.vcxproj", "{17E97A94-213D-413B-A2EB-0164CEEFDEFC}" +EndProject Global GlobalSection(SolutionConfigurationPlatforms) = preSolution Debug|x64 = Debug|x64 @@ -620,6 +628,22 @@ Global {6E278B7D-E928-4151-8613-08E91FC6D4D5}.Debug|x64.Build.0 = Debug|x64 {6E278B7D-E928-4151-8613-08E91FC6D4D5}.Release|x64.ActiveCfg = Release|x64 {6E278B7D-E928-4151-8613-08E91FC6D4D5}.Release|x64.Build.0 = Release|x64 + {65A27F2B-C070-4FDD-93EC-0DFEA35C2E53}.Debug|x64.ActiveCfg = Debug|x64 + {65A27F2B-C070-4FDD-93EC-0DFEA35C2E53}.Debug|x64.Build.0 = Debug|x64 + {65A27F2B-C070-4FDD-93EC-0DFEA35C2E53}.Release|x64.ActiveCfg = Release|x64 + {65A27F2B-C070-4FDD-93EC-0DFEA35C2E53}.Release|x64.Build.0 = Release|x64 + {99A25D0A-93FE-47F2-8223-7313E53E7951}.Debug|x64.ActiveCfg = Debug|x64 + {99A25D0A-93FE-47F2-8223-7313E53E7951}.Debug|x64.Build.0 = Debug|x64 + {99A25D0A-93FE-47F2-8223-7313E53E7951}.Release|x64.ActiveCfg = Release|x64 + {99A25D0A-93FE-47F2-8223-7313E53E7951}.Release|x64.Build.0 = Release|x64 + {E92723FC-411A-4656-9C0F-88D5D9F01EBD}.Debug|x64.ActiveCfg = Debug|x64 + {E92723FC-411A-4656-9C0F-88D5D9F01EBD}.Debug|x64.Build.0 = Debug|x64 + {E92723FC-411A-4656-9C0F-88D5D9F01EBD}.Release|x64.ActiveCfg = Release|x64 + {E92723FC-411A-4656-9C0F-88D5D9F01EBD}.Release|x64.Build.0 = Release|x64 + {17E97A94-213D-413B-A2EB-0164CEEFDEFC}.Debug|x64.ActiveCfg = Debug|x64 + {17E97A94-213D-413B-A2EB-0164CEEFDEFC}.Debug|x64.Build.0 = Debug|x64 + {17E97A94-213D-413B-A2EB-0164CEEFDEFC}.Release|x64.ActiveCfg = Release|x64 + {17E97A94-213D-413B-A2EB-0164CEEFDEFC}.Release|x64.Build.0 = Release|x64 EndGlobalSection GlobalSection(SolutionProperties) = preSolution HideSolutionNode = FALSE @@ -736,6 +760,10 @@ Global {0F437FDF-5F2B-4028-A816-FC1A2ACA51B1} = {F0B0FD83-2B22-47F8-92B1-7A5ED88B8B5E} {EC8FA476-A120-469B-BB48-DA4E0B3E50AD} = {06DEE87C-F773-49A8-A856-8CB55BDFED6D} {6E278B7D-E928-4151-8613-08E91FC6D4D5} = {9254BAD9-FDFC-4645-B2C8-EEB42F1F069D} + {65A27F2B-C070-4FDD-93EC-0DFEA35C2E53} = {8B7AD0F4-4288-4ACF-9980-3C500A00EF31} + {99A25D0A-93FE-47F2-8223-7313E53E7951} = {F0B0FD83-2B22-47F8-92B1-7A5ED88B8B5E} + {E92723FC-411A-4656-9C0F-88D5D9F01EBD} = {F0B0FD83-2B22-47F8-92B1-7A5ED88B8B5E} + {17E97A94-213D-413B-A2EB-0164CEEFDEFC} = {8B7AD0F4-4288-4ACF-9980-3C500A00EF31} EndGlobalSection GlobalSection(ExtensibilityGlobals) = postSolution SolutionGuid = {90580497-38BF-428E-A951-6EC6CFC68193} diff --git a/ROCm-Examples-VS2022.sln b/ROCm-Examples-VS2022.sln index bb61a7a1a..a4605714e 100644 --- a/ROCm-Examples-VS2022.sln +++ b/ROCm-Examples-VS2022.sln @@ -230,6 +230,14 @@ Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "spmm_vs2022", "Libraries\ro EndProject Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "monte_carlo_pi_vs2022", "Applications\monte_carlo_pi\monte_carlo_pi_vs2022.vcxproj", "{107AC26F-A20D-4B25-81DE-AFCDCDECBFFC}" EndProject +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "gtsv_vs2022", "Libraries\rocSPARSE\preconditioner\gtsv\gtsv_vs2022.vcxproj", "{3F19C0A7-CD2E-438C-88B1-9AF2F8B63B8E}" +EndProject +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "csritsv_vs2022", "Libraries\rocSPARSE\level_2\csritsv\csritsv_vs2022.vcxproj", "{DC1DF216-BC97-4797-8EA7-8DDCC38DFDCF}" +EndProject +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "spitsv_vs2022", "Libraries\rocSPARSE\level_2\spitsv\spitsv_vs2022.vcxproj", "{A987BF4A-988D-410A-B3EF-1140AEA10960}" +EndProject +Project("{8BC9CEB8-8B4A-11D0-8D11-00A0C91BC942}") = "gpsv_vs2022", "Libraries\rocSPARSE\preconditioner\gpsv\gpsv_vs2022.vcxproj", "{65DD89E3-AB8C-4EAE-B0AB-65FD1B120DC6}" +EndProject Global GlobalSection(SolutionConfigurationPlatforms) = preSolution Debug|x64 = Debug|x64 @@ -620,6 +628,22 @@ Global {107AC26F-A20D-4B25-81DE-AFCDCDECBFFC}.Debug|x64.Build.0 = Debug|x64 {107AC26F-A20D-4B25-81DE-AFCDCDECBFFC}.Release|x64.ActiveCfg = Release|x64 {107AC26F-A20D-4B25-81DE-AFCDCDECBFFC}.Release|x64.Build.0 = Release|x64 + {3F19C0A7-CD2E-438C-88B1-9AF2F8B63B8E}.Debug|x64.ActiveCfg = Debug|x64 + {3F19C0A7-CD2E-438C-88B1-9AF2F8B63B8E}.Debug|x64.Build.0 = Debug|x64 + {3F19C0A7-CD2E-438C-88B1-9AF2F8B63B8E}.Release|x64.ActiveCfg = Release|x64 + {3F19C0A7-CD2E-438C-88B1-9AF2F8B63B8E}.Release|x64.Build.0 = Release|x64 + {DC1DF216-BC97-4797-8EA7-8DDCC38DFDCF}.Debug|x64.ActiveCfg = Debug|x64 + {DC1DF216-BC97-4797-8EA7-8DDCC38DFDCF}.Debug|x64.Build.0 = Debug|x64 + {DC1DF216-BC97-4797-8EA7-8DDCC38DFDCF}.Release|x64.ActiveCfg = Release|x64 + {DC1DF216-BC97-4797-8EA7-8DDCC38DFDCF}.Release|x64.Build.0 = Release|x64 + {A987BF4A-988D-410A-B3EF-1140AEA10960}.Debug|x64.ActiveCfg = Debug|x64 + {A987BF4A-988D-410A-B3EF-1140AEA10960}.Debug|x64.Build.0 = Debug|x64 + {A987BF4A-988D-410A-B3EF-1140AEA10960}.Release|x64.ActiveCfg = Release|x64 + {A987BF4A-988D-410A-B3EF-1140AEA10960}.Release|x64.Build.0 = Release|x64 + {65DD89E3-AB8C-4EAE-B0AB-65FD1B120DC6}.Debug|x64.ActiveCfg = Debug|x64 + {65DD89E3-AB8C-4EAE-B0AB-65FD1B120DC6}.Debug|x64.Build.0 = Debug|x64 + {65DD89E3-AB8C-4EAE-B0AB-65FD1B120DC6}.Release|x64.ActiveCfg = Release|x64 + {65DD89E3-AB8C-4EAE-B0AB-65FD1B120DC6}.Release|x64.Build.0 = Release|x64 EndGlobalSection GlobalSection(SolutionProperties) = preSolution HideSolutionNode = FALSE @@ -735,6 +759,10 @@ Global {D32D396C-4B52-4AAC-AC5A-21CC99207E32} = {F91F4254-0ADD-4955-BDFE-53CB4EDBF601} {A6919683-9E28-400A-8910-1BB207B29C4D} = {7EDDB5A2-7601-435F-AEDB-30EBC68D19C9} {107AC26F-A20D-4B25-81DE-AFCDCDECBFFC} = {C735FFA9-12E1-4BEF-87B2-8891A3006505} + {3F19C0A7-CD2E-438C-88B1-9AF2F8B63B8E} = {0AFB7E3F-4173-4F47-A068-17CAB93DA563} + {DC1DF216-BC97-4797-8EA7-8DDCC38DFDCF} = {F91F4254-0ADD-4955-BDFE-53CB4EDBF601} + {A987BF4A-988D-410A-B3EF-1140AEA10960} = {F91F4254-0ADD-4955-BDFE-53CB4EDBF601} + {65DD89E3-AB8C-4EAE-B0AB-65FD1B120DC6} = {0AFB7E3F-4173-4F47-A068-17CAB93DA563} EndGlobalSection GlobalSection(ExtensibilityGlobals) = postSolution SolutionGuid = {D648FD37-D8CB-4EA5-8445-38BEF36F6736}