Skip to content

Commit

Permalink
[SYCL] Support online_compiler::compile compiled with pre-C++11 ABI (…
Browse files Browse the repository at this point in the history 
…#16269)

This is an updated version of #16179.

Run-time ABI mismatch resulting in a segmentation fault was uncovered
why enabling extra testing in #16235,
this new version fixes that.
  • Loading branch information
@aelovikov-intel
aelovikov-intel authored Dec 5, 2024
1 parent 6fd5143 commit 83ee235
Show file tree
Hide file tree
Showing 5 changed files with 99 additions and 39 deletions.
13 changes: 9 additions & 4 deletions sycl/include/sycl/detail/string_view.hpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -37,14 +37,19 @@ class string_view {

const char *data() const noexcept { return str; }

friend bool operator==(const string_view &lhs,
std::string_view rhs) noexcept {
friend bool operator==(string_view lhs, std::string_view rhs) noexcept {
return rhs == lhs.data();
}
friend bool operator==(std::string_view lhs,
const string_view &rhs) noexcept {
friend bool operator==(std::string_view lhs, string_view rhs) noexcept {
return lhs == rhs.data();
}

friend bool operator!=(string_view lhs, std::string_view rhs) noexcept {
return rhs != lhs.data();
}
friend bool operator!=(std::string_view lhs, string_view rhs) noexcept {
return lhs != rhs.data();
}
};

} // namespace detail
Expand Down
54 changes: 49 additions & 5 deletions sycl/include/sycl/ext/intel/experimental/online_compiler.hpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -18,6 +18,9 @@
namespace sycl {
inline namespace _V1 {
namespace ext::intel::experimental {
namespace detail {
using namespace sycl::detail;
}

using byte = unsigned char;

Expand Down Expand Up @@ -81,6 +84,30 @@ class __SYCL2020_DEPRECATED(
"experimental online_compiler is being deprecated. See "
"'sycl_ext_oneapi_kernel_compiler.asciidoc' instead for new kernel "
"compiler extension to kernel_bundle implementation.") online_compiler {
#if __INTEL_PREVIEW_BREAKING_CHANGES
// Refactor this during next ABI Breaking window. We have an `std::string`
// data member so cannot be accessing `this` when crossing ABI boundary.
#endif
__SYCL_EXPORT static std::vector<byte>
compile_impl(detail::string_view Src,
const std::vector<detail::string_view> &Options,
std::pair<int, int> OutputFormatVersion,
sycl::info::device_type DeviceType, device_arch DeviceArch,
bool Is64Bit, detail::string_view DeviceStepping,
void *&CompileToSPIRVHandle, void *&FreeSPIRVOutputsHandle);

std::vector<byte> compile_impl(const std::string &Source,
const std::vector<std::string> &UserArgs) {
std::vector<sycl::detail::string_view> Args;
for (auto &&Arg : UserArgs)
Args.emplace_back(Arg);

return compile_impl(std::string_view{Source}, Args, OutputFormatVersion,
DeviceType, DeviceArch, Is64Bit,
std::string_view{DeviceStepping}, CompileToSPIRVHandle,
FreeSPIRVOutputsHandle);
}

public:
/// Constructs online compiler which can target any device and produces
/// given compiled code format. Produces 64-bit device code.
Expand Down Expand Up @@ -196,9 +223,17 @@ class __SYCL2020_DEPRECATED(
/// OpenCL JIT compiler options must be supported.
template <>
template <>
__SYCL_EXPORT std::vector<byte>
online_compiler<source_language::opencl_c>::compile(
const std::string &src, const std::vector<std::string> &options);
#if !defined(__SYCL_ONLINE_COMPILER_CPP) || \
defined(__INTEL_PREVIEW_BREAKING_CHANGES)
inline
#else
__SYCL_EXPORT
#endif
std::vector<byte>
online_compiler<source_language::opencl_c>::compile(
const std::string &src, const std::vector<std::string> &options) {
return compile_impl(src, options);
}

/// Compiles the given OpenCL source. May throw \c online_compile_error.
/// @param src - contents of the source.
Expand All @@ -214,8 +249,17 @@ online_compiler<source_language::opencl_c>::compile(const std::string &src) {
/// @param options - compilation options (implementation defined).
template <>
template <>
__SYCL_EXPORT std::vector<byte> online_compiler<source_language::cm>::compile(
const std::string &src, const std::vector<std::string> &options);
#if !defined(__SYCL_ONLINE_COMPILER_CPP) || \
defined(__INTEL_PREVIEW_BREAKING_CHANGES)
inline
#else
__SYCL_EXPORT
#endif
std::vector<byte>
online_compiler<source_language::cm>::compile(
const std::string &src, const std::vector<std::string> &options) {
return compile_impl(src, options);
}

/// Compiles the given CM source \p src.
template <>
Expand Down
67 changes: 37 additions & 30 deletions sycl/source/detail/online_compiler/online_compiler.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -6,6 +6,8 @@
//
//===----------------------------------------------------------------------===//

#define __SYCL_ONLINE_COMPILER_CPP

#include <sycl/detail/os_util.hpp>
#include <sycl/detail/ur.hpp>
#include <sycl/ext/intel/experimental/online_compiler.hpp>
Expand All @@ -19,9 +21,11 @@ inline namespace _V1 {
namespace ext::intel::experimental {
namespace detail {

using namespace sycl::detail;

static std::vector<const char *>
prepareOclocArgs(sycl::info::device_type DeviceType, device_arch DeviceArch,
bool Is64Bit, const std::string &DeviceStepping,
bool Is64Bit, string_view DeviceStepping,
const std::string &UserArgs) {
std::vector<const char *> Args = {"ocloc", "-q", "-spv_only", "-device"};

Expand Down Expand Up @@ -54,7 +58,7 @@ prepareOclocArgs(sycl::info::device_type DeviceType, device_arch DeviceArch,

if (DeviceStepping != "") {
Args.push_back("-revision_id");
Args.push_back(DeviceStepping.c_str());
Args.push_back(DeviceStepping.data());
}

Args.push_back(Is64Bit ? "-64" : "-32");
Expand Down Expand Up @@ -82,11 +86,11 @@ prepareOclocArgs(sycl::info::device_type DeviceType, device_arch DeviceArch,
/// allocated during the compilation.
/// @param UserArgs - User's options to ocloc compiler.
static std::vector<byte>
compileToSPIRV(const std::string &Source, sycl::info::device_type DeviceType,
device_arch DeviceArch, bool Is64Bit,
const std::string &DeviceStepping, void *&CompileToSPIRVHandle,
void *&FreeSPIRVOutputsHandle,
compileToSPIRV(string_view Src, sycl::info::device_type DeviceType,
device_arch DeviceArch, bool Is64Bit, string_view DeviceStepping,
void *&CompileToSPIRVHandle, void *&FreeSPIRVOutputsHandle,
const std::vector<std::string> &UserArgs) {
std::string Source{Src.data()};

if (!CompileToSPIRVHandle) {
#ifdef __SYCL_RT_OS_WINDOWS
Expand Down Expand Up @@ -198,11 +202,12 @@ compileToSPIRV(const std::string &Source, sycl::info::device_type DeviceType,
}
} // namespace detail

template <>
template <>
__SYCL_EXPORT std::vector<byte>
online_compiler<source_language::opencl_c>::compile(
const std::string &Source, const std::vector<std::string> &UserArgs) {
template <source_language Lang>
__SYCL_EXPORT std::vector<byte> online_compiler<Lang>::compile_impl(
detail::string_view Src, const std::vector<detail::string_view> &Options,
std::pair<int, int> OutputFormatVersion, sycl::info::device_type DeviceType,
device_arch DeviceArch, bool Is64Bit, detail::string_view DeviceStepping,
void *&CompileToSPIRVHandle, void *&FreeSPIRVOutputsHandle) {

if (OutputFormatVersion != std::pair<int, int>{0, 0}) {
std::string Version = std::to_string(OutputFormatVersion.first) + ", " +
Expand All @@ -211,29 +216,31 @@ online_compiler<source_language::opencl_c>::compile(
Version + ") is not supported yet");
}

return detail::compileToSPIRV(Source, DeviceType, DeviceArch, Is64Bit,
DeviceStepping, CompileToSPIRVHandle,
FreeSPIRVOutputsHandle, UserArgs);
}

template <>
template <>
__SYCL_EXPORT std::vector<byte> online_compiler<source_language::cm>::compile(
const std::string &Source, const std::vector<std::string> &UserArgs) {
std::vector<std::string> UserArgs;
for (auto &&Opt : Options)
UserArgs.emplace_back(Opt.data());

if (OutputFormatVersion != std::pair<int, int>{0, 0}) {
std::string Version = std::to_string(OutputFormatVersion.first) + ", " +
std::to_string(OutputFormatVersion.second);
throw online_compile_error(std::string("The output format version (") +
Version + ") is not supported yet");
}
if constexpr (Lang == source_language::cm)
UserArgs.push_back("-cmc");

std::vector<std::string> CMUserArgs = UserArgs;
CMUserArgs.push_back("-cmc");
return detail::compileToSPIRV(Source, DeviceType, DeviceArch, Is64Bit,
return detail::compileToSPIRV(Src, DeviceType, DeviceArch, Is64Bit,
DeviceStepping, CompileToSPIRVHandle,
FreeSPIRVOutputsHandle, CMUserArgs);
FreeSPIRVOutputsHandle, UserArgs);
}

template __SYCL_EXPORT std::vector<byte>
online_compiler<source_language::opencl_c>::compile_impl(
detail::string_view Src, const std::vector<detail::string_view> &Options,
std::pair<int, int> OutputFormatVersion, sycl::info::device_type DeviceType,
device_arch DeviceArch, bool Is64Bit, detail::string_view DeviceStepping,
void *&CompileToSPIRVHandle, void *&FreeSPIRVOutputsHandle);

template __SYCL_EXPORT std::vector<byte>
online_compiler<source_language::cm>::compile_impl(
detail::string_view Src, const std::vector<detail::string_view> &Options,
std::pair<int, int> OutputFormatVersion, sycl::info::device_type DeviceType,
device_arch DeviceArch, bool Is64Bit, detail::string_view DeviceStepping,
void *&CompileToSPIRVHandle, void *&FreeSPIRVOutputsHandle);
} // namespace ext::intel::experimental

namespace ext {
Expand Down
2 changes: 2 additions & 0 deletions sycl/test/abi/sycl_symbols_linux.dump
View file
Original file line number Diff line number Diff line change
Expand Up @@ -2985,7 +2985,9 @@ _ZN4sycl3_V121__isgreaterequal_implEdd
_ZN4sycl3_V121__isgreaterequal_implEff
_ZN4sycl3_V122accelerator_selector_vERKNS0_6deviceE
_ZN4sycl3_V128verifyUSMAllocatorPropertiesERKNS0_13property_listE
_ZN4sycl3_V13ext5intel12experimental15online_compilerILNS3_15source_languageE0EE12compile_implENS0_6detail11string_viewERKSt6vectorIS8_SaIS8_EESt4pairIiiENS0_4info11device_typeENS3_11device_archEbS8_RPvSK_
_ZN4sycl3_V13ext5intel12experimental15online_compilerILNS3_15source_languageE0EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISE_EEEEES8_IhSaIhEERKSE_DpRKT_
_ZN4sycl3_V13ext5intel12experimental15online_compilerILNS3_15source_languageE1EE12compile_implENS0_6detail11string_viewERKSt6vectorIS8_SaIS8_EESt4pairIiiENS0_4info11device_typeENS3_11device_archEbS8_RPvSK_
_ZN4sycl3_V13ext5intel12experimental15online_compilerILNS3_15source_languageE1EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISE_EEEEES8_IhSaIhEERKSE_DpRKT_
_ZN4sycl3_V13ext5intel12experimental9pipe_base13get_pipe_nameB5cxx11EPKv
_ZN4sycl3_V13ext5intel12experimental9pipe_base17wait_non_blockingERKNS0_5eventE
Expand Down
2 changes: 2 additions & 0 deletions sycl/test/abi/sycl_symbols_windows.dump
View file
Original file line number Diff line number Diff line change
Expand Up @@ -3743,6 +3743,8 @@
?category@exception@_V1@sycl@@QEBAAEBVerror_category@std@@XZ
?clearArgs@handler@_V1@sycl@@AEAAXXZ
?code@exception@_V1@sycl@@QEBAAEBVerror_code@std@@XZ
?compile_impl@?$online_compiler@$00@experimental@intel@ext@_V1@sycl@@CA?AV?$vector@EV?$allocator@E@std@@@std@@Vstring_view@detail@56@AEBV?$vector@Vstring_view@detail@_V1@sycl@@V?$allocator@Vstring_view@detail@_V1@sycl@@@std@@@8@U?$pair@HH@8@W4device_type@info@56@Vdevice_arch@23456@_N0AEAPEAX6@Z
?compile_impl@?$online_compiler@$0A@@experimental@intel@ext@_V1@sycl@@CA?AV?$vector@EV?$allocator@E@std@@@std@@Vstring_view@detail@56@AEBV?$vector@Vstring_view@detail@_V1@sycl@@V?$allocator@Vstring_view@detail@_V1@sycl@@@std@@@8@U?$pair@HH@8@W4device_type@info@56@Vdevice_arch@23456@_N0AEAPEAX6@Z
?compile_impl@detail@_V1@sycl@@YA?AV?$shared_ptr@Vkernel_bundle_impl@detail@_V1@sycl@@@std@@AEBV?$kernel_bundle@$0A@@23@AEBV?$vector@Vdevice@_V1@sycl@@V?$allocator@Vdevice@_V1@sycl@@@std@@@5@AEBVproperty_list@23@@Z
?complete_fusion@fusion_wrapper@experimental@codeplay@ext@_V1@sycl@@QEAA?AVevent@56@AEBVproperty_list@56@@Z
?computeFallbackKernelBounds@handler@_V1@sycl@@AEAA?AV?$id@$01@23@_K0@Z
Expand Down

0 comments on commit 83ee235

Please sign in to comment.