[SYCL] Implement discard_events extension (#5026)

Review documentation in #4922

Author: alexanderfle

sycl/include/CL/sycl/detail/memory_manager.hpp

@@ -12,6 +12,7 @@
 #include <CL/sycl/detail/cl.h>
 #include <CL/sycl/detail/export.hpp>
 #include <CL/sycl/detail/sycl_mem_obj_i.hpp>
+#include <CL/sycl/id.hpp>
 #include <CL/sycl/property_list.hpp>
 #include <CL/sycl/range.hpp>
 
@@ -140,18 +141,43 @@ class __SYCL_EXPORT MemoryManager {
                     void *MappedPtr, std::vector<RT::PiEvent> DepEvents,
                     RT::PiEvent &OutEvent);
 
+  static void copy_usm(const void *SrcMem, QueueImplPtr Queue, size_t Len,
+                       void *DstMem, std::vector<RT::PiEvent> DepEvents,
+                       RT::PiEvent *OutEvent);
+
+  __SYCL_DEPRECATED("copy_usm() accepting PiEvent& is deprecated, use "
+                    "copy_usm() accepting PiEvent* instead")
   static void copy_usm(const void *SrcMem, QueueImplPtr Queue, size_t Len,
                        void *DstMem, std::vector<RT::PiEvent> DepEvents,
                        RT::PiEvent &OutEvent);
 
+  static void fill_usm(void *DstMem, QueueImplPtr Queue, size_t Len,
+                       int Pattern, std::vector<RT::PiEvent> DepEvents,
+                       RT::PiEvent *OutEvent);
+
+  __SYCL_DEPRECATED("fill_usm() accepting PiEvent& is deprecated, use "
+                    "fill_usm() accepting PiEvent* instead")
   static void fill_usm(void *DstMem, QueueImplPtr Queue, size_t Len,
                        int Pattern, std::vector<RT::PiEvent> DepEvents,
                        RT::PiEvent &OutEvent);
 
+  static void prefetch_usm(void *Ptr, QueueImplPtr Queue, size_t Len,
+                           std::vector<RT::PiEvent> DepEvents,
+                           RT::PiEvent *OutEvent);
+
+  __SYCL_DEPRECATED("prefetch_usm() accepting PiEvent& is deprecated, use "
+                    "prefetch_usm() accepting PiEvent* instead")
   static void prefetch_usm(void *Ptr, QueueImplPtr Queue, size_t Len,
                            std::vector<RT::PiEvent> DepEvents,
                            RT::PiEvent &OutEvent);
 
+  static void advise_usm(const void *Ptr, QueueImplPtr Queue, size_t Len,
+                         pi_mem_advice Advice,
+                         std::vector<RT::PiEvent> DepEvents,
+                         RT::PiEvent *OutEvent);
+
+  __SYCL_DEPRECATED("advise_usm() accepting PiEvent& is deprecated, use "
+                    "advise_usm() accepting PiEvent* instead")
   static void advise_usm(const void *Ptr, QueueImplPtr Queue, size_t Len,
                          pi_mem_advice Advice,
                          std::vector<RT::PiEvent> DepEvents,

sycl/include/CL/sycl/detail/property_helper.hpp

@@ -33,8 +33,9 @@ enum DataLessPropKind {
   UsePrimaryContext = 6,
   InitializeToIdentity = 7,
   UseDefaultStream = 8,
+  DiscardEvents = 9,
   // Indicates the last known dataless property.
-  LastKnownDataLessPropKind = 8,
+  LastKnownDataLessPropKind = 9,
   // Exceeding 32 may cause ABI breaking change on some of OSes.
   DataLessPropKindSize = 32
 };

sycl/include/CL/sycl/feature_test.hpp.in

@@ -35,6 +35,7 @@ namespace sycl {
 #define SYCL_EXT_ONEAPI_MATRIX 2
 #endif
 #define SYCL_EXT_ONEAPI_ASSERT 1
+#define SYCL_EXT_ONEAPI_DISCARD_QUEUE_EVENTS 1
 #define SYCL_EXT_ONEAPI_ENQUEUE_BARRIER 1
 #define SYCL_EXT_ONEAPI_FREE_FUNCTION_QUERIES 1
 #define SYCL_EXT_ONEAPI_GROUP_ALGORITHMS 1

sycl/include/CL/sycl/handler.hpp

@@ -1364,18 +1364,12 @@ class __SYCL_EXPORT handler {
   /// Registers event dependencies on this command group.
   ///
   /// \param Event is a valid SYCL event to wait on.
-  void depends_on(event Event) {
-    MEvents.push_back(detail::getSyclObjImpl(Event));
-  }
+  void depends_on(event Event);
 
   /// Registers event dependencies on this command group.
   ///
   /// \param Events is a vector of valid SYCL events to wait on.
-  void depends_on(const std::vector<event> &Events) {
-    for (const event &Event : Events) {
-      MEvents.push_back(detail::getSyclObjImpl(Event));
-    }
-  }
+  void depends_on(const std::vector<event> &Events);
 
   template <typename T>
   using remove_cv_ref_t =

sycl/include/CL/sycl/info/info_desc.hpp

@@ -288,7 +288,10 @@ enum class event : cl_event_info {
 enum class event_command_status : cl_int {
   submitted = CL_SUBMITTED,
   running = CL_RUNNING,
-  complete = CL_COMPLETE
+  complete = CL_COMPLETE,
+  // Since all BE values are positive, it is safe to use a negative value If you
+  // add other ext_oneapi values
+  ext_oneapi_unknown = -1
 };
 
 enum class event_profiling : cl_profiling_info {

sycl/include/CL/sycl/properties/queue_properties.hpp

@@ -23,6 +23,14 @@ class enable_profiling
 
 namespace ext {
 namespace oneapi {
+
+namespace property {
+namespace queue {
+class discard_events : public ::cl::sycl::detail::DataLessProperty<
+                           ::cl::sycl::detail::DiscardEvents> {};
+} // namespace queue
+} // namespace property
+
 namespace cuda {
 namespace property {
 namespace queue {
@@ -52,6 +60,9 @@ template <> struct is_property<property::queue::in_order> : std::true_type {};
 template <>
 struct is_property<property::queue::enable_profiling> : std::true_type {};
 template <>
+struct is_property<ext::oneapi::property::queue::discard_events>
+    : std::true_type {};
+template <>
 struct is_property<property::queue::cuda::use_default_stream> : std::true_type {
 };
 template <>
@@ -64,6 +75,9 @@ template <>
 struct is_property_of<property::queue::enable_profiling, queue>
     : std::true_type {};
 template <>
+struct is_property_of<ext::oneapi::property::queue::discard_events, queue>
+    : std::true_type {};
+template <>
 struct is_property_of<property::queue::cuda::use_default_stream, queue>
     : std::true_type {};
 template <>

sycl/include/CL/sycl/queue.hpp

@@ -251,8 +251,6 @@ class __SYCL_EXPORT queue {
   template <typename T> event submit(T CGF _CODELOCPARAM(&CodeLoc)) {
     _CODELOCARG(&CodeLoc);
 
-    event Event;
-
 #if __SYCL_USE_FALLBACK_ASSERT
     if (!is_host()) {
       auto PostProcess = [this, &CodeLoc](bool IsKernel, bool KernelUsesAssert,
@@ -268,14 +266,14 @@ class __SYCL_EXPORT queue {
         }
       };
 
-      Event = submit_impl_and_postprocess(CGF, CodeLoc, PostProcess);
+      auto Event = submit_impl_and_postprocess(CGF, CodeLoc, PostProcess);
+      return discard_or_return(Event);
     } else
 #endif // __SYCL_USE_FALLBACK_ASSERT
     {
-      Event = submit_impl(CGF, CodeLoc);
+      auto Event = submit_impl(CGF, CodeLoc);
+      return discard_or_return(Event);
     }
-
-    return Event;
   }
 
   /// Submits a command group function object to the queue, in order to be
@@ -293,8 +291,6 @@ class __SYCL_EXPORT queue {
   event submit(T CGF, queue &SecondaryQueue _CODELOCPARAM(&CodeLoc)) {
     _CODELOCARG(&CodeLoc);
 
-    event Event;
-
 #if __SYCL_USE_FALLBACK_ASSERT
     if (!is_host()) {
       auto PostProcess = [this, &SecondaryQueue, &CodeLoc](
@@ -315,15 +311,15 @@ class __SYCL_EXPORT queue {
         }
       };
 
-      Event = submit_impl_and_postprocess(CGF, SecondaryQueue, CodeLoc,
-                                          PostProcess);
+      auto Event = submit_impl_and_postprocess(CGF, SecondaryQueue, CodeLoc,
+                                               PostProcess);
+      return discard_or_return(Event);
     } else
 #endif // __SYCL_USE_FALLBACK_ASSERT
     {
-      Event = submit_impl(CGF, SecondaryQueue, CodeLoc);
+      auto Event = submit_impl(CGF, SecondaryQueue, CodeLoc);
+      return discard_or_return(Event);
     }
-
-    return Event;
   }
 
   /// Prevents any commands submitted afterward to this queue from executing
@@ -1089,6 +1085,10 @@ class __SYCL_EXPORT queue {
   event submit_impl(std::function<void(handler &)> CGH, queue secondQueue,
                     const detail::code_location &CodeLoc);
 
+  /// Checks if the event needs to be discarded and if so, discards it and
+  /// returns a discarded event. Otherwise, it returns input event.
+  event discard_or_return(const event &Event);
+
   // Function to postprocess submitted command
   // Arguments:
   // bool IsKernel - true if the submitted command was kernel, false otherwise

sycl/source/detail/event_impl.cpp

@@ -56,6 +56,11 @@ void event_impl::waitInternal() const {
     return;
   }
 
+  if (MState == HES_Discarded)
+    throw sycl::exception(
+        make_error_code(errc::invalid),
+        "waitInternal method cannot be used for a discarded event.");
+
   while (MState != HES_Complete)
     ;
 }
@@ -93,7 +98,8 @@ void event_impl::setContextImpl(const ContextImplPtr &Context) {
   MState = HES_NotComplete;
 }
 
-event_impl::event_impl() : MIsFlushed(true), MState(HES_Complete) {}
+event_impl::event_impl(HostEventState State)
+    : MIsFlushed(true), MState(State) {}
 
 event_impl::event_impl(RT::PiEvent Event, const context &SyclContext)
     : MEvent(Event), MContext(detail::getSyclObjImpl(SyclContext)),
@@ -188,6 +194,10 @@ void event_impl::instrumentationEpilog(void *TelemetryEvent,
 
 void event_impl::wait(
     std::shared_ptr<cl::sycl::detail::event_impl> Self) const {
+  if (MState == HES_Discarded)
+    throw sycl::exception(make_error_code(errc::invalid),
+                          "wait method cannot be used for a discarded event.");
+
 #ifdef XPTI_ENABLE_INSTRUMENTATION
   void *TelemetryEvent = nullptr;
   uint64_t IId;
@@ -304,6 +314,9 @@ template <> cl_uint event_impl::get_info<info::event::reference_count>() const {
 template <>
 info::event_command_status
 event_impl::get_info<info::event::command_execution_status>() const {
+  if (MState == HES_Discarded)
+    return info::event_command_status::ext_oneapi_unknown;
+
   if (!MHostEvent && MEvent) {
     return get_event_info<info::event::command_execution_status>::get(
         this->getHandleRef(), this->getPlugin());
@@ -333,6 +346,11 @@ pi_native_handle event_impl::getNative() const {
 }
 
 std::vector<EventImplPtr> event_impl::getWaitList() {
+  if (MState == HES_Discarded)
+    throw sycl::exception(
+        make_error_code(errc::invalid),
+        "get_wait_list() cannot be used for a discarded event.");
+
   std::lock_guard<std::mutex> Lock(MMutex);
 
   std::vector<EventImplPtr> Result;

sycl/source/detail/event_impl.hpp

@@ -32,10 +32,16 @@ using EventImplPtr = std::shared_ptr<cl::sycl::detail::event_impl>;
 
 class event_impl {
 public:
+  enum HostEventState : int {
+    HES_NotComplete = 0,
+    HES_Complete,
+    HES_Discarded
+  };
+
   /// Constructs a ready SYCL event.
   ///
   /// If the constructed SYCL event is waited on it will complete immediately.
-  event_impl();
+  event_impl(HostEventState State = HES_Complete);
   /// Constructs an event instance from a plug-in event handle.
   ///
   /// The SyclContext must match the plug-in context associated with the
@@ -190,6 +196,11 @@ class event_impl {
   /// Cleans dependencies of this event_impl
   void cleanupDependencyEvents();
 
+  /// Checks if this event is discarded by SYCL implementation.
+  ///
+  /// \return true if this event is discarded.
+  bool isDiscarded() const { return MState == HES_Discarded; }
+
 private:
   // When instrumentation is enabled emits trace event for event wait begin and
   // returns the telemetry event generated for the wait
@@ -215,8 +226,6 @@ class event_impl {
   /// the queue to the device.
   std::atomic<bool> MIsFlushed = false;
 
-  enum HostEventState : int { HES_NotComplete = 0, HES_Complete };
-
   // State of host event. Employed only for host events and event with no
   // backend's representation (e.g. alloca). Used values are listed in
   // HostEventState enum.