-
Notifications
You must be signed in to change notification settings - Fork 407
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
Label all SYCL fences #4258
Label all SYCL fences #4258
Conversation
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.
Looks fine by me
6233fc4
to
5a7263e
Compare
|
||
return m_scratchFlags; | ||
} | ||
|
||
template <typename WAT> | ||
void SYCLInternal::fence_helper(WAT& wat, const std::string& name, |
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.
What does "wat" stand for?
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 think @nliber meant it as an abbreviation for "wait and throw", implying that the template argument should be a type that implements wait_and_throw()
.
I was moving it here to escape circular header dependencies. We can discuss if not inlining this is problematic or not, of course.
try { | ||
(*queue)->wait_and_throw(); | ||
} catch (sycl::exception const& e) { | ||
Kokkos::Impl::throw_runtime_exception( | ||
std::string("There was a synchronous SYCL error:\n") += | ||
e.what()); | ||
} |
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.
Why are you not using SYCLInternal::fence_helper
here?
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.
#4258 (comment) says
[...] Also, the static fence can't really use the lowest level function anymore since we don't store the instance id with the queues. I decided to just copy the fence implementation for this case duplicating a few lines of code. Note that this case was treated in the Profiling interface separately already.
Previously, the final level calling
fence
functions forSYCL
didn't have labels. This pull request shifts the respectiveProfiling
calls down.This requires that
USMObjectMem
needs another identifier for theinstance_id
to report its fences. Also, the static fence can't really use the lowest level function anymore since we don't store the instance id with thequeues
. I decided to just copy the fence implementation for this case duplicating a few lines of code. Note that this case was treated in theProfiling
interface separately already.I also noticed that we use
Kokkos::SYCL
orKokkos::Experimental::SYCL
rather inconsistently in strings so I went for the latter everywhere since the majority of cases already usedKokkos::Experimental::SYCL
.