Skip to content

[SYCL] Rename intel_reqd_sub_group_size attribute metadata #7713

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

Closed
wants to merge 4 commits into from

Conversation

smanna12
Copy link
Contributor

@smanna12 smanna12 commented Dec 8, 2022

This commit renames metadata nodes introduced by the intel_reqd_sub_group_size attribute feature design document from using intel_ prefix to use sycl_ prefix.

Addresses #6824

Signed-off-by: Soumi Manna [email protected]

This commit renames metadata nodes introduced by the intel_reqd_sub_group_size
attribute feature design document from using intel_ prefix to use sycl_ prefix.

Addresses intel#6824

Signed-off-by: Soumi Manna <[email protected]>
@smanna12
Copy link
Contributor Author

smanna12 commented Dec 8, 2022

Follow-up comments from @steffenlarsen #6822 (comment)

Signed-off-by: Soumi Manna <[email protected]>
Signed-off-by: Soumi Manna <[email protected]>
Signed-off-by: Soumi Manna <[email protected]>
@smanna12 smanna12 marked this pull request as ready for review December 9, 2022 02:30
@steffenlarsen
Copy link
Contributor

I believe we will need the SPIR-V translator to also correctly consume the renamed metadata. @MrSidims - Can you confirm?

@@ -9,7 +9,7 @@ kernel __attribute__((vec_type_hint(uint4))) __attribute__((work_group_size_hint
// CHECK: define {{(dso_local )?}}spir_kernel void @kernel2(i32 {{[^%]*}}%a) {{[^{]+}} !vec_type_hint ![[MD3:[0-9]+]] !work_group_size_hint ![[MD4:[0-9]+]]

kernel __attribute__((intel_reqd_sub_group_size(8))) void kernel3(int a) {}
// CHECK: define {{(dso_local )?}}spir_kernel void @kernel3(i32 {{[^%]*}}%a) {{[^{]+}} !intel_reqd_sub_group_size ![[MD5:[0-9]+]]
// CHECK: define {{(dso_local )?}}spir_kernel void @kernel3(i32 {{[^%]*}}%a) {{[^{]+}} !sycl_reqd_sub_group_size ![[MD5:[0-9]+]]
Copy link
Contributor

Choose a reason for hiding this comment

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

I'm concerned about this change. This is an existing OpenCL C codegen test, why do we change the metadata name there? OpenCL flow is not being customized in our SYCL/DPC++ compiler as I understand it, so such change should go to upstream first, I suppose.

In OpenCL world, required sub-group size is an Intel extension and the translator is currently designed to only recognize intel_* metadata. By changing its name we will effectively break support for the attribute. See corresponding test and related commit in the translator repo.

Copy link
Contributor

Choose a reason for hiding this comment

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

I believe this change is a result of this discussion - #6804 (comment)

It isn't clear to me if the consensus was to actually modify this metadata. I don't see why we need to since this is an accepted Intel extension for openCL, and it exists outside of DPCPP. I understand the need to be consistent across various SYCL IR metadata but if SYCL is reusing existing constructs, I don't think we need to rename them. If we choose to do so, I agree with @AlexeySachkov that it should be done upstream.

I also can't recall why we have sycl::reqd_sub_group_size , intel::reqd_sub_group_size, intel_reqd_sub_group_size and intel::sub_group_size versions of this attribute. Can someone please explain?

@elizabethandrews
Copy link
Contributor

IMO what is important right now is to identify all the intel_ LLVM IR metadata that we generate for SYCL and see if we can/should rename it to sycl_. We already renamed metadata related to optional device features but the answer might not be as straightforward for other cases. If we want consistency and clarity, IMO this should investigation should be done first.

@MrSidims
Copy link
Contributor

MrSidims commented Jan 9, 2023

I believe we will need the SPIR-V translator to also correctly consume the renamed metadata. @MrSidims - Can you confirm?

Sorry for a long response, don't hesitate to ping me :)

Yes, it would require the appropriate change. But I have the same concern as Alexey.

@smanna12 smanna12 closed this Dec 24, 2023
@smanna12 smanna12 deleted the Rename_Attr_Metadata branch December 24, 2023 16:03
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.

5 participants