[SYCLNATIVECPU] Re-enable host device

sycl/doc/FAQ.md

@@ -16,7 +16,13 @@ compiler provides you with both host and device side compilation. Another
 requirement for code offloading to specialized devices is a compatible OpenCL
 runtime. Our [Get Started Guide](GetStartedGuide.md) will help you
 set up a proper environment. To learn more about using the DPC++ compiler,
-please refer to [Users Manual](UsersManual.md).
+please refer to [Users Manual](UsersManual.md). If using a special compiler
+is not an option for you and/or you would like to experiment without offloading
+code to non-host devices, you can exploit SYCL's host device feature. This
+gives you the ability to use any C++17 compiler. You will need to link your
+application with the DPC++ Runtime library and provide a path to the SYCL
+headers directory. Please, refer to your compiler manual to learn about
+specific build options.
 
 ### Q: How are DPC++ compilation phases different from those of a usual C++ compiler? Can I customize this flow for my applications?
 **A:** Due to the fact that both host and device code need to be compiled and
@@ -57,11 +63,10 @@ design/CompilerAndRuntimeDesign.md) document.
 
 ## Using applications built with DPC++
 
-### Q: What happens if I run my SYCL application on a machine without a supported backend?
-**A:** If a SYCL application is unable to find a suitable device, the SYCL
-objects (`sycl::context`, `sycl::queue`, etc.) or the SYCL device selector (See
-[SYCL 2020 specificaion 4.6.1.1](https://registry.khronos.org/SYCL/specs/sycl-2020/html/sycl-2020.html#sec:device-selection))
-will throw a `sycl::exception` with `errc::runtime` error code.
+### Q: What happens if I run my application on a machine without OpenCL?
+**A:** If you use the default SYCL device selector (or any other selector that
+allows host device), then a fallback to the host device will take place.
+Otherwise, an exception will be thrown.
 
 
 ## Common issues
@@ -137,6 +142,12 @@ SPIR-V, a portable intermediate representation format. It is a core feature of
 OpenCL 2.1, so any device, capable of OpenCL 2.1, should be supported.
 Otherwise, your OpenCL device must support `cl_khr_il_program` extension.
 
+Apart from that, there's also the so-called host device, which can be used when
+no suitable OpenCL device is found. The host device will use your host CPU as
+the offload target for kernel execution. Since the device code is also compiled
+for the host CPU and no JIT is required, you can easily use any classic C++
+debugging tools of your choice for the host device code.
+
 Furthermore, developers can extend capabilities of the DPC++ Runtime to
 non-OpenCL devices by writing correspondent plugins. To learn more, please
 check out our [Plugin Interface Guide](design/PluginInterface.md).

sycl/doc/GetStartedGuide.md

@@ -699,7 +699,7 @@ more. To find available options, execute:
 
 The `simple-sycl-app.exe` application doesn't specify SYCL device for
 execution, so SYCL runtime will use `default_selector` logic to select one
-of accelerators available in the system.
+of accelerators available in the system or SYCL host device.
 In this case, the behavior of the `default_selector` can be altered
 using the `ONEAPI_DEVICE_SELECTOR` environment variable, setting `cuda:*` forces
 the usage of the CUDA backend (if available), `hip:*` forces
@@ -711,6 +711,9 @@ ONEAPI_DEVICE_SELECTOR=cuda:* ./simple-sycl-app-cuda.exe
 ```
 
 The default is the OpenCL backend if available.
+If there are no OpenCL or CUDA devices available, the SYCL host device is used.
+The SYCL host device executes the SYCL application directly in the host,
+without using any low-level API.
 
 **NOTE**: `nvptx64-nvidia-cuda` is usable with `-fsycl-targets`
 if clang was built with the cmake option `SYCL_ENABLE_PLUGINS=cuda`.
@@ -850,6 +853,7 @@ which contains all the symbols required.
 
 * DPC++ device compiler fails if the same kernel was used in different
   translation units.
+* SYCL host device is not fully supported.
 * SYCL 2020 support work is in progress.
 * 32-bit host/target is not supported.
 * DPC++ works only with OpenCL low level runtimes which support out-of-order

sycl/include/sycl/backend_types.hpp

@@ -19,7 +19,7 @@ namespace sycl {
 __SYCL_INLINE_VER_NAMESPACE(_V1) {
 
 enum class backend : char {
-  host __SYCL2020_DEPRECATED("'host' backend is no longer supported") = 0,
+  host = 0,
   opencl = 1,
   ext_oneapi_level_zero = 2,
   level_zero __SYCL2020_DEPRECATED("use 'ext_oneapi_level_zero' instead") =

sycl/include/sycl/context.hpp

@@ -202,8 +202,6 @@ class __SYCL_EXPORT context : public detail::OwnerLessBase<context> {
   /// Checks if this context is a SYCL host context.
   ///
   /// \return true if this context is a SYCL host context.
-  __SYCL2020_DEPRECATED(
-      "is_host() is deprecated as the host device is no longer supported.")
   bool is_host() const;
 
   /// Returns the backend associated with this context.

sycl/include/sycl/detail/device_filter.hpp

@@ -93,6 +93,7 @@ class device_filter_list {
   bool backendCompatible(backend Backend);
   bool deviceTypeCompatible(info::device_type DeviceType);
   bool deviceNumberCompatible(int DeviceNum);
+  bool containsHost();
   friend std::ostream &operator<<(std::ostream &Out,
                                   const device_filter_list &List);
 };

sycl/include/sycl/device.hpp

@@ -100,8 +100,6 @@ class __SYCL_EXPORT device : public detail::OwnerLessBase<device> {
   /// Check if device is a host device
   ///
   /// \return true if SYCL device is a host device
-  __SYCL2020_DEPRECATED(
-      "is_host() is deprecated as the host device is no longer supported.")
   bool is_host() const;
 
   /// Check if device is a CPU device

sycl/include/sycl/device_selector.hpp

@@ -96,7 +96,7 @@ __SYCL2020_DEPRECATED("Use the callable sycl::accelerator_selector_v instead.")
 ///
 /// \ingroup sycl_api_dev_sel
 class __SYCL_EXPORT
-__SYCL2020_DEPRECATED("Host device is no longer supported.") host_selector
+__SYCL2020_DEPRECATED("Use a callable function instead.") host_selector
     : public device_selector {
 public:
   int operator()(const device &dev) const override;

sycl/include/sycl/event.hpp

@@ -69,8 +69,6 @@ class __SYCL_EXPORT event : public detail::OwnerLessBase<event> {
   /// Checks if this event is a SYCL host event.
   ///
   /// \return true if this event is a SYCL host event.
-  __SYCL2020_DEPRECATED(
-      "is_host() is deprecated as the host device is no longer supported.")
   bool is_host() const;
 
   /// Return the list of events that this event waits for.

sycl/include/sycl/handler.hpp

@@ -421,7 +421,6 @@ class __SYCL_EXPORT handler {
 
   ~handler() = default;
 
-  // TODO: Private and unusued. Remove when ABI break is allowed.
   bool is_host() { return MIsHost; }
 
 #ifdef __SYCL_DEVICE_ONLY__
@@ -510,10 +509,16 @@ class __SYCL_EXPORT handler {
                        sizeof(sampler), ArgIndex);
   }
 
-  // TODO: Unusued. Remove when ABI break is allowed.
   void verifyKernelInvoc(const kernel &Kernel) {
-    std::ignore = Kernel;
-    return;
+    if (is_host()) {
+      throw invalid_object_error(
+          "This kernel invocation method cannot be used on the host",
+          PI_ERROR_INVALID_DEVICE);
+    }
+    if (Kernel.is_host()) {
+      throw invalid_object_error("Invalid kernel type, OpenCL expected",
+                                 PI_ERROR_INVALID_KERNEL);
+    }
   }
 
   /* The kernel passed to StoreLambda can take an id, an item or an nd_item as
@@ -1103,6 +1108,7 @@ class __SYCL_EXPORT handler {
   template <int Dims>
   void parallel_for_impl(range<Dims> NumWorkItems, kernel Kernel) {
     throwIfActionIsCreated();
+    verifyKernelInvoc(Kernel);
     MKernel = detail::getSyclObjImpl(std::move(Kernel));
     detail::checkValueRange<Dims>(NumWorkItems);
     MNDRDesc.set(std::move(NumWorkItems));
@@ -1747,6 +1753,7 @@ class __SYCL_EXPORT handler {
   /// \param Kernel is a SYCL kernel object.
   void single_task(kernel Kernel) {
     throwIfActionIsCreated();
+    verifyKernelInvoc(Kernel);
     // Ignore any set kernel bundles and use the one associated with the kernel
     setHandlerKernelBundle(Kernel);
     // No need to check if range is out of INT_MAX limits as it's compile-time
@@ -1783,6 +1790,7 @@ class __SYCL_EXPORT handler {
   void parallel_for(range<Dims> NumWorkItems, id<Dims> WorkItemOffset,
                     kernel Kernel) {
     throwIfActionIsCreated();
+    verifyKernelInvoc(Kernel);
     MKernel = detail::getSyclObjImpl(std::move(Kernel));
     detail::checkValueRange<Dims>(NumWorkItems, WorkItemOffset);
     MNDRDesc.set(std::move(NumWorkItems), std::move(WorkItemOffset));
@@ -1801,6 +1809,7 @@ class __SYCL_EXPORT handler {
   /// \param Kernel is a SYCL kernel function.
   template <int Dims> void parallel_for(nd_range<Dims> NDRange, kernel Kernel) {
     throwIfActionIsCreated();
+    verifyKernelInvoc(Kernel);
     MKernel = detail::getSyclObjImpl(std::move(Kernel));
     detail::checkValueRange<Dims>(NDRange);
     MNDRDesc.set(std::move(NDRange));

sycl/include/sycl/kernel.hpp

@@ -107,8 +107,6 @@ class __SYCL_EXPORT kernel : public detail::OwnerLessBase<kernel> {
   /// Check if the associated SYCL context is a SYCL host context.
   ///
   /// \return true if this SYCL kernel is a host kernel.
-  __SYCL2020_DEPRECATED(
-      "is_host() is deprecated as the host device is no longer supported.")
   bool is_host() const;
 
   /// Get the context that this kernel is defined for.

sycl/include/sycl/platform.hpp

@@ -108,8 +108,6 @@ class __SYCL_EXPORT platform : public detail::OwnerLessBase<platform> {
   /// Checks if this SYCL platform is a host platform.
   ///
   /// \return true if this SYCL platform is a host platform.
-  __SYCL2020_DEPRECATED(
-      "is_host() is deprecated as the host device is no longer supported.")
   bool is_host() const;
 
   /// Returns all SYCL devices associated with this platform.

sycl/include/sycl/queue.hpp

@@ -462,8 +462,6 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
   device get_device() const;
 
   /// \return true if this queue is a SYCL host queue.
-  __SYCL2020_DEPRECATED(
-      "is_host() is deprecated as the host device is no longer supported.")
   bool is_host() const;
 
   /// Queries SYCL queue for information.
@@ -488,24 +486,28 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
     _CODELOCARG(&CodeLoc);
     detail::tls_code_loc_t TlsCodeLocCapture(CodeLoc);
 #if __SYCL_USE_FALLBACK_ASSERT
-    auto PostProcess = [this, &CodeLoc](bool IsKernel, bool KernelUsesAssert,
-                                        event &E) {
-      if (IsKernel && !device_has(aspect::ext_oneapi_native_assert) &&
-          KernelUsesAssert && !device_has(aspect::accelerator)) {
-        // __devicelib_assert_fail isn't supported by Device-side Runtime
-        // Linking against fallback impl of __devicelib_assert_fail is
-        // performed by program manager class
-        // Fallback assert isn't supported for FPGA
-        submitAssertCapture(*this, E, /* SecondaryQueue = */ nullptr, CodeLoc);
-      }
-    };
-
-    auto Event = submit_impl_and_postprocess(CGF, CodeLoc, PostProcess);
-    return discard_or_return(Event);
-#else
-    auto Event = submit_impl(CGF, CodeLoc);
-    return discard_or_return(Event);
+    if (!is_host()) {
+      auto PostProcess = [this, &CodeLoc](bool IsKernel, bool KernelUsesAssert,
+                                          event &E) {
+        if (IsKernel && !device_has(aspect::ext_oneapi_native_assert) &&
+            KernelUsesAssert && !device_has(aspect::accelerator)) {
+          // __devicelib_assert_fail isn't supported by Device-side Runtime
+          // Linking against fallback impl of __devicelib_assert_fail is
+          // performed by program manager class
+          // Fallback assert isn't supported for FPGA
+          submitAssertCapture(*this, E, /* SecondaryQueue = */ nullptr,
+                              CodeLoc);
+        }
+      };
+
+      auto Event = submit_impl_and_postprocess(CGF, CodeLoc, PostProcess);
+      return discard_or_return(Event);
+    } else
 #endif // __SYCL_USE_FALLBACK_ASSERT
+    {
+      auto Event = submit_impl(CGF, CodeLoc);
+      return discard_or_return(Event);
+    }
   }
 
   /// Submits a command group function object to the queue, in order to be
@@ -524,27 +526,34 @@ class __SYCL_EXPORT queue : public detail::OwnerLessBase<queue> {
     _CODELOCARG(&CodeLoc);
     detail::tls_code_loc_t TlsCodeLocCapture(CodeLoc);
 #if __SYCL_USE_FALLBACK_ASSERT
-    auto PostProcess = [this, &SecondaryQueue, &CodeLoc](
-                           bool IsKernel, bool KernelUsesAssert, event &E) {
-      if (IsKernel && !device_has(aspect::ext_oneapi_native_assert) &&
-          KernelUsesAssert && !device_has(aspect::accelerator)) {
-        // Only secondary queues on devices need to be added to the assert
-        // capture.
-        // __devicelib_assert_fail isn't supported by Device-side Runtime
-        // Linking against fallback impl of __devicelib_assert_fail is
-        // performed by program manager class
-        // Fallback assert isn't supported for FPGA
-        submitAssertCapture(*this, E, &SecondaryQueue, CodeLoc);
-      }
-    };
-
-    auto Event =
-        submit_impl_and_postprocess(CGF, SecondaryQueue, CodeLoc, PostProcess);
-    return discard_or_return(Event);
-#else
-    auto Event = submit_impl(CGF, SecondaryQueue, CodeLoc);
-    return discard_or_return(Event);
+    if (!is_host()) {
+      auto PostProcess = [this, &SecondaryQueue, &CodeLoc](
+                             bool IsKernel, bool KernelUsesAssert, event &E) {
+        if (IsKernel && !device_has(aspect::ext_oneapi_native_assert) &&
+            KernelUsesAssert && !device_has(aspect::accelerator)) {
+          // Only secondary queues on devices need to be added to the assert
+          // capture.
+          // TODO: Handle case where primary queue is host but the secondary
+          // queue is not.
+          queue *DeviceSecondaryQueue =
+              SecondaryQueue.is_host() ? nullptr : &SecondaryQueue;
+          // __devicelib_assert_fail isn't supported by Device-side Runtime
+          // Linking against fallback impl of __devicelib_assert_fail is
+          // performed by program manager class
+          // Fallback assert isn't supported for FPGA
+          submitAssertCapture(*this, E, DeviceSecondaryQueue, CodeLoc);
+        }
+      };
+
+      auto Event = submit_impl_and_postprocess(CGF, SecondaryQueue, CodeLoc,
+                                               PostProcess);
+      return discard_or_return(Event);
+    } else
 #endif // __SYCL_USE_FALLBACK_ASSERT
+    {
+      auto Event = submit_impl(CGF, SecondaryQueue, CodeLoc);
+      return discard_or_return(Event);
+    }
   }
 
   /// Prevents any commands submitted afterward to this queue from executing

sycl/source/CMakeLists.txt

@@ -172,6 +172,7 @@ set(SYCL_SOURCES
     "detail/error_handling/error_handling.cpp"
     "detail/event_impl.cpp"
     "detail/filter_selector_impl.cpp"
+    "detail/force_device.cpp"
     "detail/fusion/fusion_wrapper.cpp"
     "detail/fusion/fusion_wrapper_impl.cpp"
     "detail/global_handler.cpp"

sycl/source/context.cpp

@@ -124,11 +124,7 @@ context::get_info() const {
 
 cl_context context::get() const { return impl->get(); }
 
-bool context::is_host() const {
-  bool IsHost = impl->is_host();
-  assert(!IsHost && "context::is_host should not be called in implementation.");
-  return IsHost;
-}
+bool context::is_host() const { return impl->is_host(); }
 
 backend context::get_backend() const noexcept { return getImplBackend(impl); }
 

sycl/source/detail/backend_impl.hpp

@@ -7,16 +7,17 @@
 //===----------------------------------------------------------------------===//
 
 #pragma once
-#include <cassert>
 #include <sycl/backend_types.hpp>
 
 namespace sycl {
 __SYCL_INLINE_VER_NAMESPACE(_V1) {
 namespace detail {
 
 template <class T> backend getImplBackend(const T &Impl) {
-  assert(!Impl->is_host() && "Cannot get the backend for host.");
+  if (Impl->is_host())
+    return backend::host;
   return Impl->getPlugin().getBackend();
+
 }
 
 } // namespace detail