Skip to content

[SYCL] Implement sycl_ext_oneapi_fp8 for CRI#21568

Merged
KornevNikita merged 101 commits into
intel:syclfrom
dklochkov-emb:sycl-ext-one-api-fp8-new-arch
Jun 12, 2026
Merged

[SYCL] Implement sycl_ext_oneapi_fp8 for CRI#21568
KornevNikita merged 101 commits into
intel:syclfrom
dklochkov-emb:sycl-ext-one-api-fp8-new-arch

Conversation

@dklochkov-emb

@dklochkov-emb dklochkov-emb commented Mar 19, 2026

Copy link
Copy Markdown
Contributor

This PR adds FP8 types for CRI docs.
It adds the first part which includes:

  1. Initial implementation
  2. Unit tests which check expected behavior and values
  3. Unit tests which check the fact of builtin calls.
  4. E2E tests of all 3 data types: fp8_e4m3, fp8_e5m2 and fp8_e8m0

@dklochkov-emb dklochkov-emb requested a review from a team as a code owner March 19, 2026 13:48
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
@dklochkov-emb

dklochkov-emb commented Mar 31, 2026

Copy link
Copy Markdown
Contributor Author

@intel/llvm-reviewers-runtime please, review

Copilot AI left a comment

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

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

Pull request overview

This PR introduces initial host-side implementations of OneAPI experimental FP8 types (E4M3, E5M2, E8M0) and adds unit tests validating conversions/encodings plus tests that verify the expected SPIR-V builtin call paths (via mocks) for device-only code paths.

Changes:

  • Added sycl::ext::oneapi::experimental FP8 type definitions and conversion logic in a new public header.
  • Added unit tests for value/encoding behavior for fp8 E4M3, E5M2, and E8M0 on CPU.
  • Added mocked builtin implementations and tests to confirm which builtins are invoked from constructors/conversions.

Reviewed changes

Copilot reviewed 8 out of 8 changed files in this pull request and generated 8 comments.

Show a summary per file
File Description
sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Adds the FP8 type implementations, conversion helpers, and builtin hooks.
sycl/unittests/Extensions/fp8/fp8_e4m3.cpp Adds host unit tests for E4M3 encoding/conversion behavior.
sycl/unittests/Extensions/fp8/fp8_e5m2.cpp Adds host unit tests for E5M2 encoding/conversion behavior.
sycl/unittests/Extensions/fp8/fp8_e8m0.cpp Adds host unit tests for E8M0 encoding/conversion behavior.
sycl/unittests/Extensions/fp8/builtin_mocks.hpp Provides mock builtin symbols and counters to validate builtin call paths.
sycl/unittests/Extensions/fp8/builtin_call_tests.cpp Adds tests asserting the expected builtin calls occur (via the mocks).
sycl/unittests/Extensions/fp8/CMakeLists.txt Registers the new FP8 unit test target and sources.
sycl/unittests/Extensions/CMakeLists.txt Adds the new fp8 unit test subdirectory to the Extensions suite.

Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/unittests/Extensions/fp8/fp8_e8m0.cpp

@gmlueck gmlueck left a comment

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

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

Here are my comments so far. I haven't looked at E8M0 yet, or at any of the CPU conversions.

Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
@dm-vodopyanov

Copy link
Copy Markdown
Contributor

@dklochkov-emb can you please provide status for all failed fp8-related tests in pre-commit?

BTW, I restarted some of the checks due to CI issue, so need to wait a bit for the completion.

@dm-vodopyanov dm-vodopyanov left a comment

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

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

Overall, LGTM

Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp
@dklochkov-emb dklochkov-emb requested a review from gmlueck June 11, 2026 13:09
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp Outdated
Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp
@sys-ce-bb

Copy link
Copy Markdown
Contributor

@intel/llvm-gatekeepers please consider merging

Comment thread sycl/include/sycl/ext/oneapi/experimental/float_8bit/types.hpp
@dklochkov-emb

Copy link
Copy Markdown
Contributor Author

@intel/llvm-gatekeepers Please, merge it

@KornevNikita KornevNikita merged commit b09acde into intel:sycl Jun 12, 2026
33 checks passed
@sarnex

sarnex commented Jun 12, 2026

Copy link
Copy Markdown
Contributor

@dklochkov-emb We are seeing postcommit CI failures in check-sycl that seem related

GTEST_OUTPUT=json:/__w/llvm/llvm/build/tools/sycl/unittests/Extensions/fp8/./FP8TypesTests-Non_Preview_Tests-SYCL-Unit-313267-102-106.json GTEST_SHUFFLE=0 GTEST_TOTAL_SHARDS=106 GTEST_SHARD_INDEX=102 /__w/llvm/llvm/build/tools/sycl/unittests/Extensions/fp8/./FP8TypesTests-Non_Preview_Tests
--

Script:
--
/__w/llvm/llvm/build/tools/sycl/unittests/Extensions/fp8/./FP8TypesTests-Non_Preview_Tests --gtest_filter=FP8E8M0Test.MarrayBFloat16ToEvenRounding
--
/__w/llvm/llvm/src/sycl/unittests/Extensions/fp8/fp8_e8m0.cpp:488: Failure
Death test: { fp8_e8m0_x2 value(in, rounding::to_even); (void)value; }
    Result: failed to die.
 Error msg:
[  DEATH   ] 


/__w/llvm/llvm/src/sycl/unittests/Extensions/fp8/fp8_e8m0.cpp:488
Death test: { fp8_e8m0_x2 value(in, rounding::to_even); (void)value; }
    Result: failed to die.
 Error msg:
[  DEATH   ] 

https://github.com/intel/llvm/actions/runs/27421671974/job/81048588785

Can you please fix ASAP and if not possible revert? Thanks

@dklochkov-emb

Copy link
Copy Markdown
Contributor Author

@sarnex I temporarily disabled tests while bug is not fixed, PR:
#22309

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.