Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

DPC++ executor #648

Merged
merged 8 commits into from
Nov 4, 2020
Merged

DPC++ executor #648

merged 8 commits into from
Nov 4, 2020

Conversation

tcojean
Copy link
Member

@tcojean tcojean commented Oct 14, 2020

This is a first version of a DPC++ executor.

Everything is working under Intel, and tests are passing. Here are some
major changes:

  • Upgrade GTEST to a more recent version which allows using GTEST_SKIP()
    to skip selected tests.
  • Add a new Dpcpp executor. Use it in all gko::Operation and
    wherever appropriate.
  • Add the general architecture with all kernels set as GKO_NOT_IMPLEMENTED
  • Cope with unsupported copies between devices by going through the
    master executors. This (should) allow to copy from an AMD GPU to a
    NVIDIA GPU, or from Intel to AMD/NVIDIA (untested).
  • Implement a few simple kernels in the components and stop/criterion
    to ensure kernel execution works properly.
  • Implement DPC++ specific executor tests.
  • Circumvent some issues with abs and sqrt by using the sycl
    specific functions when appropriate. Note that it's still needed to
    namespace abs (and probably sqrt as well) calls inside DPC++
    kernels, otherwise the default (std) version is still used.

There are still a few things I want to do:

  • Add the DPC++ executor to all examples/benchmarks/documentation.
  • Generate documentation and check that it is complete.
  • Check that copying through master works (e.g. for CUDA and HIP).
  • Make Windows MSVC work with the new gtest.
  • Try to add an Intel CPU (hopefully GPU) CI job.

@tcojean tcojean added is:new-feature A request or implementation of a feature that does not exist yet. reg:build This is related to the build system. mod:core This is related to the core module. 1:ST:WIP This PR is a work in progress. Not ready for review. mod:dpcpp This is related to the DPC++ module. labels Oct 14, 2020
@tcojean tcojean self-assigned this Oct 14, 2020
@tcojean tcojean force-pushed the dpcpp_executor branch 2 times, most recently from 321b069 to db9fbf1 Compare October 14, 2020 16:25
@tcojean tcojean added the plat:intel This is related to the Intel compilers. label Oct 14, 2020
@tcojean tcojean force-pushed the dpcpp_executor branch 2 times, most recently from df03dd9 to cc532f2 Compare October 14, 2020 16:48
CMakeLists.txt Outdated Show resolved Hide resolved
@tcojean tcojean force-pushed the dpcpp_executor branch 2 times, most recently from 6fd8391 to 66ca1bf Compare October 15, 2020 09:27
@codecov
Copy link

codecov bot commented Oct 15, 2020

Codecov Report

Merging #648 into develop will decrease coverage by 0.08%.
The diff coverage is 64.28%.

Impacted file tree graph

@@             Coverage Diff             @@
##           develop     #648      +/-   ##
===========================================
- Coverage    92.92%   92.83%   -0.09%     
===========================================
  Files          320      322       +2     
  Lines        22833    22904      +71     
===========================================
+ Hits         21217    21263      +46     
- Misses        1616     1641      +25     
Impacted Files Coverage Δ
core/base/executor.cpp 15.38% <0.00%> (-2.80%) ⬇️
core/device_hooks/cuda_hooks.cpp 51.72% <0.00%> (-3.84%) ⬇️
core/device_hooks/hip_hooks.cpp 51.72% <0.00%> (-3.84%) ⬇️
core/test/base/array.cpp 100.00% <ø> (ø)
core/test/base/combination.cpp 86.36% <ø> (ø)
core/test/base/composition.cpp 77.77% <ø> (ø)
core/test/base/iterator_factory.cpp 100.00% <ø> (ø)
core/test/base/mtx_io.cpp 97.45% <ø> (ø)
core/test/factorization/par_ilu.cpp 92.85% <ø> (ø)
core/test/log/convergence.cpp 100.00% <ø> (ø)
... and 86 more

Continue to review full report at Codecov.

Legend - Click here to learn more
Δ = absolute <relative> (impact), ø = not affected, ? = missing data
Powered by Codecov. Last update 9caff82...a5cc5d2. Read the comment docs.

@tcojean tcojean force-pushed the dpcpp_executor branch 2 times, most recently from e8873b9 to 2325ee3 Compare October 15, 2020 16:18
@tcojean
Copy link
Member Author

tcojean commented Oct 15, 2020

A few comments:

@tcojean tcojean force-pushed the dpcpp_executor branch 6 times, most recently from 541ae8b to 66caf9e Compare October 17, 2020 14:14
@tcojean tcojean added 1:ST:ready-for-review This PR is ready for review and removed 1:ST:WIP This PR is a work in progress. Not ready for review. labels Oct 19, 2020
Copy link
Member

@yhmtsai yhmtsai left a comment

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

an issue is found when compiling it

dpcpp/factorization/factorization_kernels.dp.cpp Outdated Show resolved Hide resolved
@tcojean tcojean force-pushed the dpcpp_executor branch 3 times, most recently from 66eeab2 to 9a522b1 Compare October 20, 2020 18:09
Copy link
Member

@pratikvn pratikvn left a comment

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

LGTM!

cmake/create_test.cmake Show resolved Hide resolved
cuda/CMakeLists.txt Show resolved Hide resolved
doc/examples/examples.hpp.in Outdated Show resolved Hide resolved
include/ginkgo/core/base/executor.hpp Outdated Show resolved Hide resolved
Copy link
Member

@thoasm thoasm left a comment

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

LGTM!

CMakeLists.txt Outdated Show resolved Hide resolved
core/device_hooks/dpcpp_hooks.cpp Show resolved Hide resolved
core/device_hooks/dpcpp_hooks.cpp Outdated Show resolved Hide resolved
core/test/base/executor.cpp Outdated Show resolved Hide resolved
core/test/base/executor.cpp Outdated Show resolved Hide resolved
Copy link
Member

@yhmtsai yhmtsai left a comment

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

LGTM in general.
minor format and something would like to give a try

Comment on lines +55 to +60
{"accelerator", sycl::info::device_type::accelerator},
{"all", sycl::info::device_type::all},
{"cpu", sycl::info::device_type::cpu},
{"gpu", sycl::info::device_type::gpu}};
Copy link
Member

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

maybe use enum to declare them?

Copy link
Member Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I would rather keep using strings particularly since I can also be case insensitive (which I now added, actually).

dpcpp/base/executor.dp.cpp Outdated Show resolved Hide resolved
}
max_workgroup_size_ =
device.get_info<sycl::info::device::max_work_group_size>();
auto *queue = new sycl::queue{device, sycl::property::queue::in_order{}};
Copy link
Member

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Could you write some comments about in_order here?

dpcpp/base/executor.dp.cpp Show resolved Hide resolved
dpcpp/components/prefix_sum.dp.cpp Outdated Show resolved Hide resolved
@@ -151,15 +165,16 @@ class ExecutorBase;
* OMP
* CUDA(0)
* HIP(0)
* DPC++(0)
Copy link
Member

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

should we use DPC++ or DPCPP?

Copy link
Member Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Well that's a good question, so far I use in documentation always DPC++ I believe, and only during the code I use DPCPP since we cannot use the + symbol where we want. With the code of this example, it will produce DPC++(0) indeed.

@@ -908,7 +913,11 @@ GKO_INLINE GKO_ATTRIBUTES constexpr xstd::enable_if_t<is_complex_s<T>::value,
remove_complex<T>>
abs(const T &x)
{
#ifdef CL_SYCL_LANGUAGE_VERSION
Copy link
Member

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Could you add some todo here? like DPCPP currently needs to use OPENCL to support complex multiplication.

set(GINKGO_HAS_HIP OFF)
find_package(OpenMP)
include(CheckLanguage)
check_language(CUDA)
try_compile(GKO_CAN_COMPILE_DPCPP ${PROJECT_BINARY_DIR}/dpcpp
SOURCES ${PROJECT_SOURCE_DIR}/dpcpp/test_dpcpp.dp.cpp
CXX_STANDARD 17)
Copy link
Member

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

it does not use dpct header, so it might be compiled with std14.

Copy link
Member Author

@tcojean tcojean Oct 26, 2020

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I think this should also be the case for DPC++ itself. See https://spec.oneapi.com/versions/latest/elements/dpcpp/source/index.html

${PROJECT_BINARY_DIR} ${CMAKE_CURRENT_BINARY_DIR})
string(REPLACE "/" "_" TEST_TARGET_NAME "${REL_BINARY_DIR}/${test_name}")
add_executable(${TEST_TARGET_NAME} ${test_name}.dp.cpp)
target_compile_features("${TEST_TARGET_NAME}" PUBLIC cxx_std_17)
Copy link
Member

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

In the dpct beta update 10, they describe something related to std17 and std::max_element.
Maybe dpct does not require std17 anymore.
Because we do not use dpct header currently, could you try std14?

Note. I saw the DPCT beta update 10 release note yesterday, but I can not find it anymore.

Copy link
Member

@upsj upsj left a comment

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

LGTM!

core/devices/dpcpp/CMakeLists.txt Show resolved Hide resolved
dpcpp/components/csr_spgeam.dp.hpp Outdated Show resolved Hide resolved
dpcpp/components/fill_array.dp.cpp Outdated Show resolved Hide resolved
dpcpp/components/format_conversion.dp.hpp Outdated Show resolved Hide resolved
dpcpp/components/precision_conversion.dp.cpp Outdated Show resolved Hide resolved
dpcpp/matrix/csr_kernels.dp.cpp Outdated Show resolved Hide resolved
dpcpp/stop/criterion_kernels.dp.cpp Outdated Show resolved Hide resolved
dpcpp/components/absolute_array.dp.cpp Outdated Show resolved Hide resolved
dpcpp/components/absolute_array.dp.cpp Outdated Show resolved Hide resolved
dpcpp/base/executor.dp.cpp Show resolved Hide resolved
@tcojean tcojean force-pushed the dpcpp_executor branch 5 times, most recently from 74a7a31 to ea97fb7 Compare October 28, 2020 15:47
@tcojean tcojean added 1:ST:ready-to-merge This PR is ready to merge. and removed 1:ST:ready-for-review This PR is ready for review labels Oct 29, 2020
tcojean and others added 8 commits November 4, 2020 10:55
+ `TYPED_TEST_CASE` is deprecated for the newer `TYPED_TEST_SUITE`
  version.
+ We can now use `GTEST_SKIP()` to skip specific tests.
+ Add a new `Dpcpp` executor. Use it in all `gko::Operation` and
  wherever appropriate.
+ Add the general architecture with all kernels set as `GKO_NOT_IMPLEMENTED`
+ Cope with unsupported copies between devices by going through the
  master executors. This (should) allow to copy from an AMD GPU to a
  NVIDIA GPU, or from Intel to AMD/NVIDIA (untested).
+ Implement a few simple kernels in the components and `stop/criterion`
  to ensure kernel execution works properly.
+ Implement DPC++ specific executor tests.
+ Circumvent some issues with `abs` and `sqrt` my using the sycl
  specific functions when appropriate. Note that it's still needed to
  namespace `abs` (and probably `sqrt` as well) calls inside DPC++
  kernels, otherwise the default (std) version is still used.
`using T = x;` is banned inside any gtest class used in a `TYPED_TEST`
fashion because gtest also defines a `T` of its own and somehow `MSVC`
confuses the two.
+ Fix some documentation issues.
+ Fix some spacing issues.
+ Do not duplicate `GKO_DECLARE_EXECUTOR_FRIEND` macro.

Co-authored-by: Pratik Nayak <[email protected]>
Co-authored-by: Thomas Grützmacher <[email protected]>
Co-authored-by: Yuhsiang M. Tsai <[email protected]>
+ Add some environment variable output to `get_info.cmake`.
+ Fix formatting of `GKO_NOT_IMPLEMENTED` declarations.
+ Allow lowercase in the `DPC++` strings passed to executors.
+ Fix kernels storing SYCL `id` into `int` types.
+ Add some extra comments to tricky parts of the code.
+ Improve some other code formatting issues.

Co-authored-by: Tobias Ribizel <[email protected]>
Co-authored-by: Yuhsiang M. Tsai <[email protected]>
@sonarcloud
Copy link

sonarcloud bot commented Nov 4, 2020

Kudos, SonarCloud Quality Gate passed!

Bug A 0 Bugs
Vulnerability A 0 Vulnerabilities (and Security Hotspot 0 Security Hotspots to review)
Code Smell A 67 Code Smells

17.5% 17.5% Coverage
1.4% 1.4% Duplication

warning The version of Java (1.8.0_121) you have used to run this analysis is deprecated and we will stop accepting it accepting it soon.Please update to at least Java 11.
Read more here

@tcojean tcojean merged commit 99c2004 into develop Nov 4, 2020
@tcojean tcojean deleted the dpcpp_executor branch November 4, 2020 16:46
tcojean added a commit that referenced this pull request Aug 20, 2021
Ginkgo release 1.4.0

The Ginkgo team is proud to announce the new Ginkgo minor release 1.4.0. This
release brings most of the Ginkgo functionality to the Intel DPC++ ecosystem
which enables Intel-GPU and CPU execution. The only Ginkgo features which have
not been ported yet are some preconditioners.

Ginkgo's mixed-precision support is greatly enhanced thanks to:
1. The new Accessor concept, which allows writing kernels featuring on-the-fly
memory compression, among other features. The accessor can be used as
header-only, see the [accessor BLAS benchmarks repository](https://github.com/ginkgo-project/accessor-BLAS/tree/develop) as a usage example.
2. All LinOps now transparently support mixed-precision execution. By default,
this is done through a temporary copy which may have a performance impact but
already allows mixed-precision research.

Native mixed-precision ELL kernels are implemented which do not see this cost.
The accessor is also leveraged in a new CB-GMRES solver which allows for
performance improvements by compressing the Krylov basis vectors. Many other
features have been added to Ginkgo, such as reordering support, a new IDR
solver, Incomplete Cholesky preconditioner, matrix assembly support (only CPU
for now), machine topology information, and more!

Supported systems and requirements:
+ For all platforms, cmake 3.13+
+ C++14 compliant compiler
+ Linux and MacOS
  + gcc: 5.3+, 6.3+, 7.3+, all versions after 8.1+
  + clang: 3.9+
  + Intel compiler: 2018+
  + Apple LLVM: 8.0+
  + CUDA module: CUDA 9.0+
  + HIP module: ROCm 3.5+
  + DPC++ module: Intel OneAPI 2021.3. Set the CXX compiler to `dpcpp`.
+ Windows
  + MinGW and Cygwin: gcc 5.3+, 6.3+, 7.3+, all versions after 8.1+
  + Microsoft Visual Studio: VS 2019
  + CUDA module: CUDA 9.0+, Microsoft Visual Studio
  + OpenMP module: MinGW or Cygwin.


Algorithm and important feature additions:
+ Add a new DPC++ Executor for SYCL execution and other base utilities
  [#648](#648), [#661](#661), [#757](#757), [#832](#832)
+ Port matrix formats, solvers and related kernels to DPC++. For some kernels,
  also make use of a shared kernel implementation for all executors (except
  Reference). [#710](#710), [#799](#799), [#779](#779), [#733](#733), [#844](#844), [#843](#843), [#789](#789), [#845](#845), [#849](#849), [#855](#855), [#856](#856)
+ Add accessors which allow multi-precision kernels, among other things.
  [#643](#643), [#708](#708)
+ Add support for mixed precision operations through apply in all LinOps. [#677](#677)
+ Add incomplete Cholesky factorizations and preconditioners as well as some
  improvements to ILU. [#672](#672), [#837](#837), [#846](#846)
+ Add an AMGX implementation and kernels on all devices but DPC++.
  [#528](#528), [#695](#695), [#860](#860)
+ Add a new mixed-precision capability solver, Compressed Basis GMRES
  (CB-GMRES). [#693](#693), [#763](#763)
+ Add the IDR(s) solver. [#620](#620)
+ Add a new fixed-size block CSR matrix format (for the Reference executor).
  [#671](#671), [#730](#730)
+ Add native mixed-precision support to the ELL format. [#717](#717), [#780](#780)
+ Add Reverse Cuthill-McKee reordering [#500](#500), [#649](#649)
+ Add matrix assembly support on CPUs. [#644](#644)
+ Extends ISAI from triangular to general and spd matrices. [#690](#690)

Other additions:
+ Add the possibility to apply real matrices to complex vectors.
  [#655](#655), [#658](#658)
+ Add functions to compute the absolute of a matrix format. [#636](#636)
+ Add symmetric permutation and improve existing permutations.
  [#684](#684), [#657](#657), [#663](#663)
+ Add a MachineTopology class with HWLOC support [#554](#554), [#697](#697)
+ Add an implicit residual norm criterion. [#702](#702), [#818](#818), [#850](#850)
+ Row-major accessor is generalized to more than 2 dimensions and a new
  "block column-major" accessor has been added. [#707](#707)
+ Add an heat equation example. [#698](#698), [#706](#706)
+ Add ccache support in CMake and CI. [#725](#725), [#739](#739)
+ Allow tuning and benchmarking variables non intrusively. [#692](#692)
+ Add triangular solver benchmark [#664](#664)
+ Add benchmarks for BLAS operations [#772](#772), [#829](#829)
+ Add support for different precisions and consistent index types in benchmarks.
  [#675](#675), [#828](#828)
+ Add a Github bot system to facilitate development and PR management.
  [#667](#667), [#674](#674), [#689](#689), [#853](#853)
+ Add Intel (DPC++) CI support and enable CI on HPC systems. [#736](#736), [#751](#751), [#781](#781)
+ Add ssh debugging for Github Actions CI. [#749](#749)
+ Add pipeline segmentation for better CI speed. [#737](#737)


Changes:
+ Add a Scalar Jacobi specialization and kernels. [#808](#808), [#834](#834), [#854](#854)
+ Add implicit residual log for solvers and benchmarks. [#714](#714)
+ Change handling of the conjugate in the dense dot product. [#755](#755)
+ Improved Dense stride handling. [#774](#774)
+ Multiple improvements to the OpenMP kernels performance, including COO,
an exclusive prefix sum, and more. [#703](#703), [#765](#765), [#740](#740)
+ Allow specialization of submatrix and other dense creation functions in solvers. [#718](#718)
+ Improved Identity constructor and treatment of rectangular matrices. [#646](#646)
+ Allow CUDA/HIP executors to select allocation mode. [#758](#758)
+ Check if executors share the same memory. [#670](#670)
+ Improve test install and smoke testing support. [#721](#721)
+ Update the JOSS paper citation and add publications in the documentation.
  [#629](#629), [#724](#724)
+ Improve the version output. [#806](#806)
+ Add some utilities for dim and span. [#821](#821)
+ Improved solver and preconditioner benchmarks. [#660](#660)
+ Improve benchmark timing and output. [#669](#669), [#791](#791), [#801](#801), [#812](#812)


Fixes:
+ Sorting fix for the Jacobi preconditioner. [#659](#659)
+ Also log the first residual norm in CGS [#735](#735)
+ Fix BiCG and HIP CSR to work with complex matrices. [#651](#651)
+ Fix Coo SpMV on strided vectors. [#807](#807)
+ Fix segfault of extract_diagonal, add short-and-fat test. [#769](#769)
+ Fix device_reset issue by moving counter/mutex to device. [#810](#810)
+ Fix `EnableLogging` superclass. [#841](#841)
+ Support ROCm 4.1.x and breaking HIP_PLATFORM changes. [#726](#726)
+ Decreased test size for a few device tests. [#742](#742)
+ Fix multiple issues with our CMake HIP and RPATH setup.
  [#712](#712), [#745](#745), [#709](#709)
+ Cleanup our CMake installation step. [#713](#713)
+ Various simplification and fixes to the Windows CMake setup. [#720](#720), [#785](#785)
+ Simplify third-party integration. [#786](#786)
+ Improve Ginkgo device arch flags management. [#696](#696)
+ Other fixes and improvements to the CMake setup.
  [#685](#685), [#792](#792), [#705](#705), [#836](#836)
+ Clarification of dense norm documentation [#784](#784)
+ Various development tools fixes and improvements [#738](#738), [#830](#830), [#840](#840)
+ Make multiple operators/constructors explicit. [#650](#650), [#761](#761)
+ Fix some issues, memory leaks and warnings found by MSVC.
  [#666](#666), [#731](#731)
+ Improved solver memory estimates and consistent iteration counts [#691](#691)
+ Various logger improvements and fixes [#728](#728), [#743](#743), [#754](#754)
+ Fix for ForwardIterator requirements in iterator_factory. [#665](#665)
+ Various benchmark fixes. [#647](#647), [#673](#673), [#722](#722)
+ Various CI fixes and improvements. [#642](#642), [#641](#641), [#795](#795), [#783](#783), [#793](#793), [#852](#852)


Related PR: #857
tcojean added a commit that referenced this pull request Aug 23, 2021
Release 1.4.0 to master

The Ginkgo team is proud to announce the new Ginkgo minor release 1.4.0. This
release brings most of the Ginkgo functionality to the Intel DPC++ ecosystem
which enables Intel-GPU and CPU execution. The only Ginkgo features which have
not been ported yet are some preconditioners.

Ginkgo's mixed-precision support is greatly enhanced thanks to:
1. The new Accessor concept, which allows writing kernels featuring on-the-fly
memory compression, among other features. The accessor can be used as
header-only, see the [accessor BLAS benchmarks repository](https://github.com/ginkgo-project/accessor-BLAS/tree/develop) as a usage example.
2. All LinOps now transparently support mixed-precision execution. By default,
this is done through a temporary copy which may have a performance impact but
already allows mixed-precision research.

Native mixed-precision ELL kernels are implemented which do not see this cost.
The accessor is also leveraged in a new CB-GMRES solver which allows for
performance improvements by compressing the Krylov basis vectors. Many other
features have been added to Ginkgo, such as reordering support, a new IDR
solver, Incomplete Cholesky preconditioner, matrix assembly support (only CPU
for now), machine topology information, and more!

Supported systems and requirements:
+ For all platforms, cmake 3.13+
+ C++14 compliant compiler
+ Linux and MacOS
  + gcc: 5.3+, 6.3+, 7.3+, all versions after 8.1+
  + clang: 3.9+
  + Intel compiler: 2018+
  + Apple LLVM: 8.0+
  + CUDA module: CUDA 9.0+
  + HIP module: ROCm 3.5+
  + DPC++ module: Intel OneAPI 2021.3. Set the CXX compiler to `dpcpp`.
+ Windows
  + MinGW and Cygwin: gcc 5.3+, 6.3+, 7.3+, all versions after 8.1+
  + Microsoft Visual Studio: VS 2019
  + CUDA module: CUDA 9.0+, Microsoft Visual Studio
  + OpenMP module: MinGW or Cygwin.


Algorithm and important feature additions:
+ Add a new DPC++ Executor for SYCL execution and other base utilities
  [#648](#648), [#661](#661), [#757](#757), [#832](#832)
+ Port matrix formats, solvers and related kernels to DPC++. For some kernels,
  also make use of a shared kernel implementation for all executors (except
  Reference). [#710](#710), [#799](#799), [#779](#779), [#733](#733), [#844](#844), [#843](#843), [#789](#789), [#845](#845), [#849](#849), [#855](#855), [#856](#856)
+ Add accessors which allow multi-precision kernels, among other things.
  [#643](#643), [#708](#708)
+ Add support for mixed precision operations through apply in all LinOps. [#677](#677)
+ Add incomplete Cholesky factorizations and preconditioners as well as some
  improvements to ILU. [#672](#672), [#837](#837), [#846](#846)
+ Add an AMGX implementation and kernels on all devices but DPC++.
  [#528](#528), [#695](#695), [#860](#860)
+ Add a new mixed-precision capability solver, Compressed Basis GMRES
  (CB-GMRES). [#693](#693), [#763](#763)
+ Add the IDR(s) solver. [#620](#620)
+ Add a new fixed-size block CSR matrix format (for the Reference executor).
  [#671](#671), [#730](#730)
+ Add native mixed-precision support to the ELL format. [#717](#717), [#780](#780)
+ Add Reverse Cuthill-McKee reordering [#500](#500), [#649](#649)
+ Add matrix assembly support on CPUs. [#644](#644)
+ Extends ISAI from triangular to general and spd matrices. [#690](#690)

Other additions:
+ Add the possibility to apply real matrices to complex vectors.
  [#655](#655), [#658](#658)
+ Add functions to compute the absolute of a matrix format. [#636](#636)
+ Add symmetric permutation and improve existing permutations.
  [#684](#684), [#657](#657), [#663](#663)
+ Add a MachineTopology class with HWLOC support [#554](#554), [#697](#697)
+ Add an implicit residual norm criterion. [#702](#702), [#818](#818), [#850](#850)
+ Row-major accessor is generalized to more than 2 dimensions and a new
  "block column-major" accessor has been added. [#707](#707)
+ Add an heat equation example. [#698](#698), [#706](#706)
+ Add ccache support in CMake and CI. [#725](#725), [#739](#739)
+ Allow tuning and benchmarking variables non intrusively. [#692](#692)
+ Add triangular solver benchmark [#664](#664)
+ Add benchmarks for BLAS operations [#772](#772), [#829](#829)
+ Add support for different precisions and consistent index types in benchmarks.
  [#675](#675), [#828](#828)
+ Add a Github bot system to facilitate development and PR management.
  [#667](#667), [#674](#674), [#689](#689), [#853](#853)
+ Add Intel (DPC++) CI support and enable CI on HPC systems. [#736](#736), [#751](#751), [#781](#781)
+ Add ssh debugging for Github Actions CI. [#749](#749)
+ Add pipeline segmentation for better CI speed. [#737](#737)


Changes:
+ Add a Scalar Jacobi specialization and kernels. [#808](#808), [#834](#834), [#854](#854)
+ Add implicit residual log for solvers and benchmarks. [#714](#714)
+ Change handling of the conjugate in the dense dot product. [#755](#755)
+ Improved Dense stride handling. [#774](#774)
+ Multiple improvements to the OpenMP kernels performance, including COO,
an exclusive prefix sum, and more. [#703](#703), [#765](#765), [#740](#740)
+ Allow specialization of submatrix and other dense creation functions in solvers. [#718](#718)
+ Improved Identity constructor and treatment of rectangular matrices. [#646](#646)
+ Allow CUDA/HIP executors to select allocation mode. [#758](#758)
+ Check if executors share the same memory. [#670](#670)
+ Improve test install and smoke testing support. [#721](#721)
+ Update the JOSS paper citation and add publications in the documentation.
  [#629](#629), [#724](#724)
+ Improve the version output. [#806](#806)
+ Add some utilities for dim and span. [#821](#821)
+ Improved solver and preconditioner benchmarks. [#660](#660)
+ Improve benchmark timing and output. [#669](#669), [#791](#791), [#801](#801), [#812](#812)


Fixes:
+ Sorting fix for the Jacobi preconditioner. [#659](#659)
+ Also log the first residual norm in CGS [#735](#735)
+ Fix BiCG and HIP CSR to work with complex matrices. [#651](#651)
+ Fix Coo SpMV on strided vectors. [#807](#807)
+ Fix segfault of extract_diagonal, add short-and-fat test. [#769](#769)
+ Fix device_reset issue by moving counter/mutex to device. [#810](#810)
+ Fix `EnableLogging` superclass. [#841](#841)
+ Support ROCm 4.1.x and breaking HIP_PLATFORM changes. [#726](#726)
+ Decreased test size for a few device tests. [#742](#742)
+ Fix multiple issues with our CMake HIP and RPATH setup.
  [#712](#712), [#745](#745), [#709](#709)
+ Cleanup our CMake installation step. [#713](#713)
+ Various simplification and fixes to the Windows CMake setup. [#720](#720), [#785](#785)
+ Simplify third-party integration. [#786](#786)
+ Improve Ginkgo device arch flags management. [#696](#696)
+ Other fixes and improvements to the CMake setup.
  [#685](#685), [#792](#792), [#705](#705), [#836](#836)
+ Clarification of dense norm documentation [#784](#784)
+ Various development tools fixes and improvements [#738](#738), [#830](#830), [#840](#840)
+ Make multiple operators/constructors explicit. [#650](#650), [#761](#761)
+ Fix some issues, memory leaks and warnings found by MSVC.
  [#666](#666), [#731](#731)
+ Improved solver memory estimates and consistent iteration counts [#691](#691)
+ Various logger improvements and fixes [#728](#728), [#743](#743), [#754](#754)
+ Fix for ForwardIterator requirements in iterator_factory. [#665](#665)
+ Various benchmark fixes. [#647](#647), [#673](#673), [#722](#722)
+ Various CI fixes and improvements. [#642](#642), [#641](#641), [#795](#795), [#783](#783), [#793](#793), [#852](#852)

Related PR: #866
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Labels
1:ST:ready-to-merge This PR is ready to merge. is:new-feature A request or implementation of a feature that does not exist yet. mod:core This is related to the core module. mod:dpcpp This is related to the DPC++ module. plat:intel This is related to the Intel compilers. reg:build This is related to the build system.
Projects
None yet
Development

Successfully merging this pull request may close these issues.

None yet

5 participants