Skip to content

[SYCL][E2E] Ensure lowering of llvm.bitreverse for 2/4-bit scalars is functionally correct #13359

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 6 commits into from
May 22, 2024
Merged

[SYCL][E2E] Ensure lowering of llvm.bitreverse for 2/4-bit scalars is functionally correct #13359

merged 6 commits into from
May 22, 2024

Conversation

LU-JOHN
Copy link
Contributor

@LU-JOHN LU-JOHN commented Apr 11, 2024

Ensure that lowering of llvm.bitreverse.i2/i4 by llvm-spirv is functionally correct.

@LU-JOHN LU-JOHN requested a review from a team as a code owner April 11, 2024 01:05
@LU-JOHN LU-JOHN requested a review from againull April 11, 2024 01:05
@LU-JOHN LU-JOHN marked this pull request as draft April 11, 2024 01:05
@LU-JOHN LU-JOHN marked this pull request as ready for review May 10, 2024 16:05
@LU-JOHN LU-JOHN temporarily deployed to WindowsCILock May 10, 2024 16:21 — with GitHub Actions Inactive
@againull
Copy link
Contributor

@LU-JOHN Introduced test fails in pre-commit:

2024-05-10T17:40:26.8989941Z env ONEAPI_DEVICE_SELECTOR=level_zero:gpu  /__w/llvm/llvm/build-e2e/LLVMIntrinsicLowering/Output/bitreverse.cpp.tmp.out
2024-05-10T17:40:26.8990827Z # executed command: env ONEAPI_DEVICE_SELECTOR=level_zero:gpu /__w/llvm/llvm/build-e2e/LLVMIntrinsicLowering/Output/bitreverse.cpp.tmp.out
2024-05-10T17:40:26.8991396Z # .---command stderr------------
2024-05-10T17:40:26.8991920Z # | terminate called after throwing an instance of 'sycl::_V1::compile_program_error'
2024-05-10T17:40:26.8992343Z # |   what():  The program was built for 1 devices
2024-05-10T17:40:26.8992718Z # | Build program log for 'Intel(R) Iris(R) Xe Graphics':
2024-05-10T17:40:26.8993054Z # |  -11 (PI_ERROR_BUILD_PROGRAM_FAILURE)
2024-05-10T17:40:26.8993319Z # `-----------------------------
2024-05-10T17:40:26.8993583Z # error: command failed with exit status: -6

@@ -2,6 +2,10 @@

// UNSUPPORTED: hip || cuda

// TODO: Remove XFAIL after fixing
// https://github.com/intel/intel-graphics-compiler/issues/330
// XFAIL: gpu-intel-gen12
Copy link
Contributor

Choose a reason for hiding this comment

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

Is this test passing on other GPUs (like pvc, dg2)?

Copy link
Contributor Author

@LU-JOHN LU-JOHN May 17, 2024

Choose a reason for hiding this comment

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

On commit 1ad567a3f7f47920549bb58d6b433b7af869c7dd, it was only failing on:

SYCL Pre Commit on Windows / e2e / Intel GEN12 Graphics with Level Zero
SYCL Pre Commit on Linux / test (Intel, ["Linux", "gen12"], ghcr.io/intel/llvm/ubuntu2204_intel_drivers

Everything is passing now with the latest commit d7b722ec0d1eaa22dbfaceb17d1525f4418738f5.

Copy link
Contributor

Choose a reason for hiding this comment

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

We test only on Gen12 in pre-commit testing, so because you marked the test with XFAIL: gpu-intel-gen12, it passes now on Gen12. Unfortunately, it doesn't mean that test passes on other hardware. So it may easily fail on DG2 testing in post-commit.

Copy link
Contributor

Choose a reason for hiding this comment

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

Also, does it make sense to land this PR since IGC has some issues with these instructions?
Basically, because of this you disable testing of all other builtins (for other types).

Maybe it worth outlining this builtings to a separate test? And maybe even check only llvm-spirv translation for now (not e2e)?

Copy link
Contributor Author

Choose a reason for hiding this comment

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

Test has a segmentation fault on pvc machine and dg2 machine.

Copy link
Contributor Author

Choose a reason for hiding this comment

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

2/4-bit bitreverse testing moved to new file.

@againull againull self-requested a review May 17, 2024 19:26
Copy link
Contributor

@againull againull left a comment

Choose a reason for hiding this comment

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

Marking as requested changes for now to avoid accidental merge, just because some issues need to be addressed first.

Lu, John added 6 commits May 21, 2024 12:55
@LU-JOHN LU-JOHN temporarily deployed to WindowsCILock May 21, 2024 21:35 — with GitHub Actions Inactive
@LU-JOHN LU-JOHN temporarily deployed to WindowsCILock May 21, 2024 21:48 — with GitHub Actions Inactive
@againull againull merged commit 9b4bf7c into intel:sycl May 22, 2024
13 checks passed
@steffenlarsen
Copy link
Contributor

Post-commit failure should be addressed by #13889.

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