Skip to content

[SYCL][NATIVECPU] Support specialization constants on Native CPU #14446

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.

Sign up for GitHub

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
Sep 16, 2024
Merged

[SYCL][NATIVECPU] Support specialization constants on Native CPU #14446

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
6 commits
Select commit Hold shift + click to select a range
ee7c4ba
Initial support for spec constants on Native CPU
PietroGhg Jun 3, 2024
6a18136
Mark kernel-bundle-api.cpp test unsupported
PietroGhg Jun 26, 2024
dd32fd0
Link to SYCLLowerIR
PietroGhg Jul 4, 2024
21c278c
Formatting
PietroGhg Jul 5, 2024
fe5edca
Use DefaultValue to compute md
PietroGhg Jul 9, 2024
d01810c
Update __SYCL_DEVICE_BINARY_TARGET_NATIVE_CPU
PietroGhg Aug 14, 2024
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
3 changes: 2 additions & 1 deletion llvm/lib/SYCLLowerIR/SpecConstants.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -951,8 +951,9 @@ PreservedAnalyses SpecConstantsPass::run(Module &M,
updatePaddingInLastMDNode(Ctx, SCMetadata, Padding);
}

auto *DefValTy = DefaultValue->getType();
SCMetadata[SymID] = generateSpecConstantMetadata(
M, SymID, SCTy, NextID, /* is native spec constant */ false);
M, SymID, DefValTy, NextID, /* is native spec constant */ false);

++NextID.ID;
NextOffset += Size;
Expand Down
1 change: 1 addition & 0 deletions llvm/lib/SYCLNativeCPUUtils/CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -13,6 +13,7 @@ add_llvm_component_library(LLVMSYCLNativeCPUUtils
Core
Support
Passes
SYCLLowerIR
Target
TargetParser
TransformUtils
Expand Down
2 changes: 2 additions & 0 deletions llvm/lib/SYCLNativeCPUUtils/PipelineSYCLNativeCPU.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -14,6 +14,7 @@
#include "llvm/SYCLLowerIR/ConvertToMuxBuiltinsSYCLNativeCPU.h"
#include "llvm/SYCLLowerIR/PrepareSYCLNativeCPU.h"
#include "llvm/SYCLLowerIR/RenameKernelSYCLNativeCPU.h"
#include "llvm/SYCLLowerIR/SpecConstants.h"
#include "llvm/SYCLLowerIR/UtilsSYCLNativeCPU.h"
#include "llvm/Support/CommandLine.h"

Expand Down Expand Up @@ -60,6 +61,7 @@ static cl::opt<bool>
void llvm::sycl::utils::addSYCLNativeCPUBackendPasses(
llvm::ModulePassManager &MPM, ModuleAnalysisManager &MAM,
OptimizationLevel OptLevel) {
MPM.addPass(SpecConstantsPass(SpecConstantsPass::HandlingMode::emulation));
MPM.addPass(ConvertToMuxBuiltinsSYCLNativeCPUPass());
#ifdef NATIVECPU_USE_OCK
MPM.addPass(compiler::utils::TransferKernelMetadataPass());
Expand Down
14 changes: 7 additions & 7 deletions sycl/cmake/modules/FetchUnifiedRuntime.cmake
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -117,13 +117,13 @@ if(SYCL_UR_USE_FETCH_CONTENT)
endfunction()

set(UNIFIED_RUNTIME_REPO "https://github.com/oneapi-src/unified-runtime.git")
# commit f31160dea6d142014f441bc4ca5e58e48827490e
# Merge: 2bbe9526 64068799
# Author: Piotr Balcer <piotr.balcer@intel.com>
# Date: Thu Sep 12 14:19:48 2024 +0200
# Merge pull request #2083 from kswiecicki/xpti-init-fix
# Fix XPTI initialization bug
set(UNIFIED_RUNTIME_TAG f31160dea6d142014f441bc4ca5e58e48827490e)
# commit fa9ebe7bd3d9bd11dd5ea8a59eff12f5746411d3
# Merge: 92638b2a 9eb1c74f
# Author: Omar Ahmed <omar.ahmed@codeplay.com>
# Date: Fri Sep 13 14:44:27 2024 +0100
# Merge pull request #1821 from PietroGhg/pietro/native_cpu_specconstants
# [NATIVECPU] Initial support for spec constants on Native CPU
set(UNIFIED_RUNTIME_TAG fa9ebe7bd3d9bd11dd5ea8a59eff12f5746411d3)

set(UMF_BUILD_EXAMPLES OFF CACHE INTERNAL "EXAMPLES")
# Due to the use of dependentloadflag and no installer for UMF and hwloc we need
Expand Down
2 changes: 2 additions & 0 deletions sycl/source/kernel_bundle.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -329,6 +329,8 @@ bool is_compatible(const std::vector<kernel_id> &KernelIDs, const device &Dev) {
return BE == sycl::backend::ext_oneapi_cuda;
} else if (strcmp(Target, __SYCL_DEVICE_BINARY_TARGET_AMDGCN) == 0) {
return BE == sycl::backend::ext_oneapi_hip;
} else if (strcmp(Target, __SYCL_DEVICE_BINARY_TARGET_NATIVE_CPU) == 0) {
return BE == sycl::backend::ext_oneapi_native_cpu;
}

return false;
Expand Down
1 change: 1 addition & 0 deletions sycl/test-e2e/SpecConstants/2020/kernel-bundle-api.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -11,6 +11,7 @@
// RUN: %{run} %t.out
//
// UNSUPPORTED: hip
// UNSUPPORTED: native_cpu
Copy link
Contributor

Choose a reason for hiding this comment

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

Why is this test being marked as unsupported on native CPU if this PR is intended to add support for this feature? Does it mean that the test passed and this patch introduces a regression?

Copy link
Contributor Author

Choose a reason for hiding this comment

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

We run Native CPU e2e tests downstream, we weren't running the tests in test-e2e/SpecConstants before this PR. With this PR all the tests in test-e2e/SpecConstants are passing, besides this one, which is also marked as UNSUPPORTED for other targets that use emulated SC, so this is not a regression.


#include <cstdlib>
#include <iostream>
Expand Down
Loading