From 2fb6362cfcac78c2b74d0eafd184e24e1d02b718 Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Mon, 30 Aug 2021 20:52:14 +0300 Subject: [PATCH 01/10] [SYCL] Align some extensions with SYCL 2020 This patch 1. aligns these extensions with SYCL 2020 [section #6 in the spec]: - Enqueue barrier [SYCL_EXT_INTEL_ENQUEUE_BARRIER] - Level Zero backend [SYCL_EXT_ONEAPI_BACKEND_LEVEL_ZERO] - Local memory [SYCL_EXT_ONEAPI_LOCAL_MEMORY] - mem_channel property [SYCL_EXT_INTEL_MEM_CHANNEL_PROPERTY] - USM address spaces [SYCL_EXT_INTEL_USM_ADDRESS_SPACES] 2. deprecates these extensions: - sycl::detail::bit_cast [SYCL_INTEL_bitcast] 3. changes the location of these extensions: - sycl::ext::intel::online_compiler moves to sycl::ext::intel::experimental. sycl::ext::intel::online_compiler is deprecated. --- sycl/doc/CompilerAndRuntimeDesign.md | 4 +- .../EnqueueBarrier/enqueue_barrier.asciidoc | 57 ++- .../LevelZeroBackend/LevelZeroBackend.md | 13 +- ..._property.asciidoc => MemChannel.asciidoc} | 26 +- .../usm_address_spaces.asciidoc | 19 +- sycl/include/CL/sycl/access/access.hpp | 6 +- sycl/include/CL/sycl/backend_types.hpp | 3 +- sycl/include/CL/sycl/bit_cast.hpp | 2 +- sycl/include/CL/sycl/feature_test.hpp | 5 + sycl/include/CL/sycl/handler.hpp | 19 +- .../CL/sycl/properties/buffer_properties.hpp | 10 +- sycl/include/CL/sycl/queue.hpp | 37 +- .../intel/experimental/online_compiler.hpp | 233 ++++++++++++ .../sycl/ext/intel/online_compiler.hpp | 358 +----------------- .../ext/oneapi}/backend/level_zero.hpp | 8 + .../sycl/ext/oneapi/group_local_memory.hpp | 4 +- sycl/source/CMakeLists.txt | 6 +- sycl/source/abi_replacements_linux.txt | 12 + sycl/source/backend/level_zero.cpp | 4 + .../online_compiler/online_compiler.cpp | 26 +- sycl/source/handler.cpp | 7 +- sycl/test/abi/sycl_symbols_linux.dump | 11 + sycl/test/warnings/sycl_2020_deprecations.cpp | 2 +- 23 files changed, 454 insertions(+), 418 deletions(-) rename sycl/doc/extensions/MemChannel/{SYCL_INTEL_mem_channel_property.asciidoc => MemChannel.asciidoc} (65%) create mode 100644 sycl/include/sycl/ext/intel/experimental/online_compiler.hpp rename sycl/include/{CL/sycl => sycl/ext/oneapi}/backend/level_zero.hpp (97%) create mode 100644 sycl/source/abi_replacements_linux.txt diff --git a/sycl/doc/CompilerAndRuntimeDesign.md b/sycl/doc/CompilerAndRuntimeDesign.md index f55c5df905aa2..29cb464e594fe 100644 --- a/sycl/doc/CompilerAndRuntimeDesign.md +++ b/sycl/doc/CompilerAndRuntimeDesign.md @@ -918,8 +918,8 @@ space attributes in SYCL mode: | Address space attribute | SYCL address_space enumeration | |-------------------------|--------------------------------| | `__attribute__((opencl_global))` | global_space, constant_space | -| `__attribute__((opencl_global_host))` | global_host_space | -| `__attribute__((opencl_global_device))` | global_device_space | +| `__attribute__((opencl_global_host))` | ext_intel_global_host_space | +| `__attribute__((opencl_global_device))` | ext_intel_global_device_space | | `__attribute__((opencl_local))` | local_space | | `__attribute__((opencl_private))` | private_space | | `__attribute__((opencl_constant))` | N/A diff --git a/sycl/doc/extensions/EnqueueBarrier/enqueue_barrier.asciidoc b/sycl/doc/extensions/EnqueueBarrier/enqueue_barrier.asciidoc index 76cd475c27bd0..15350727b03b0 100644 --- a/sycl/doc/extensions/EnqueueBarrier/enqueue_barrier.asciidoc +++ b/sycl/doc/extensions/EnqueueBarrier/enqueue_barrier.asciidoc @@ -1,4 +1,4 @@ -= SYCL_INTEL_enqueue_barrier += SYCL_EXT_INTEL_ENQUEUE_BARRIER :source-highlighter: coderay :coderay-linenums-mode: table @@ -50,6 +50,22 @@ Revision: 1 == Contact Please open an issue in the https://github.com/intel/llvm/tree/sycl/sycl/doc/extensions/[extensions repository] +== Feature Test Macro + +This extension provides a feature-test macro as described in the core SYCL +specification section 6.3.3 "Feature test macros". Therefore, an +implementation supporting this extension must predefine the macro +`SYCL_EXT_INTEL_ENQUEUE_BARRIER` to one of the values defined in the table below. +Applications can test for the existence of this macro to determine if the +implementation supports this feature, or applications can test the macro's +value to determine which of the extension's APIs the implementation supports. + +[%header,cols="1,5"] +|=== +|Value |Description +|1 |Initial extension version. Base features are supported. +|=== + == Dependencies This extension is written against the SYCL 1.2.1 specification, Revision v1.2.1-6. @@ -75,9 +91,9 @@ two new members to the `queue` class: [grid="rows"] [options="header"] |======================================== -|*handler::barrier*|*queue::submit_barrier* -|`void barrier()` | `event submit_barrier()` -|`void barrier( const vector_class &waitList )` | `event submit_barrier( const vector_class &waitList )` +|*handler::ext_intel_barrier*|*queue::ext_intel_submit_barrier* +|`void ext_intel_barrier()` | `event ext_intel_submit_barrier()` +|`void ext_intel_barrier( const vector_class &waitList )` | `event ext_intel_submit_barrier( const vector_class &waitList )` |======================================== The first variant of the barrier takes no parameters, and waits for all previously submitted commands to the queue to enter the `info::event_command_status::complete` state before any command later submitted to the same queue is allowed to execute. A second variant of the barrier accepts a list of events, with the behavior that no commands submitted to the same queue after barrier submission may execute until all events in the `waitList` have entered the `info::event_command_status::complete` state. Both variants are non-blocking from the host program perspective, in that they do not wait for the barrier conditions to have been met before returning. @@ -93,7 +109,7 @@ Some forms of the new barrier methods return an `event`, which can be used to pe CG4 doesn't execute until all previous command groups submitted to the same queue (CG1, CG2, CG3) have entered the completed state. -==== 1. Using `handler::barrier()`: +==== 1. Using `handler::ext_intel_barrier()`: [source,c++,NoName,linenums] ---- @@ -109,7 +125,7 @@ Queue.submit([&](cl::sycl::handler& cgh) { }); Queue.submit([&](cl::sycl::handler& cgh) { - cgh.barrier(); + cgh.ext_intel_barrier(); }); Queue.submit([&](cl::sycl::handler& cgh) { @@ -118,7 +134,7 @@ Queue.submit([&](cl::sycl::handler& cgh) { ... ---- -==== 2. Using `queue::submit_barrier()`: +==== 2. Using `queue::ext_intel_submit_barrier()`: [source,c++,NoName,linenums] ---- @@ -133,7 +149,7 @@ Queue.submit([&](cl::sycl::handler& cgh) { // CG3 }); -Queue.submit_barrier(); +Queue.ext_intel_submit_barrier(); Queue.submit([&](cl::sycl::handler& cgh) { // CG4 @@ -146,7 +162,7 @@ Queue.submit([&](cl::sycl::handler& cgh) { CG3 requires CG1 (in Queue1) and CG2 (in Queue2) to have completed before it (CG3) begins execution. -==== 1. Using `handler::barrier()`: +==== 1. Using `handler::ext_intel_barrier()`: [source,c++,NoName,linenums] ---- @@ -160,7 +176,7 @@ auto event_barrier2 = Queue2.submit([&](cl::sycl::handler& cgh) { }); Queue3.submit([&](cl::sycl::handler& cgh) { - cgh.barrier( vector_class{event_barrier1, event_barrier2} ); + cgh.ext_intel_barrier( vector_class{event_barrier1, event_barrier2} ); }); Queue3.submit([&](cl::sycl::handler& cgh) { @@ -169,7 +185,7 @@ Queue3.submit([&](cl::sycl::handler& cgh) { ... ---- -==== 2. Using `queue::submit_barrier()`: +==== 2. Using `queue::ext_intel_submit_barrier()`: [source,c++,NoName,linenums] ---- @@ -182,7 +198,7 @@ auto event_barrier2 = Queue2.submit([&](cl::sycl::handler& cgh) { // CG2 }); -Queue3.submit_barrier( vector_class{event_barrier1, event_barrier2} ); +Queue3.ext_intel_submit_barrier( vector_class{event_barrier1, event_barrier2} ); Queue3.submit([&](cl::sycl::handler& cgh) { // CG3 @@ -211,9 +227,9 @@ void wait(); template event submit(T cgf, const queue &secondaryQueue); -event submit_barrier(); +event ext_intel_submit_barrier(); -event submit_barrier( const vector_class &waitList ); +event ext_intel_submit_barrier( const vector_class &waitList ); void wait(); ... @@ -225,8 +241,8 @@ void wait(); [options="header"] |======================================== |*Member functions*|*Description* -|`event submit_barrier()` | Same effect as submitting a `handler::barrier()` within a command group to this `queue`. The returned event enters the `info::event_command_status::complete` state when all events that the barrier is dependent on (implicitly from all previously submitted commands to the same queue) have entered the `info::event_command_status::complete` state. -|`event submit_barrier( const vector_class &waitList )` | Same effect as submitting a `handler:barrier( const vector_class &waitList )` within a command group to this `queue`. The returned event enters the `info::event_command_status::complete` state when all events that the barrier is dependent on (explicitly from `waitList`) have entered the `info::event_command_status::complete` state. +|`event ext_intel_submit_barrier()` | Same effect as submitting a `handler::ext_intel_barrier()` within a command group to this `queue`. The returned event enters the `info::event_command_status::complete` state when all events that the barrier is dependent on (implicitly from all previously submitted commands to the same queue) have entered the `info::event_command_status::complete` state. +|`event ext_intel_submit_barrier( const vector_class &waitList )` | Same effect as submitting a `handler:ext_intel_barrier( const vector_class &waitList )` within a command group to this `queue`. The returned event enters the `info::event_command_status::complete` state when all events that the barrier is dependent on (explicitly from `waitList`) have entered the `info::event_command_status::complete` state. |======================================== @@ -261,9 +277,9 @@ void fill(accessor dest, const T& src); template void fill(accessor dest, const T& src); -void barrier(); +void ext_intel_barrier(); -void barrier( const vector_class &waitList ); +void ext_intel_barrier( const vector_class &waitList ); }; ... @@ -284,8 +300,8 @@ Barriers can be created by two members of the `handler` class that force synchro [options="header"] |======================================== |*Member functions*|*Description* -|`void barrier()` | Prevents any commands submitted afterward to this queue from executing until all commands previously submitted to this queue have entered the `info::event_command_status::complete` state. -|`void barrier( const vector_class &waitList` ) | Prevents any commands submitted afterward to this queue from executing until all events in `waitList` have entered the `info::event_command_status::complete` state. If `waitList` is empty, then the barrier has no effect. +|`void ext_intel_barrier()` | Prevents any commands submitted afterward to this queue from executing until all commands previously submitted to this queue have entered the `info::event_command_status::complete` state. +|`void ext_intel_barrier( const vector_class &waitList` ) | Prevents any commands submitted afterward to this queue from executing until all events in `waitList` have entered the `info::event_command_status::complete` state. If `waitList` is empty, then the barrier has no effect. |======================================== == References @@ -303,6 +319,7 @@ None. |======================================== |Rev|Date|Author|Changes |1|2020-02-26|Ye Ting|*Initial public release* +|2|2021-08-30|Dmitry Vodopyanov|*Updated according to SYCL 2020 reqs for extensions* |======================================== //************************************************************************ diff --git a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md index ce8b00d2ef513..1d24612606912 100644 --- a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md +++ b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md @@ -23,7 +23,7 @@ The Level-Zero backend is added to the cl::sycl::backend enumeration: ``` C++ enum class backend { // ... - level_zero, + ext_oneapi_level_zero, // ... }; ``` @@ -55,7 +55,7 @@ and they must be included in the order shown: ``` C++ #include "level_zero/ze_api.h" - #include "sycl/backend/level_zero.hpp" + #include "sycl/ext/oneapi/backend/level_zero.hpp" ``` ### 4.1 Mapping of SYCL objects to Level-Zero handles @@ -71,7 +71,7 @@ These SYCL objects encapsulate the corresponding Level-Zero handles: ### 4.2 Obtaining of native Level-Zero handles from SYCL objects -The ```get_native()``` member function is how a raw native Level-Zero handle can be obtained +The ```get_native()``` member function is how a raw native Level-Zero handle can be obtained for a specific SYCL object. It is currently supported for SYCL ```platform```, ```device```, ```context```, ```queue```, ```event``` and ```program``` classes. There is also a free-function defined in ```cl::sycl``` namespace that can be used instead of the member function: ``` C++ @@ -81,7 +81,7 @@ auto get_native(const SyclObjectT &Obj) -> ``` ### 4.3 Construct a SYCL object from a Level-Zero handle -The following free functions defined in the ```cl::sycl::level_zero``` namespace allow an application to create +The following free functions defined in the ```cl::sycl::ext::oneapi::level_zero``` namespace allow an application to create a SYCL object that encapsulates a corresponding Level-Zero object: | Level-Zero interoperability function |Description| @@ -103,11 +103,15 @@ some interoperability API supports overriding this behavior and keep the ownersh Use this enumeration for explicit specification of the ownership: ``` C++ namespace sycl { +namespace ext { +namespace oneapi { namespace level_zero { enum class ownership { transfer, keep }; } // namespace level_zero +} // namespace oneapi +} // namespace ext } // namespace sycl ``` @@ -193,3 +197,4 @@ struct free_memory { |3|2021-04-13|James Brodman|Free Memory Query |4|2021-07-06|Rehana Begam|Introduced explicit ownership for queue |5|2021-07-25|Sergey Maslov|Introduced SYCL interop for events +|6|2021-08-30|Dmitry Vodopyanov|Updated according to SYCL 2020 reqs for extensions diff --git a/sycl/doc/extensions/MemChannel/SYCL_INTEL_mem_channel_property.asciidoc b/sycl/doc/extensions/MemChannel/MemChannel.asciidoc similarity index 65% rename from sycl/doc/extensions/MemChannel/SYCL_INTEL_mem_channel_property.asciidoc rename to sycl/doc/extensions/MemChannel/MemChannel.asciidoc index 15b309851043f..dfedebf986766 100644 --- a/sycl/doc/extensions/MemChannel/SYCL_INTEL_mem_channel_property.asciidoc +++ b/sycl/doc/extensions/MemChannel/MemChannel.asciidoc @@ -1,4 +1,4 @@ -= SYCL_INTEL_mem_channel_property += SYCL_EXT_INTEL_MEM_CHANNEL_PROPERTY == Introduction NOTE: Khronos(R) is a registered trademark and SYCL(TM) and SPIR(TM) are trademarks of The Khronos Group Inc. OpenCL(TM) is a trademark of Apple Inc. used by permission by Khronos. @@ -31,6 +31,22 @@ This extension is written against the SYCL 2020 provisional specification, Revis The use of this extension requires a target that supports cl_intel_mem_channel_property or equivalent if OpenCL is used as the underlying device runtime. +== Feature Test Macro + +This extension provides a feature-test macro as described in the core SYCL +specification section 6.3.3 "Feature test macros". Therefore, an +implementation supporting this extension must predefine the macro +`SYCL_EXT_INTEL_MEM_CHANNEL_PROPERTY` to one of the values defined in the table below. +Applications can test for the existence of this macro to determine if the +implementation supports this feature, or applications can test the macro's +value to determine which of the extension's APIs the implementation supports. + +[%header,cols="1,5"] +|=== +|Value |Description +|1 |Initial extension version. Base features are supported. +|=== + == Overview On some targets manual assignment of buffers to memory regions can improve memory bandwidth. This extension adds a buffer property to indicate in which memory channel a particular buffer should be allocated. This information is an optimization hint to the runtime and thus it is legal to ignore. @@ -45,7 +61,7 @@ Add a new property to Table 4.33: Properties supported by the SYCL buffer class [options="header"] |=== | Property | Description -| property::buffer::mem_channel | The `mem_channel` property is a hint to the SYCL runtime that the buffer should be stored in a particular memory channel provided to the property. +| property::buffer::ext_intel_mem_channel | The `ext_intel_mem_channel` property is a hint to the SYCL runtime that the buffer should be stored in a particular memory channel provided to the property. |=== -- @@ -55,7 +71,7 @@ Add a new constructor to Table 4.34: Constructors of the buffer property classes [options="header"] |=== | Constructor | Description -| property::buffer::mem_channel::mem_channel(cl_uint channel) | Constructs a SYCL `mem_channel` property instance with the specified channel ID. The range of valid values depends on the target and is implementation defined. Invalid values do not need to result in an error as the property is only a hint. +| property::buffer::ext_intel_mem_channel::ext_intel_mem_channel(cl_uint channel) | Constructs a SYCL `ext_intel_mem_channel` property instance with the specified channel ID. The range of valid values depends on the target and is implementation defined. Invalid values do not need to result in an error as the property is only a hint. |=== -- @@ -65,7 +81,7 @@ Add a new member function to Table 4.35: Member functions of the buffer property [options="header"] |=== | Member function | Description -| cl_uint property::buffer::mem_channel::get_channel() const | Returns the cl_uint which was specified when constructing this SYCL `mem_channel` property. +| cl_uint property::buffer::ext_intel_mem_channel::get_channel() const | Returns the cl_uint which was specified when constructing this SYCL `ext_intel_mem_channel` property. |=== -- @@ -107,4 +123,6 @@ Add an entry for the new aspect to Table 4.20: Device aspects defined by the cor |======================================== |Rev|Date|Author|Changes |1|2020-10-26|Joe Garvey|*Initial public draft* +|2|2021-08-30|Dmitry Vodopyanov|*Updated according to SYCL 2020 reqs for extensions* + |======================================== diff --git a/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc b/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc index 73e9de4758904..e00de9e430db3 100644 --- a/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc +++ b/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc @@ -1,4 +1,4 @@ -= SYCL_INTEL_usm_address_spaces += SYCL_EXT_INTEL_USM_ADDRESS_SPACES == Introduction This extension introduces two new address spaces and their corresponding multi_ptr specializations. @@ -36,6 +36,22 @@ This extension is written against the SYCL 1.2.1 specification, Revision 7. It If SPIR-V is used by the implementation, this extension also requires support for the SPV_INTEL_usm_storage_classes SPIR-V extension. +== Feature Test Macro + +This extension provides a feature-test macro as described in the core SYCL +specification section 6.3.3 "Feature test macros". Therefore, an +implementation supporting this extension must predefine the macro +`SYCL_EXT_INTEL_USM_ADDRESS_SPACES` to one of the values defined in the table below. +Applications can test for the existence of this macro to determine if the +implementation supports this feature, or applications can test the macro's +value to determine which of the extension's APIs the implementation supports. + +[%header,cols="1,5"] +|=== +|Value |Description +|1 |Initial extension version. Base features are supported. +|=== + == Overview This extension adds two new address spaces: device and host that are subsets of the global address space. @@ -121,4 +137,5 @@ using host_ptr = multi_ptr |======================================== |Rev|Date|Author|Changes |A|2020-06-18|Joe Garvey|Initial public draft +|B|2021-08-30|Dmitry Vodopyanov|Updated according to SYCL 2020 reqs for extensions |======================================== diff --git a/sycl/include/CL/sycl/access/access.hpp b/sycl/include/CL/sycl/access/access.hpp index 8f4b9a92804e8..0ae4f058a6de7 100644 --- a/sycl/include/CL/sycl/access/access.hpp +++ b/sycl/include/CL/sycl/access/access.hpp @@ -46,8 +46,10 @@ enum class address_space : int { global_space = 1, constant_space = 2, local_space = 3, - global_device_space = 4, - global_host_space = 5 + ext_intel_global_device_space = 4, + ext_intel_host_device_space = 5, + global_device_space __SYCL2020_DEPRECATED("use ext_intel_global_device_space instead") = ext_intel_global_device_space, + global_host_space __SYCL2020_DEPRECATED("use ext_intel_host_device_space instead") = ext_intel_host_device_space, }; } // namespace access diff --git a/sycl/include/CL/sycl/backend_types.hpp b/sycl/include/CL/sycl/backend_types.hpp index 0652adb14c2eb..ba150edf738f3 100644 --- a/sycl/include/CL/sycl/backend_types.hpp +++ b/sycl/include/CL/sycl/backend_types.hpp @@ -21,7 +21,8 @@ namespace sycl { enum class backend : char { host = 0, opencl = 1, - level_zero = 2, + ext_oneapi_level_zero = 2, + level_zero __SYCL2020_DEPRECATED("use ext_oneapi_level_zero instead") = ext_oneapi_level_zero, cuda = 3, all = 4, esimd_cpu = 5, diff --git a/sycl/include/CL/sycl/bit_cast.hpp b/sycl/include/CL/sycl/bit_cast.hpp index 2a042c92811bd..289c9a97e8ab8 100644 --- a/sycl/include/CL/sycl/bit_cast.hpp +++ b/sycl/include/CL/sycl/bit_cast.hpp @@ -22,7 +22,6 @@ namespace detail { inline void memcpy(void *Dst, const void *Src, std::size_t Size); } -// sycl::bit_cast ( no longer sycl::detail::bit_cast ) template #if __cpp_lib_bit_cast || __has_builtin(__builtin_bit_cast) constexpr @@ -54,6 +53,7 @@ constexpr namespace detail { template +__SYCL2020_DEPRECATED("use sycl::bit_cast instead") #if __cpp_lib_bit_cast || __has_builtin(__builtin_bit_cast) constexpr #endif diff --git a/sycl/include/CL/sycl/feature_test.hpp b/sycl/include/CL/sycl/feature_test.hpp index 4625cfa06fed7..a36e743f914e9 100644 --- a/sycl/include/CL/sycl/feature_test.hpp +++ b/sycl/include/CL/sycl/feature_test.hpp @@ -24,6 +24,11 @@ namespace sycl { #define SYCL_EXT_ONEAPI_MATRIX 2 #endif #define SYCL_EXT_INTEL_BF16_CONVERSION 1 +#define SYCL_EXT_INTEL_ENQUEUE_BARRIER 1 +#define SYCL_EXT_INTEL_MEM_CHANNEL_PROPERTY 1 +#define SYCL_EXT_INTEL_USM_ADDRESS_SPACES 1 +#define SYCL_EXT_ONEAPI_BACKEND_LEVEL_ZERO 1 + } // namespace sycl } // __SYCL_INLINE_NAMESPACE(cl) diff --git a/sycl/include/CL/sycl/handler.hpp b/sycl/include/CL/sycl/handler.hpp index b5881d3a2b4d4..ce16f12ac7660 100644 --- a/sycl/include/CL/sycl/handler.hpp +++ b/sycl/include/CL/sycl/handler.hpp @@ -2281,17 +2281,34 @@ class __SYCL_EXPORT handler { /// Prevents any commands submitted afterward to this queue from executing /// until all commands previously submitted to this queue have entered the /// complete state. - void barrier() { + void ext_intel_barrier() { throwIfActionIsCreated(); setType(detail::CG::Barrier); } + /// Prevents any commands submitted afterward to this queue from executing + /// until all commands previously submitted to this queue have entered the + /// complete state. + __SYCL2020_DEPRECATED("use ext_intel_barrier instead") + void barrier() { + ext_intel_barrier(); + } + + /// Prevents any commands submitted afterward to this queue from executing + /// until all events in WaitList have entered the complete state. If WaitList + /// is empty, then the barrier has no effect. + /// + /// \param WaitList is a vector of valid SYCL events that need to complete + /// before barrier command can be executed. + void ext_intel_barrier(const std::vector &WaitList); + /// Prevents any commands submitted afterward to this queue from executing /// until all events in WaitList have entered the complete state. If WaitList /// is empty, then the barrier has no effect. /// /// \param WaitList is a vector of valid SYCL events that need to complete /// before barrier command can be executed. + __SYCL2020_DEPRECATED("use ext_intel_barrier instead") void barrier(const std::vector &WaitList); /// Copies data from one memory region to another, both pointed by diff --git a/sycl/include/CL/sycl/properties/buffer_properties.hpp b/sycl/include/CL/sycl/properties/buffer_properties.hpp index aed2b3e8f21c6..a5b1e4a929472 100644 --- a/sycl/include/CL/sycl/properties/buffer_properties.hpp +++ b/sycl/include/CL/sycl/properties/buffer_properties.hpp @@ -40,16 +40,22 @@ class context_bound sycl::context MCtx; }; -class mem_channel : public detail::PropertyWithData< +class ext_intel_mem_channel : public detail::PropertyWithData< detail::PropWithDataKind::BufferMemChannel> { public: - mem_channel(uint32_t Channel) : MChannel(Channel) {} + ext_intel_mem_channel(uint32_t Channel) : MChannel(Channel) {} uint32_t get_channel() const { return MChannel; } private: uint32_t MChannel; }; +class __SYCL2020_DEPRECATED("use ext_intel_mem_channel instead") mem_channel + : public ext_intel_mem_channel { +public: + mem_channel(uint32_t Channel) : ext_intel_mem_channel(Channel) {} +}; + } // namespace buffer } // namespace property diff --git a/sycl/include/CL/sycl/queue.hpp b/sycl/include/CL/sycl/queue.hpp index 413bc49f29186..309a9aa217a01 100644 --- a/sycl/include/CL/sycl/queue.hpp +++ b/sycl/include/CL/sycl/queue.hpp @@ -251,8 +251,35 @@ class __SYCL_EXPORT queue { /// \param CodeLoc is the code location of the submit call (default argument) /// \return a SYCL event object, which corresponds to the queue the command /// group is being enqueued on. - event submit_barrier(_CODELOCONLYPARAM(&CodeLoc)) { - return submit([=](handler &CGH) { CGH.barrier(); } _CODELOCFW(CodeLoc)); + event ext_intel_submit_barrier(_CODELOCONLYPARAM(&CodeLoc)) { + return submit([=](handler &CGH) { CGH.ext_intel_barrier(); } _CODELOCFW(CodeLoc)); + } + + /// Prevents any commands submitted afterward to this queue from executing + /// until all commands previously submitted to this queue have entered the + /// complete state. + /// + /// \param CodeLoc is the code location of the submit call (default argument) + /// \return a SYCL event object, which corresponds to the queue the command + /// group is being enqueued on. + __SYCL2020_DEPRECATED("use ext_intel_submit_barrier() instead") + event submit_barrier() { + return ext_intel_submit_barrier(); + } + + /// Prevents any commands submitted afterward to this queue from executing + /// until all events in WaitList have entered the complete state. If WaitList + /// is empty, then ext_intel_submit_barrier has no effect. + /// + /// \param WaitList is a vector of valid SYCL events that need to complete + /// before barrier command can be executed. + /// \param CodeLoc is the code location of the submit call (default argument) + /// \return a SYCL event object, which corresponds to the queue the command + /// group is being enqueued on. + event ext_intel_submit_barrier( + const std::vector &WaitList _CODELOCPARAM(&CodeLoc)) { + return submit( + [=](handler &CGH) { CGH.ext_intel_barrier(WaitList); } _CODELOCFW(CodeLoc)); } /// Prevents any commands submitted afterward to this queue from executing @@ -264,10 +291,10 @@ class __SYCL_EXPORT queue { /// \param CodeLoc is the code location of the submit call (default argument) /// \return a SYCL event object, which corresponds to the queue the command /// group is being enqueued on. + __SYCL2020_DEPRECATED("use ext_intel_submit_barrier() instead") event - submit_barrier(const std::vector &WaitList _CODELOCPARAM(&CodeLoc)) { - return submit( - [=](handler &CGH) { CGH.barrier(WaitList); } _CODELOCFW(CodeLoc)); + submit_barrier(const std::vector &WaitList) { + return ext_intel_submit_barrier(WaitList); } /// Performs a blocking wait for the completion of all enqueued tasks in the diff --git a/sycl/include/sycl/ext/intel/experimental/online_compiler.hpp b/sycl/include/sycl/ext/intel/experimental/online_compiler.hpp new file mode 100644 index 0000000000000..267bad4aee484 --- /dev/null +++ b/sycl/include/sycl/ext/intel/experimental/online_compiler.hpp @@ -0,0 +1,233 @@ +//===------- 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 // for __SYCL_INLINE_NAMESPACE +#include // for __SYCL_EXPORT +#include + +#include +#include + +__SYCL_INLINE_NAMESPACE(cl) { +namespace sycl { +namespace ext { +namespace intel { +namespace experimental { + +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) {} + + 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 + }; + + 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(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 online_compiler { +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 <> +__SYCL_EXPORT std::vector +online_compiler::compile( + const std::string &src, const std::vector &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 <> +__SYCL_EXPORT std::vector online_compiler::compile( + const std::string &src, const std::vector &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 experimental +} // namespace intel +} // namespace ext + +namespace ext { +namespace __SYCL2020_DEPRECATED("use 'ext::intel::experimental' instead") intel { + using namespace ext::intel::experimental; +} // namespace intel +} // namespace ext + +namespace __SYCL2020_DEPRECATED("use 'ext::intel::experimental' instead") INTEL { + using namespace ext::intel::experimental; +} // namespace INTEL +} // namespace sycl +} // __SYCL_INLINE_NAMESPACE(cl) diff --git a/sycl/include/sycl/ext/intel/online_compiler.hpp b/sycl/include/sycl/ext/intel/online_compiler.hpp index ef104f3dd9e97..38161eeabb7d8 100644 --- a/sycl/include/sycl/ext/intel/online_compiler.hpp +++ b/sycl/include/sycl/ext/intel/online_compiler.hpp @@ -8,358 +8,10 @@ #pragma once -#include // for __SYCL_INLINE_NAMESPACE -#include // for __SYCL_EXPORT -#include +#include -#include -#include +__SYCL_WARNING( + "sycl/ext/intel/online_compiler.hpp usage is deprecated, include " + "sycl/ext/intel/experimental/online_compiler.hpp instead") -__SYCL_INLINE_NAMESPACE(cl) { -namespace sycl { -namespace ext { -namespace intel { - -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) {} - - 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 - }; - - 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(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 online_compiler { -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 <> -__SYCL_EXPORT std::vector -online_compiler::compile( - const std::string &src, const std::vector &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 <> -__SYCL_EXPORT std::vector online_compiler::compile( - const std::string &src, const std::vector &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 intel -} // namespace ext - -namespace __SYCL2020_DEPRECATED("use 'ext::intel' instead") INTEL { - using namespace ext::intel; - - /// Designates a source language for the online compiler. - enum class source_language { opencl_c = 0, cm = 1 }; - - template class online_compiler { - 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(ext::intel::compiled_code_format fmt = - ext::intel::compiled_code_format::spir_v) { - MOnlineCompiler = - ext::intel::online_compiler<(ext::intel::source_language)Lang>(fmt); - } - - /// 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, - ext::intel::device_arch arch, - ext::intel::compiled_code_format fmt = - ext::intel::compiled_code_format::spir_v) { - MOnlineCompiler = - ext::intel::online_compiler<(ext::intel::source_language)Lang>( - dev_type, arch, fmt); - } - - /// 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 &device) { - MOnlineCompiler = - ext::intel::online_compiler<(ext::intel::source_language)Lang>( - device); - } - - /// 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(ext::intel::compiled_code_format fmt) { - MOnlineCompiler.setOutputFormat(fmt); - return *this; - } - - /// Sets the compiled code format version of the compilation target and - /// returns *this. - online_compiler &setOutputFormatVersion(int major, int minor) { - MOnlineCompiler.setOutputFormatVersion(major, minor); - return *this; - } - - /// Sets the device type of the compilation target and returns *this. - online_compiler &setTargetDeviceType(sycl::info::device_type type) { - MOnlineCompiler.setTargetDeviceType(type); - return *this; - } - - /// Sets the device architecture of the compilation target and returns - /// *this. - online_compiler &setTargetDeviceArch(device_arch arch) { - MOnlineCompiler.setTargetDeviceArch(arch); - return *this; - } - - /// Makes the compilation target 32-bit and returns *this. - online_compiler &set32bitTarget() { - MOnlineCompiler.set32bitTarget(); - return *this; - }; - - /// Makes the compilation target 64-bit and returns *this. - online_compiler &set64bitTarget() { - MOnlineCompiler.set64bitTarget(); - return *this; - }; - - /// Sets implementation-defined target device stepping of the compilation - /// target and returns *this. - online_compiler &setTargetDeviceStepping(const std::string &id) { - MOnlineCompiler.setTargetDeviceStepping(id); - return *this; - } - - private: - ext::intel::online_compiler<(ext::intel::source_language)Lang> - MOnlineCompiler; - }; - - // 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 <> - __SYCL_EXPORT std::vector - online_compiler::compile( - const std::string &src, const std::vector &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 MOnlineCompiler.compile(src); - } - - /// Compiles the given CM source \p src. - /// @param src - contents of the source. - /// @param options - compilation options (implementation defined). - template <> - template <> - __SYCL_EXPORT std::vector online_compiler::compile( - const std::string &src, const std::vector &options); - - /// Compiles the given CM source \p src. - template <> - template <> - std::vector online_compiler::compile( - const std::string &src) { - return MOnlineCompiler.compile(src); - } - -} // namespace INTEL -} // namespace sycl -} // __SYCL_INLINE_NAMESPACE(cl) +#include \ No newline at end of file diff --git a/sycl/include/CL/sycl/backend/level_zero.hpp b/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp similarity index 97% rename from sycl/include/CL/sycl/backend/level_zero.hpp rename to sycl/include/sycl/ext/oneapi/backend/level_zero.hpp index 08e8d48afee63..8f76286d2f23d 100644 --- a/sycl/include/CL/sycl/backend/level_zero.hpp +++ b/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp @@ -77,6 +77,8 @@ template <> struct InteropFeatureSupportMap { }; } // namespace detail +namespace ext { +namespace oneapi { namespace level_zero { // Since Level-Zero is not doing any reference counting itself, we have to @@ -164,5 +166,11 @@ T make(const context &Context, } } // namespace level_zero +} // namespace oneapi +} // namespace ext + +namespace __SYCL2020_DEPRECATED("use 'ext::oneapi::level_zero' instead") level_zero { + using ext::oneapi::level_zero; +} } // namespace sycl } // __SYCL_INLINE_NAMESPACE(cl) diff --git a/sycl/include/sycl/ext/oneapi/group_local_memory.hpp b/sycl/include/sycl/ext/oneapi/group_local_memory.hpp index 1f37c29c5bec1..89aba9bf4a436 100644 --- a/sycl/include/sycl/ext/oneapi/group_local_memory.hpp +++ b/sycl/include/sycl/ext/oneapi/group_local_memory.hpp @@ -35,7 +35,7 @@ std::enable_if_t::value && return reinterpret_cast<__attribute__((opencl_local)) T *>(AllocatedMem); #else throw feature_not_supported( - "SYCL_INTEL_local_memory extension is not supported on host device", + "SYCL_EXT_ONEAPI_LOCAL_MEMORY extension is not supported on host device", PI_INVALID_OPERATION); #endif } @@ -60,7 +60,7 @@ std::enable_if_t::value && // Silence unused variable warning [&args...] {}(); throw feature_not_supported( - "SYCL_INTEL_local_memory extension is not supported on host device", + "SYCL_EXT_ONEAPI_LOCAL_MEMORY extension is not supported on host device", PI_INVALID_OPERATION); #endif } diff --git a/sycl/source/CMakeLists.txt b/sycl/source/CMakeLists.txt index fd6e120107c92..d1d31b3f4cc29 100644 --- a/sycl/source/CMakeLists.txt +++ b/sycl/source/CMakeLists.txt @@ -50,8 +50,12 @@ function(add_sycl_rt_library LIB_NAME) target_compile_options(${LIB_OBJ_NAME} PUBLIC -fvisibility=hidden -fvisibility-inlines-hidden) set(linker_script "${CMAKE_CURRENT_SOURCE_DIR}/ld-version-script.txt") + set(abi_linker_script "${CMAKE_CURRENT_SOURCE_DIR}/abi_replacements_linux.txt") target_link_libraries( - ${LIB_NAME} PRIVATE "-Wl,--version-script=${linker_script}") + ${LIB_NAME} PRIVATE "-Wl,${abi_linker_script}") + set_target_properties(${LIB_NAME} PROPERTIES LINK_DEPENDS ${abi_linker_script}) + target_link_libraries( + ${LIB_NAME} PRIVATE "-Wl,--version-script=${linker_script}") set_target_properties(${LIB_NAME} PROPERTIES LINK_DEPENDS ${linker_script}) if (SYCL_ENABLE_XPTI_TRACING) target_link_libraries(${LIB_NAME} PRIVATE dl) diff --git a/sycl/source/abi_replacements_linux.txt b/sycl/source/abi_replacements_linux.txt new file mode 100644 index 0000000000000..edf56b2ebe733 --- /dev/null +++ b/sycl/source/abi_replacements_linux.txt @@ -0,0 +1,12 @@ +_ZN2cl4sycl10level_zero10make_queueERKNS0_7contextEmb = _ZN2cl4sycl3ext6oneapi10level_zero10make_queueERKNS0_7contextEmb; +_ZN2cl4sycl10level_zero10make_queueERKNS0_7contextEm = _ZN2cl4sycl3ext6oneapi10level_zero10make_queueERKNS0_7contextEm; +_ZN2cl4sycl10level_zero12make_programERKNS0_7contextEm = _ZN2cl4sycl3ext6oneapi10level_zero12make_programERKNS0_7contextEm; +_ZN2cl4sycl10level_zero12make_contextERKSt6vectorINS0_6deviceESaIS3_EEmb = _ZN2cl4sycl3ext6oneapi10level_zero12make_contextERKSt6vectorINS0_6deviceESaIS5_EEmb; +_ZN2cl4sycl10level_zero12make_contextERKSt6vectorINS0_6deviceESaIS3_EEm = _ZN2cl4sycl3ext6oneapi10level_zero12make_contextERKSt6vectorINS0_6deviceESaIS5_EEm; +_ZN2cl4sycl10level_zero11make_deviceERKNS0_8platformEm = _ZN2cl4sycl3ext6oneapi10level_zero11make_deviceERKNS0_8platformEm; +_ZN2cl4sycl10level_zero10make_eventERKNS0_7contextEmb = _ZN2cl4sycl3ext6oneapi10level_zero10make_eventERKNS0_7contextEmb; +_ZN2cl4sycl10level_zero13make_platformEm = _ZN2cl4sycl3ext6oneapi10level_zero13make_platformEm; +_ZN2cl4sycl5INTEL15online_compilerILNS1_15source_languageE0EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISC_EEEEES6_IhSaIhEERKSC_DpRKT_ = _ZN2cl4sycl3ext5intel12experimental15online_compilerILNS3_15source_languageE0EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISE_EEEEES8_IhSaIhEERKSE_DpRKT_; +_ZN2cl4sycl5INTEL15online_compilerILNS1_15source_languageE1EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISC_EEEEES6_IhSaIhEERKSC_DpRKT_ = _ZN2cl4sycl3ext5intel12experimental15online_compilerILNS3_15source_languageE1EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISE_EEEEES8_IhSaIhEERKSE_DpRKT_; +_ZN2cl4sycl3ext5intel15online_compilerILNS2_15source_languageE1EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISD_EEEEES7_IhSaIhEERKSD_DpRKT_ = _ZN2cl4sycl3ext5intel12experimental15online_compilerILNS3_15source_languageE1EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISE_EEEEES8_IhSaIhEERKSE_DpRKT_; +_ZN2cl4sycl3ext5intel15online_compilerILNS2_15source_languageE0EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISD_EEEEES7_IhSaIhEERKSD_DpRKT_ = _ZN2cl4sycl3ext5intel12experimental15online_compilerILNS3_15source_languageE0EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISE_EEEEES8_IhSaIhEERKSE_DpRKT_; \ No newline at end of file diff --git a/sycl/source/backend/level_zero.cpp b/sycl/source/backend/level_zero.cpp index 6463cc6e20bf2..fe10236415ec0 100644 --- a/sycl/source/backend/level_zero.cpp +++ b/sycl/source/backend/level_zero.cpp @@ -15,6 +15,8 @@ __SYCL_INLINE_NAMESPACE(cl) { namespace sycl { +namespace ext { +namespace oneapi { namespace level_zero { using namespace detail; @@ -103,5 +105,7 @@ __SYCL_EXPORT event make_event(const context &Context, } } // namespace level_zero +} // namespace oneapi +} // namespace ext } // namespace sycl } // __SYCL_INLINE_NAMESPACE(cl) diff --git a/sycl/source/detail/online_compiler/online_compiler.cpp b/sycl/source/detail/online_compiler/online_compiler.cpp index 2c304f602af8d..14a50d7049664 100644 --- a/sycl/source/detail/online_compiler/online_compiler.cpp +++ b/sycl/source/detail/online_compiler/online_compiler.cpp @@ -18,6 +18,7 @@ __SYCL_INLINE_NAMESPACE(cl) { namespace sycl { namespace ext { namespace intel { +namespace experimental { namespace detail { static std::vector @@ -229,27 +230,18 @@ __SYCL_EXPORT std::vector online_compiler::compile( DeviceStepping, CompileToSPIRVHandle, FreeSPIRVOutputsHandle, CMUserArgs); } - +} // namespace experimental } // namespace intel } // namespace ext -namespace __SYCL2020_DEPRECATED("use 'ext::intel' instead") INTEL { - using namespace ext::intel; - - template <> - template <> - __SYCL_EXPORT std::vector - online_compiler::compile( - const std::string &Source, const std::vector &UserArgs) { - return MOnlineCompiler.compile(Source, UserArgs); - } +namespace ext { +namespace __SYCL2020_DEPRECATED("use 'ext::intel::experimental' instead") intel { + using namespace ext::intel::experimental; +} // namespace intel +} // namespace ext - template <> - template <> - __SYCL_EXPORT std::vector online_compiler::compile( - const std::string &Source, const std::vector &UserArgs) { - return MOnlineCompiler.compile(Source, UserArgs); - } +namespace __SYCL2020_DEPRECATED("use 'ext::intel::experimental' instead") INTEL { + using namespace ext::intel::experimental; } // namespace INTEL } // namespace sycl } // __SYCL_INLINE_NAMESPACE(cl) diff --git a/sycl/source/handler.cpp b/sycl/source/handler.cpp index 72e9cf42e0b27..b3c1d3c7b3365 100644 --- a/sycl/source/handler.cpp +++ b/sycl/source/handler.cpp @@ -488,7 +488,7 @@ std::string handler::getKernelName() { return MKernel->get_info(); } -void handler::barrier(const std::vector &WaitList) { +void handler::ext_intel_barrier(const std::vector &WaitList) { throwIfActionIsCreated(); MCGType = detail::CG::BarrierWaitlist; MEventsWaitWithBarrier.resize(WaitList.size()); @@ -497,6 +497,11 @@ void handler::barrier(const std::vector &WaitList) { [](const event &Event) { return detail::getSyclObjImpl(Event); }); } +__SYCL2020_DEPRECATED("use ext_intel_barrier() instead") +void handler::barrier(const std::vector &WaitList) { + handler::ext_intel_barrier(WaitList); +} + using namespace sycl::detail; bool handler::DisableRangeRounding() { return SYCLConfig::get(); diff --git a/sycl/test/abi/sycl_symbols_linux.dump b/sycl/test/abi/sycl_symbols_linux.dump index d94751483d76d..dadde84ee500c 100644 --- a/sycl/test/abi/sycl_symbols_linux.dump +++ b/sycl/test/abi/sycl_symbols_linux.dump @@ -3633,8 +3633,18 @@ _ZN2cl4sycl20aligned_alloc_sharedEmmRKNS0_5queueE _ZN2cl4sycl20aligned_alloc_sharedEmmRKNS0_5queueERKNS0_13property_listE _ZN2cl4sycl20aligned_alloc_sharedEmmRKNS0_6deviceERKNS0_7contextE _ZN2cl4sycl20aligned_alloc_sharedEmmRKNS0_6deviceERKNS0_7contextERKNS0_13property_listE +_ZN2cl4sycl3ext5intel12experimental15online_compilerILNS3_15source_languageE0EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISE_EEEEES8_IhSaIhEERKSE_DpRKT_ +_ZN2cl4sycl3ext5intel12experimental15online_compilerILNS3_15source_languageE1EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISE_EEEEES8_IhSaIhEERKSE_DpRKT_ _ZN2cl4sycl3ext5intel15online_compilerILNS2_15source_languageE0EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISD_EEEEES7_IhSaIhEERKSD_DpRKT_ _ZN2cl4sycl3ext5intel15online_compilerILNS2_15source_languageE1EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISD_EEEEES7_IhSaIhEERKSD_DpRKT_ +_ZN2cl4sycl3ext6oneapi10level_zero10make_eventERKNS0_7contextEmb +_ZN2cl4sycl3ext6oneapi10level_zero10make_queueERKNS0_7contextEm +_ZN2cl4sycl3ext6oneapi10level_zero10make_queueERKNS0_7contextEmb +_ZN2cl4sycl3ext6oneapi10level_zero11make_deviceERKNS0_8platformEm +_ZN2cl4sycl3ext6oneapi10level_zero12make_contextERKSt6vectorINS0_6deviceESaIS5_EEm +_ZN2cl4sycl3ext6oneapi10level_zero12make_contextERKSt6vectorINS0_6deviceESaIS5_EEmb +_ZN2cl4sycl3ext6oneapi10level_zero12make_programERKNS0_7contextEm +_ZN2cl4sycl3ext6oneapi10level_zero13make_platformEm _ZN2cl4sycl3ext6oneapi15filter_selectorC1ERKNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEE _ZN2cl4sycl3ext6oneapi15filter_selectorC2ERKNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEE _ZN2cl4sycl3ext6oneapi6detail16reduGetMaxWGSizeESt10shared_ptrINS0_6detail10queue_implEEm @@ -3899,6 +3909,7 @@ _ZN2cl4sycl7handler10mem_adviseEPKvmi _ZN2cl4sycl7handler10processArgEPvRKNS0_6detail19kernel_param_kind_tEimRmb _ZN2cl4sycl7handler10processArgEPvRKNS0_6detail19kernel_param_kind_tEimRmbb _ZN2cl4sycl7handler13getKernelNameB5cxx11Ev +_ZN2cl4sycl7handler17ext_intel_barrierERKSt6vectorINS0_5eventESaIS3_EE _ZN2cl4sycl7handler18RangeRoundingTraceEv _ZN2cl4sycl7handler18extractArgsAndReqsEv _ZN2cl4sycl7handler20DisableRangeRoundingEv diff --git a/sycl/test/warnings/sycl_2020_deprecations.cpp b/sycl/test/warnings/sycl_2020_deprecations.cpp index b244dcc1da164..43265f85580ca 100644 --- a/sycl/test/warnings/sycl_2020_deprecations.cpp +++ b/sycl/test/warnings/sycl_2020_deprecations.cpp @@ -148,7 +148,7 @@ int main() { sycl::ONEAPI::atomic_fence(sycl::ONEAPI::memory_order::relaxed, sycl::ONEAPI::memory_scope::work_group); - // expected-warning@+1{{'INTEL' is deprecated: use 'ext::intel' instead}} + // expected-warning@+1{{'INTEL' is deprecated: use 'ext::intel::experimental' instead}} auto SL = sycl::INTEL::source_language::opencl_c; (void)SL; From 3fbc2d265d9928c24cef162ef4e384af284777d6 Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Mon, 30 Aug 2021 23:23:47 +0300 Subject: [PATCH 02/10] Fix pre-commit checks --- sycl/include/CL/sycl/access/access.hpp | 8 +++- sycl/include/CL/sycl/backend_types.hpp | 3 +- sycl/include/CL/sycl/bit_cast.hpp | 5 +-- sycl/include/CL/sycl/feature_test.hpp | 1 - sycl/include/CL/sycl/handler.hpp | 8 ++-- .../CL/sycl/properties/buffer_properties.hpp | 4 +- sycl/include/CL/sycl/queue.hpp | 22 ++++++---- .../intel/experimental/online_compiler.hpp | 6 ++- .../sycl/ext/intel/online_compiler.hpp | 2 +- .../sycl/ext/oneapi/backend/level_zero.hpp | 5 ++- sycl/source/abi_replacements_windows.cpp | 42 +++++++++++++++++++ .../online_compiler/online_compiler.cpp | 8 ++-- sycl/source/handler.cpp | 2 +- sycl/test/abi/sycl_symbols_windows.dump | 14 +++++++ .../on-device/basic_tests/interop/traits.cpp | 2 +- 15 files changed, 99 insertions(+), 33 deletions(-) diff --git a/sycl/include/CL/sycl/access/access.hpp b/sycl/include/CL/sycl/access/access.hpp index 0ae4f058a6de7..cba9308111cc0 100644 --- a/sycl/include/CL/sycl/access/access.hpp +++ b/sycl/include/CL/sycl/access/access.hpp @@ -48,8 +48,12 @@ enum class address_space : int { local_space = 3, ext_intel_global_device_space = 4, ext_intel_host_device_space = 5, - global_device_space __SYCL2020_DEPRECATED("use ext_intel_global_device_space instead") = ext_intel_global_device_space, - global_host_space __SYCL2020_DEPRECATED("use ext_intel_host_device_space instead") = ext_intel_host_device_space, + global_device_space __SYCL2020_DEPRECATED( + "use 'ext_intel_global_device_space' instead") = + ext_intel_global_device_space, + global_host_space __SYCL2020_DEPRECATED( + "use 'ext_intel_host_device_space' instead") = + ext_intel_host_device_space, }; } // namespace access diff --git a/sycl/include/CL/sycl/backend_types.hpp b/sycl/include/CL/sycl/backend_types.hpp index ba150edf738f3..28c05a2ad215d 100644 --- a/sycl/include/CL/sycl/backend_types.hpp +++ b/sycl/include/CL/sycl/backend_types.hpp @@ -22,7 +22,8 @@ enum class backend : char { host = 0, opencl = 1, ext_oneapi_level_zero = 2, - level_zero __SYCL2020_DEPRECATED("use ext_oneapi_level_zero instead") = ext_oneapi_level_zero, + level_zero __SYCL2020_DEPRECATED("use 'ext_oneapi_level_zero' instead") = + ext_oneapi_level_zero, cuda = 3, all = 4, esimd_cpu = 5, diff --git a/sycl/include/CL/sycl/bit_cast.hpp b/sycl/include/CL/sycl/bit_cast.hpp index 289c9a97e8ab8..6fe0b85dfe637 100644 --- a/sycl/include/CL/sycl/bit_cast.hpp +++ b/sycl/include/CL/sycl/bit_cast.hpp @@ -53,12 +53,11 @@ constexpr namespace detail { template -__SYCL2020_DEPRECATED("use sycl::bit_cast instead") +__SYCL2020_DEPRECATED("use 'sycl::bit_cast' instead") #if __cpp_lib_bit_cast || __has_builtin(__builtin_bit_cast) constexpr #endif - To - bit_cast(const From &from) noexcept { + To bit_cast(const From &from) noexcept { return sycl::bit_cast(from); } } // namespace detail diff --git a/sycl/include/CL/sycl/feature_test.hpp b/sycl/include/CL/sycl/feature_test.hpp index a36e743f914e9..ef7c55c4946b9 100644 --- a/sycl/include/CL/sycl/feature_test.hpp +++ b/sycl/include/CL/sycl/feature_test.hpp @@ -29,6 +29,5 @@ namespace sycl { #define SYCL_EXT_INTEL_USM_ADDRESS_SPACES 1 #define SYCL_EXT_ONEAPI_BACKEND_LEVEL_ZERO 1 - } // namespace sycl } // __SYCL_INLINE_NAMESPACE(cl) diff --git a/sycl/include/CL/sycl/handler.hpp b/sycl/include/CL/sycl/handler.hpp index ce16f12ac7660..bc60902a65bfc 100644 --- a/sycl/include/CL/sycl/handler.hpp +++ b/sycl/include/CL/sycl/handler.hpp @@ -2289,10 +2289,8 @@ class __SYCL_EXPORT handler { /// Prevents any commands submitted afterward to this queue from executing /// until all commands previously submitted to this queue have entered the /// complete state. - __SYCL2020_DEPRECATED("use ext_intel_barrier instead") - void barrier() { - ext_intel_barrier(); - } + __SYCL2020_DEPRECATED("use 'ext_intel_barrier' instead") + void barrier() { ext_intel_barrier(); } /// Prevents any commands submitted afterward to this queue from executing /// until all events in WaitList have entered the complete state. If WaitList @@ -2308,7 +2306,7 @@ class __SYCL_EXPORT handler { /// /// \param WaitList is a vector of valid SYCL events that need to complete /// before barrier command can be executed. - __SYCL2020_DEPRECATED("use ext_intel_barrier instead") + __SYCL2020_DEPRECATED("use 'ext_intel_barrier' instead") void barrier(const std::vector &WaitList); /// Copies data from one memory region to another, both pointed by diff --git a/sycl/include/CL/sycl/properties/buffer_properties.hpp b/sycl/include/CL/sycl/properties/buffer_properties.hpp index a5b1e4a929472..19f2a3355decc 100644 --- a/sycl/include/CL/sycl/properties/buffer_properties.hpp +++ b/sycl/include/CL/sycl/properties/buffer_properties.hpp @@ -41,7 +41,7 @@ class context_bound }; class ext_intel_mem_channel : public detail::PropertyWithData< - detail::PropWithDataKind::BufferMemChannel> { + detail::PropWithDataKind::BufferMemChannel> { public: ext_intel_mem_channel(uint32_t Channel) : MChannel(Channel) {} uint32_t get_channel() const { return MChannel; } @@ -50,7 +50,7 @@ class ext_intel_mem_channel : public detail::PropertyWithData< uint32_t MChannel; }; -class __SYCL2020_DEPRECATED("use ext_intel_mem_channel instead") mem_channel +class __SYCL2020_DEPRECATED("use 'ext_intel_mem_channel' instead") mem_channel : public ext_intel_mem_channel { public: mem_channel(uint32_t Channel) : ext_intel_mem_channel(Channel) {} diff --git a/sycl/include/CL/sycl/queue.hpp b/sycl/include/CL/sycl/queue.hpp index 309a9aa217a01..f7367cace346c 100644 --- a/sycl/include/CL/sycl/queue.hpp +++ b/sycl/include/CL/sycl/queue.hpp @@ -252,7 +252,8 @@ class __SYCL_EXPORT queue { /// \return a SYCL event object, which corresponds to the queue the command /// group is being enqueued on. event ext_intel_submit_barrier(_CODELOCONLYPARAM(&CodeLoc)) { - return submit([=](handler &CGH) { CGH.ext_intel_barrier(); } _CODELOCFW(CodeLoc)); + return submit( + [=](handler &CGH) { CGH.ext_intel_barrier(); } _CODELOCFW(CodeLoc)); } /// Prevents any commands submitted afterward to this queue from executing @@ -262,9 +263,10 @@ class __SYCL_EXPORT queue { /// \param CodeLoc is the code location of the submit call (default argument) /// \return a SYCL event object, which corresponds to the queue the command /// group is being enqueued on. - __SYCL2020_DEPRECATED("use ext_intel_submit_barrier() instead") - event submit_barrier() { - return ext_intel_submit_barrier(); + __SYCL2020_DEPRECATED("use 'ext_intel_submit_barrier' instead") + event submit_barrier(_CODELOCONLYPARAM(&CodeLoc)) { + _CODELOCARG(&CodeLoc); + return ext_intel_submit_barrier(CodeLoc); } /// Prevents any commands submitted afterward to this queue from executing @@ -278,8 +280,9 @@ class __SYCL_EXPORT queue { /// group is being enqueued on. event ext_intel_submit_barrier( const std::vector &WaitList _CODELOCPARAM(&CodeLoc)) { - return submit( - [=](handler &CGH) { CGH.ext_intel_barrier(WaitList); } _CODELOCFW(CodeLoc)); + return submit([=](handler &CGH) { + CGH.ext_intel_barrier(WaitList); + } _CODELOCFW(CodeLoc)); } /// Prevents any commands submitted afterward to this queue from executing @@ -291,10 +294,11 @@ class __SYCL_EXPORT queue { /// \param CodeLoc is the code location of the submit call (default argument) /// \return a SYCL event object, which corresponds to the queue the command /// group is being enqueued on. - __SYCL2020_DEPRECATED("use ext_intel_submit_barrier() instead") + __SYCL2020_DEPRECATED("use 'ext_intel_submit_barrier' instead") event - submit_barrier(const std::vector &WaitList) { - return ext_intel_submit_barrier(WaitList); + submit_barrier(const std::vector &WaitList _CODELOCPARAM(&CodeLoc)) { + _CODELOCARG(&CodeLoc); + return ext_intel_submit_barrier(WaitList, CodeLoc); } /// Performs a blocking wait for the completion of all enqueued tasks in the diff --git a/sycl/include/sycl/ext/intel/experimental/online_compiler.hpp b/sycl/include/sycl/ext/intel/experimental/online_compiler.hpp index 267bad4aee484..d4f436307cb83 100644 --- a/sycl/include/sycl/ext/intel/experimental/online_compiler.hpp +++ b/sycl/include/sycl/ext/intel/experimental/online_compiler.hpp @@ -221,12 +221,14 @@ online_compiler::compile(const std::string &src) { } // namespace ext namespace ext { -namespace __SYCL2020_DEPRECATED("use 'ext::intel::experimental' instead") intel { +namespace __SYCL2020_DEPRECATED( + "use 'ext::intel::experimental' instead") intel { using namespace ext::intel::experimental; } // namespace intel } // namespace ext -namespace __SYCL2020_DEPRECATED("use 'ext::intel::experimental' instead") INTEL { +namespace __SYCL2020_DEPRECATED( + "use 'ext::intel::experimental' instead") INTEL { using namespace ext::intel::experimental; } // namespace INTEL } // namespace sycl diff --git a/sycl/include/sycl/ext/intel/online_compiler.hpp b/sycl/include/sycl/ext/intel/online_compiler.hpp index 38161eeabb7d8..d7212bb1b26ff 100644 --- a/sycl/include/sycl/ext/intel/online_compiler.hpp +++ b/sycl/include/sycl/ext/intel/online_compiler.hpp @@ -14,4 +14,4 @@ __SYCL_WARNING( "sycl/ext/intel/online_compiler.hpp usage is deprecated, include " "sycl/ext/intel/experimental/online_compiler.hpp instead") -#include \ No newline at end of file +#include diff --git a/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp b/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp index 8f76286d2f23d..a8ab2af5aa186 100644 --- a/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp +++ b/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp @@ -169,8 +169,9 @@ T make(const context &Context, } // namespace oneapi } // namespace ext -namespace __SYCL2020_DEPRECATED("use 'ext::oneapi::level_zero' instead") level_zero { - using ext::oneapi::level_zero; +namespace __SYCL2020_DEPRECATED("use 'ext::oneapi::level_zero' instead") + level_zero { + using namespace ext::oneapi::level_zero; } } // namespace sycl } // __SYCL_INLINE_NAMESPACE(cl) diff --git a/sycl/source/abi_replacements_windows.cpp b/sycl/source/abi_replacements_windows.cpp index 35f782a38064a..ab4c9fa223e94 100644 --- a/sycl/source/abi_replacements_windows.cpp +++ b/sycl/source/abi_replacements_windows.cpp @@ -7,3 +7,45 @@ #pragma comment( \ linker, \ "/export:?accessGlobalOffset@stream_impl@detail@sycl@cl@@QEAA?AV?$accessor@I$00$0EAF@$0HNO@$0A@V?$accessor_property_list@$$V@ONEAPI@sycl@cl@@@34@AEAVhandler@34@@Z=?accessGlobalOffset@stream_impl@detail@sycl@cl@@QEAA?AV?$accessor@I$00$0EAF@$0HNO@$0A@V?$accessor_property_list@$$V@oneapi@ext@sycl@cl@@@34@AEAVhandler@34@@Z") +#pragma comment( \ + linker, \ + "/export:?make_context@level_zero@sycl@cl@@YA?AVcontext@23@AEBV?$vector@Vdevice@sycl@cl@@V?$allocator@Vdevice@sycl@cl@@@std@@@std@@_K_N@Z=?make_context@level_zero@oneapi@ext@sycl@cl@@YA?AVcontext@45@AEBV?$vector@Vdevice@sycl@cl@@V?$allocator@Vdevice@sycl@cl@@@std@@@std@@_K_N@Z") +#pragma comment( \ + linker, \ + "/export:?make_event@level_zero@sycl@cl@@YA?AVevent@23@AEBVcontext@23@_K_N@Z=?make_event@level_zero@oneapi@ext@sycl@cl@@YA?AVevent@45@AEBVcontext@45@_K_N@Z") +#pragma comment( \ + linker, \ + "/export:?make_queue@level_zero@sycl@cl@@YA?AVqueue@23@AEBVcontext@23@_K@Z=?make_queue@level_zero@oneapi@ext@sycl@cl@@YA?AVqueue@45@AEBVcontext@45@_K@Z") +#pragma comment( \ + linker, \ + "/export:??$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@intel@ext@sycl@cl@@QEAA?AV?$vector@EV?$allocator@E@std@@@std@@AEBV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@6@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@@6@@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@$00@experimental@intel@ext@sycl@cl@@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") +#pragma comment( \ + linker, \ + "/export:?make_device@level_zero@sycl@cl@@YA?AVdevice@23@AEBVplatform@23@_K@Z=?make_device@level_zero@oneapi@ext@sycl@cl@@YA?AVdevice@45@AEBVplatform@45@_K@Z") +#pragma comment( \ + linker, \ + "/export:?make_platform@level_zero@sycl@cl@@YA?AVplatform@23@_K@Z=?make_platform@level_zero@oneapi@ext@sycl@cl@@YA?AVplatform@45@_K@Z") +#pragma comment( \ + linker, \ + "/export:??$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@@INTEL@sycl@cl@@QEAA?AV?$vector@EV?$allocator@E@std@@@std@@AEBV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@5@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@@5@@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@sycl@cl@@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") +#pragma comment( \ + linker, \ + "/export:?submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBV?$vector@Vevent@sycl@cl@@V?$allocator@Vevent@sycl@cl@@@std@@@std@@AEBUcode_location@detail@23@@Z=?ext_intel_submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBV?$vector@Vevent@sycl@cl@@V?$allocator@Vevent@sycl@cl@@@std@@@std@@AEBUcode_location@detail@23@@Z") +#pragma comment( \ + linker, \ + "/export:?make_queue@level_zero@sycl@cl@@YA?AVqueue@23@AEBVcontext@23@_K_N@Z=?make_queue@level_zero@oneapi@ext@sycl@cl@@YA?AVqueue@45@AEBVcontext@45@_K_N@Z") +#pragma comment( \ + linker, \ + "/export:?make_program@level_zero@sycl@cl@@YA?AVprogram@23@AEBVcontext@23@_K@Z=?make_program@level_zero@oneapi@ext@sycl@cl@@YA?AVprogram@45@AEBVcontext@45@_K@Z") +#pragma comment( \ + linker, \ + "/export:??$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@@intel@ext@sycl@cl@@QEAA?AV?$vector@EV?$allocator@E@std@@@std@@AEBV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@6@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@@6@@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@sycl@cl@@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") +#pragma comment( \ + linker, \ + "/export:?make_context@level_zero@sycl@cl@@YA?AVcontext@23@AEBV?$vector@Vdevice@sycl@cl@@V?$allocator@Vdevice@sycl@cl@@@std@@@std@@_K@Z=?make_context@level_zero@oneapi@ext@sycl@cl@@YA?AVcontext@45@AEBV?$vector@Vdevice@sycl@cl@@V?$allocator@Vdevice@sycl@cl@@@std@@@std@@_K@Z") +#pragma comment( \ + linker, \ + "/export:??$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@INTEL@sycl@cl@@QEAA?AV?$vector@EV?$allocator@E@std@@@std@@AEBV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@5@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@@5@@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@$00@experimental@intel@ext@sycl@cl@@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") +#pragma comment( \ + linker, \ + "/export:?submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBUcode_location@detail@23@@Z=?ext_intel_submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBUcode_location@detail@23@@Z") diff --git a/sycl/source/detail/online_compiler/online_compiler.cpp b/sycl/source/detail/online_compiler/online_compiler.cpp index 14a50d7049664..96a1641750bc3 100644 --- a/sycl/source/detail/online_compiler/online_compiler.cpp +++ b/sycl/source/detail/online_compiler/online_compiler.cpp @@ -8,7 +8,7 @@ #include #include -#include +#include #include @@ -235,12 +235,14 @@ __SYCL_EXPORT std::vector online_compiler::compile( } // namespace ext namespace ext { -namespace __SYCL2020_DEPRECATED("use 'ext::intel::experimental' instead") intel { +namespace __SYCL2020_DEPRECATED( + "use 'ext::intel::experimental' instead") intel { using namespace ext::intel::experimental; } // namespace intel } // namespace ext -namespace __SYCL2020_DEPRECATED("use 'ext::intel::experimental' instead") INTEL { +namespace __SYCL2020_DEPRECATED( + "use 'ext::intel::experimental' instead") INTEL { using namespace ext::intel::experimental; } // namespace INTEL } // namespace sycl diff --git a/sycl/source/handler.cpp b/sycl/source/handler.cpp index b3c1d3c7b3365..6fedb20cb2cb5 100644 --- a/sycl/source/handler.cpp +++ b/sycl/source/handler.cpp @@ -497,7 +497,7 @@ void handler::ext_intel_barrier(const std::vector &WaitList) { [](const event &Event) { return detail::getSyclObjImpl(Event); }); } -__SYCL2020_DEPRECATED("use ext_intel_barrier() instead") +__SYCL2020_DEPRECATED("use 'ext_intel_barrier' instead") void handler::barrier(const std::vector &WaitList) { handler::ext_intel_barrier(WaitList); } diff --git a/sycl/test/abi/sycl_symbols_windows.dump b/sycl/test/abi/sycl_symbols_windows.dump index 81161742f0910..8a57f82de37e2 100644 --- a/sycl/test/abi/sycl_symbols_windows.dump +++ b/sycl/test/abi/sycl_symbols_windows.dump @@ -8,8 +8,10 @@ # 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@INTEL@sycl@cl@@QEAA?AV?$vector@EV?$allocator@E@std@@@std@@AEBV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@5@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@@5@@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@$00@experimental@intel@ext@sycl@cl@@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@$00@intel@ext@sycl@cl@@QEAA?AV?$vector@EV?$allocator@E@std@@@std@@AEBV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@6@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@@6@@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@@INTEL@sycl@cl@@QEAA?AV?$vector@EV?$allocator@E@std@@@std@@AEBV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@5@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@@5@@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@sycl@cl@@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@@intel@ext@sycl@cl@@QEAA?AV?$vector@EV?$allocator@E@std@@@std@@AEBV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@6@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@@6@@Z ??$create_sub_devices@$0BAIG@@device@sycl@cl@@QEBA?AV?$vector@Vdevice@sycl@cl@@V?$allocator@Vdevice@sycl@cl@@@std@@@std@@_K@Z ??$create_sub_devices@$0BAIH@@device@sycl@cl@@QEBA?AV?$vector@Vdevice@sycl@cl@@V?$allocator@Vdevice@sycl@cl@@@std@@@std@@AEBV?$vector@_KV?$allocator@_K@std@@@4@@Z @@ -1738,6 +1740,10 @@ ?expm1@__host_std@cl@@YA?AVhalf@half_impl@detail@sycl@2@V34562@@Z ?expm1@__host_std@cl@@YAMM@Z ?expm1@__host_std@cl@@YANN@Z +?ext_intel_barrier@handler@sycl@cl@@QEAAXAEBV?$vector@Vevent@sycl@cl@@V?$allocator@Vevent@sycl@cl@@@std@@@std@@@Z +?ext_intel_barrier@handler@sycl@cl@@QEAAXXZ +?ext_intel_submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBUcode_location@detail@23@@Z +?ext_intel_submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBV?$vector@Vevent@sycl@cl@@V?$allocator@Vevent@sycl@cl@@@std@@@std@@AEBUcode_location@detail@23@@Z ?extractArgsAndReqs@handler@sycl@cl@@AEAAXXZ ?extractArgsAndReqsFromLambda@handler@sycl@cl@@AEAAXPEAD_KPEBUkernel_param_desc_t@detail@23@@Z ?extractArgsAndReqsFromLambda@handler@sycl@cl@@AEAAXPEAD_KPEBUkernel_param_desc_t@detail@23@_N@Z @@ -2542,25 +2548,33 @@ ?mad@__host_std@cl@@YANNNN@Z ?makeDir@OSUtil@detail@sycl@cl@@SAHPEBD@Z ?make_context@detail@sycl@cl@@YA?AVcontext@23@_KAEBV?$function@$$A6AXVexception_list@sycl@cl@@@Z@std@@W4backend@23@@Z +?make_context@level_zero@oneapi@ext@sycl@cl@@YA?AVcontext@45@AEBV?$vector@Vdevice@sycl@cl@@V?$allocator@Vdevice@sycl@cl@@@std@@@std@@_K@Z +?make_context@level_zero@oneapi@ext@sycl@cl@@YA?AVcontext@45@AEBV?$vector@Vdevice@sycl@cl@@V?$allocator@Vdevice@sycl@cl@@@std@@@std@@_K_N@Z ?make_context@level_zero@sycl@cl@@YA?AVcontext@23@AEBV?$vector@Vdevice@sycl@cl@@V?$allocator@Vdevice@sycl@cl@@@std@@@std@@_K@Z ?make_context@level_zero@sycl@cl@@YA?AVcontext@23@AEBV?$vector@Vdevice@sycl@cl@@V?$allocator@Vdevice@sycl@cl@@@std@@@std@@_K_N@Z ?make_context@opencl@sycl@cl@@YA?AVcontext@23@_K@Z ?make_device@detail@sycl@cl@@YA?AVdevice@23@_KW4backend@23@@Z +?make_device@level_zero@oneapi@ext@sycl@cl@@YA?AVdevice@45@AEBVplatform@45@_K@Z ?make_device@level_zero@sycl@cl@@YA?AVdevice@23@AEBVplatform@23@_K@Z ?make_device@opencl@sycl@cl@@YA?AVdevice@23@_K@Z ?make_error_code@sycl@cl@@YA?AVerror_code@std@@W4errc@12@@Z ?make_event@detail@sycl@cl@@YA?AVevent@23@_KAEBVcontext@23@W4backend@23@@Z ?make_event@detail@sycl@cl@@YA?AVevent@23@_KAEBVcontext@23@_NW4backend@23@@Z +?make_event@level_zero@oneapi@ext@sycl@cl@@YA?AVevent@45@AEBVcontext@45@_K_N@Z ?make_event@level_zero@sycl@cl@@YA?AVevent@23@AEBVcontext@23@_K_N@Z ?make_kernel@detail@sycl@cl@@YA?AVkernel@23@_KAEBVcontext@23@W4backend@23@@Z ?make_kernel_bundle@detail@sycl@cl@@YA?AV?$shared_ptr@Vkernel_bundle_impl@detail@sycl@cl@@@std@@_KAEBVcontext@23@W4bundle_state@23@W4backend@23@@Z ?make_platform@detail@sycl@cl@@YA?AVplatform@23@_KW4backend@23@@Z +?make_platform@level_zero@oneapi@ext@sycl@cl@@YA?AVplatform@45@_K@Z ?make_platform@level_zero@sycl@cl@@YA?AVplatform@23@_K@Z ?make_platform@opencl@sycl@cl@@YA?AVplatform@23@_K@Z +?make_program@level_zero@oneapi@ext@sycl@cl@@YA?AVprogram@45@AEBVcontext@45@_K@Z ?make_program@level_zero@sycl@cl@@YA?AVprogram@23@AEBVcontext@23@_K@Z ?make_program@opencl@sycl@cl@@YA?AVprogram@23@AEBVcontext@23@_K@Z ?make_queue@detail@sycl@cl@@YA?AVqueue@23@_KAEBVcontext@23@AEBV?$function@$$A6AXVexception_list@sycl@cl@@@Z@std@@W4backend@23@@Z ?make_queue@detail@sycl@cl@@YA?AVqueue@23@_KAEBVcontext@23@_NAEBV?$function@$$A6AXVexception_list@sycl@cl@@@Z@std@@W4backend@23@@Z +?make_queue@level_zero@oneapi@ext@sycl@cl@@YA?AVqueue@45@AEBVcontext@45@_K@Z +?make_queue@level_zero@oneapi@ext@sycl@cl@@YA?AVqueue@45@AEBVcontext@45@_K_N@Z ?make_queue@level_zero@sycl@cl@@YA?AVqueue@23@AEBVcontext@23@_K@Z ?make_queue@level_zero@sycl@cl@@YA?AVqueue@23@AEBVcontext@23@_K_N@Z ?make_queue@opencl@sycl@cl@@YA?AVqueue@23@AEBVcontext@23@_K@Z diff --git a/sycl/test/on-device/basic_tests/interop/traits.cpp b/sycl/test/on-device/basic_tests/interop/traits.cpp index 92c824c6961b2..d8669788038a8 100644 --- a/sycl/test/on-device/basic_tests/interop/traits.cpp +++ b/sycl/test/on-device/basic_tests/interop/traits.cpp @@ -13,7 +13,7 @@ constexpr auto Backend = sycl::backend::opencl; #ifdef USE_L0 #include -#include +#include constexpr auto Backend = sycl::backend::level_zero; #endif From bd5bec5119bb79601aef686db35d69af88c558d1 Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Tue, 31 Aug 2021 15:14:00 +0300 Subject: [PATCH 03/10] Address CR comment --- sycl/include/CL/sycl/backend/level_zero.hpp | 16 +++++++++++++ sycl/test/warnings/sycl_2020_deprecations.cpp | 23 +++++++++++++++++++ 2 files changed, 39 insertions(+) create mode 100644 sycl/include/CL/sycl/backend/level_zero.hpp diff --git a/sycl/include/CL/sycl/backend/level_zero.hpp b/sycl/include/CL/sycl/backend/level_zero.hpp new file mode 100644 index 0000000000000..8d808fec3e5ee --- /dev/null +++ b/sycl/include/CL/sycl/backend/level_zero.hpp @@ -0,0 +1,16 @@ +//==--------- level_zero.hpp - SYCL Level-Zero backend ---------------------==// +// +// 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 + +__SYCL_WARNING("CL/sycl/backend/level_zero.hpp usage is deprecated, include " + "sycl/ext/oneapi/backend/level_zero.hpp instead") + +#include diff --git a/sycl/test/warnings/sycl_2020_deprecations.cpp b/sycl/test/warnings/sycl_2020_deprecations.cpp index 43265f85580ca..6aab31242b819 100644 --- a/sycl/test/warnings/sycl_2020_deprecations.cpp +++ b/sycl/test/warnings/sycl_2020_deprecations.cpp @@ -152,5 +152,28 @@ int main() { auto SL = sycl::INTEL::source_language::opencl_c; (void)SL; + // expected-warning@+1{{'intel' is deprecated: use 'ext::intel::experimental' instead}} + auto SLExtIntel = sycl::ext::intel::source_language::opencl_c; + (void)SLExtIntel; + + // expected-warning@+1{{'level_zero' is deprecated: use 'ext_oneapi_level_zero' instead}} + auto LevelZeroBackend = sycl::backend::level_zero; + (void)LevelZeroBackend; + + sycl::half Val = 1.0f; + // expected-warning@+1{{'bit_cast' is deprecated: use 'sycl::bit_cast' instead}} + auto BitCastRes = sycl::detail::bit_cast(Val); + (void)BitCastRes; + + // expected-warning@+1{{'submit_barrier' is deprecated: use 'ext_intel_submit_barrier' instead}} + Queue.submit_barrier(); + + // expected-warning@+1{{'barrier' is deprecated: use 'ext_intel_barrier' instead}} + Queue.submit([&](sycl::handler &CGH) { CGH.barrier(); }); + + // expected-warning@+1{{'mem_channel' is deprecated: use 'ext_intel_mem_channel' instead}} + sycl::property_list MemChannelProp{sycl::property::buffer::mem_channel(2)}; + (void)MemChannelProp; + return 0; } From 8315b0138ef73f2a7e19a320c57ea8d4e32b695f Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Wed, 1 Sep 2021 16:28:05 +0300 Subject: [PATCH 04/10] Update USMAddressSpaces spec: device/host_space -> ext_intel_global_device/host_space --- .../USMAddressSpaces/usm_address_spaces.asciidoc | 14 +++++++------- 1 file changed, 7 insertions(+), 7 deletions(-) diff --git a/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc b/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc index e00de9e430db3..a5f28a2d038db 100644 --- a/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc +++ b/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc @@ -87,15 +87,15 @@ enum class address_space : int { local_space, constant_space, private_space, - device_space, - host_space + ext_intel_global_device_space, + ext_intel_global_host_space }; ``` Add the following new conversion operator: ```c++ // Explicit conversion to global_space -// Only available if Space == address_space::device_space || Space == address_space::host_space +// Only available if Space == address_space::ext_intel_global_device_space || Space == address_space::ext_intel_global_host_space explicit operator multi_ptr() const; ``` @@ -109,12 +109,12 @@ a| ```c++ template +ext_intel_global_device_space> template multi_ptr( accessor) -``` | Constructs a multi_ptr from an accessor of access::target::global_buffer. +``` | Constructs a multi_ptr from an accessor of access::target::global_buffer. |=== -- @@ -123,10 +123,10 @@ device_space> Add device_ptr and host_ptr aliases to the list of multi_ptr aliases as follows: ```c++ template -using device_ptr = multi_ptr +using device_ptr = multi_ptr template -using host_ptr = multi_ptr +using host_ptr = multi_ptr ``` == Revision History From 8a83962d83b4e3d4bb38de25550112b7b497f210 Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Wed, 1 Sep 2021 16:44:50 +0300 Subject: [PATCH 05/10] Update name strings in extension specs --- sycl/doc/extensions/EnqueueBarrier/enqueue_barrier.asciidoc | 2 +- .../doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc | 2 +- 2 files changed, 2 insertions(+), 2 deletions(-) diff --git a/sycl/doc/extensions/EnqueueBarrier/enqueue_barrier.asciidoc b/sycl/doc/extensions/EnqueueBarrier/enqueue_barrier.asciidoc index 15350727b03b0..20d23e5cca8fd 100644 --- a/sycl/doc/extensions/EnqueueBarrier/enqueue_barrier.asciidoc +++ b/sycl/doc/extensions/EnqueueBarrier/enqueue_barrier.asciidoc @@ -28,7 +28,7 @@ This document presents a series of changes proposed for a future version of the == Name Strings -+SYCL_INTEL_enqueue_barrier+ ++SYCL_EXT_INTEL_ENQUEUE_BARRIER+ == Notice diff --git a/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc b/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc index a5f28a2d038db..2ddd6cc158870 100644 --- a/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc +++ b/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc @@ -12,7 +12,7 @@ NOTE: This document is better viewed when rendered as html with asciidoctor. Gi This document describes an extension to the SYCL USM extension that adds new explicit address spaces for the possible locations that USM pointers can be allocated. Users can create pointers that point into these address spaces explicitly in order to pass additional information to their compiler so as to enable optimizations. == Name Strings -+SYCL_INTEL_usm_address_spaces+ ++SYCL_EXT_INTEL_USM_ADDRESS_SPACES+ == Notice Copyright (c) 2020 Intel Corporation. All rights reserved. From 8047c6ff82bb4519e2e8cbda1f817526ee5e915a Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Fri, 3 Sep 2021 21:12:36 +0300 Subject: [PATCH 06/10] Applied CR comments --- .../EnqueueBarrier/enqueue_barrier.asciidoc | 71 +++++++++---------- .../LevelZeroBackend/LevelZeroBackend.md | 1 - .../extensions/MemChannel/MemChannel.asciidoc | 10 +-- .../usm_address_spaces.asciidoc | 30 ++++---- sycl/include/CL/sycl/feature_test.hpp | 2 +- sycl/include/CL/sycl/handler.hpp | 4 +- sycl/include/CL/sycl/queue.hpp | 14 ++-- sycl/source/abi_replacements_windows.cpp | 4 +- sycl/source/handler.cpp | 5 -- sycl/test/abi/sycl_symbols_windows.dump | 4 +- 10 files changed, 66 insertions(+), 79 deletions(-) diff --git a/sycl/doc/extensions/EnqueueBarrier/enqueue_barrier.asciidoc b/sycl/doc/extensions/EnqueueBarrier/enqueue_barrier.asciidoc index 20d23e5cca8fd..9b2b7789f196a 100644 --- a/sycl/doc/extensions/EnqueueBarrier/enqueue_barrier.asciidoc +++ b/sycl/doc/extensions/EnqueueBarrier/enqueue_barrier.asciidoc @@ -1,4 +1,4 @@ -= SYCL_EXT_INTEL_ENQUEUE_BARRIER += SYCL_EXT_ONEAPI_ENQUEUE_BARRIER :source-highlighter: coderay :coderay-linenums-mode: table @@ -25,11 +25,6 @@ NOTE: This document is better viewed when rendered as html with asciidoctor. Gi This document presents a series of changes proposed for a future version of the SYCL Specification. The goal of this proposal is to provide non-blocking APIs that provide synchronization on SYCL command queue for programmers. - -== Name Strings - -+SYCL_EXT_INTEL_ENQUEUE_BARRIER+ - == Notice Copyright (c) 2019-2020 Intel Corporation. All rights reserved. @@ -45,7 +40,7 @@ Because the interfaces defined by this specification are not final and are subje == Version Built On: {docdate} + -Revision: 1 +Revision: 2 == Contact Please open an issue in the https://github.com/intel/llvm/tree/sycl/sycl/doc/extensions/[extensions repository] @@ -55,7 +50,7 @@ Please open an issue in the https://github.com/intel/llvm/tree/sycl/sycl/doc/ext This extension provides a feature-test macro as described in the core SYCL specification section 6.3.3 "Feature test macros". Therefore, an implementation supporting this extension must predefine the macro -`SYCL_EXT_INTEL_ENQUEUE_BARRIER` to one of the values defined in the table below. +`SYCL_EXT_ONEAPI_ENQUEUE_BARRIER` to one of the values defined in the table below. Applications can test for the existence of this macro to determine if the implementation supports this feature, or applications can test the macro's value to determine which of the extension's APIs the implementation supports. @@ -68,12 +63,12 @@ value to determine which of the extension's APIs the implementation supports. == Dependencies -This extension is written against the SYCL 1.2.1 specification, Revision v1.2.1-6. +This extension is written against the SYCL 2020 specification, revision 3. == Overview -SYCL 1.2.1 defines a graph-based task execution model, based on kernels or explicit memory operations submitted to out-of-order queues. Dependencies between these kernels are represented by -accessors that form data dependence edges in the execution graph. The USM extension <> doesn't have accessors, so instead solves +SYCL 2020 defines a graph-based task execution model, based on kernels or explicit memory operations submitted to out-of-order queues. Dependencies between these kernels are represented by +accessors that form data dependence edges in the execution graph. Unified Shared Memory (USM) doesn't have accessors, so instead solves this by defining `handler::depends_on` methods to specify event-based control dependencies between command groups. There are situations where defining dependencies based on events is more explicit than desired or required by an application. For instance, the user may know that a given task depends on all previously submitted tasks. Instead of explicitly adding all the required depends_on calls, the user could express this intent via a single call, making the program more concise and explicit. @@ -91,9 +86,9 @@ two new members to the `queue` class: [grid="rows"] [options="header"] |======================================== -|*handler::ext_intel_barrier*|*queue::ext_intel_submit_barrier* -|`void ext_intel_barrier()` | `event ext_intel_submit_barrier()` -|`void ext_intel_barrier( const vector_class &waitList )` | `event ext_intel_submit_barrier( const vector_class &waitList )` +|*handler::ext_intel_barrier*|*queue::ext_oneapi_submit_barrier* +|`void ext_intel_barrier()` | `event ext_oneapi_submit_barrier()` +|`void ext_intel_barrier( const vector_class &waitList )` | `event ext_oneapi_submit_barrier( const vector_class &waitList )` |======================================== The first variant of the barrier takes no parameters, and waits for all previously submitted commands to the queue to enter the `info::event_command_status::complete` state before any command later submitted to the same queue is allowed to execute. A second variant of the barrier accepts a list of events, with the behavior that no commands submitted to the same queue after barrier submission may execute until all events in the `waitList` have entered the `info::event_command_status::complete` state. Both variants are non-blocking from the host program perspective, in that they do not wait for the barrier conditions to have been met before returning. @@ -134,7 +129,7 @@ Queue.submit([&](cl::sycl::handler& cgh) { ... ---- -==== 2. Using `queue::ext_intel_submit_barrier()`: +==== 2. Using `queue::ext_oneapi_submit_barrier()`: [source,c++,NoName,linenums] ---- @@ -149,7 +144,7 @@ Queue.submit([&](cl::sycl::handler& cgh) { // CG3 }); -Queue.ext_intel_submit_barrier(); +Queue.ext_oneapi_submit_barrier(); Queue.submit([&](cl::sycl::handler& cgh) { // CG4 @@ -185,7 +180,7 @@ Queue3.submit([&](cl::sycl::handler& cgh) { ... ---- -==== 2. Using `queue::ext_intel_submit_barrier()`: +==== 2. Using `queue::ext_oneapi_submit_barrier()`: [source,c++,NoName,linenums] ---- @@ -198,7 +193,7 @@ auto event_barrier2 = Queue2.submit([&](cl::sycl::handler& cgh) { // CG2 }); -Queue3.ext_intel_submit_barrier( vector_class{event_barrier1, event_barrier2} ); +Queue3.ext_oneapi_submit_barrier( vector_class{event_barrier1, event_barrier2} ); Queue3.submit([&](cl::sycl::handler& cgh) { // CG3 @@ -227,44 +222,45 @@ void wait(); template event submit(T cgf, const queue &secondaryQueue); -event ext_intel_submit_barrier(); +event ext_oneapi_submit_barrier(); -event ext_intel_submit_barrier( const vector_class &waitList ); +event ext_oneapi_submit_barrier( const vector_class &waitList ); void wait(); ... ---- -=== Add rows to Table 4.22 +=== Add rows to Table 28 [cols="70,300"] [grid="rows"] [options="header"] |======================================== |*Member functions*|*Description* -|`event ext_intel_submit_barrier()` | Same effect as submitting a `handler::ext_intel_barrier()` within a command group to this `queue`. The returned event enters the `info::event_command_status::complete` state when all events that the barrier is dependent on (implicitly from all previously submitted commands to the same queue) have entered the `info::event_command_status::complete` state. -|`event ext_intel_submit_barrier( const vector_class &waitList )` | Same effect as submitting a `handler:ext_intel_barrier( const vector_class &waitList )` within a command group to this `queue`. The returned event enters the `info::event_command_status::complete` state when all events that the barrier is dependent on (explicitly from `waitList`) have entered the `info::event_command_status::complete` state. +|`event ext_oneapi_submit_barrier()` | Same effect as submitting a `handler::ext_intel_barrier()` within a command group to this `queue`. The returned event enters the `info::event_command_status::complete` state when all events that the barrier is dependent on (implicitly from all previously submitted commands to the same queue) have entered the `info::event_command_status::complete` state. +|`event ext_oneapi_submit_barrier( const vector_class &waitList )` | Same effect as submitting a `handler:ext_intel_barrier( const vector_class &waitList )` within a command group to this `queue`. The returned event enters the `info::event_command_status::complete` state when all events that the barrier is dependent on (explicitly from `waitList`) have entered the `info::event_command_status::complete` state. |======================================== -=== Modify Section 4.8.2 +=== Modify Section 4.9.3 ==== Change first sentence from: -A command group scope in SYCL, as it is defined in Section 3.4.1, consists of a single kernel or explicit memory -operation (handler methods such as copy, update_host, fill), together with its requirements. +The member functions and objects defined in this scope will define the requirements for the kernel execution or +explicit memory operation, and will be used by the SYCL runtime to evaluate if the operation is ready for execution. ==== To: -A command group scope in SYCL, as it is defined in Section 3.4.1, consists of a single kernel, explicit memory -operation (handler methods such as copy, update_host, fill) or barrier, together with its requirements. +The member functions and objects defined in this scope will define the requirements for the kernel execution, +explicit memory operation or barrier, and will be used by the SYCL runtime to evaluate if the operation is ready for execution. + -=== Modify part of Section 4.8.3 +=== Modify part of Section 4.9.4 *Change from:* [source,c++,NoName,linenums] ---- ... -template -void fill(accessor dest, const T& src); +template +void fill(void *ptr, const T &pattern, size_t count); }; ... @@ -274,8 +270,8 @@ void fill(accessor dest, const T& src); [source,c++,NoName,linenums] ---- ... -template -void fill(accessor dest, const T& src); +template +void fill(void *ptr, const T &pattern, size_t count); void ext_intel_barrier(); @@ -285,15 +281,15 @@ void ext_intel_barrier( const vector_class &waitList ); ... ---- -=== Add a new section between Section 4.8.6 and 4.8.7 +=== Add a new section between Section 4.9.4 and 4.9.5 -4.8.X SYCL functions for enqueued synchronization barriers +4.9.X SYCL functions for enqueued synchronization barriers Barriers may be submitted to a queue, with the effect that they prevent later operations submitted to the same queue from executing until the barrier wait conditions have been satisfied. The wait conditions can be explicitly described by `waitList` or implicitly from all previously submitted commands to the same queue. There are no constraints on the context from which queues may participate in the `waitList`. Enqueued barriers do not block host program execution, but instead form additional dependence edges with the execution task graph. Barriers can be created by two members of the `handler` class that force synchronization on the SYCL command queue. The first variant of the `handler` barrier (`handler::barrier()`) takes no parameters, and waits for all previously submitted commands to the queue to enter the `info::event_command_status::complete` state before any command later submitted to the same queue is allowed to execute. The second variant of the `handler` barrier (`handler::barrier( const vector_class &waitList )`) accepts a list of events, with the behavior that no commands submitted to the same queue after barrier submission may execute until all events in the waitList have entered the `info::event_command_status::complete` state. -=== Add a new table in the new section between 4.8.6 and 4.8.7: Member functions of the handler class. +=== Add a new table in the new section between 4.9.4 and 4.9.5: Member functions of the handler class. [cols="70,300"] [grid="rows"] @@ -304,9 +300,6 @@ Barriers can be created by two members of the `handler` class that force synchro |`void ext_intel_barrier( const vector_class &waitList` ) | Prevents any commands submitted afterward to this queue from executing until all events in `waitList` have entered the `info::event_command_status::complete` state. If `waitList` is empty, then the barrier has no effect. |======================================== -== References -1. [[usmlink]]https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/USM/USM.adoc - == Issues None. diff --git a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md index 1d24612606912..55ccf83219f1e 100644 --- a/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md +++ b/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md @@ -9,7 +9,6 @@ The currently supported targets are all Intel GPUs starting with Gen9. NOTE: This specification is a draft. While describing the currently implemented behaviors it is known to be not complete nor exhaustive. We shall continue to add more information, e.g. explain general mapping of SYCL programming model to Level-Zero API. - It will also be gradually changing to a SYCL-2020 conforming implementation. ## 2. Prerequisites diff --git a/sycl/doc/extensions/MemChannel/MemChannel.asciidoc b/sycl/doc/extensions/MemChannel/MemChannel.asciidoc index dfedebf986766..343e4404a5c1a 100644 --- a/sycl/doc/extensions/MemChannel/MemChannel.asciidoc +++ b/sycl/doc/extensions/MemChannel/MemChannel.asciidoc @@ -23,11 +23,11 @@ Because the interfaces defined by this specification are not final and are subje == Version Built On: {docdate} + -Revision: 1 +Revision: 2 == Dependencies -This extension is written against the SYCL 2020 provisional specification, Revision 1. +This extension is written against the SYCL 2020 specification, Revision 3. The use of this extension requires a target that supports cl_intel_mem_channel_property or equivalent if OpenCL is used as the underlying device runtime. @@ -65,7 +65,7 @@ Add a new property to Table 4.33: Properties supported by the SYCL buffer class |=== -- -Add a new constructor to Table 4.34: Constructors of the buffer property classes as follows: +Add a new constructor to Table 41: Constructors of the buffer property classes as follows: -- [options="header"] @@ -75,7 +75,7 @@ Add a new constructor to Table 4.34: Constructors of the buffer property classes |=== -- -Add a new member function to Table 4.35: Member functions of the buffer property classes as follows: +Add a new member function to Table 42: Member functions of the buffer property classes as follows: -- [options="header"] @@ -103,7 +103,7 @@ enum class aspect { } // namespace sycl ``` -Add an entry for the new aspect to Table 4.20: Device aspects defined by the core SYCL specification: +Add an entry for the new aspect to Table 26: Device aspects defined by the core SYCL specification: -- [options="header"] diff --git a/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc b/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc index 2ddd6cc158870..5eda13b1b93dd 100644 --- a/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc +++ b/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc @@ -11,9 +11,6 @@ NOTE: Khronos(R) is a registered trademark and SYCL(TM) and SPIR(TM) are tradema NOTE: This document is better viewed when rendered as html with asciidoctor. GitHub does not render image icons. This document describes an extension to the SYCL USM extension that adds new explicit address spaces for the possible locations that USM pointers can be allocated. Users can create pointers that point into these address spaces explicitly in order to pass additional information to their compiler so as to enable optimizations. -== Name Strings -+SYCL_EXT_INTEL_USM_ADDRESS_SPACES+ - == Notice Copyright (c) 2020 Intel Corporation. All rights reserved. @@ -28,11 +25,11 @@ Because the interfaces defined by this specification are not final and are subje == Version Built On: {docdate} + -Revision: 1 +Revision: B == Dependencies -This extension is written against the SYCL 1.2.1 specification, Revision 7. It requires the Unified Shared Memory SYCL proposal. +This extension is written against the SYCL 2020 specification, Revision 3. If SPIR-V is used by the implementation, this extension also requires support for the SPV_INTEL_usm_storage_classes SPIR-V extension. @@ -61,9 +58,9 @@ The goal of this division of the global address space is to enable users to expl While automatic address space inference is often possible for accessors, it is harder for USM pointers as it requires inter-procedural optimization with the host code. This additional information can be particularly beneficial on FPGA targets where knowing that a pointer only ever accesses host or device memory can allow compilers to produce more area efficient memory-accessing hardware. -== Modifications to the SYCL Specification, Version 1.2.1 revision 7 +== Modifications to the SYCL Specification, Version 2020 revision 3 -=== Section 3.5.2 SYCL Device Memory Model +=== Section 3.8.2 SYCL Device Memory Model Add to the end of the definition of global memory: Global memory is a virtual address space which overlaps the device and host address spaces. @@ -74,21 +71,22 @@ Add two new memory regions as follows: *Host memory* is a sub-region of global memory. USM pointers allocated with the host alloc type reside in this address space. -=== Section 3.5.2.1 Access to memory +=== Section 3.8.2.1 Access to memory -In the second last paragraph, add cl::sycl::device_ptr and cl::sycl::host_ptr to the list of explicit pointer classes. +In the second last paragraph, add sycl::device_ptr and sycl::host_ptr to the list of explicit pointer classes. === Section 4.7.7.1 Multi-pointer Class In the overview of the multi_ptr class replace the address_space enum with the following: ```c++ enum class address_space : int { - global_space, - local_space, - constant_space, - private_space, - ext_intel_global_device_space, - ext_intel_global_host_space + global_space, + local_space, + constant_space, // Deprecated in SYCL 2020 + private_space, + generic_space, + ext_intel_global_device_space, + ext_intel_global_host_space }; ``` @@ -99,7 +97,7 @@ Add the following new conversion operator: explicit operator multi_ptr() const; ``` -Add a new row to Table 4.54: Constructors of the SYCL multi_ptr class template, as follows: +Add a new row to Table 91: Constructors of the SYCL multi_ptr class template, as follows: -- [options="header"] diff --git a/sycl/include/CL/sycl/feature_test.hpp b/sycl/include/CL/sycl/feature_test.hpp index ef7c55c4946b9..08f51244adc90 100644 --- a/sycl/include/CL/sycl/feature_test.hpp +++ b/sycl/include/CL/sycl/feature_test.hpp @@ -24,7 +24,7 @@ namespace sycl { #define SYCL_EXT_ONEAPI_MATRIX 2 #endif #define SYCL_EXT_INTEL_BF16_CONVERSION 1 -#define SYCL_EXT_INTEL_ENQUEUE_BARRIER 1 +#define SYCL_EXT_ONEAPI_ENQUEUE_BARRIER 1 #define SYCL_EXT_INTEL_MEM_CHANNEL_PROPERTY 1 #define SYCL_EXT_INTEL_USM_ADDRESS_SPACES 1 #define SYCL_EXT_ONEAPI_BACKEND_LEVEL_ZERO 1 diff --git a/sycl/include/CL/sycl/handler.hpp b/sycl/include/CL/sycl/handler.hpp index bc60902a65bfc..dcd873d34e324 100644 --- a/sycl/include/CL/sycl/handler.hpp +++ b/sycl/include/CL/sycl/handler.hpp @@ -2307,7 +2307,9 @@ class __SYCL_EXPORT handler { /// \param WaitList is a vector of valid SYCL events that need to complete /// before barrier command can be executed. __SYCL2020_DEPRECATED("use 'ext_intel_barrier' instead") - void barrier(const std::vector &WaitList); + void barrier(const std::vector &WaitList) { + ext_intel_barrier(WaitList); + } /// Copies data from one memory region to another, both pointed by /// USM pointers. diff --git a/sycl/include/CL/sycl/queue.hpp b/sycl/include/CL/sycl/queue.hpp index f7367cace346c..ea405a56d6344 100644 --- a/sycl/include/CL/sycl/queue.hpp +++ b/sycl/include/CL/sycl/queue.hpp @@ -251,7 +251,7 @@ class __SYCL_EXPORT queue { /// \param CodeLoc is the code location of the submit call (default argument) /// \return a SYCL event object, which corresponds to the queue the command /// group is being enqueued on. - event ext_intel_submit_barrier(_CODELOCONLYPARAM(&CodeLoc)) { + event ext_oneapi_submit_barrier(_CODELOCONLYPARAM(&CodeLoc)) { return submit( [=](handler &CGH) { CGH.ext_intel_barrier(); } _CODELOCFW(CodeLoc)); } @@ -263,22 +263,22 @@ class __SYCL_EXPORT queue { /// \param CodeLoc is the code location of the submit call (default argument) /// \return a SYCL event object, which corresponds to the queue the command /// group is being enqueued on. - __SYCL2020_DEPRECATED("use 'ext_intel_submit_barrier' instead") + __SYCL2020_DEPRECATED("use 'ext_oneapi_submit_barrier' instead") event submit_barrier(_CODELOCONLYPARAM(&CodeLoc)) { _CODELOCARG(&CodeLoc); - return ext_intel_submit_barrier(CodeLoc); + return ext_oneapi_submit_barrier(CodeLoc); } /// Prevents any commands submitted afterward to this queue from executing /// until all events in WaitList have entered the complete state. If WaitList - /// is empty, then ext_intel_submit_barrier has no effect. + /// is empty, then ext_oneapi_submit_barrier has no effect. /// /// \param WaitList is a vector of valid SYCL events that need to complete /// before barrier command can be executed. /// \param CodeLoc is the code location of the submit call (default argument) /// \return a SYCL event object, which corresponds to the queue the command /// group is being enqueued on. - event ext_intel_submit_barrier( + event ext_oneapi_submit_barrier( const std::vector &WaitList _CODELOCPARAM(&CodeLoc)) { return submit([=](handler &CGH) { CGH.ext_intel_barrier(WaitList); @@ -294,11 +294,11 @@ class __SYCL_EXPORT queue { /// \param CodeLoc is the code location of the submit call (default argument) /// \return a SYCL event object, which corresponds to the queue the command /// group is being enqueued on. - __SYCL2020_DEPRECATED("use 'ext_intel_submit_barrier' instead") + __SYCL2020_DEPRECATED("use 'ext_oneapi_submit_barrier' instead") event submit_barrier(const std::vector &WaitList _CODELOCPARAM(&CodeLoc)) { _CODELOCARG(&CodeLoc); - return ext_intel_submit_barrier(WaitList, CodeLoc); + return ext_oneapi_submit_barrier(WaitList, CodeLoc); } /// Performs a blocking wait for the completion of all enqueued tasks in the diff --git a/sycl/source/abi_replacements_windows.cpp b/sycl/source/abi_replacements_windows.cpp index ab4c9fa223e94..609954225e91e 100644 --- a/sycl/source/abi_replacements_windows.cpp +++ b/sycl/source/abi_replacements_windows.cpp @@ -30,7 +30,7 @@ "/export:??$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@@INTEL@sycl@cl@@QEAA?AV?$vector@EV?$allocator@E@std@@@std@@AEBV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@5@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@@5@@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@sycl@cl@@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") #pragma comment( \ linker, \ - "/export:?submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBV?$vector@Vevent@sycl@cl@@V?$allocator@Vevent@sycl@cl@@@std@@@std@@AEBUcode_location@detail@23@@Z=?ext_intel_submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBV?$vector@Vevent@sycl@cl@@V?$allocator@Vevent@sycl@cl@@@std@@@std@@AEBUcode_location@detail@23@@Z") + "/export:?submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBV?$vector@Vevent@sycl@cl@@V?$allocator@Vevent@sycl@cl@@@std@@@std@@AEBUcode_location@detail@23@@Z=?ext_oneapi_submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBV?$vector@Vevent@sycl@cl@@V?$allocator@Vevent@sycl@cl@@@std@@@std@@AEBUcode_location@detail@23@@Z") #pragma comment( \ linker, \ "/export:?make_queue@level_zero@sycl@cl@@YA?AVqueue@23@AEBVcontext@23@_K_N@Z=?make_queue@level_zero@oneapi@ext@sycl@cl@@YA?AVqueue@45@AEBVcontext@45@_K_N@Z") @@ -48,4 +48,4 @@ "/export:??$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@INTEL@sycl@cl@@QEAA?AV?$vector@EV?$allocator@E@std@@@std@@AEBV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@5@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@@5@@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@$00@experimental@intel@ext@sycl@cl@@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") #pragma comment( \ linker, \ - "/export:?submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBUcode_location@detail@23@@Z=?ext_intel_submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBUcode_location@detail@23@@Z") + "/export:?submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBUcode_location@detail@23@@Z=?ext_oneapi_submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBUcode_location@detail@23@@Z") diff --git a/sycl/source/handler.cpp b/sycl/source/handler.cpp index 6fedb20cb2cb5..aefeece16d503 100644 --- a/sycl/source/handler.cpp +++ b/sycl/source/handler.cpp @@ -497,11 +497,6 @@ void handler::ext_intel_barrier(const std::vector &WaitList) { [](const event &Event) { return detail::getSyclObjImpl(Event); }); } -__SYCL2020_DEPRECATED("use 'ext_intel_barrier' instead") -void handler::barrier(const std::vector &WaitList) { - handler::ext_intel_barrier(WaitList); -} - using namespace sycl::detail; bool handler::DisableRangeRounding() { return SYCLConfig::get(); diff --git a/sycl/test/abi/sycl_symbols_windows.dump b/sycl/test/abi/sycl_symbols_windows.dump index b2c238c592408..72d8a01f6b2a4 100644 --- a/sycl/test/abi/sycl_symbols_windows.dump +++ b/sycl/test/abi/sycl_symbols_windows.dump @@ -1742,8 +1742,8 @@ ?expm1@__host_std@cl@@YANN@Z ?ext_intel_barrier@handler@sycl@cl@@QEAAXAEBV?$vector@Vevent@sycl@cl@@V?$allocator@Vevent@sycl@cl@@@std@@@std@@@Z ?ext_intel_barrier@handler@sycl@cl@@QEAAXXZ -?ext_intel_submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBUcode_location@detail@23@@Z -?ext_intel_submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBV?$vector@Vevent@sycl@cl@@V?$allocator@Vevent@sycl@cl@@@std@@@std@@AEBUcode_location@detail@23@@Z +?ext_oneapi_submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBUcode_location@detail@23@@Z +?ext_oneapi_submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBV?$vector@Vevent@sycl@cl@@V?$allocator@Vevent@sycl@cl@@@std@@@std@@AEBUcode_location@detail@23@@Z ?extractArgsAndReqs@handler@sycl@cl@@AEAAXXZ ?extractArgsAndReqsFromLambda@handler@sycl@cl@@AEAAXPEAD_KPEBUkernel_param_desc_t@detail@23@@Z ?extractArgsAndReqsFromLambda@handler@sycl@cl@@AEAAXPEAD_KPEBUkernel_param_desc_t@detail@23@_N@Z From d849467ac1826aaee7997d97850825c5182c6c0c Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Mon, 6 Sep 2021 15:14:40 +0300 Subject: [PATCH 07/10] Apply comments + fix pre-commit --- .../EnqueueBarrier/enqueue_barrier.asciidoc | 26 +++++++------- .../extensions/MemChannel/MemChannel.asciidoc | 36 +++++-------------- .../SYCL_ONEAPI_dot_accumulate.asciidoc | 8 ++--- .../usm_address_spaces.asciidoc | 6 ++-- .../SYCL_INTEL_buffer_location.asciidoc | 4 +-- sycl/include/CL/sycl/feature_test.hpp | 1 - sycl/include/CL/sycl/handler.hpp | 14 ++++---- .../CL/sycl/properties/buffer_properties.hpp | 12 ++----- sycl/include/CL/sycl/queue.hpp | 4 +-- sycl/source/handler.cpp | 7 +++- sycl/test/abi/sycl_symbols_linux.dump | 2 +- sycl/test/abi/sycl_symbols_windows.dump | 4 +-- sycl/test/warnings/sycl_2020_deprecations.cpp | 8 ++--- 13 files changed, 53 insertions(+), 79 deletions(-) diff --git a/sycl/doc/extensions/EnqueueBarrier/enqueue_barrier.asciidoc b/sycl/doc/extensions/EnqueueBarrier/enqueue_barrier.asciidoc index 9b2b7789f196a..880313dae5af3 100644 --- a/sycl/doc/extensions/EnqueueBarrier/enqueue_barrier.asciidoc +++ b/sycl/doc/extensions/EnqueueBarrier/enqueue_barrier.asciidoc @@ -86,9 +86,9 @@ two new members to the `queue` class: [grid="rows"] [options="header"] |======================================== -|*handler::ext_intel_barrier*|*queue::ext_oneapi_submit_barrier* -|`void ext_intel_barrier()` | `event ext_oneapi_submit_barrier()` -|`void ext_intel_barrier( const vector_class &waitList )` | `event ext_oneapi_submit_barrier( const vector_class &waitList )` +|*handler::ext_oneapi_barrier*|*queue::ext_oneapi_submit_barrier* +|`void ext_oneapi_barrier()` | `event ext_oneapi_submit_barrier()` +|`void ext_oneapi_barrier( const vector_class &waitList )` | `event ext_oneapi_submit_barrier( const vector_class &waitList )` |======================================== The first variant of the barrier takes no parameters, and waits for all previously submitted commands to the queue to enter the `info::event_command_status::complete` state before any command later submitted to the same queue is allowed to execute. A second variant of the barrier accepts a list of events, with the behavior that no commands submitted to the same queue after barrier submission may execute until all events in the `waitList` have entered the `info::event_command_status::complete` state. Both variants are non-blocking from the host program perspective, in that they do not wait for the barrier conditions to have been met before returning. @@ -104,7 +104,7 @@ Some forms of the new barrier methods return an `event`, which can be used to pe CG4 doesn't execute until all previous command groups submitted to the same queue (CG1, CG2, CG3) have entered the completed state. -==== 1. Using `handler::ext_intel_barrier()`: +==== 1. Using `handler::ext_oneapi_barrier()`: [source,c++,NoName,linenums] ---- @@ -120,7 +120,7 @@ Queue.submit([&](cl::sycl::handler& cgh) { }); Queue.submit([&](cl::sycl::handler& cgh) { - cgh.ext_intel_barrier(); + cgh.ext_oneapi_barrier(); }); Queue.submit([&](cl::sycl::handler& cgh) { @@ -157,7 +157,7 @@ Queue.submit([&](cl::sycl::handler& cgh) { CG3 requires CG1 (in Queue1) and CG2 (in Queue2) to have completed before it (CG3) begins execution. -==== 1. Using `handler::ext_intel_barrier()`: +==== 1. Using `handler::ext_oneapi_barrier()`: [source,c++,NoName,linenums] ---- @@ -171,7 +171,7 @@ auto event_barrier2 = Queue2.submit([&](cl::sycl::handler& cgh) { }); Queue3.submit([&](cl::sycl::handler& cgh) { - cgh.ext_intel_barrier( vector_class{event_barrier1, event_barrier2} ); + cgh.ext_oneapi_barrier( vector_class{event_barrier1, event_barrier2} ); }); Queue3.submit([&](cl::sycl::handler& cgh) { @@ -236,8 +236,8 @@ void wait(); [options="header"] |======================================== |*Member functions*|*Description* -|`event ext_oneapi_submit_barrier()` | Same effect as submitting a `handler::ext_intel_barrier()` within a command group to this `queue`. The returned event enters the `info::event_command_status::complete` state when all events that the barrier is dependent on (implicitly from all previously submitted commands to the same queue) have entered the `info::event_command_status::complete` state. -|`event ext_oneapi_submit_barrier( const vector_class &waitList )` | Same effect as submitting a `handler:ext_intel_barrier( const vector_class &waitList )` within a command group to this `queue`. The returned event enters the `info::event_command_status::complete` state when all events that the barrier is dependent on (explicitly from `waitList`) have entered the `info::event_command_status::complete` state. +|`event ext_oneapi_submit_barrier()` | Same effect as submitting a `handler::ext_oneapi_barrier()` within a command group to this `queue`. The returned event enters the `info::event_command_status::complete` state when all events that the barrier is dependent on (implicitly from all previously submitted commands to the same queue) have entered the `info::event_command_status::complete` state. +|`event ext_oneapi_submit_barrier( const vector_class &waitList )` | Same effect as submitting a `handler:ext_oneapi_barrier( const vector_class &waitList )` within a command group to this `queue`. The returned event enters the `info::event_command_status::complete` state when all events that the barrier is dependent on (explicitly from `waitList`) have entered the `info::event_command_status::complete` state. |======================================== @@ -273,9 +273,9 @@ void fill(void *ptr, const T &pattern, size_t count); template void fill(void *ptr, const T &pattern, size_t count); -void ext_intel_barrier(); +void ext_oneapi_barrier(); -void ext_intel_barrier( const vector_class &waitList ); +void ext_oneapi_barrier( const vector_class &waitList ); }; ... @@ -296,8 +296,8 @@ Barriers can be created by two members of the `handler` class that force synchro [options="header"] |======================================== |*Member functions*|*Description* -|`void ext_intel_barrier()` | Prevents any commands submitted afterward to this queue from executing until all commands previously submitted to this queue have entered the `info::event_command_status::complete` state. -|`void ext_intel_barrier( const vector_class &waitList` ) | Prevents any commands submitted afterward to this queue from executing until all events in `waitList` have entered the `info::event_command_status::complete` state. If `waitList` is empty, then the barrier has no effect. +|`void ext_oneapi_barrier()` | Prevents any commands submitted afterward to this queue from executing until all commands previously submitted to this queue have entered the `info::event_command_status::complete` state. +|`void ext_oneapi_barrier( const vector_class &waitList` ) | Prevents any commands submitted afterward to this queue from executing until all events in `waitList` have entered the `info::event_command_status::complete` state. If `waitList` is empty, then the barrier has no effect. |======================================== == Issues diff --git a/sycl/doc/extensions/MemChannel/MemChannel.asciidoc b/sycl/doc/extensions/MemChannel/MemChannel.asciidoc index 343e4404a5c1a..15b309851043f 100644 --- a/sycl/doc/extensions/MemChannel/MemChannel.asciidoc +++ b/sycl/doc/extensions/MemChannel/MemChannel.asciidoc @@ -1,4 +1,4 @@ -= SYCL_EXT_INTEL_MEM_CHANNEL_PROPERTY += SYCL_INTEL_mem_channel_property == Introduction NOTE: Khronos(R) is a registered trademark and SYCL(TM) and SPIR(TM) are trademarks of The Khronos Group Inc. OpenCL(TM) is a trademark of Apple Inc. used by permission by Khronos. @@ -23,30 +23,14 @@ Because the interfaces defined by this specification are not final and are subje == Version Built On: {docdate} + -Revision: 2 +Revision: 1 == Dependencies -This extension is written against the SYCL 2020 specification, Revision 3. +This extension is written against the SYCL 2020 provisional specification, Revision 1. The use of this extension requires a target that supports cl_intel_mem_channel_property or equivalent if OpenCL is used as the underlying device runtime. -== Feature Test Macro - -This extension provides a feature-test macro as described in the core SYCL -specification section 6.3.3 "Feature test macros". Therefore, an -implementation supporting this extension must predefine the macro -`SYCL_EXT_INTEL_MEM_CHANNEL_PROPERTY` to one of the values defined in the table below. -Applications can test for the existence of this macro to determine if the -implementation supports this feature, or applications can test the macro's -value to determine which of the extension's APIs the implementation supports. - -[%header,cols="1,5"] -|=== -|Value |Description -|1 |Initial extension version. Base features are supported. -|=== - == Overview On some targets manual assignment of buffers to memory regions can improve memory bandwidth. This extension adds a buffer property to indicate in which memory channel a particular buffer should be allocated. This information is an optimization hint to the runtime and thus it is legal to ignore. @@ -61,27 +45,27 @@ Add a new property to Table 4.33: Properties supported by the SYCL buffer class [options="header"] |=== | Property | Description -| property::buffer::ext_intel_mem_channel | The `ext_intel_mem_channel` property is a hint to the SYCL runtime that the buffer should be stored in a particular memory channel provided to the property. +| property::buffer::mem_channel | The `mem_channel` property is a hint to the SYCL runtime that the buffer should be stored in a particular memory channel provided to the property. |=== -- -Add a new constructor to Table 41: Constructors of the buffer property classes as follows: +Add a new constructor to Table 4.34: Constructors of the buffer property classes as follows: -- [options="header"] |=== | Constructor | Description -| property::buffer::ext_intel_mem_channel::ext_intel_mem_channel(cl_uint channel) | Constructs a SYCL `ext_intel_mem_channel` property instance with the specified channel ID. The range of valid values depends on the target and is implementation defined. Invalid values do not need to result in an error as the property is only a hint. +| property::buffer::mem_channel::mem_channel(cl_uint channel) | Constructs a SYCL `mem_channel` property instance with the specified channel ID. The range of valid values depends on the target and is implementation defined. Invalid values do not need to result in an error as the property is only a hint. |=== -- -Add a new member function to Table 42: Member functions of the buffer property classes as follows: +Add a new member function to Table 4.35: Member functions of the buffer property classes as follows: -- [options="header"] |=== | Member function | Description -| cl_uint property::buffer::ext_intel_mem_channel::get_channel() const | Returns the cl_uint which was specified when constructing this SYCL `ext_intel_mem_channel` property. +| cl_uint property::buffer::mem_channel::get_channel() const | Returns the cl_uint which was specified when constructing this SYCL `mem_channel` property. |=== -- @@ -103,7 +87,7 @@ enum class aspect { } // namespace sycl ``` -Add an entry for the new aspect to Table 26: Device aspects defined by the core SYCL specification: +Add an entry for the new aspect to Table 4.20: Device aspects defined by the core SYCL specification: -- [options="header"] @@ -123,6 +107,4 @@ Add an entry for the new aspect to Table 26: Device aspects defined by the core |======================================== |Rev|Date|Author|Changes |1|2020-10-26|Joe Garvey|*Initial public draft* -|2|2021-08-30|Dmitry Vodopyanov|*Updated according to SYCL 2020 reqs for extensions* - |======================================== diff --git a/sycl/doc/extensions/SYCL_ONEAPI_dot_accumulate.asciidoc b/sycl/doc/extensions/SYCL_ONEAPI_dot_accumulate.asciidoc index 64471166407bf..de7622f6121ca 100755 --- a/sycl/doc/extensions/SYCL_ONEAPI_dot_accumulate.asciidoc +++ b/sycl/doc/extensions/SYCL_ONEAPI_dot_accumulate.asciidoc @@ -46,7 +46,7 @@ Because the interfaces defined by this specification are not final and are subje == Version Built On: {docdate} + -Revision: B +Revision: 3 == Contact @@ -135,9 +135,9 @@ None. [options="header"] |======================================== |Rev|Date|Author|Changes -|A|2019-12-13|Ben Ashbaugh|*Initial draft* -|B|2019-12-18|Ben Ashbaugh|Switched to standard C++ fixed width types. -|C|2020-10-26|Rajiv Deodhar|Added int32 types. +|1|2019-12-13|Ben Ashbaugh|*Initial draft* +|2|2019-12-18|Ben Ashbaugh|Switched to standard C++ fixed width types. +|3|2020-10-26|Rajiv Deodhar|Added int32 types. |======================================== //************************************************************************ diff --git a/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc b/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc index 5eda13b1b93dd..138f2874919a4 100644 --- a/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc +++ b/sycl/doc/extensions/USMAddressSpaces/usm_address_spaces.asciidoc @@ -25,7 +25,7 @@ Because the interfaces defined by this specification are not final and are subje == Version Built On: {docdate} + -Revision: B +Revision: 2 == Dependencies @@ -134,6 +134,6 @@ using host_ptr = multi_ptr &WaitList); + void ext_oneapi_barrier(const std::vector &WaitList); /// Prevents any commands submitted afterward to this queue from executing /// until all events in WaitList have entered the complete state. If WaitList @@ -2306,10 +2306,8 @@ class __SYCL_EXPORT handler { /// /// \param WaitList is a vector of valid SYCL events that need to complete /// before barrier command can be executed. - __SYCL2020_DEPRECATED("use 'ext_intel_barrier' instead") - void barrier(const std::vector &WaitList) { - ext_intel_barrier(WaitList); - } + __SYCL2020_DEPRECATED("use 'ext_oneapi_barrier' instead") + void barrier(const std::vector &WaitList); /// Copies data from one memory region to another, both pointed by /// USM pointers. diff --git a/sycl/include/CL/sycl/properties/buffer_properties.hpp b/sycl/include/CL/sycl/properties/buffer_properties.hpp index 19f2a3355decc..aed2b3e8f21c6 100644 --- a/sycl/include/CL/sycl/properties/buffer_properties.hpp +++ b/sycl/include/CL/sycl/properties/buffer_properties.hpp @@ -40,22 +40,16 @@ class context_bound sycl::context MCtx; }; -class ext_intel_mem_channel : public detail::PropertyWithData< - detail::PropWithDataKind::BufferMemChannel> { +class mem_channel : public detail::PropertyWithData< + detail::PropWithDataKind::BufferMemChannel> { public: - ext_intel_mem_channel(uint32_t Channel) : MChannel(Channel) {} + mem_channel(uint32_t Channel) : MChannel(Channel) {} uint32_t get_channel() const { return MChannel; } private: uint32_t MChannel; }; -class __SYCL2020_DEPRECATED("use 'ext_intel_mem_channel' instead") mem_channel - : public ext_intel_mem_channel { -public: - mem_channel(uint32_t Channel) : ext_intel_mem_channel(Channel) {} -}; - } // namespace buffer } // namespace property diff --git a/sycl/include/CL/sycl/queue.hpp b/sycl/include/CL/sycl/queue.hpp index ea405a56d6344..b33d983d854fc 100644 --- a/sycl/include/CL/sycl/queue.hpp +++ b/sycl/include/CL/sycl/queue.hpp @@ -253,7 +253,7 @@ class __SYCL_EXPORT queue { /// group is being enqueued on. event ext_oneapi_submit_barrier(_CODELOCONLYPARAM(&CodeLoc)) { return submit( - [=](handler &CGH) { CGH.ext_intel_barrier(); } _CODELOCFW(CodeLoc)); + [=](handler &CGH) { CGH.ext_oneapi_barrier(); } _CODELOCFW(CodeLoc)); } /// Prevents any commands submitted afterward to this queue from executing @@ -281,7 +281,7 @@ class __SYCL_EXPORT queue { event ext_oneapi_submit_barrier( const std::vector &WaitList _CODELOCPARAM(&CodeLoc)) { return submit([=](handler &CGH) { - CGH.ext_intel_barrier(WaitList); + CGH.ext_oneapi_barrier(WaitList); } _CODELOCFW(CodeLoc)); } diff --git a/sycl/source/handler.cpp b/sycl/source/handler.cpp index aefeece16d503..efc0c14baeaf7 100644 --- a/sycl/source/handler.cpp +++ b/sycl/source/handler.cpp @@ -488,7 +488,7 @@ std::string handler::getKernelName() { return MKernel->get_info(); } -void handler::ext_intel_barrier(const std::vector &WaitList) { +void handler::ext_oneapi_barrier(const std::vector &WaitList) { throwIfActionIsCreated(); MCGType = detail::CG::BarrierWaitlist; MEventsWaitWithBarrier.resize(WaitList.size()); @@ -497,6 +497,11 @@ void handler::ext_intel_barrier(const std::vector &WaitList) { [](const event &Event) { return detail::getSyclObjImpl(Event); }); } +__SYCL2020_DEPRECATED("use 'ext_oneapi_barrier' instead") +void handler::barrier(const std::vector &WaitList) { + handler::ext_oneapi_barrier(WaitList); +} + using namespace sycl::detail; bool handler::DisableRangeRounding() { return SYCLConfig::get(); diff --git a/sycl/test/abi/sycl_symbols_linux.dump b/sycl/test/abi/sycl_symbols_linux.dump index 5b1edf8048e68..2ee4b28a8249d 100644 --- a/sycl/test/abi/sycl_symbols_linux.dump +++ b/sycl/test/abi/sycl_symbols_linux.dump @@ -3911,7 +3911,7 @@ _ZN2cl4sycl7handler10mem_adviseEPKvmi _ZN2cl4sycl7handler10processArgEPvRKNS0_6detail19kernel_param_kind_tEimRmb _ZN2cl4sycl7handler10processArgEPvRKNS0_6detail19kernel_param_kind_tEimRmbb _ZN2cl4sycl7handler13getKernelNameB5cxx11Ev -_ZN2cl4sycl7handler17ext_intel_barrierERKSt6vectorINS0_5eventESaIS3_EE +_ZN2cl4sycl7handler17ext_oneapi_barrierERKSt6vectorINS0_5eventESaIS3_EE _ZN2cl4sycl7handler18RangeRoundingTraceEv _ZN2cl4sycl7handler18extractArgsAndReqsEv _ZN2cl4sycl7handler20DisableRangeRoundingEv diff --git a/sycl/test/abi/sycl_symbols_windows.dump b/sycl/test/abi/sycl_symbols_windows.dump index 72d8a01f6b2a4..1842eca884e9e 100644 --- a/sycl/test/abi/sycl_symbols_windows.dump +++ b/sycl/test/abi/sycl_symbols_windows.dump @@ -1740,8 +1740,8 @@ ?expm1@__host_std@cl@@YA?AVhalf@half_impl@detail@sycl@2@V34562@@Z ?expm1@__host_std@cl@@YAMM@Z ?expm1@__host_std@cl@@YANN@Z -?ext_intel_barrier@handler@sycl@cl@@QEAAXAEBV?$vector@Vevent@sycl@cl@@V?$allocator@Vevent@sycl@cl@@@std@@@std@@@Z -?ext_intel_barrier@handler@sycl@cl@@QEAAXXZ +?ext_oneapi_barrier@handler@sycl@cl@@QEAAXAEBV?$vector@Vevent@sycl@cl@@V?$allocator@Vevent@sycl@cl@@@std@@@std@@@Z +?ext_oneapi_barrier@handler@sycl@cl@@QEAAXXZ ?ext_oneapi_submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBUcode_location@detail@23@@Z ?ext_oneapi_submit_barrier@queue@sycl@cl@@QEAA?AVevent@23@AEBV?$vector@Vevent@sycl@cl@@V?$allocator@Vevent@sycl@cl@@@std@@@std@@AEBUcode_location@detail@23@@Z ?extractArgsAndReqs@handler@sycl@cl@@AEAAXXZ diff --git a/sycl/test/warnings/sycl_2020_deprecations.cpp b/sycl/test/warnings/sycl_2020_deprecations.cpp index 6aab31242b819..5a20b34f3e5e7 100644 --- a/sycl/test/warnings/sycl_2020_deprecations.cpp +++ b/sycl/test/warnings/sycl_2020_deprecations.cpp @@ -165,15 +165,11 @@ int main() { auto BitCastRes = sycl::detail::bit_cast(Val); (void)BitCastRes; - // expected-warning@+1{{'submit_barrier' is deprecated: use 'ext_intel_submit_barrier' instead}} + // expected-warning@+1{{'submit_barrier' is deprecated: use 'ext_oneapi_submit_barrier' instead}} Queue.submit_barrier(); - // expected-warning@+1{{'barrier' is deprecated: use 'ext_intel_barrier' instead}} + // expected-warning@+1{{'barrier' is deprecated: use 'ext_oneapi_barrier' instead}} Queue.submit([&](sycl::handler &CGH) { CGH.barrier(); }); - // expected-warning@+1{{'mem_channel' is deprecated: use 'ext_intel_mem_channel' instead}} - sycl::property_list MemChannelProp{sycl::property::buffer::mem_channel(2)}; - (void)MemChannelProp; - return 0; } From c063dd367c74529b225a6d67ccd5184a35d95403 Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Mon, 6 Sep 2021 16:39:40 +0300 Subject: [PATCH 08/10] Fix name of the new exported symbol --- sycl/test/abi/sycl_symbols_linux.dump | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/test/abi/sycl_symbols_linux.dump b/sycl/test/abi/sycl_symbols_linux.dump index 2ee4b28a8249d..1dd86ae0a0ff0 100644 --- a/sycl/test/abi/sycl_symbols_linux.dump +++ b/sycl/test/abi/sycl_symbols_linux.dump @@ -3911,7 +3911,7 @@ _ZN2cl4sycl7handler10mem_adviseEPKvmi _ZN2cl4sycl7handler10processArgEPvRKNS0_6detail19kernel_param_kind_tEimRmb _ZN2cl4sycl7handler10processArgEPvRKNS0_6detail19kernel_param_kind_tEimRmbb _ZN2cl4sycl7handler13getKernelNameB5cxx11Ev -_ZN2cl4sycl7handler17ext_oneapi_barrierERKSt6vectorINS0_5eventESaIS3_EE +_ZN2cl4sycl7handler18ext_oneapi_barrierERKSt6vectorINS0_5eventESaIS3_EE _ZN2cl4sycl7handler18RangeRoundingTraceEv _ZN2cl4sycl7handler18extractArgsAndReqsEv _ZN2cl4sycl7handler20DisableRangeRoundingEv From 1c801e87c3a5ee25cecaa3a3029ca0c7de3f6fcc Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Tue, 7 Sep 2021 16:49:39 +0300 Subject: [PATCH 09/10] Fix a test after resolved merge conflict --- .../sycl/ext/oneapi/backend/level_zero.hpp | 6 +-- .../basic_tests/interop-level-zero-2020.cpp | 52 +++++++++++-------- 2 files changed, 32 insertions(+), 26 deletions(-) diff --git a/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp b/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp index f67ea7e2f8add..cb41cc34447f3 100644 --- a/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp +++ b/sycl/include/sycl/ext/oneapi/backend/level_zero.hpp @@ -78,7 +78,7 @@ template <> struct BackendInput { using type = struct { interop::type NativeHandle; std::vector DeviceList; - level_zero::ownership Ownership; + ext::oneapi::level_zero::ownership Ownership; }; }; @@ -191,10 +191,10 @@ template <> context make_context( const backend_input_t &BackendObject, const async_handler &Handler) { - return level_zero::make_context( + return ext::oneapi::level_zero::make_context( BackendObject.DeviceList, detail::pi::cast(BackendObject.NativeHandle), - BackendObject.Ownership == level_zero::ownership::keep); + BackendObject.Ownership == ext::oneapi::level_zero::ownership::keep); } namespace __SYCL2020_DEPRECATED("use 'ext::oneapi::level_zero' instead") diff --git a/sycl/test/basic_tests/interop-level-zero-2020.cpp b/sycl/test/basic_tests/interop-level-zero-2020.cpp index eace5afa7d8cf..4e40829cffca3 100644 --- a/sycl/test/basic_tests/interop-level-zero-2020.cpp +++ b/sycl/test/basic_tests/interop-level-zero-2020.cpp @@ -41,9 +41,11 @@ int main() { // return_type is used when retrieving the backend specific native object from // a SYCL object. See the relevant backend specification for details. - backend_traits::return_type ZeDriver; - backend_traits::return_type ZeDevice; - backend_traits::return_type ZeContext; + backend_traits::return_type + ZeDriver; + backend_traits::return_type ZeDevice; + backend_traits::return_type + ZeContext; // 4.5.1.2 For each SYCL runtime class T which supports SYCL application // interoperability, a specialization of get_native must be defined, which @@ -52,20 +54,20 @@ int main() { // application interoperability. The lifetime of the object returned are // backend-defined and specified in the backend specification. - ZeDriver = get_native(Platform); - ZeDevice = get_native(Device); - ZeContext = get_native(Context); + ZeDriver = get_native(Platform); + ZeDevice = get_native(Device); + ZeContext = get_native(Context); // Check deprecated // expected-warning@+2 {{'get_native' is deprecated: Use SYCL-2020 sycl::get_native free function}} - // expected-warning@+1 {{'get_native' is deprecated: Use SYCL-2020 sycl::get_native free function}} - ZeDriver = Platform.get_native(); + // expected-warning@+1 {{'get_native' is deprecated: Use SYCL-2020 sycl::get_native free function}} + ZeDriver = Platform.get_native(); // expected-warning@+2 {{'get_native' is deprecated: Use SYCL-2020 sycl::get_native free function}} - // expected-warning@+1 {{'get_native' is deprecated: Use SYCL-2020 sycl::get_native free function}} - ZeDevice = Device.get_native(); + // expected-warning@+1 {{'get_native' is deprecated: Use SYCL-2020 sycl::get_native free function}} + ZeDevice = Device.get_native(); // expected-warning@+2 {{'get_native' is deprecated: Use SYCL-2020 sycl::get_native free function}} - // expected-warning@+1 {{'get_native' is deprecated: Use SYCL-2020 sycl::get_native free function}} - ZeContext = Context.get_native(); + // expected-warning@+1 {{'get_native' is deprecated: Use SYCL-2020 sycl::get_native free function}} + ZeContext = Context.get_native(); // 4.5.1.1 For each SYCL runtime class T which supports SYCL application // interoperability with the SYCL backend, a specialization of input_type must @@ -83,27 +85,31 @@ int main() { // behavior of these template functions is defined by the SYCL backend // specification document. - backend_input_t InteropPlatformInput{ZeDriver}; + backend_input_t + InteropPlatformInput{ZeDriver}; platform InteropPlatform = - make_platform(InteropPlatformInput); + make_platform(InteropPlatformInput); - backend_input_t InteropDeviceInput{ZeDevice}; - device InteropDevice = make_device(InteropDeviceInput); + backend_input_t InteropDeviceInput{ + ZeDevice}; + device InteropDevice = + make_device(InteropDeviceInput); - backend_input_t InteropContextInput{ + backend_input_t InteropContextInput{ ZeContext, std::vector(1, InteropDevice), - level_zero::ownership::keep}; + ext::oneapi::level_zero::ownership::keep}; context InteropContext = - make_context(InteropContextInput); + make_context(InteropContextInput); // Check deprecated // expected-warning@+1 {{'make' is deprecated: Use SYCL-2020 sycl::make_platform free function}} - auto P = level_zero::make(ZeDriver); + auto P = ext::oneapi::level_zero::make(ZeDriver); // expected-warning@+1 {{'make' is deprecated: Use SYCL-2020 sycl::make_device free function}} - auto D = level_zero::make(P, ZeDevice); + auto D = ext::oneapi::level_zero::make(P, ZeDevice); // expected-warning@+1 {{'make' is deprecated: Use SYCL-2020 sycl::make_context free function}} - auto C = level_zero::make(std::vector(1, D), ZeContext, - level_zero::ownership::keep); + auto C = ext::oneapi::level_zero::make( + std::vector(1, D), ZeContext, + ext::oneapi::level_zero::ownership::keep); return 0; } From 198aacaf7a134a0979f2147d2b5711747593e618 Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Tue, 7 Sep 2021 21:10:22 +0300 Subject: [PATCH 10/10] Return back some mods for MemChannel spec --- .../extensions/MemChannel/MemChannel.asciidoc | 27 +++++++++++++++---- sycl/include/CL/sycl/feature_test.hpp | 1 + 2 files changed, 23 insertions(+), 5 deletions(-) diff --git a/sycl/doc/extensions/MemChannel/MemChannel.asciidoc b/sycl/doc/extensions/MemChannel/MemChannel.asciidoc index 15b309851043f..50a0258a32973 100644 --- a/sycl/doc/extensions/MemChannel/MemChannel.asciidoc +++ b/sycl/doc/extensions/MemChannel/MemChannel.asciidoc @@ -1,4 +1,4 @@ -= SYCL_INTEL_mem_channel_property += SYCL_EXT_INTEL_MEM_CHANNEL_PROPERTY == Introduction NOTE: Khronos(R) is a registered trademark and SYCL(TM) and SPIR(TM) are trademarks of The Khronos Group Inc. OpenCL(TM) is a trademark of Apple Inc. used by permission by Khronos. @@ -23,14 +23,30 @@ Because the interfaces defined by this specification are not final and are subje == Version Built On: {docdate} + -Revision: 1 +Revision: 2 == Dependencies -This extension is written against the SYCL 2020 provisional specification, Revision 1. +This extension is written against the SYCL 2020 specification, Revision 3. The use of this extension requires a target that supports cl_intel_mem_channel_property or equivalent if OpenCL is used as the underlying device runtime. +== Feature Test Macro + +This extension provides a feature-test macro as described in the core SYCL +specification section 6.3.3 "Feature test macros". Therefore, an +implementation supporting this extension must predefine the macro +`SYCL_EXT_INTEL_MEM_CHANNEL_PROPERTY` to one of the values defined in the table below. +Applications can test for the existence of this macro to determine if the +implementation supports this feature, or applications can test the macro's +value to determine which of the extension's APIs the implementation supports. + +[%header,cols="1,5"] +|=== +|Value |Description +|1 |Initial extension version. Base features are supported. +|=== + == Overview On some targets manual assignment of buffers to memory regions can improve memory bandwidth. This extension adds a buffer property to indicate in which memory channel a particular buffer should be allocated. This information is an optimization hint to the runtime and thus it is legal to ignore. @@ -59,7 +75,7 @@ Add a new constructor to Table 4.34: Constructors of the buffer property classes |=== -- -Add a new member function to Table 4.35: Member functions of the buffer property classes as follows: +Add a new member function to Table 42: Member functions of the buffer property classes as follows: -- [options="header"] @@ -87,7 +103,7 @@ enum class aspect { } // namespace sycl ``` -Add an entry for the new aspect to Table 4.20: Device aspects defined by the core SYCL specification: +Add an entry for the new aspect to Table 26: Device aspects defined by the core SYCL specification: -- [options="header"] @@ -107,4 +123,5 @@ Add an entry for the new aspect to Table 4.20: Device aspects defined by the cor |======================================== |Rev|Date|Author|Changes |1|2020-10-26|Joe Garvey|*Initial public draft* +|2|2021-08-30|Dmitry Vodopyanov|*Updated according to some SYCL 2020 reqs for extensions* |======================================== diff --git a/sycl/include/CL/sycl/feature_test.hpp b/sycl/include/CL/sycl/feature_test.hpp index c2e81561f23df..08f51244adc90 100644 --- a/sycl/include/CL/sycl/feature_test.hpp +++ b/sycl/include/CL/sycl/feature_test.hpp @@ -25,6 +25,7 @@ namespace sycl { #endif #define SYCL_EXT_INTEL_BF16_CONVERSION 1 #define SYCL_EXT_ONEAPI_ENQUEUE_BARRIER 1 +#define SYCL_EXT_INTEL_MEM_CHANNEL_PROPERTY 1 #define SYCL_EXT_INTEL_USM_ADDRESS_SPACES 1 #define SYCL_EXT_ONEAPI_BACKEND_LEVEL_ZERO 1