oneAPI DPC++ Compiler 2022-09

New features

SYCL Compiler

• Added ability to enforce stateless memory accesses for ESIMD. [1811162]
• Added support for -fsycl-force-target compiler option. [1d95f2e]
• Added support for [[intel::max_reinvocation_delay]] loop attribute. [90fa5bb]
• Added support for -fsycl-huge-device-code compiler option, which allows
  linking object files larger than 2GB. [f963062]
• Added support for compiling .cu files with SYCL compiler. [e76ad72]
• Added support for assert on HIP backend. [ade1870]
• Enabled CXX standard library functions for CUDA backend. [1fe92c5]
• Implemented group collective built-in functions for more integral types. [d4933b6]

SYCL Library

• Implemented SYCL 2020 callable device selectors. [64f0db7]
• Implemented SYCL 2020 standalone device selectors. [bfc7e98]
• Added SYCL 2020 property interfaces for local_accessor, usm_allocator,
  accessor and host_accessor classes. [1136b40] [da7dcf8]
• Added support for fpga_simulator_selector. [9bef890]
• Added support for local_accessor. Deprecated target::local. [e4423ef]
• Added support for querying free device memory on Level Zero backend. [0eeef2b]
• Added support for querying free device memory on CUDA and HIP backends. [436f0d8]
• Implemented bfloat16 conversions from/to float for host. [2a383f1]
• Added support for ext::oneapi::property::queue::discard_events to
  Level Zero PI plugin. [1372120]
• Added lsc_atomic support on ESIMD emulator. [0c051a8]
• Added dpas support on ESIMD emulator. [3d506a3]
• Added C++ API for imf libdevice built-ins. [830916a]
• Implemented make_queue for CUDA backend. [89460e8]
• Implemented has_native_event and make_event for CUDA backend. [74369c8]
• Added support of CUDA XPTI tracing. [0cd0414]
• Introduced predicates for ESIMD lsc_block_store/load. [f44edce]
• Added experimental set_kernel_properties API and use_double_grf property
  for ESIMD. [9a55da5]
• Added "eager initialization" mode to Level Zero PI plugin. It might result
  in an unnecessary work done by the plugin, but ensures the fastest possible
  execution on hot and reportable paths. [c145959]
• Added full support of element wise operations on joint_matrix on CUDA
  backend including bfloat16 support. [0a1d751]
• Implemented group::get_linear_id(int) method [6e83c12]

Documentation

• Added stateful to stateless memory access conversion
  design document. [3e03f30]
• Added sycl_ext_oneapi_complex
  extension proposal. [01589da]
• Updated sycl_ext_intel_fpga_device_selector
  extension to add fpga_simulator_accessor. [9bef890]
• Added sycl_ext_intel_fpga_kernel_interface_properties extension proposal. [4b6bd14]
• Updated sycl_ext_oneapi_complex_algorithms
  extension to include sycl::complex as supported type for algorithms. [07c5b48]
• Clarified sub-group size calculation in sycl_ext_oneapi_invoke_simd extension spec. [9b33ad0]
• Updated sycl_ext_oneapi_accessor_properties
  to mark has_property API as noexcept. [7805aa3]
• Updated sycl_ext_intel_device_info
  to support querying free device memory. [0eeef2b]
• Updated sycl_ext_oneapi_matrix
  with description of new matrix features. [770f540]
• Moved sycl_ext_oneapi_invoke_simd
  extensions specification from proposed to experimental because
  implementation is available. [6bee344]

Improvements

SYCL Library

• Ensured that a correct errc thrown for an unassociated placeholder
  accessor. [4f9935a]
• Removed dependency on OpenCL ICD Loader from the runtime. [90e8b5e]
• Added support for ZEBIN format to persistent caching mechanism. [34dcf83]
• Added identification mechanism for binaries in newer ZEBIN format. [f4dee54]
• Switched to use struct information descriptors in accordance with SYCL 2020.
  Removed some deprecated information queries. [b3cbda5]
• Updated kernel_device_specific::max_sub_group_size query to match SYCL 2020
  spec. Deprecated the old variant. [7842d05]
• Deprecated SYCL 1.2.1 device selectors. [c058380]
• Improved error messages reported for unsupported device partitioning. [1c9ddba]
• Made device and platform default to default_selector_v. [b32dd41]
• Deprecated address_space::constant_space. [351b123]
• Marked sycl::exception::has_context as noexcept. [ad923c9]
• Improved range reductions performance on CPU. [3323da6]
• Made sycl::exception nothrow copy constructible. [289e33d]
• Marked has_property methods as noexcept. [417b5a2]
• Improved sycl::event::get_profiling_info exception message when event is
  default constructed. [2e86cd4]
• Added a diagnostic (in form of static_assert) about kernel lambda size
  mismatch between host and device. [d278c67] [ec179b7] [f417a88]
• Updated pipes class to throw exceptions if used on host. [eab2969]
• Updated ESIMD Emulator PI plugin to report support for cl_khr_fp64
  extension. [398571a]
• Updated Level Zero plugin to prefer copy engine for memory read/write
  operations. [65c3ea2]
• Optimized some memory transfers. [92d35cd]
• Enabled event caching in Level Zero PI plugin. [a41b33c]
• Optimized some reductions with parallel_for accepting sycl::range
  for discrete GPUs. [c22a5d3]
• Improved performance of event synchronization on CUDA backend. [c4f326a]
• Added ability to use descendent devices of context members within that
  context. Not supported with OpenCL backend yet. [a0c8c50] [78a483c]
• Added support for querying atomic64 device capability with HIP backend. [cb190fc]
• Enabled FTZ operations for CUDA/PTX backend via
  -fcuda-flush-denormals-to-zero. [e8e7ae8]
• Improved error message about incorrect kernel argument types with CUDA backend. [2542e6a]
• Limited allowed argument types for rol/ror ESIMD functions to better
  represent HW capabilities. [b05f256]
• Implemented mem_advise reset and managed memory checks for CUDA backend. [fe18839]
• Added concurrent memory check to mem_advise on CUDA backend. [33746d8]
• Enabled multiple HIP streams per SYCL queue. [e0c40a9]
• Implemented lazy mechanism of setting context for default-constructed events. [ed92c4c]
• Improved performance for multi-dimensional accessors with multiple accesses
  in a kernel. [7c58b9a]

SYCL Compiler

• Increased max _Bitint size to 4096 for FPGA target. [db5f72a] [3f06cad]
• Removed deprecation message for [[intel::disable_loop_pipelining]] attribute. [07201f5]
• Allowed __builtin_assume_aligned to be called from device code. [24937ea]
• Improved link step performance when per_kernel device code split is used. [84de9d6]
• Added support for SYCL_EXTERNAL on device_global variables. [8b958f6]
• Updated __builtin_intel_fpga_mem to accept more parameters. [231338d]
• Updated ivdep attribute to allow safelen = 0. [558b3ba]
• Improved linking with sycl.lib on Windows. [404d281]
• Implemented more diagnostics about incorrect device_global usages. [1265721]
• Improved library resolution for libsycl.so. [4ce19d6]
• Improved diagnostics when linking with mismatched objects. [0e0202e]
• Added a warning for floating-point size changes after implicit conversions. [e4f5d55]
• Made invoke_simd convert its argument to appropriate types. [038764f]

Documentation

• Removed explicit cl namespace references. [433ea5c]
• Added a short guideline on using CMake with SYCL compiler. [fa603c3]

Bug fixes

SYCL Library

• Fixed a compilation issue where it wasn't possible to pass an initializer list
  for dependency events vector in queue shortcuts with offset
  parameter. [f4f83d9]
• Fixed sycl::get_pointer_device throwing an exception when it passed a
  descendent device (sub-device) instead of a root device. [26d5d98]
• Fixed memory leak happening when kernel bundles are linked. [980677d]
• Fixed USM free throwing an exception when it passed a context created for
  a descendent device. [c49d494]
• Fixed accessor's CTAD for g++ host compiler. [57aabe7]
• Fixed a compilation issue when using multi-dimensional accessor's subscript
  operator. [22e3fc5]
• Fixed "definition with the same mangled name" error happening when used
  multiple buffer reductions in a kernel. [a0a4d72]
• Fixed a compilation issue with SYCL math built-ins when GCC < 11.1 is used as
  a host compiler. [c786894]
• Fixed a compilation issue with SYCL math built-ins (such as sycl::modf,
  for example) not accepting pointers to half. [e286166]
• Fixed an issues with reductions when MSVC is used as host compiler. [94c4b80]
• Fixed a compilation issue when fully specialized sycl::span is initialized
  from an array. [2b50820]
• Fixed a crash in...

DPC++ daily 2022-10-22

sycl-nightly/20221022

[NFC][SYCL] Minor refactor around KernelPropertiesUnpacker (#7125)

DPC++ daily 2022-10-21

[SYCL] Implement sycl_ext_intel_device_architecture for AOT (#7008)

Spec:
https://github.com/intel/llvm/blob/7f2b17ed5b54271fb115cad02d40bae98b2c0454/sycl/doc/design/DeviceIf.md
Tests: https://github.com/intel/llvm-test-suite/pull/1322

DPC++ daily 2022-10-20

[CI][NFC] Fix more github action warnings (#7090)

Continuation of #7055

There will be more patches, e.g. it's appeared that download-artifact is
not
fixed yet: https://github.com/actions/download-artifact/issues/185

DPC++ daily 2022-10-19

[SYCL] Fix properties_kernel_negative for Windows (#7110)

Windows uses long long for uint62_t while the expected notes in
properties_kernel_negative assumed long was the corresponding type. This
commit makes the notes assume either long or long long.

Signed-off-by: Larsen, Steffen <steffen.larsen@intel.com>

DPC++ daily 2022-10-18

Cherry-pick of "Add SPIR-V 1.4 checks" (#6899)

One of the patches missing in intel/llvm

Original commit:

https://github.com/KhronosGroup/SPIRV-LLVM-Translator/commit/c5b3c8e3283b

Signed-off-by: Sidorov, Dmitry <dmitry.sidorov@intel.com>

DPC++ daily 2022-10-17

[LIBCLC] Fix mangling in atomic functions (#7057)

Hard coding the substitution for `AtomicFAddEXT` meant that for the case
without AS specified, it was pushing the `MemorySemanticFlag` inside
`Scope` namespace creating:
```cpp
__spirv_AtomicFAddEXT(double*, __spv::Scope::Flag, __spv::Scope::MemorySemanticsMask::Flag, double)
```

For the `Increment`, the number of chars for the mangled name `_Z24` was
correct, but the typo in the op name was resulting in the pointer being
lost, resulting in the first argument being just a value:
```cpp
__spirv_AtomicIncrementP(unsigned long AS1, __spv::Scope::Flag, __spv::Scope::MemorySemanticsMask::Flag)
```

DPC++ daily 2022-10-16

[SYCL] Fix ESIMD_EMULATOR being picked as default device (#6870)

This commit attempts to fix the problem that occurs when the default
device selector picks the ESIMD_EMULATOR when other devices are
available.

DPC++ daily 2022-10-15

[SYCL] Fix ESIMD_EMULATOR being picked as default device (#6870)

This commit attempts to fix the problem that occurs when the default
device selector picks the ESIMD_EMULATOR when other devices are
available.

DPC++ daily 2022-10-14

[SYCL][NFC] Fix a C string size bug in buffer location unit test (#7060)

Increase the size returned by redefinedDeviceGetInfo by one to leave
space for the null terminator.