Skip to content

[SYCL][ESIMD] Move rdtsc function out of experimental namespace #13417

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

Merged
merged 2 commits into from
Apr 17, 2024

Conversation

fineg74
Copy link
Contributor

@fineg74 fineg74 commented Apr 16, 2024

No description provided.

@fineg74 fineg74 marked this pull request as ready for review April 16, 2024 17:54
@fineg74 fineg74 requested a review from a team as a code owner April 16, 2024 17:54
detail::is_esimd_scalar<TArg>::value,
TRes>
ESIMD_NODEBUG
ESIMD_INLINE std::enable_if_t<detail::is_esimd_scalar<TRes>::value &&
Copy link
Contributor

Choose a reason for hiding this comment

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

nit: there are some whitespace changes of this pattern that in my opinion decrease readability, i would expect we only need to touch the rdtsc function and not these

Copy link
Contributor Author

Choose a reason for hiding this comment

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

Fixed

@@ -1302,6 +1301,13 @@ __ESIMD_API uint32_t subb(uint32_t &borrow, uint32_t src0, uint32_t src1) {
return Res[0];
}

/// rdtsc - get the value of timestamp counter.
/// \return the current value of timestamp counter
Copy link
Contributor

Choose a reason for hiding this comment

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

I think it's @return based on other functions

Copy link
Contributor Author

Choose a reason for hiding this comment

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

Fixed

@@ -41,9 +41,9 @@ int test_rdtsc() {
auto Kernel = ([=](sycl::nd_item<1> ndi) [[intel::sycl_explicit_simd]] {
using namespace sycl::ext::intel::esimd;
auto Idx = ndi.get_global_id(0);
uint64_t StartCounter = sycl::ext::intel::experimental::esimd::rdtsc();
Copy link
Contributor

Choose a reason for hiding this comment

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

nit: maybe we call both experimental and non-experimental based on Idx % 2 or something

Copy link
Contributor Author

Choose a reason for hiding this comment

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

Done

Copy link
Contributor

@v-klochkov v-klochkov left a comment

Choose a reason for hiding this comment

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

Looks good to me.
Can you please check if this extension is ready for usage, maybe create a copy/draft of this PR with OpReadClockKHR

Copy link
Contributor

@v-klochkov v-klochkov left a comment

Choose a reason for hiding this comment

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

The change is correct as written now, and the experiment with using SPIRV instead of GenX intrinsic can be done in a separate PR.

@v-klochkov v-klochkov merged commit a0d8f01 into intel:sycl Apr 17, 2024
@fineg74 fineg74 deleted the rdtsc branch April 18, 2024 01:04
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Labels
None yet
Projects
None yet
Development

Successfully merging this pull request may close these issues.

3 participants