-
Notifications
You must be signed in to change notification settings - Fork 578
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
Tpetra: Allow enabling SYCL backend in Tpetra #9086
Conversation
Status Flag 'Pre-Test Inspection' - - This Pull Request Requires Inspection... The code must be inspected by a member of the Team before Testing/Merging |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Thanks, @masterleinad , for these changes.
I looked at the Tpetra changes only, and request a few changes or explanations.
It would be good for @brian-kelley to review as well.
@@ -284,7 +284,7 @@ struct UnpackCrsMatrixAndCombineFunctor { | |||
|
|||
if (expected_num_bytes > num_bytes) | |||
{ | |||
printf( | |||
KOKKOS_IMPL_DO_NOT_USE_PRINTF( |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
I prefer not to use KOKKOS_IMPL details in Tpetra. Is there a problem with printf and SYCL? What does KOKKOS_IMPL_DO_NOT_USE_PRINTF do? How would one know to use this macro rather than printf?
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Oh, I should have pointed this out directly. SYCL
doesn't support printf
on the device. In Kokkos
we are defining a workaround to make this work regardless with the intention of only using it internally (such that it could be removed at any point). The alternative is to just guard these places with #ifndef KOKKOS_ENABLE_SYCL
or similar.
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
yes there is a problem with printf in SYCL. Namely that it isn't supported currently and you have to call sycl::printf or something like that ...
We are talking to various compiler folks to get that changed, and thus didn't want to add an official portability functionality like KOKKOS_PRINTF yet. But without changing it this guy won't compile/work with SYCL right now.
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Probably doing #ifndef KOKKOS_ENABLE_SYCL around the printfs is better than using a Kokkos implementation detail. Plus, it documents why we are doing something different from printf -- that is, that SYCL alone doesn't like it. When Kokkos adds an official portable KOKKOS_PRINTF, we will adopt it.
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Fixed in 45fefc7
(#9086).
@@ -471,7 +471,8 @@ int main (int argc, char *argv[]) | |||
{ | |||
TimeMonitor timerMultiVectorFill(*TimeMonitor::getNewTimer("4) MultiVectorFill")); | |||
|
|||
auto value = X->getLocalView<typename exec_space::memory_space>(Tpetra::Access::OverwriteAll); | |||
using device_type = typename tpetra_multivector_type::node_type::device_type; | |||
auto value = X->getLocalView<device_type>(Tpetra::Access::OverwriteAll); |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
OK, but prefer use of getLocalViewDevice(Tpetra::Access::OverwriteAll).
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
See 3221242
(#9086).
Status Flag 'Pre-Test Inspection' - - This Pull Request Requires Inspection... The code must be inspected by a member of the Team before Testing/Merging |
Btw. we are looking into making all the cahnges necessary for Kokkos/KokkosKernels part of the 3.4.1 patch release. So we wouldn't necessarily want you to go ahead and merge this before that is out and in Trilinos, and which point the changes are more or less confined to Tpetra. |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
The only issue I see is the CommBufferMemorySpace thing is templated on execution space so the specialization should use SYCL, not the SYCLSharedUSMSpace. HIP is wrong above, I found that yesterday and will fix in another PR.
@@ -132,6 +135,14 @@ namespace DefaultTypes { | |||
}; | |||
#endif | |||
|
|||
#ifdef KOKKOS_ENABLE_SYCL | |||
template<> | |||
struct CommBufferMemorySpace<Kokkos::Experimental::SYCLSharedUSMSpace> |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
This should actually be specialized on SYCL the execution space. The HIPHostPinnedSpace above is wrong, it should be HIP.
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Fixed in 68f57e3
(#9086).
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Thanks for your explanations. I resolved most of my conversations.
I'd still like to see KOKKOS_IMPL_DO_NOT_USE_PRINTF replaced with #ifndef KOKKOS_ENABLE_SYCL.
@@ -284,7 +284,7 @@ struct UnpackCrsMatrixAndCombineFunctor { | |||
|
|||
if (expected_num_bytes > num_bytes) | |||
{ | |||
printf( | |||
KOKKOS_IMPL_DO_NOT_USE_PRINTF( |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Probably doing #ifndef KOKKOS_ENABLE_SYCL around the printfs is better than using a Kokkos implementation detail. Plus, it documents why we are doing something different from printf -- that is, that SYCL alone doesn't like it. When Kokkos adds an official portable KOKKOS_PRINTF, we will adopt it.
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Thanks for making that change.
Status Flag 'Pre-Test Inspection' - SUCCESS: The last commit to this Pull Request has been INSPECTED AND APPROVED by [ brian-kelley ]! |
Status Flag 'Pull Request AutoTester' - Testing Jenkins Projects: Pull Request Auto Testing STARTING (click to expand)Build InformationTest Name: Trilinos_pullrequest_gcc_8.3.0
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_gcc_7.2.0_serial
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_gcc_7.2.0_debug
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_intel_17.0.1
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_cuda_10.1.105
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_clang_10.0.0
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_cuda_10.1.105_uvm_off
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_python_3
Jenkins Parameters
Using Repos:
Pull Request Author: masterleinad |
Status Flag 'Pull Request AutoTester' - Jenkins Testing: all Jobs PASSED Pull Request Auto Testing has PASSED (click to expand)Build InformationTest Name: Trilinos_pullrequest_gcc_8.3.0
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_gcc_7.2.0_serial
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_gcc_7.2.0_debug
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_intel_17.0.1
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_cuda_10.1.105
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_clang_10.0.0
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_cuda_10.1.105_uvm_off
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_python_3
Jenkins Parameters
|
Status Flag 'Pre-Merge Inspection' - SUCCESS: The last commit to this Pull Request has been INSPECTED AND APPROVED by [ brian-kelley kddevin ]! |
Status Flag 'Pull Request AutoTester' - AutoMerge IS ENABLED, but the Label AT: AUTOMERGE is not set. Either set Label AT: AUTOMERGE or manually merge the PR... |
packages/tpetra/CMakeLists.txt
Outdated
@@ -529,6 +529,29 @@ ELSE () # NOT Tpetra_INST_HIP | |||
MESSAGE(STATUS "NOTE: Kokkos::HIP is ON (the CMake option Kokkos_ENABLE_HIP is ON), but the corresponding Tpetra Node type is disabled. If you want to enable instantiation and use of Kokkos::HIP in Tpetra, please also set the CMake option Tpetra_INST_HIP:BOOL=ON. If you use the Kokkos::HIP version of Tpetra without doing this, you will get link errors!") | |||
ENDIF () | |||
ENDIF () # Tpetra_INST_HIP | |||
|
|||
# Kokkos::SYCL (Kokkos::Compat::KokkosSYCLWrapperNode) | |||
GLOBAL_SET(HAVE_TPETRA_INST_SYCL_DEFAULT OFF) |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
It would make sense to move this statement further up around line 365 and you might want to add an ASSERT_DEFINED(Kokkos_ENABLE_SYCL)
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
See a2f73a1
(#9086).
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
I just had a small comment regarding the top CMakeList.txt in tpetra, it's nothing big but we should try to be uniform if possible.
This reverts commit b918d95.
I rebased to resolve the merge conflicts and fixed the resulting problems (making sure that all the tests pass again on Intel GPUs). |
Status Flag 'Pre-Test Inspection' - - This Pull Request Requires Inspection... The code must be inspected by a member of the Team before Testing/Merging |
Status Flag 'Pre-Test Inspection' - SUCCESS: The last commit to this Pull Request has been INSPECTED AND APPROVED by [ brian-kelley ]! |
Status Flag 'Pull Request AutoTester' - Testing Jenkins Projects: Pull Request Auto Testing STARTING (click to expand)Build InformationTest Name: Trilinos_pullrequest_gcc_8.3.0
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_gcc_7.2.0_serial
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_gcc_7.2.0_debug
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_intel_17.0.1
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_cuda_10.1.105
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_cuda_10.1.105_uvm_off
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_clang_10.0.0
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_python_3
Jenkins Parameters
Using Repos:
Pull Request Author: masterleinad |
Status Flag 'Pull Request AutoTester' - Jenkins Testing: all Jobs PASSED Pull Request Auto Testing has PASSED (click to expand)Build InformationTest Name: Trilinos_pullrequest_gcc_8.3.0
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_gcc_7.2.0_serial
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_gcc_7.2.0_debug
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_intel_17.0.1
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_cuda_10.1.105
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_cuda_10.1.105_uvm_off
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_clang_10.0.0
Jenkins Parameters
Build InformationTest Name: Trilinos_pullrequest_python_3
Jenkins Parameters
|
Status Flag 'Pre-Merge Inspection' - SUCCESS: The last commit to this Pull Request has been INSPECTED AND APPROVED by [ brian-kelley ]! |
Status Flag 'Pull Request AutoTester' - AutoMerge IS ENABLED, but the Label AT: AUTOMERGE is not set. Either set Label AT: AUTOMERGE or manually merge the PR... |
2 similar comments
Status Flag 'Pull Request AutoTester' - AutoMerge IS ENABLED, but the Label AT: AUTOMERGE is not set. Either set Label AT: AUTOMERGE or manually merge the PR... |
Status Flag 'Pull Request AutoTester' - AutoMerge IS ENABLED, but the Label AT: AUTOMERGE is not set. Either set Label AT: AUTOMERGE or manually merge the PR... |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Thanks, @masterleinad
Status Flag 'Pull Request AutoTester' - AutoMerge IS ENABLED, but the Label AT: AUTOMERGE is not set. Either set Label AT: AUTOMERGE or manually merge the PR... |
Thanks @masterleinad for taking care of this, @crtrott and I are also OK with the state of the PR so I am merging it now. |
@trilinos/tpetra
Motivation
This pull request allows enabling Kokkos' SYCL backend in
Tpetra
to allowTrilinos
running on Intel GPUs.Currently, it needs kokkos/kokkos#4012, kokkos/kokkos#4007, kokkos/kokkos#3983, kokkos/kokkos#3998 and kokkos/kokkos-kernels#959 (all relevant changes also contained in this pull request) to build
Trilinos
withTpetra
enabled and running the unit tests. At the moment,ImbalancedRowMatrix
,assembleElement
, andMatMat
are failing and have been disabled using preprocessor macros and marking the relevant places with
FIXME_SYCL
.Testing
Tested on a CPU and on a Intel GPU passing all tests (with some disabled as described above).