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

[SYCL] Use std::array as storage for sycl::vec on device #14130

Merged
merged 22 commits into from
Jun 12, 2024

Conversation

uditagarwal97
Copy link
Contributor

@uditagarwal97 uditagarwal97 commented Jun 10, 2024

Replaces #13270

Changing the storage to std::array instead of Clang's extension fixes strict ansi-aliasing violation and simplifies device code.

@uditagarwal97 uditagarwal97 self-assigned this Jun 10, 2024
@uditagarwal97 uditagarwal97 changed the title Vec refac 4 [SYCL] Use std::array as storage for syc::vec on device Jun 10, 2024
@uditagarwal97 uditagarwal97 changed the title [SYCL] Use std::array as storage for syc::vec on device [SYCL] Use std::array as storage for sycl::vec on device Jun 10, 2024
@uditagarwal97 uditagarwal97 marked this pull request as ready for review June 11, 2024 13:36
@uditagarwal97 uditagarwal97 requested a review from a team as a code owner June 11, 2024 13:36
sycl/include/sycl/detail/vector_arith.hpp Show resolved Hide resolved
sycl/include/sycl/vector_preview.hpp Outdated Show resolved Hide resolved
sycl/include/sycl/vector_preview.hpp Outdated Show resolved Hide resolved
sycl/include/sycl/vector_preview.hpp Outdated Show resolved Hide resolved
sycl/include/sycl/vector_preview.hpp Outdated Show resolved Hide resolved
// CHECK-NEXT: [[EXTRACTVEC_I_I_I:%.*]] = shufflevector <4 x i16> [[LOADVEC4_I_I_I]], <4 x i16> poison, <3 x i32> <i32 0, i32 1, i32 2>
// CHECK-NEXT: [[A_ASCAST:%.*]] = addrspacecast ptr [[A]] to ptr addrspace(4)
// CHECK-NEXT: tail call void @llvm.experimental.noalias.scope.decl(metadata [[META95:![0-9]+]])
// CHECK-NEXT: store i64 0, ptr addrspace(4) [[AGG_RESULT]], align 8, !alias.scope [[META95]]
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 this store is unnecessary, but we can keep it for now.

// CHECK-NEXT: tail call void @llvm.experimental.noalias.scope.decl(metadata [[META100:![0-9]+]])
// CHECK-NEXT: call void @llvm.lifetime.start.p0(i64 4, ptr nonnull [[REF_TMP_I]])
// CHECK-NEXT: [[REF_TMP_ASCAST_I:%.*]] = addrspacecast ptr [[REF_TMP_I]] to ptr addrspace(4)
// CHECK-NEXT: tail call void @llvm.memset.p4.i64(ptr addrspace(4) noundef align 32 dereferenceable(32) [[AGG_RESULT]], i8 0, i64 32, i1 false), !alias.scope [[META100]]
Copy link
Contributor

Choose a reason for hiding this comment

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

Likewise.

@uditagarwal97
Copy link
Contributor Author

@intel/llvm-gatekeepers the PR is ready.

@aelovikov-intel aelovikov-intel merged commit e7defab into intel:sycl Jun 12, 2024
14 checks passed
@aelovikov-intel
Copy link
Contributor

I think this broke CTS in Nightly.

@uditagarwal97
Copy link
Contributor Author

I think this broke CTS in Nightly.

PR for fix: #14164

ianayl pushed a commit to ianayl/sycl that referenced this pull request Jun 13, 2024
…4130)

Replaces intel#13270

Changing the storage to std::array instead of Clang's extension fixes
strict ansi-aliasing violation and simplifies device code.
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.

2 participants