[SYCL][Docs] Add legacy SYCL 1.2.1 image aspect

sycl/doc/PreprocessorMacros.md

@@ -34,6 +34,11 @@ This file describes macros that have effect on SYCL compiler and run-time.
   Disables all deprecation warnings in SYCL runtime headers, including SYCL
   1.2.1 deprecations.
 
+- **SYCL_DISABLE_IMAGE_ASPECT_WARNING**
+
+  Disables warning diagnostic issued when calling `device::has(aspect::image)`
+  and `platform::has(aspect::image)`.
+
 - **SYCL_FALLBACK_ASSERT**
 
   Defining as non-zero enables the fallback assert feature even on devices

sycl/doc/extensions/supported/sycl_ext_intel_legacy_image.asciidoc

@@ -0,0 +1,112 @@
+= sycl_ext_intel_legacy_image
+
+:source-highlighter: coderay
+:coderay-linenums-mode: table
+
+// This section needs to be after the document title.
+:doctype: book
+:toc2:
+:toc: left
+:encoding: utf-8
+:lang: en
+:dpcpp: pass:[DPC++]
+
+// Set the default source code type in this document to C++,
+// for syntax highlighting purposes.  This is needed because
+// docbook uses c++ and html5 uses cpp.
+:language: {basebackend@docbook:c++:cpp}
+
+
+== Notice
+
+[%hardbreaks]
+Copyright (C) 2023-2023 Intel Corporation.  All rights reserved.
+
+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.
+
+
+== Contact
+
+To report problems with this extension, please open a new issue at:
+
+https://github.com/intel/llvm/issues
+
+
+== Dependencies
+
+This extension is written against the SYCL 2020 revision 6 specification.  All
+references below to the "core SYCL specification" or to section numbers in the
+SYCL specification refer to that revision.
+
+
+== Status
+
+This extension is implemented and fully supported by {dpcpp}.
+
+
+== Overview
+
+SYCL 2020 removed the SYCL 1.2.1 `image` class as well as the associated image
+accessors and `sampler` class. However, the device info query for
+`sycl::info::device::image_support` stayed in SYCL 2020 as deprecated. As the
+specification states that this query returns the same value as
+`device::has(aspect::image)`, the user can no longer query support for the SYCL
+1.2.1 images in implementations that support these.
+
+This extension adds the new aspect `sycl::aspect::ext_intel_legacy_image`
+intended for querying if a device supports SYCL 1.2.1 images.
+
+
+== Specification
+
+=== Feature test macro
+
+This extension provides a feature-test macro as described in the core SYCL
+specification.  An implementation supporting this extension must predefine the
+macro `SYCL_EXT_INTEL_LEGACY_IMAGE` 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 features the implementation
+supports.
+
+[%header,cols="1,5"]
+|===
+|Value
+|Description
+
+|1
+|Initial version of this extension.
+|===
+
+
+=== The `ext_intel_legacy_image` aspect
+
+This extension adds the following new value to the SYCL `aspect` enum:
+
+```
+namespace sycl {
+
+enum class aspect {
+  ...
+  ext_intel_legacy_image
+};
+
+} // namespace sycl
+```
+
+The new aspect has the following behaviour when queried via `device::has()`:
+
+[%header,cols="1,5"]
+|===
+|Aspect
+|Description
+
+|`aspect::ext_intel_legacy_image`
+|Indicates that the device supports SYCL 1.2.1 image accessor, as defined in
+ section 4.7.6.11 to 4.7.6.13 of the SYCL 1.2.1 Specification Rev. 7, and
+ samplers, as defined in section 4.7.8 and 4.7.9 of the SYCL 1.2.1 Specification
+ Rev. 7.
+|===
+

sycl/include/sycl/accessor.hpp

@@ -665,13 +665,6 @@ class image_accessor
   static_assert(Dimensions > 0 && Dimensions <= 3,
                 "Dimensions can be 1/2/3 for image accessor.");
 
-  template <typename Param>
-  void checkDeviceFeatureSupported(const device &Device) {
-    if (!Device.get_info<Param>())
-      throw feature_not_supported("Images are not supported by this device.",
-                                  PI_ERROR_INVALID_OPERATION);
-  }
-
 #ifdef __SYCL_DEVICE_ONLY__
 
   sycl::vec<int, Dimensions> getRangeInternal() const {
@@ -773,8 +766,12 @@ class image_accessor
         MImageCount(ImageRef.size()),
         MImgChannelOrder(ImageRef.getChannelOrder()),
         MImgChannelType(ImageRef.getChannelType()) {
-    checkDeviceFeatureSupported<info::device::image_support>(
-        getDeviceFromHandler(CommandGroupHandlerRef));
+
+    device Device = getDeviceFromHandler(CommandGroupHandlerRef);
+    if (!Device.has(aspect::ext_intel_legacy_image))
+      throw feature_not_supported(
+          "SYCL 1.2.1 images are not supported by this device.",
+          PI_ERROR_INVALID_OPERATION);
   }
 #endif
 

sycl/include/sycl/detail/defines_elementary.hpp

@@ -58,6 +58,20 @@
 #endif
 #endif // __SYCL2020_DEPRECATED
 
+#ifndef __SYCL_WARN_IMAGE_ASPECT
+#if !defined(SYCL_DISABLE_IMAGE_ASPECT_WARNING) && __has_attribute(diagnose_if)
+#define __SYCL_WARN_IMAGE_ASPECT(aspect_param)                                   \
+  __attribute__((diagnose_if(                                                    \
+      aspect_param == aspect::image,                                             \
+      "SYCL 2020 images are not supported on any devices. Consider using "       \
+      "‘aspect::ext_intel_legacy_image’ instead. Disable this warning with " \
+      "by defining SYCL_DISABLE_IMAGE_ASPECT_WARNING.",                          \
+      "warning")))
+#else
+#define __SYCL_WARN_IMAGE_ASPECT(aspect)
+#endif
+#endif // __SYCL_WARN_IMAGE_ASPECT
+
 #ifndef __SYCL_HAS_CPP_ATTRIBUTE
 #if defined(__cplusplus) && defined(__has_cpp_attribute)
 #define __SYCL_HAS_CPP_ATTRIBUTE(x) __has_cpp_attribute(x)

sycl/include/sycl/device.hpp

@@ -229,7 +229,7 @@ class __SYCL_EXPORT device : public detail::OwnerLessBase<device> {
   /// Provisional Spec.
   ///
   /// \return true if the SYCL device has the given feature.
-  bool has(aspect Aspect) const;
+  bool has(aspect Aspect) const __SYCL_WARN_IMAGE_ASPECT(Aspect);
 
 private:
   std::shared_ptr<detail::device_impl> impl;

sycl/include/sycl/device_aspect_macros.hpp

@@ -213,6 +213,11 @@
 #define __SYCL_ALL_DEVICES_HAVE_40__ 0
 #endif
 
+#ifndef __SYCL_ALL_DEVICES_HAVE_41__
+// __SYCL_ASPECT(ext_intel_legacy_image, 41)
+#define __SYCL_ALL_DEVICES_HAVE_41__ 0
+#endif
+
 #ifndef __SYCL_ANY_DEVICE_HAS_0__
 // __SYCL_ASPECT(host, 0)
 #define __SYCL_ANY_DEVICE_HAS_0__ 0
@@ -417,3 +422,8 @@
 // __SYCL_ASPECT(emulated, 40)
 #define __SYCL_ANY_DEVICE_HAS_40__ 0
 #endif
+
+#ifndef __SYCL_ANY_DEVICE_HAS_41__
+// __SYCL_ASPECT(ext_intel_legacy_image, 41)
+#define __SYCL_ANY_DEVICE_HAS_41__ 0
+#endif

sycl/include/sycl/device_aspect_traits.hpp

@@ -135,6 +135,9 @@ struct all_devices_have<aspect::ext_intel_memory_bus_width>
 template <>
 struct all_devices_have<aspect::emulated>
     : std::bool_constant<__SYCL_ALL_DEVICES_HAVE_40__> {};
+template <>
+struct all_devices_have<aspect::ext_intel_legacy_image>
+    : std::bool_constant<__SYCL_ALL_DEVICES_HAVE_41__> {};
 
 #ifdef __SYCL_ANY_DEVICE_HAS_ANY_ASPECT__
 // Special case where any_device_has is trivially true.
@@ -264,6 +267,9 @@ struct any_device_has<aspect::ext_intel_memory_bus_width>
 template <>
 struct any_device_has<aspect::emulated>
     : std::bool_constant<__SYCL_ANY_DEVICE_HAS_40__> {};
+template <>
+struct any_device_has<aspect::ext_intel_legacy_image>
+    : std::bool_constant<__SYCL_ANY_DEVICE_HAS_41__> {};
 #endif // __SYCL_ANY_DEVICE_HAS_ANY_ASPECT__
 
 template <aspect Aspect>

sycl/include/sycl/info/aspects.def

@@ -35,3 +35,4 @@ __SYCL_ASPECT(ext_intel_device_id, 37)
 __SYCL_ASPECT(ext_intel_memory_clock_rate, 38)
 __SYCL_ASPECT(ext_intel_memory_bus_width, 39)
 __SYCL_ASPECT(emulated, 40)
+__SYCL_ASPECT(ext_intel_legacy_image, 41)

sycl/include/sycl/info/device_traits_deprecated.def

@@ -1,5 +1,5 @@
 // Marked deprecated in SYCL 2020 spec
-__SYCL_PARAM_TRAITS_DEPRECATED(image_support, "deprecated in SYCL 2020, use device::has(aspect::image) instead")
+__SYCL_PARAM_TRAITS_DEPRECATED(image_support, "deprecated in SYCL 2020, use device::has(aspect::ext_intel_legacy_image) to query for SYCL 1.2.1 image support")
 __SYCL_PARAM_TRAITS_DEPRECATED(max_constant_buffer_size,"deprecated in SYCL 2020")
 __SYCL_PARAM_TRAITS_DEPRECATED(max_constant_args,"deprecated in SYCL 2020")
 __SYCL_PARAM_TRAITS_DEPRECATED(host_unified_memory,"deprecated in SYCL 2020, use device::has() with "

sycl/include/sycl/platform.hpp

@@ -149,7 +149,7 @@ class __SYCL_EXPORT platform : public detail::OwnerLessBase<platform> {
   ///
   /// \return true if all of the SYCL devices on this platform have the
   /// given feature.
-  bool has(aspect Aspect) const;
+  bool has(aspect Aspect) const __SYCL_WARN_IMAGE_ASPECT(Aspect);
 
   /// Return this platform's default context
   ///

sycl/source/detail/device_impl.cpp

@@ -442,6 +442,14 @@ bool device_impl::has(aspect Aspect) const {
             &async_barrier_supported, nullptr) == PI_SUCCESS;
     return call_successful && async_barrier_supported;
   }
+  case aspect::ext_intel_legacy_image: {
+    pi_bool legacy_image_support = PI_FALSE;
+    bool call_successful =
+        getPlugin().call_nocheck<detail::PiApiKind::piDeviceGetInfo>(
+            MDevice, PI_DEVICE_INFO_IMAGE_SUPPORT, sizeof(pi_bool),
+            &legacy_image_support, nullptr) == PI_SUCCESS;
+    return call_successful && legacy_image_support;
+  }
   default:
     throw runtime_error("This device aspect has not been implemented yet.",
                         PI_ERROR_INVALID_DEVICE);

sycl/test-e2e/Basic/image/srgba-read.cpp

@@ -97,7 +97,7 @@ int main() {
   if (D.has(aspect::ext_oneapi_srgb))
     std::cout << "aspect::ext_oneapi_srgb detected" << std::endl;
 
-  if (D.has(aspect::image)) {
+  if (D.has(aspect::ext_intel_legacy_image)) {
     // RGBA -- (normal, non-linearized)
     std::cout << "rgba -------" << std::endl;
     test_rd(image_channel_order::rgba, image_channel_type::unorm_int8);

sycl/test-e2e/DiscardEvents/discard_events_check_images.cpp

@@ -175,7 +175,7 @@ int main(int Argc, const char *Argv[]) {
   sycl::queue Q(props);
 
   auto dev = Q.get_device();
-  if (dev.has(aspect::image)) {
+  if (dev.has(aspect::ext_intel_legacy_image)) {
     if (TestType == "image") {
       std::cerr << "RunTest_ImageTest" << std::endl;
       RunTest_ImageTest(Q);

sycl/test-e2e/Sampler/basic-rw-float.cpp

@@ -84,7 +84,7 @@ int main() {
   queue Q;
   device D = Q.get_device();
 
-  if (D.has(aspect::image)) {
+  if (D.has(aspect::ext_intel_legacy_image)) {
     // the _int8 channels are one byte per channel, or four bytes per pixel (for
     // RGBA) the _int16/fp16 channels are two bytes per channel, or eight bytes
     // per pixel (for RGBA) the _int32/fp32  channels are four bytes per

sycl/test-e2e/Sampler/basic-rw.cpp

@@ -98,7 +98,7 @@ int main() {
   queue Q;
   device D = Q.get_device();
 
-  if (D.has(aspect::image)) {
+  if (D.has(aspect::ext_intel_legacy_image)) {
     // the _int8 channels are one byte per channel, or four bytes per pixel (for
     // RGBA) the _int16/fp16 channels are two bytes per channel, or eight bytes
     // per pixel (for RGBA) the _int32/fp32  channels are four bytes per

sycl/test-e2e/Sampler/normalized-clamp-linear-float.cpp

@@ -146,7 +146,7 @@ int main() {
   queue Q;
   device D = Q.get_device();
 
-  if (D.has(aspect::image)) {
+  if (D.has(aspect::ext_intel_legacy_image)) {
     // the _int8 channels are one byte per channel, or four bytes per pixel (for
     // RGBA) the _int16/fp16 channels are two bytes per channel, or eight bytes
     // per pixel (for RGBA) the _int32/fp32  channels are four bytes per

sycl/test-e2e/Sampler/normalized-clamp-nearest.cpp

@@ -127,7 +127,7 @@ int main() {
   queue Q;
   device D = Q.get_device();
 
-  if (D.has(aspect::image)) {
+  if (D.has(aspect::ext_intel_legacy_image)) {
     // the _int8 channels are one byte per channel, or four bytes per pixel (for
     // RGBA) the _int16/fp16 channels are two bytes per channel, or eight bytes
     // per pixel (for RGBA) the _int32/fp32  channels are four bytes per

sycl/test-e2e/Sampler/normalized-clampedge-linear-float.cpp

@@ -135,7 +135,7 @@ int main() {
   queue Q;
   device D = Q.get_device();
 
-  if (D.has(aspect::image)) {
+  if (D.has(aspect::ext_intel_legacy_image)) {
     // the _int8 channels are one byte per channel, or four bytes per pixel (for
     // RGBA) the _int16/fp16 channels are two bytes per channel, or eight bytes
     // per pixel (for RGBA) the _int32/fp32  channels are four bytes per

sycl/test-e2e/Sampler/normalized-clampedge-nearest.cpp

@@ -129,7 +129,7 @@ int main() {
   queue Q;
   device D = Q.get_device();
 
-  if (D.has(aspect::image)) {
+  if (D.has(aspect::ext_intel_legacy_image)) {
     // the _int8 channels are one byte per channel, or four bytes per pixel (for
     // RGBA) the _int16/fp16 channels are two bytes per channel, or eight bytes
     // per pixel (for RGBA) the _int32/fp32  channels are four bytes per

sycl/test-e2e/Sampler/normalized-mirror-linear-float.cpp

@@ -153,7 +153,7 @@ int main() {
   queue Q;
   device D = Q.get_device();
 
-  if (D.has(aspect::image)) {
+  if (D.has(aspect::ext_intel_legacy_image)) {
     // the _int8 channels are one byte per channel, or four bytes per pixel (for
     // RGBA) the _int16/fp16 channels are two bytes per channel, or eight bytes
     // per pixel (for RGBA) the _int32/fp32  channels are four bytes per

sycl/test-e2e/Sampler/normalized-mirror-nearest.cpp

@@ -151,7 +151,7 @@ int main() {
   queue Q;
   device D = Q.get_device();
 
-  if (D.has(aspect::image)) {
+  if (D.has(aspect::ext_intel_legacy_image)) {
     // the _int8 channels are one byte per channel, or four bytes per pixel (for
     // RGBA) the _int16/fp16 channels are two bytes per channel, or eight bytes
     // per pixel (for RGBA) the _int32/fp32  channels are four bytes per

sycl/test-e2e/Sampler/normalized-none-linear-float.cpp

@@ -135,7 +135,7 @@ int main() {
   queue Q;
   device D = Q.get_device();
 
-  if (D.has(aspect::image)) {
+  if (D.has(aspect::ext_intel_legacy_image)) {
     // the _int8 channels are one byte per channel, or four bytes per pixel (for
     // RGBA) the _int16/fp16 channels are two bytes per channel, or eight bytes
     // per pixel (for RGBA) the _int32/fp32  channels are four bytes per

sycl/test-e2e/Sampler/normalized-none-nearest.cpp

@@ -146,7 +146,7 @@ int main() {
   queue Q;
   device D = Q.get_device();
 
-  if (D.has(aspect::image)) {
+  if (D.has(aspect::ext_intel_legacy_image)) {
     // the _int8 channels are one byte per channel, or four bytes per pixel (for
     // RGBA) the _int16/fp16 channels are two bytes per channel, or eight bytes
     // per pixel (for RGBA) the _int32/fp32  channels are four bytes per

sycl/test-e2e/Sampler/normalized-repeat-linear-float.cpp

@@ -153,7 +153,7 @@ int main() {
   queue Q;
   device D = Q.get_device();
 
-  if (D.has(aspect::image)) {
+  if (D.has(aspect::ext_intel_legacy_image)) {
     // the _int8 channels are one byte per channel, or four bytes per pixel (for
     // RGBA) the _int16/fp16 channels are two bytes per channel, or eight bytes
     // per pixel (for RGBA) the _int32/fp32  channels are four bytes per

sycl/test-e2e/Sampler/normalized-repeat-nearest.cpp

@@ -149,7 +149,7 @@ int main() {
   queue Q;
   device D = Q.get_device();
 
-  if (D.has(aspect::image)) {
+  if (D.has(aspect::ext_intel_legacy_image)) {
     // the _int8 channels are one byte per channel, or four bytes per pixel (for
     // RGBA) the _int16/fp16 channels are two bytes per channel, or eight bytes
     // per pixel (for RGBA) the _int32/fp32  channels are four bytes per

sycl/test-e2e/Sampler/unnormalized-clamp-linear-float.cpp

@@ -129,7 +129,7 @@ int main() {
   queue Q;
   device D = Q.get_device();
 
-  if (D.has(aspect::image)) {
+  if (D.has(aspect::ext_intel_legacy_image)) {
     // the _int8 channels are one byte per channel, or four bytes per pixel (for
     // RGBA) the _int16/fp16 channels are two bytes per channel, or eight bytes
     // per pixel (for RGBA) the _int32/fp32  channels are four bytes per