From 2504ec7d501fee9cfb823024434a3d8ed1493215 Mon Sep 17 00:00:00 2001 From: Chris Perkins Date: Fri, 24 Jan 2025 13:45:46 -0800 Subject: [PATCH 1/3] removing Online Compiler. Moved ocloc_api.h to KernelCompiler directory (needed for compiling OpenCL C). Still needs Win ABI symbols --- .../sycl_ext_intel_online_compiler.asciidoc | 208 ------------- .../sycl_ext_intel_online_compiler.asciidoc | 2 - .../intel/experimental/online_compiler.hpp | 274 ------------------ .../sycl/ext/intel/online_compiler.hpp | 20 -- sycl/source/CMakeLists.txt | 1 - .../kernel_compiler_opencl.cpp | 2 +- .../ocloc_api.h | 0 .../online_compiler/online_compiler.cpp | 267 ----------------- .../OnlineCompiler/online_compiler_L0.cpp | 79 ----- .../OnlineCompiler/online_compiler_OpenCL.cpp | 113 -------- .../OnlineCompiler/online_compiler_common.hpp | 193 ------------ sycl/test/abi/sycl_abi_neutrality_test.cpp | 4 - sycl/test/abi/sycl_symbols_linux.dump | 4 - .../test/basic_tests/no_math_in_global_ns.cpp | 1 - sycl/test/warnings/sycl_2020_deprecations.cpp | 6 - 15 files changed, 1 insertion(+), 1173 deletions(-) delete mode 100644 sycl/doc/extensions/deprecated/sycl_ext_intel_online_compiler.asciidoc delete mode 100644 sycl/doc/extensions/experimental/sycl_ext_intel_online_compiler.asciidoc delete mode 100644 sycl/include/sycl/ext/intel/experimental/online_compiler.hpp delete mode 100644 sycl/include/sycl/ext/intel/online_compiler.hpp rename sycl/source/detail/{online_compiler => kernel_compiler}/ocloc_api.h (100%) delete mode 100644 sycl/source/detail/online_compiler/online_compiler.cpp delete mode 100644 sycl/test-e2e/OnlineCompiler/online_compiler_L0.cpp delete mode 100644 sycl/test-e2e/OnlineCompiler/online_compiler_OpenCL.cpp delete mode 100644 sycl/test-e2e/OnlineCompiler/online_compiler_common.hpp diff --git a/sycl/doc/extensions/deprecated/sycl_ext_intel_online_compiler.asciidoc b/sycl/doc/extensions/deprecated/sycl_ext_intel_online_compiler.asciidoc deleted file mode 100644 index 1b6fec92b033d..0000000000000 --- a/sycl/doc/extensions/deprecated/sycl_ext_intel_online_compiler.asciidoc +++ /dev/null @@ -1,208 +0,0 @@ -= SYCL Intel extension: Online Compilation -Konstantin Bobrovskii , John Pennycook -v0.1 -:source-highlighter: pygments -:icons: font -:dpcpp: pass:[DPC++] - -== Introduction -This document describes an interface for online compilation from high-level languages, such as -OpenCL, to a binary format, such as SPIR-V, loadable by SYCL backends. Unlike SYCL 2020 provisional's -OpenCL backend online compilation interface, this interface is not bound to any particular backend and does -not require available SYCL context for online compilation. - -This gives a flexibility to "cross-compile" to SPIR-V or other supported formats without any SYCL -device or context available. The online compilation API uses the `online_compiler` class to access -compilation services. Instances of the class are constructed based on a specification of the desired -compilation target passed to the constructors - such as compiled code format, target architecture, -etc. All the settings are optional, and by default the target is generic SPIR-V. - -This API is an Intel SYCL extension. - -== Status - -This extension has been deprecated. Although it is still supported in {dpcpp}, -we expect that the interfaces defined in this specification will be removed in -an upcoming {dpcpp} release. *Shipping software products should stop using -APIs defined in this specification and use an alternative instead.* - -== Online compilation API - -All online compilation API elements reside in the `sycl::INTEL` namespace. - -=== Source language specification - -Elements of the enum designate the source language: -[source,c++] ------------------ -enum class source_language { - opencl_c, // OpenCL C language - cm // Intel's C-for-Media language -}; ------------------ - -=== APIs to express compilation target characteristics - -The desired format of the compiled code: -[source,c++] ------------------ -enum class compiled_code_format { - spir_v -}; ------------------ - -Target device architecture: -[source,c++] ------------------ -class device_arch { -public: - static constexpr int any = 0; // designates an unspecified architecture - device_arch(int Val); - - // GPU architecture IDs - enum gpu { gpu_any = 1, ... }; - // CPU architecture IDs - enum cpu { cpu_any = 1, ... }; - // FPGA architecture IDs - enum fpga { fpga_any = 1, ... }; - - // Converts this architecture representation to an integer. - operator int(); -}; ------------------ - -=== Compiler API - -To compile a source, a user program must first construct an instance of the `sycl::ext::intel::online_compiler` class. Then pass the source as a `std::string` object to online compiler's `compile` function along with other relevant parameters. The `online_compiler` is templated by the source language, and the `compile` function is a variadic template function. Instantiations of the `online_compiler::compile` for different languages may have different sets of formal parameters. The `compile` function returns a binary blob - a `std::vector` - with the device code compiled according to the compilation target specification provided at online compiler construction time. - -==== Online compiler -[source,c++] ------------------ -template class online_compiler; ------------------ - -==== Compilation target specification elements. -[cols="40,60",options="header"] -|=== -|Element name and type |Description - -|`compiled_code_format` OutputFormat -|Compiled code format. - -|`std::pair` OutputFormatVersion -|Compiled code format version - a pair of "major" and "minor" components. - -|`sycl::info::device_type` DeviceType -|Target device type. - -|`device_arch` DeviceArch -|Target device architecture. - -|`bool` Is64Bit -|Whether the target device architecture is 64-bit. - -|`std::string` DeviceStepping -|Target device stepping (implementation defined). -|=== - -Online compiler construction or source compilation may be unsuccessful, in which case an instance -of `sycl::ext::intel::online_compile_error` is thrown. For example, when some of the compilation -target specification elements are not supported by the implementation, or there is a syntax error -in the source program. - - -==== `sycl::ext::intel::online_compiler` constructors. -[cols="40,60",options="header"] -|=== -|Constructor |Description - -|`online_compiler(compiled_code_format fmt = compiled_code_format::spir_v)` -| Constructs online compiler which can target any device and produces - given compiled code format. Produced device code is 64-bit. OutputFormatVersion is - implementation defined. The created compiler is "optimistic" - it assumes all applicable SYCL - device capabilities are supported by the target device(s). - -|`online_compiler( - sycl::info::device_type dev_type, - device_arch arch, - compiled_code_format fmt = compiled_code_format::spir_v)` -| Constructor version which allows to specify target device type and architecture. - -|`online_compiler(const sycl::device &dev)` -|Constructs online compiler for the target specified by given SYCL device. -|=== - -==== The compilation function - `online_compiler::compile` -It compiles given in-memory source to a binary blob. Blob format, -other parameters are set in the constructor. Specialization for each language will provide exact -signatures, which can be different for different languages.Throws `online_compile_error` if -compilation is not successful. -[source,c++] ------------------ -template - std::vector compile(const std::string &src, const Tys&... args); ------------------ - -Instantiations of the compilation function: -[source,c++] ------------------ -/// Compiles given OpenCL source. May throw \c online_compile_error. -/// @param src - contents of the source -/// @param options - compilation options (implementation defined); standard -/// OpenCL JIT compiler options must be supported -template <> -template <> -std::vector online_compiler::compile( - const std::string &src, const std::vector &options); - -/// Compiles given CM source. -template <> -template <> -std::vector online_compiler::compile( - const std::string &src); - -/// Compiles given CM source. -/// @param options - compilation options (implementation defined) -template <> -template <> -std::vector online_compiler::compile( - const std::string &src, const std::vector &options); ------------------ - -== API usage example -This example compiles an OpenCL source to a generic SPIR-V. -[source,c++] ------------------ -#include "sycl/ext/intel/online_compiler.hpp" - -#include -#include - -static const char *kernelSource = R"===( -__kernel void my_kernel(__global int *in, __global int *out) { - size_t i = get_global_id(0); - out[i] = in[i] + 1; -} -)==="; - -using namespace sycl::INTEL; - -int main(int argc, char **argv) { - online_compiler compiler; - std::vector blob; - - try { - blob = compiler.compile( - std::string(kernelSource), - std::vector { - std::string("-cl-fast-relaxed-math") - } - ); - } - catch (online_compile_error &e) { - std::cout << "compilation failed\n"; - return 1; - } - return 0; -} ------------------ diff --git a/sycl/doc/extensions/experimental/sycl_ext_intel_online_compiler.asciidoc b/sycl/doc/extensions/experimental/sycl_ext_intel_online_compiler.asciidoc deleted file mode 100644 index 525ad49f90ad1..0000000000000 --- a/sycl/doc/extensions/experimental/sycl_ext_intel_online_compiler.asciidoc +++ /dev/null @@ -1,2 +0,0 @@ -This extension has been deprecated, but the specification is still available -link:../deprecated/sycl_ext_intel_online_compiler.asciidoc[here]. \ No newline at end of file diff --git a/sycl/include/sycl/ext/intel/experimental/online_compiler.hpp b/sycl/include/sycl/ext/intel/experimental/online_compiler.hpp deleted file mode 100644 index 365265e762c86..0000000000000 --- a/sycl/include/sycl/ext/intel/experimental/online_compiler.hpp +++ /dev/null @@ -1,274 +0,0 @@ -//===------- online_compiler.hpp - Online source compilation service ------===// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// - -#pragma once - -#include -#include // for __SYCL_EXPORT -#include - -#include -#include - -namespace sycl { -inline namespace _V1 { -namespace ext::intel::experimental { -namespace detail { -using namespace sycl::detail; -} - -using byte = unsigned char; - -enum class compiled_code_format { - spir_v = 0 // the only format supported for now -}; - -class device_arch { -public: - static constexpr int any = 0; - - device_arch(int Val) : Val(Val) {} - - // TODO1: the list must be extended with a bunch of new GPUs available. - // TODO2: the list of supported GPUs grows rapidly. - // The API must allow user to define the target GPU option even if it is - // not listed in this enumerator below. - enum gpu { - gpu_any = 1, - gpu_gen9 = 2, - gpu_skl = gpu_gen9, - gpu_gen9_5 = 3, - gpu_kbl = gpu_gen9_5, - gpu_cfl = gpu_gen9_5, - gpu_gen11 = 4, - gpu_icl = gpu_gen11, - gpu_gen12 = 5, - gpu_tgl = gpu_gen12, - gpu_tgllp = gpu_gen12 - }; - - enum cpu { - cpu_any = 1, - }; - - enum fpga { - fpga_any = 1, - }; - - operator int() { return Val; } - -private: - int Val; -}; - -/// Represents an error happend during online compilation. -class online_compile_error : public sycl::exception { -public: - online_compile_error() = default; - online_compile_error(const std::string &Msg) - : sycl::exception(make_error_code(errc::invalid), Msg) {} -}; - -/// Designates a source language for the online compiler. -enum class source_language { opencl_c = 0, cm = 1 }; - -/// Represents an online compiler for the language given as template -/// parameter. -template -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 - compile_impl(detail::string_view Src, - const std::vector &Options, - std::pair OutputFormatVersion, - sycl::info::device_type DeviceType, device_arch DeviceArch, - bool Is64Bit, detail::string_view DeviceStepping, - void *&CompileToSPIRVHandle, void *&FreeSPIRVOutputsHandle); - - std::vector compile_impl(const std::string &Source, - const std::vector &UserArgs) { - std::vector 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. - /// The created compiler is "optimistic" - it assumes all applicable SYCL - /// device capabilities are supported by the target device(s). - online_compiler(compiled_code_format fmt = compiled_code_format::spir_v) - : OutputFormat(fmt), OutputFormatVersion({0, 0}), - DeviceType(sycl::info::device_type::all), DeviceArch(device_arch::any), - Is64Bit(true), DeviceStepping("") {} - - /// Constructs online compiler which targets given architecture and produces - /// given compiled code format. Produces 64-bit device code. - /// Throws online_compile_error if values of constructor arguments are - /// contradictory or not supported - e.g. if the source language is not - /// supported for given device type. - online_compiler(sycl::info::device_type dev_type, device_arch arch, - compiled_code_format fmt = compiled_code_format::spir_v) - : OutputFormat(fmt), OutputFormatVersion({0, 0}), DeviceType(dev_type), - DeviceArch(arch), Is64Bit(true), DeviceStepping("") {} - - /// Constructs online compiler for the target specified by given SYCL device. - // TODO: the initial version generates the generic code (SKL now), need - // to do additional device::info calls to determine the device by it's - // features. - online_compiler(const sycl::device &) - : OutputFormat(compiled_code_format::spir_v), OutputFormatVersion({0, 0}), - DeviceType(sycl::info::device_type::all), DeviceArch(device_arch::any), - Is64Bit(true), DeviceStepping("") {} - - /// Compiles given in-memory \c Lang source to a binary blob. Blob format, - /// other parameters are set in the constructor by the compilation target - /// specification parameters. - /// Specialization for each language will provide exact signatures, which - /// can be different for different languages. - /// Throws online_compile_error if compilation is not successful. - template - std::vector compile(const std::string &src, const Tys &...args); - - /// Sets the compiled code format of the compilation target and returns *this. - online_compiler &setOutputFormat(compiled_code_format fmt) { - OutputFormat = fmt; - return *this; - } - - /// Sets the compiled code format version of the compilation target and - /// returns *this. - online_compiler &setOutputFormatVersion(int major, int minor) { - OutputFormatVersion = {major, minor}; - return *this; - } - - /// Sets the device type of the compilation target and returns *this. - online_compiler &setTargetDeviceType(sycl::info::device_type type) { - DeviceType = type; - return *this; - } - - /// Sets the device architecture of the compilation target and returns *this. - online_compiler &setTargetDeviceArch(device_arch arch) { - DeviceArch = arch; - return *this; - } - - /// Makes the compilation target 32-bit and returns *this. - online_compiler &set32bitTarget() { - Is64Bit = false; - return *this; - }; - - /// Makes the compilation target 64-bit and returns *this. - online_compiler &set64bitTarget() { - Is64Bit = true; - return *this; - }; - - /// Sets implementation-defined target device stepping of the compilation - /// target and returns *this. - online_compiler &setTargetDeviceStepping(const std::string &id) { - DeviceStepping = id; - return *this; - } - -private: - /// Compiled code format. - compiled_code_format OutputFormat; - - /// Compiled code format version - a pair of "major" and "minor" components - std::pair OutputFormatVersion; - - /// Target device type - sycl::info::device_type DeviceType; - - /// Target device architecture - device_arch DeviceArch; - - /// Whether the target device architecture is 64-bit - bool Is64Bit; - - /// Target device stepping (implementation defined) - std::string DeviceStepping; - - /// Handles to helper functions used by the implementation. - void *CompileToSPIRVHandle = nullptr; - void *FreeSPIRVOutputsHandle = nullptr; -}; - -// Specializations of the online_compiler class and 'compile' function for -// particular languages and parameter types. - -/// Compiles the given OpenCL source. May throw \c online_compile_error. -/// @param src - contents of the source. -/// @param options - compilation options (implementation defined); standard -/// OpenCL JIT compiler options must be supported. -template <> -template <> -#if !defined(__SYCL_ONLINE_COMPILER_CPP) || \ - defined(__INTEL_PREVIEW_BREAKING_CHANGES) -inline -#else -__SYCL_EXPORT -#endif - std::vector - online_compiler::compile( - const std::string &src, const std::vector &options) { - return compile_impl(src, options); -} - -/// Compiles the given OpenCL source. May throw \c online_compile_error. -/// @param src - contents of the source. -template <> -template <> -std::vector -online_compiler::compile(const std::string &src) { - return compile(src, std::vector{}); -} - -/// Compiles the given CM source \p src. -/// @param src - contents of the source. -/// @param options - compilation options (implementation defined). -template <> -template <> -#if !defined(__SYCL_ONLINE_COMPILER_CPP) || \ - defined(__INTEL_PREVIEW_BREAKING_CHANGES) -inline -#else -__SYCL_EXPORT -#endif - std::vector - online_compiler::compile( - const std::string &src, const std::vector &options) { - return compile_impl(src, options); -} - -/// Compiles the given CM source \p src. -template <> -template <> -std::vector -online_compiler::compile(const std::string &src) { - return compile(src, std::vector{}); -} - -} // namespace ext::intel::experimental -} // namespace _V1 -} // namespace sycl diff --git a/sycl/include/sycl/ext/intel/online_compiler.hpp b/sycl/include/sycl/ext/intel/online_compiler.hpp deleted file mode 100644 index 7cc4a35cde6b2..0000000000000 --- a/sycl/include/sycl/ext/intel/online_compiler.hpp +++ /dev/null @@ -1,20 +0,0 @@ -//===------- online_compiler.hpp - Online source compilation service ------===// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// - -#pragma once - -#include - -#if !defined(_MSC_VER) || defined(__clang__) -// MSVC doesn't support #warning and we cannot use other methods to report a -// warning from inside a system header (which SYCL is considered to be). -#warning sycl/ext/intel/online_compiler.hpp usage is deprecated, \ -include sycl/ext/intel/experimental/online_compiler.hpp instead -#endif - -#include diff --git a/sycl/source/CMakeLists.txt b/sycl/source/CMakeLists.txt index d9f8801d45ba5..b60e70662f058 100644 --- a/sycl/source/CMakeLists.txt +++ b/sycl/source/CMakeLists.txt @@ -282,7 +282,6 @@ set(SYCL_COMMON_SOURCES "detail/platform_impl.cpp" "detail/program_manager/program_manager.cpp" "detail/queue_impl.cpp" - "detail/online_compiler/online_compiler.cpp" "detail/os_util.cpp" "detail/persistent_device_code_cache.cpp" "detail/platform_util.cpp" diff --git a/sycl/source/detail/kernel_compiler/kernel_compiler_opencl.cpp b/sycl/source/detail/kernel_compiler/kernel_compiler_opencl.cpp index 5452bf40795dd..63907ff913dca 100644 --- a/sycl/source/detail/kernel_compiler/kernel_compiler_opencl.cpp +++ b/sycl/source/detail/kernel_compiler/kernel_compiler_opencl.cpp @@ -11,8 +11,8 @@ #include "kernel_compiler_opencl.hpp" -#include "../online_compiler/ocloc_api.h" #include "../split_string.hpp" +#include "ocloc_api.h" #include // strlen #include // for std::function diff --git a/sycl/source/detail/online_compiler/ocloc_api.h b/sycl/source/detail/kernel_compiler/ocloc_api.h similarity index 100% rename from sycl/source/detail/online_compiler/ocloc_api.h rename to sycl/source/detail/kernel_compiler/ocloc_api.h diff --git a/sycl/source/detail/online_compiler/online_compiler.cpp b/sycl/source/detail/online_compiler/online_compiler.cpp deleted file mode 100644 index 138fc880edb92..0000000000000 --- a/sycl/source/detail/online_compiler/online_compiler.cpp +++ /dev/null @@ -1,267 +0,0 @@ -//==----------- online_compiler.cpp ----------------------------------------==// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// - -#define __SYCL_ONLINE_COMPILER_CPP - -#include -#include -#include - -#include - -#include "ocloc_api.h" - -namespace sycl { -inline namespace _V1 { -namespace ext::intel::experimental { -namespace detail { - -using namespace sycl::detail; - -static std::vector -prepareOclocArgs(sycl::info::device_type DeviceType, device_arch DeviceArch, - bool Is64Bit, string_view DeviceStepping, - const std::string &UserArgs) { - std::vector Args = {"ocloc", "-q", "-spv_only", "-device"}; - - if (DeviceType == sycl::info::device_type::gpu) { - switch (DeviceArch) { - case device_arch::gpu_gen9: - Args.push_back("skl"); - break; - - case device_arch::gpu_gen9_5: - Args.push_back("cfl"); - break; - - case device_arch::gpu_gen11: - Args.push_back("icllp"); - break; - - case device_arch::gpu_gen12: - Args.push_back("tgllp"); - break; - - default: - Args.push_back("tgllp"); - } - } else { - // TODO: change that to generic device when ocloc adds support for it. - // For now "tgllp" is used as the option supported on all known GPU RT. - Args.push_back("tgllp"); - } - - if (DeviceStepping != "") { - Args.push_back("-revision_id"); - Args.push_back(DeviceStepping.data()); - } - - Args.push_back(Is64Bit ? "-64" : "-32"); - - if (UserArgs != "") { - Args.push_back("-options"); - Args.push_back(UserArgs.c_str()); - } - - return Args; -} - -/// Compiles the given source \p Source to SPIR-V IL and returns IL as a vector -/// of bytes. -/// @param Source - Either OpenCL or CM source code. -/// @param DeviceType - SYCL device type, e.g. cpu, gpu, accelerator, etc. -/// @param DeviceArch - More detailed info on the target device architecture. -/// @param Is64Bit - If set to true, specifies the 64-bit architecture. -/// Otherwise, 32-bit is assumed. -/// @param DeviceStepping - implementation specific target device stepping. -/// @param CompileToSPIRVHandle - Output parameter. It is set to the address -/// of the library function doing the compilation. -/// @param FreeSPIRVOutputsHandle - Output parameter. It is set to the address -/// of the library function freeing memory -/// allocated during the compilation. -/// @param UserArgs - User's options to ocloc compiler. -static std::vector -compileToSPIRV(string_view Src, sycl::info::device_type DeviceType, - device_arch DeviceArch, bool Is64Bit, string_view DeviceStepping, - void *&CompileToSPIRVHandle, void *&FreeSPIRVOutputsHandle, - const std::vector &UserArgs) { - std::string Source{Src.data()}; - - if (!CompileToSPIRVHandle) { -#ifdef __SYCL_RT_OS_WINDOWS - static const std::string OclocLibraryName = "ocloc64.dll"; -#else - static const std::string OclocLibraryName = "libocloc.so"; -#endif - auto CustomDeleter = [](void *StoredPtr) { - if (!StoredPtr) - return; - std::ignore = sycl::detail::ur::unloadOsLibrary(StoredPtr); - }; - std::unique_ptr OclocLibrary( - sycl::detail::ur::loadOsLibrary(OclocLibraryName), CustomDeleter); - if (!OclocLibrary) - throw online_compile_error("Cannot load ocloc library: " + - OclocLibraryName); - void *OclocVersionHandle = sycl::detail::ur::getOsLibraryFuncAddress( - OclocLibrary.get(), "oclocVersion"); - // The initial versions of ocloc library did not have the oclocVersion() - // function. Those versions had the same API as the first version of ocloc - // library having that oclocVersion() function. - int LoadedVersion = ocloc_version_t::OCLOC_VERSION_1_0; - if (OclocVersionHandle) { - decltype(::oclocVersion) *OclocVersionFunc = - reinterpret_cast(OclocVersionHandle); - LoadedVersion = OclocVersionFunc(); - } - // The loaded library with version (A.B) is compatible with expected API/ABI - // version (X.Y) used here if A == B and B >= Y. - int LoadedVersionMajor = LoadedVersion >> 16; - int LoadedVersionMinor = LoadedVersion & 0xffff; - int CurrentVersionMajor = ocloc_version_t::OCLOC_VERSION_CURRENT >> 16; - int CurrentVersionMinor = ocloc_version_t::OCLOC_VERSION_CURRENT & 0xffff; - if (LoadedVersionMajor != CurrentVersionMajor || - LoadedVersionMinor < CurrentVersionMinor) - throw online_compile_error( - std::string("Found incompatible version of ocloc library: (") + - std::to_string(LoadedVersionMajor) + "." + - std::to_string(LoadedVersionMinor) + - "). The supported versions are (" + - std::to_string(CurrentVersionMajor) + - ".N), where (N >= " + std::to_string(CurrentVersionMinor) + ")."); - - CompileToSPIRVHandle = sycl::detail::ur::getOsLibraryFuncAddress( - OclocLibrary.get(), "oclocInvoke"); - if (!CompileToSPIRVHandle) - throw online_compile_error("Cannot load oclocInvoke() function"); - FreeSPIRVOutputsHandle = sycl::detail::ur::getOsLibraryFuncAddress( - OclocLibrary.get(), "oclocFreeOutput"); - if (!FreeSPIRVOutputsHandle) { - CompileToSPIRVHandle = NULL; - throw online_compile_error("Cannot load oclocFreeOutput() function"); - } - OclocLibrary.release(); - } - - std::string CombinedUserArgs; - for (const auto &UserArg : UserArgs) { - if (UserArg == "") - continue; - if (CombinedUserArgs != "") - CombinedUserArgs = CombinedUserArgs + " " + UserArg; - else - CombinedUserArgs = UserArg; - } - std::vector Args = detail::prepareOclocArgs( - DeviceType, DeviceArch, Is64Bit, DeviceStepping, CombinedUserArgs); - - uint32_t NumOutputs = 0; - byte **Outputs = nullptr; - uint64_t *OutputLengths = nullptr; - char **OutputNames = nullptr; - - const byte *Sources[] = {reinterpret_cast(Source.c_str())}; - const char *SourceName = "main.cl"; - const uint64_t SourceLengths[] = {Source.length() + 1}; - - Args.push_back("-file"); - Args.push_back(SourceName); - - decltype(::oclocInvoke) *OclocInvokeFunc = - reinterpret_cast(CompileToSPIRVHandle); - int CompileError = - OclocInvokeFunc(Args.size(), Args.data(), 1, Sources, SourceLengths, - &SourceName, 0, nullptr, nullptr, nullptr, &NumOutputs, - &Outputs, &OutputLengths, &OutputNames); - - std::vector SpirV; - std::string CompileLog; - for (uint32_t I = 0; I < NumOutputs; I++) { - size_t NameLen = strlen(OutputNames[I]); - if (NameLen >= 4 && strstr(OutputNames[I], ".spv") != nullptr && - Outputs[I] != nullptr) { - assert(SpirV.size() == 0 && "More than one SPIR-V output found."); - SpirV = std::vector(Outputs[I], Outputs[I] + OutputLengths[I]); - } else if (!strcmp(OutputNames[I], "stdout.log")) { - CompileLog = std::string(reinterpret_cast(Outputs[I])); - } - } - - // Try to free memory before reporting possible error. - decltype(::oclocFreeOutput) *OclocFreeOutputFunc = - reinterpret_cast(FreeSPIRVOutputsHandle); - int MemFreeError = - OclocFreeOutputFunc(&NumOutputs, &Outputs, &OutputLengths, &OutputNames); - - if (CompileError) - throw online_compile_error("ocloc reported compilation errors: {\n" + - CompileLog + "\n}"); - if (SpirV.empty()) - throw online_compile_error( - "Unexpected output: ocloc did not return SPIR-V"); - if (MemFreeError) - throw online_compile_error("ocloc cannot safely free resources"); - - return SpirV; -} -} // namespace detail - -template -__SYCL_EXPORT std::vector online_compiler::compile_impl( - detail::string_view Src, const std::vector &Options, - std::pair OutputFormatVersion, sycl::info::device_type DeviceType, - device_arch DeviceArch, bool Is64Bit, detail::string_view DeviceStepping, - void *&CompileToSPIRVHandle, void *&FreeSPIRVOutputsHandle) { - - if (OutputFormatVersion != std::pair{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"); - } - - std::vector UserArgs; - for (auto &&Opt : Options) - UserArgs.emplace_back(Opt.data()); - - if constexpr (Lang == source_language::cm) - UserArgs.push_back("-cmc"); - - return detail::compileToSPIRV(Src, DeviceType, DeviceArch, Is64Bit, - DeviceStepping, CompileToSPIRVHandle, - FreeSPIRVOutputsHandle, UserArgs); -} - -template __SYCL_EXPORT std::vector -online_compiler::compile_impl( - detail::string_view Src, const std::vector &Options, - std::pair OutputFormatVersion, sycl::info::device_type DeviceType, - device_arch DeviceArch, bool Is64Bit, detail::string_view DeviceStepping, - void *&CompileToSPIRVHandle, void *&FreeSPIRVOutputsHandle); - -template __SYCL_EXPORT std::vector -online_compiler::compile_impl( - detail::string_view Src, const std::vector &Options, - std::pair 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 { -namespace __SYCL2020_DEPRECATED( - "use 'ext::intel::experimental' instead") intel { -using namespace ext::intel::experimental; -} -} // namespace ext - -namespace __SYCL2020_DEPRECATED( - "use 'ext::intel::experimental' instead") INTEL { -using namespace ext::intel::experimental; -} -} // namespace _V1 -} // namespace sycl diff --git a/sycl/test-e2e/OnlineCompiler/online_compiler_L0.cpp b/sycl/test-e2e/OnlineCompiler/online_compiler_L0.cpp deleted file mode 100644 index 4de91a66941aa..0000000000000 --- a/sycl/test-e2e/OnlineCompiler/online_compiler_L0.cpp +++ /dev/null @@ -1,79 +0,0 @@ -// REQUIRES: level_zero, level_zero_dev_kit, cm-compiler -// XFAIL: gpu -// XFAIL-TRACKER: https://github.com/intel/llvm/issues/16406 -// RUN: %{build} -Wno-error=deprecated-declarations -DRUN_KERNELS %level_zero_options -o %t.out -// RUN: %{run} %t.out - -// This test checks ext::intel feature class online_compiler for Level-Zero. -// All Level-Zero specific code is kept here and the common part that can be -// re-used by other backends is kept in online_compiler_common.hpp file. - -#include -#include - -#include - -// clang-format off -#include -#include -// clang-format on - -using byte = unsigned char; - -#ifdef RUN_KERNELS -bool testSupported(sycl::queue &Queue) { - return Queue.get_backend() == sycl::backend::ext_oneapi_level_zero; -} - -sycl::kernel getSYCLKernelWithIL(sycl::queue &Queue, - const std::vector &IL) { - - ze_module_desc_t ZeModuleDesc = {}; - ZeModuleDesc.format = ZE_MODULE_FORMAT_IL_SPIRV; - ZeModuleDesc.inputSize = IL.size(); - ZeModuleDesc.pInputModule = IL.data(); - ZeModuleDesc.pBuildFlags = ""; - ZeModuleDesc.pConstants = nullptr; - - sycl::context Context = Queue.get_context(); - sycl::device Device = Queue.get_device(); - auto ZeDevice = - sycl::get_native(Device); - auto ZeContext = - sycl::get_native(Context); - - ze_module_build_log_handle_t ZeBuildLog; - ze_module_handle_t ZeModule; - ze_result_t ZeResult = zeModuleCreate(ZeContext, ZeDevice, &ZeModuleDesc, - &ZeModule, &ZeBuildLog); - assert(ZeResult == ZE_RESULT_SUCCESS); - - ze_kernel_handle_t ZeKernel = nullptr; - - ze_kernel_desc_t ZeKernelDesc{ZE_STRUCTURE_TYPE_KERNEL_DESC, nullptr, 0, - "my_kernel"}; - ZeResult = zeKernelCreate(ZeModule, &ZeKernelDesc, &ZeKernel); - assert(ZeResult == ZE_RESULT_SUCCESS); - sycl::kernel_bundle SyclKB = - sycl::make_kernel_bundle( - {ZeModule, sycl::ext::oneapi::level_zero::ownership::keep}, Context); - - auto Kernel = sycl::make_kernel( - {SyclKB, ZeKernel, sycl::ext::oneapi::level_zero::ownership::keep}, - Context); - - // Should not throw an exception - try { - auto num_args = Kernel.get_info(); - (void)num_args; - } catch (sycl::exception &e) { - assert(false && "Using \"info::kernel::num_args\" query for valid kernel " - "should not throw an exception."); - } - - return Kernel; -} -#endif // RUN_KERNELS - -#include "online_compiler_common.hpp" diff --git a/sycl/test-e2e/OnlineCompiler/online_compiler_OpenCL.cpp b/sycl/test-e2e/OnlineCompiler/online_compiler_OpenCL.cpp deleted file mode 100644 index b0023426f0631..0000000000000 --- a/sycl/test-e2e/OnlineCompiler/online_compiler_OpenCL.cpp +++ /dev/null @@ -1,113 +0,0 @@ -// REQUIRES: opencl, opencl_icd, cm-compiler -// XFAIL: gpu || cpu || accelerator -// XFAIL-TRACKER: https://github.com/intel/llvm/issues/16406 -// RUN: %{build} -Wno-error=deprecated-declarations -DRUN_KERNELS %opencl_lib -o %t.out -// RUN: %{run} %t.out - -// This test checks ext::intel feature class online_compiler for OpenCL. -// All OpenCL specific code is kept here and the common part that can be -// re-used by other backends is kept in online_compiler_common.hpp file. - -#include -#include -#include -#include - -#include - -using byte = unsigned char; - -#ifdef RUN_KERNELS -std::tuple GetOCLVersion(sycl::device Device) { - cl_int Err; - cl_device_id ClDevice = sycl::get_native(Device); - - size_t VersionSize = 0; - Err = clGetDeviceInfo(ClDevice, CL_DEVICE_VERSION, 0, nullptr, &VersionSize); - assert(Err == CL_SUCCESS); - - std::string Version(VersionSize, '\0'); - Err = clGetDeviceInfo(ClDevice, CL_DEVICE_VERSION, VersionSize, - Version.data(), nullptr); - assert(Err == CL_SUCCESS); - - std::string_view Prefix = "OpenCL "; - size_t VersionBegin = Version.find_first_of(" "); - size_t VersionEnd = Version.find_first_of(" ", VersionBegin + 1); - size_t VersionSeparator = Version.find_first_of(".", VersionBegin + 1); - - bool HaveOCLPrefix = - std::equal(Prefix.begin(), Prefix.end(), Version.begin()); - - assert(HaveOCLPrefix && VersionBegin != std::string::npos && - VersionEnd != std::string::npos && - VersionSeparator != std::string::npos); - - std::string VersionMajor{Version.begin() + VersionBegin + 1, - Version.begin() + VersionSeparator}; - std::string VersionMinor{Version.begin() + VersionSeparator + 1, - Version.begin() + VersionEnd}; - - unsigned long OCLMajor = strtoul(VersionMajor.c_str(), nullptr, 10); - unsigned long OCLMinor = strtoul(VersionMinor.c_str(), nullptr, 10); - - assert(OCLMajor > 0 && (OCLMajor > 2 || OCLMinor <= 2) && - OCLMajor != UINT_MAX && OCLMinor != UINT_MAX); - - return std::make_tuple(OCLMajor, OCLMinor); -} - -bool testSupported(sycl::queue &Queue) { - if (Queue.get_backend() != sycl::backend::opencl) - return false; - - sycl::device Device = Queue.get_device(); - auto [OCLMajor, OCLMinor] = GetOCLVersion(Device); - - // Creating a program from IL is only supported on >=2.1 or if - // cl_khr_il_program is supported on the device. - return (OCLMajor == 2 && OCLMinor >= 1) || OCLMajor > 2 || - Device.has_extension("cl_khr_il_program"); -} - -sycl::kernel getSYCLKernelWithIL(sycl::queue &Queue, - const std::vector &IL) { - sycl::context Context = Queue.get_context(); - - cl_int Err = 0; - cl_program ClProgram = 0; - - sycl::device Device = Queue.get_device(); - auto [OCLMajor, OCLMinor] = GetOCLVersion(Device); - if ((OCLMajor == 2 && OCLMinor >= 1) || OCLMajor > 2) { - // clCreateProgramWithIL is supported if OCL version >=2.1. - ClProgram = - clCreateProgramWithIL(sycl::get_native(Context), - IL.data(), IL.size(), &Err); - } else { - // Fall back to using extension function for building IR. - using ApiFuncT = - cl_program(CL_API_CALL *)(cl_context, const void *, size_t, cl_int *); - ApiFuncT FuncPtr = - reinterpret_cast(clGetExtensionFunctionAddressForPlatform( - sycl::get_native(Context.get_platform()), - "clCreateProgramWithILKHR")); - - assert(FuncPtr != nullptr); - - ClProgram = FuncPtr(sycl::get_native(Context), - IL.data(), IL.size(), &Err); - } - assert(Err == CL_SUCCESS); - - Err = clBuildProgram(ClProgram, 0, nullptr, nullptr, nullptr, nullptr); - assert(Err == CL_SUCCESS); - - cl_kernel ClKernel = clCreateKernel(ClProgram, "my_kernel", &Err); - assert(Err == CL_SUCCESS); - - return sycl::make_kernel(ClKernel, Context); -} -#endif // RUN_KERNELS - -#include "online_compiler_common.hpp" diff --git a/sycl/test-e2e/OnlineCompiler/online_compiler_common.hpp b/sycl/test-e2e/OnlineCompiler/online_compiler_common.hpp deleted file mode 100644 index b585126f95674..0000000000000 --- a/sycl/test-e2e/OnlineCompiler/online_compiler_common.hpp +++ /dev/null @@ -1,193 +0,0 @@ -#include -#include - -#include -#include - -auto constexpr CLSource = R"===( -__kernel void my_kernel(__global int *in, __global int *out) { - size_t i = get_global_id(0); - out[i] = in[i]*2 + 100; -} -)==="; - -auto constexpr CLSourceSyntaxError = R"===( -__kernel void my_kernel(__global int *in, __global int *out) { - syntax error here - size_t i = get_global_id(0); - out[i] = in[i]*2 + 100; -} -)==="; - -auto constexpr CMSource = R"===( -extern "C" -void cm_kernel() { -} -)==="; - -using namespace sycl::ext::intel; - -#ifdef RUN_KERNELS -void testSyclKernel(sycl::queue &Q, sycl::kernel Kernel) { - std::cout << "Run the kernel now:\n"; - constexpr int N = 4; - int InputArray[N] = {0, 1, 2, 3}; - int OutputArray[N] = {}; - - sycl::buffer InputBuf(InputArray, sycl::range<1>(N)); - sycl::buffer OutputBuf(OutputArray, sycl::range<1>(N)); - - Q.submit([&](sycl::handler &CGH) { - CGH.set_arg(0, InputBuf.get_access(CGH)); - CGH.set_arg(1, OutputBuf.get_access(CGH)); - CGH.parallel_for(sycl::range<1>{N}, Kernel); - }); - - auto Out = OutputBuf.get_host_access(); - for (int I = 0; I < N; I++) - std::cout << I << "*2 + 100 = " << Out[I] << "\n"; -} -#endif // RUN_KERNELS - -int main(int argc, char **argv) { - sycl::queue Q; - sycl::device Device = Q.get_device(); - -#ifdef RUN_KERNELS - if (!testSupported(Q)) { - std::cout << "Building for IL is not supported. Skipping!" << std::endl; - return 0; - } -#endif - - { // Compile and run a trivial OpenCL kernel. - std::cout << "Test case1\n"; - sycl::ext::intel::experimental::online_compiler< - sycl::ext::intel::experimental::source_language::opencl_c> - Compiler; - std::vector IL; - try { - IL = Compiler.compile( - CLSource, - // Pass two options to check that more than one is accepted. - std::vector{"-cl-fast-relaxed-math", - "-cl-finite-math-only"}); - std::cout << "IL size = " << IL.size() << "\n"; - assert(IL.size() > 0 && "Unexpected IL size"); - } catch (sycl::exception &e) { - std::cout << "Compilation to IL failed: " << e.what() << "\n"; - return 1; - } -#ifdef RUN_KERNELS - testSyclKernel(Q, getSYCLKernelWithIL(Q, IL)); -#endif // RUN_KERNELS - } - - { // Compile and run a trivial OpenCL kernel using online_compiler() - // constructor accepting SYCL device. - std::cout << "Test case2\n"; - sycl::ext::intel::experimental::online_compiler< - sycl::ext::intel::experimental::source_language::opencl_c> - Compiler(Device); - std::vector IL; - try { - IL = Compiler.compile(CLSource); - std::cout << "IL size = " << IL.size() << "\n"; - assert(IL.size() > 0 && "Unexpected IL size"); - } catch (sycl::exception &e) { - std::cout << "Compilation to IL failed: " << e.what() << "\n"; - return 1; - } -#ifdef RUN_KERNELS - testSyclKernel(Q, getSYCLKernelWithIL(Q, IL)); -#endif // RUN_KERNELS - } - - // TODO: this test is temporarily turned off because CI buildbots do not set - // PATHs to clangFEWrapper library properly. - { // Compile a trivial CM kernel. - std::cout << "Test case3\n"; - sycl::ext::intel::experimental::online_compiler< - sycl::ext::intel::experimental::source_language::cm> - Compiler; - try { - std::vector IL = Compiler.compile(CMSource); - - std::cout << "IL size = " << IL.size() << "\n"; - assert(IL.size() > 0 && "Unexpected IL size"); - } catch (sycl::exception &e) { - std::cout << "Compilation to IL failed: " << e.what() << "\n"; - return 1; - } - } - - { // Compile a source with syntax errors. - std::cout << "Test case4\n"; - sycl::ext::intel::experimental::online_compiler< - sycl::ext::intel::experimental::source_language::opencl_c> - Compiler; - std::vector IL; - bool TestPassed = false; - try { - IL = Compiler.compile(CLSourceSyntaxError); - } catch (sycl::exception &e) { - std::string Msg = e.what(); - if (Msg.find("syntax error here") != std::string::npos) - TestPassed = true; - else - std::cerr << "Unexpected exception: " << Msg << "\n"; - } - assert(TestPassed && "Failed to throw an exception for syntax error"); - if (!TestPassed) - return 1; - } - - { // Compile a good CL source using unrecognized compilation options. - std::cout << "Test case5\n"; - sycl::ext::intel::experimental::online_compiler< - sycl::ext::intel::experimental::source_language::opencl_c> - Compiler; - std::vector IL; - bool TestPassed = false; - try { - IL = Compiler.compile(CLSource, - // Intentionally use incorrect option. - std::vector{"WRONG_OPTION"}); - } catch (sycl::exception &e) { - std::string Msg = e.what(); - if (Msg.find("WRONG_OPTION") != std::string::npos) - TestPassed = true; - else - std::cerr << "Unexpected exception: " << Msg << "\n"; - } - assert(TestPassed && - "Failed to throw an exception for unrecognized option"); - if (!TestPassed) - return 1; - } - - { // Try compiling CM source with OpenCL compiler. - std::cout << "Test case6\n"; - sycl::ext::intel::experimental::online_compiler< - sycl::ext::intel::experimental::source_language::opencl_c> - Compiler; - std::vector IL; - bool TestPassed = false; - try { - // Intentionally pass CMSource instead of CLSource. - IL = Compiler.compile(CMSource); - } catch (sycl::exception &e) { - std::string Msg = e.what(); - if (Msg.find("error: expected identifier or '('") != std::string::npos) - TestPassed = true; - else - std::cerr << "Unexpected exception: " << Msg << "\n"; - } - assert(TestPassed && "Failed to throw an exception for wrong program"); - if (!TestPassed) - return 1; - } - - std::cout << "\nAll test cases passed.\n"; - return 0; -} diff --git a/sycl/test/abi/sycl_abi_neutrality_test.cpp b/sycl/test/abi/sycl_abi_neutrality_test.cpp index 6920e60031dba..d1a4a8df9bd6c 100644 --- a/sycl/test/abi/sycl_abi_neutrality_test.cpp +++ b/sycl/test/abi/sycl_abi_neutrality_test.cpp @@ -19,10 +19,6 @@ // old entry points. Others were exported unnecessarily but only actually used // inside DSO, yet we have to keep the entry points as well. -// https://github.com/intel/llvm/pull/16179 -// CHECK:_ZN4sycl3_V13ext5intel12experimental15online_compilerILNS3_15source_languageE0EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISE_EEEEES8_IhSaIhEERKSE_DpRKT_ -// CHECK:_ZN4sycl3_V13ext5intel12experimental15online_compilerILNS3_15source_languageE1EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISE_EEEEES8_IhSaIhEERKSE_DpRKT_ -// // https://github.com/intel/llvm/pull/16178 // CHECK:_ZN4sycl3_V13ext5intel12experimental9pipe_base13get_pipe_nameB5cxx11EPKv // diff --git a/sycl/test/abi/sycl_symbols_linux.dump b/sycl/test/abi/sycl_symbols_linux.dump index 98dee24572890..224dd494a6258 100644 --- a/sycl/test/abi/sycl_symbols_linux.dump +++ b/sycl/test/abi/sycl_symbols_linux.dump @@ -2985,10 +2985,6 @@ _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 _ZN4sycl3_V13ext5intel12experimental9pipe_base18get_pipe_name_implEPKv diff --git a/sycl/test/basic_tests/no_math_in_global_ns.cpp b/sycl/test/basic_tests/no_math_in_global_ns.cpp index c8f582ea45f59..25408c526c7aa 100644 --- a/sycl/test/basic_tests/no_math_in_global_ns.cpp +++ b/sycl/test/basic_tests/no_math_in_global_ns.cpp @@ -22,7 +22,6 @@ #include #include #include -#include #include using namespace sycl; diff --git a/sycl/test/warnings/sycl_2020_deprecations.cpp b/sycl/test/warnings/sycl_2020_deprecations.cpp index 6fd61a0a1137e..d3af96fcf1cc5 100644 --- a/sycl/test/warnings/sycl_2020_deprecations.cpp +++ b/sycl/test/warnings/sycl_2020_deprecations.cpp @@ -2,7 +2,6 @@ // expected-warning@CL/sycl.hpp:* {{CL/sycl.hpp is deprecated, use sycl/sycl.hpp}} #include -#include int main() { cl_context ClCtx; @@ -88,11 +87,6 @@ int main() { // expected-warning@+1{{'exception' is deprecated: The version of an exception constructor which takes no arguments is deprecated.}} sycl::exception ex; - // expected-warning@+1{{'online_compiler' is deprecated}} - sycl::ext::intel::experimental::online_compiler< - sycl::ext::intel::experimental::source_language::opencl_c> - oc(Device); - Queue.submit([](sycl::handler &CGH) { // expected-warning@+3{{'nd_range' is deprecated: offsets are deprecated in SYCL2020}} // expected-warning@+2{{'nd_range' is deprecated: offsets are deprecated in SYCL2020}} From 565c208bbe11eb781378a00038586ead04a08a7e Mon Sep 17 00:00:00 2001 From: "Perkins, Chris" Date: Fri, 24 Jan 2025 14:42:46 -0800 Subject: [PATCH 2/3] windows symbols updated --- sycl/test/abi/sycl_symbols_windows.dump | 30 +++++++++++-------------- 1 file changed, 13 insertions(+), 17 deletions(-) diff --git a/sycl/test/abi/sycl_symbols_windows.dump b/sycl/test/abi/sycl_symbols_windows.dump index edc11b37a071f..da02292f0028b 100644 --- a/sycl/test/abi/sycl_symbols_windows.dump +++ b/sycl/test/abi/sycl_symbols_windows.dump @@ -7,8 +7,6 @@ # REQUIRES: windows # UNSUPPORTED: libcxx -??$compile@V?$vector@V?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@V?$allocator@V?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@@2@@std@@@?$online_compiler@$00@experimental@intel@ext@_V1@sycl@@QEAA?AV?$vector@EV?$allocator@E@std@@@std@@AEBV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@7@AEBV?$vector@V?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@V?$allocator@V?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@@2@@7@@Z -??$compile@V?$vector@V?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@V?$allocator@V?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@@2@@std@@@?$online_compiler@$0A@@experimental@intel@ext@_V1@sycl@@QEAA?AV?$vector@EV?$allocator@E@std@@@std@@AEBV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@7@AEBV?$vector@V?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@V?$allocator@V?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@@2@@7@@Z ??$create_sub_devices@$0BAIG@@device@_V1@sycl@@QEBA?AV?$vector@Vdevice@_V1@sycl@@V?$allocator@Vdevice@_V1@sycl@@@std@@@std@@_K@Z ??$create_sub_devices@$0BAIH@@device@_V1@sycl@@QEBA?AV?$vector@Vdevice@_V1@sycl@@V?$allocator@Vdevice@_V1@sycl@@@std@@@std@@AEBV?$vector@_KV?$allocator@_K@std@@@4@@Z ??$create_sub_devices@$0BAII@@device@_V1@sycl@@QEBA?AV?$vector@Vdevice@_V1@sycl@@V?$allocator@Vdevice@_V1@sycl@@@std@@@std@@W4partition_affinity_domain@info@12@@Z @@ -282,9 +280,9 @@ ??0SampledImageAccessorBaseHost@detail@_V1@sycl@@QEAA@$$QEAV0123@@Z ??0SampledImageAccessorBaseHost@detail@_V1@sycl@@QEAA@AEBV0123@@Z ??0SampledImageAccessorBaseHost@detail@_V1@sycl@@QEAA@V?$range@$02@23@PEAXHHV?$id@$02@23@W4image_channel_type@23@W4image_channel_order@23@Uimage_sampler@23@AEBVproperty_list@23@@Z -??0SubmissionInfo@detail@_V1@sycl@@QEAA@XZ -??0SubmissionInfo@detail@_V1@sycl@@QEAA@AEBV0123@@Z ??0SubmissionInfo@detail@_V1@sycl@@QEAA@$$QEAV0123@@Z +??0SubmissionInfo@detail@_V1@sycl@@QEAA@AEBV0123@@Z +??0SubmissionInfo@detail@_V1@sycl@@QEAA@XZ ??0UnsampledImageAccessorBaseHost@detail@_V1@sycl@@IEAA@AEBV?$shared_ptr@VUnsampledImageAccessorImplHost@detail@_V1@sycl@@@std@@@Z ??0UnsampledImageAccessorBaseHost@detail@_V1@sycl@@QEAA@$$QEAV0123@@Z ??0UnsampledImageAccessorBaseHost@detail@_V1@sycl@@QEAA@AEBV0123@@Z @@ -330,18 +328,12 @@ ??0device_image_plain@detail@_V1@sycl@@QEAA@AEBV?$shared_ptr@Vdevice_image_impl@detail@_V1@sycl@@@std@@@Z ??0device_selector@_V1@sycl@@QEAA@AEBV012@@Z ??0device_selector@_V1@sycl@@QEAA@XZ +??0dynamic_command_group@experimental@oneapi@ext@_V1@sycl@@QEAA@$$QEAV012345@@Z +??0dynamic_command_group@experimental@oneapi@ext@_V1@sycl@@QEAA@AEBV012345@@Z +??0dynamic_command_group@experimental@oneapi@ext@_V1@sycl@@QEAA@AEBV?$command_graph@$0A@@12345@AEBV?$vector@V?$function@$$A6AXAEAVhandler@_V1@sycl@@@Z@std@@V?$allocator@V?$function@$$A6AXAEAVhandler@_V1@sycl@@@Z@std@@@2@@std@@@Z ??0dynamic_parameter_base@detail@experimental@oneapi@ext@_V1@sycl@@QEAA@$$QEAV0123456@@Z ??0dynamic_parameter_base@detail@experimental@oneapi@ext@_V1@sycl@@QEAA@AEBV0123456@@Z ??0dynamic_parameter_base@detail@experimental@oneapi@ext@_V1@sycl@@QEAA@V?$command_graph@$0A@@23456@_KPEBX@Z -?get_active_index@dynamic_command_group@experimental@oneapi@ext@_V1@sycl@@QEBA_KXZ -??4dynamic_command_group@experimental@oneapi@ext@_V1@sycl@@QEAAAEAV012345@$$QEAV012345@@Z -?addImpl@modifiable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@IEAA?AVnode@34567@AEAVdynamic_command_group@34567@AEBV?$vector@Vnode@experimental@oneapi@ext@_V1@sycl@@V?$allocator@Vnode@experimental@oneapi@ext@_V1@sycl@@@std@@@std@@@Z -??0dynamic_command_group@experimental@oneapi@ext@_V1@sycl@@QEAA@AEBV?$command_graph@$0A@@12345@AEBV?$vector@V?$function@$$A6AXAEAVhandler@_V1@sycl@@@Z@std@@V?$allocator@V?$function@$$A6AXAEAVhandler@_V1@sycl@@@Z@std@@@2@@std@@@Z -??0dynamic_command_group@experimental@oneapi@ext@_V1@sycl@@QEAA@$$QEAV012345@@Z -??0dynamic_command_group@experimental@oneapi@ext@_V1@sycl@@QEAA@AEBV012345@@Z -?set_active_index@dynamic_command_group@experimental@oneapi@ext@_V1@sycl@@QEAAX_K@Z -??1dynamic_command_group@experimental@oneapi@ext@_V1@sycl@@QEAA@XZ -??4dynamic_command_group@experimental@oneapi@ext@_V1@sycl@@QEAAAEAV012345@AEBV012345@@Z ??0event@_V1@sycl@@AEAA@V?$shared_ptr@Vevent_impl@detail@_V1@sycl@@@std@@@Z ??0event@_V1@sycl@@QEAA@$$QEAV012@@Z ??0event@_V1@sycl@@QEAA@AEBV012@@Z @@ -479,6 +471,7 @@ ??1device@_V1@sycl@@QEAA@XZ ??1device_image_plain@detail@_V1@sycl@@QEAA@XZ ??1device_selector@_V1@sycl@@UEAA@XZ +??1dynamic_command_group@experimental@oneapi@ext@_V1@sycl@@QEAA@XZ ??1dynamic_parameter_base@detail@experimental@oneapi@ext@_V1@sycl@@QEAA@XZ ??1event@_V1@sycl@@QEAA@XZ ??1exception@_V1@sycl@@UEAA@XZ @@ -556,6 +549,8 @@ ??4device_image_plain@detail@_V1@sycl@@QEAAAEAV0123@$$QEAV0123@@Z ??4device_image_plain@detail@_V1@sycl@@QEAAAEAV0123@AEBV0123@@Z ??4device_selector@_V1@sycl@@QEAAAEAV012@AEBV012@@Z +??4dynamic_command_group@experimental@oneapi@ext@_V1@sycl@@QEAAAEAV012345@$$QEAV012345@@Z +??4dynamic_command_group@experimental@oneapi@ext@_V1@sycl@@QEAAAEAV012345@AEBV012345@@Z ??4dynamic_parameter_base@detail@experimental@oneapi@ext@_V1@sycl@@QEAAAEAV0123456@$$QEAV0123456@@Z ??4dynamic_parameter_base@detail@experimental@oneapi@ext@_V1@sycl@@QEAAAEAV0123456@AEBV0123456@@Z ??4event@_V1@sycl@@QEAAAEAV012@$$QEAV012@@Z @@ -654,8 +649,8 @@ ?GDBMethodsAnchor@UnsampledImageAccessorBaseHost@detail@_V1@sycl@@IEAAXXZ ?GetRangeRoundingSettings@handler@_V1@sycl@@AEAAXAEA_K00@Z ?HasAssociatedAccessor@handler@_V1@sycl@@AEBA_NPEAVAccessorImplHost@detail@23@W4target@access@23@@Z -?PostProcessorFunc@SubmissionInfo@detail@_V1@sycl@@QEBAAEBV?$optional@V?$function@$$A6AX_N0AEAVevent@_V1@sycl@@@Z@std@@@234@XZ ?PostProcessorFunc@SubmissionInfo@detail@_V1@sycl@@QEAAAEAV?$optional@V?$function@$$A6AX_N0AEAVevent@_V1@sycl@@@Z@std@@@234@XZ +?PostProcessorFunc@SubmissionInfo@detail@_V1@sycl@@QEBAAEBV?$optional@V?$function@$$A6AX_N0AEAVevent@_V1@sycl@@@Z@std@@@234@XZ ?PushBack@exception_list@_V1@sycl@@AEAAX$$QEAVexception_ptr@std@@@Z ?PushBack@exception_list@_V1@sycl@@AEAAXAEBVexception_ptr@std@@@Z ?RangeRoundingTrace@handler@_V1@sycl@@AEAA_NXZ @@ -3707,6 +3702,7 @@ ?addHostAccessorAndWait@detail@_V1@sycl@@YAXPEAVAccessorImplHost@123@@Z ?addHostSampledImageAccessorAndWait@detail@_V1@sycl@@YAXPEAVSampledImageAccessorImplHost@123@@Z ?addHostUnsampledImageAccessorAndWait@detail@_V1@sycl@@YAXPEAVUnsampledImageAccessorImplHost@123@@Z +?addImpl@modifiable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@IEAA?AVnode@34567@AEAVdynamic_command_group@34567@AEBV?$vector@Vnode@experimental@oneapi@ext@_V1@sycl@@V?$allocator@Vnode@experimental@oneapi@ext@_V1@sycl@@@std@@@std@@@Z ?addImpl@modifiable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@IEAA?AVnode@34567@AEBV?$vector@Vnode@experimental@oneapi@ext@_V1@sycl@@V?$allocator@Vnode@experimental@oneapi@ext@_V1@sycl@@@std@@@std@@@Z ?addImpl@modifiable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@IEAA?AVnode@34567@V?$function@$$A6AXAEAVhandler@_V1@sycl@@@Z@std@@AEBV?$vector@Vnode@experimental@oneapi@ext@_V1@sycl@@V?$allocator@Vnode@experimental@oneapi@ext@_V1@sycl@@@std@@@std@@@Z ?addLifetimeSharedPtrStorage@handler@_V1@sycl@@AEAAXV?$shared_ptr@$$CBX@std@@@Z @@ -3752,8 +3748,6 @@ ?checkNodePropertiesAndThrow@modifiable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@KAXAEBVproperty_list@67@@Z ?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 @@ -3847,8 +3841,8 @@ ?ext_oneapi_get_graph@queue@_V1@sycl@@QEBA?AV?$command_graph@$0A@@experimental@oneapi@ext@23@XZ ?ext_oneapi_get_kernel@kernel_bundle_plain@detail@_V1@sycl@@AEAA?AVkernel@34@Vstring_view@234@@Z ?ext_oneapi_get_kernel@kernel_bundle_plain@detail@_V1@sycl@@QEAA?AVkernel@34@AEBV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@@Z -?ext_oneapi_get_last_event_impl@queue@_V1@sycl@@AEBA?AV?$optional@Vevent@_V1@sycl@@@detail@23@XZ ?ext_oneapi_get_last_event@queue@_V1@sycl@@QEBA?AV?$optional@Vevent@_V1@sycl@@@std@@XZ +?ext_oneapi_get_last_event_impl@queue@_V1@sycl@@AEBA?AV?$optional@Vevent@_V1@sycl@@@detail@23@XZ ?ext_oneapi_get_state@queue@_V1@sycl@@QEBA?AW4queue_state@experimental@oneapi@ext@23@XZ ?ext_oneapi_graph@handler@_V1@sycl@@QEAAXV?$command_graph@$00@experimental@oneapi@ext@23@@Z ?ext_oneapi_graph@queue@_V1@sycl@@QEAA?AVevent@23@V?$command_graph@$00@experimental@oneapi@ext@23@AEBUcode_location@detail@23@@Z @@ -4014,6 +4008,7 @@ ?getType@handler@_V1@sycl@@AEBA?AW4CGType@detail@23@XZ ?getValueFromDynamicParameter@detail@_V1@sycl@@YAPEAXAEAVdynamic_parameter_base@1experimental@oneapi@ext@23@@Z ?get_access_mode@experimental@oneapi@ext@_V1@sycl@@YA?AW4address_access_mode@12345@PEBX_KAEBVcontext@45@@Z +?get_active_index@dynamic_command_group@experimental@oneapi@ext@_V1@sycl@@QEBA_KXZ ?get_addressing_mode@sampler@_V1@sycl@@QEBA?AW4addressing_mode@23@XZ ?get_allocator_internal@buffer_plain@detail@_V1@sycl@@IEBAAEBV?$unique_ptr@VSYCLMemObjAllocator@detail@_V1@sycl@@U?$default_delete@VSYCLMemObjAllocator@detail@_V1@sycl@@@std@@@std@@XZ ?get_allocator_internal@image_plain@detail@_V1@sycl@@IEBAAEBV?$unique_ptr@VSYCLMemObjAllocator@detail@_V1@sycl@@U?$default_delete@VSYCLMemObjAllocator@detail@_V1@sycl@@@std@@@std@@XZ @@ -4274,6 +4269,7 @@ ?setType@handler@_V1@sycl@@AEAAXW4CGType@detail@23@@Z ?setUserFacingNodeType@handler@_V1@sycl@@AEAAXW4node_type@experimental@oneapi@ext@23@@Z ?set_access_mode@experimental@oneapi@ext@_V1@sycl@@YAXPEBX_KW4address_access_mode@12345@AEBVcontext@45@@Z +?set_active_index@dynamic_command_group@experimental@oneapi@ext@_V1@sycl@@QEAAX_K@Z ?set_arg@handler@_V1@sycl@@QEAAXH$$QEAVraw_kernel_arg@experimental@oneapi@ext@23@@Z ?set_final_data_internal@buffer_plain@detail@_V1@sycl@@IEAAXAEBV?$function@$$A6AXAEBV?$function@$$A6AXPEAX@Z@std@@@Z@std@@@Z ?set_final_data_internal@buffer_plain@detail@_V1@sycl@@IEAAXXZ From e154c5805f12d2723672c66cec6dcc21ebd03379 Mon Sep 17 00:00:00 2001 From: Chris Perkins Date: Fri, 24 Jan 2025 15:57:58 -0800 Subject: [PATCH 3/3] adding overlooked test --- sycl/test/warnings/deprecated_headers.cpp | 5 +---- 1 file changed, 1 insertion(+), 4 deletions(-) diff --git a/sycl/test/warnings/deprecated_headers.cpp b/sycl/test/warnings/deprecated_headers.cpp index 1f6e8bc440877..726ea322d2a92 100644 --- a/sycl/test/warnings/deprecated_headers.cpp +++ b/sycl/test/warnings/deprecated_headers.cpp @@ -5,7 +5,4 @@ #include // expected-warning@sycl/backend/level_zero.hpp:16 {{sycl/backend/level_zero.hpp usage is deprecated, include sycl/ext/oneapi/backend/level_zero.hpp instead}} -#include - -// expected-warning@sycl/ext/intel/online_compiler.hpp:16 {{sycl/ext/intel/online_compiler.hpp usage is deprecated, include sycl/ext/intel/experimental/online_compiler.hpp instead}} -#include +#include \ No newline at end of file