[SYCL] Implementation of discard_events

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,39 @@ 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);
+
+  // TODO: Delete this function when ABI breaking changes are allowed.
   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);
+
+  // TODO: Delete this function when ABI breaking changes are allowed.
   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);
+
+  // TODO: Delete this function when ABI breaking changes are allowed.
   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);
+
+  // TODO: Delete this function when ABI breaking changes are allowed.
   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

@@ -1351,6 +1351,12 @@ class __SYCL_EXPORT handler {
   ///
   /// \param Event is a valid SYCL event to wait on.
   void depends_on(event Event) {
+    if (info::event_command_status::ext_oneapi_unknown ==
+        Event.get_info<info::event::command_execution_status>()) {
+      throw sycl::exception(make_error_code(errc::invalid),
+                            "Queue operation cannot depend on invalid event.");
+    }
+
     MEvents.push_back(detail::getSyclObjImpl(Event));
   }
 
@@ -1359,6 +1365,12 @@ class __SYCL_EXPORT handler {
   /// \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) {
+      if (info::event_command_status::ext_oneapi_unknown ==
+          Event.get_info<info::event::command_execution_status>()) {
+        throw sycl::exception(
+            make_error_code(errc::invalid),
+            "Queue operation cannot depend on invalid event.");
+      }
       MEvents.push_back(detail::getSyclObjImpl(Event));
     }
   }

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

@@ -287,7 +287,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,
+  ext_oneapi_unknown =
+      submitted + running + complete +
+      1 // TODO: a more elegant way to ensure that the unique value is here
 };
 
 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,
@@ -267,14 +265,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
@@ -292,8 +290,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](
@@ -313,15 +309,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
@@ -1087,6 +1083,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 an invalid 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_Invalid)
+    throw sycl::exception(
+        make_error_code(errc::invalid),
+        "waitInternal method cannot be used for an invalid event.");
+
   while (MState != HES_Complete)
     ;
 }
@@ -93,7 +98,7 @@ void event_impl::setContextImpl(const ContextImplPtr &Context) {
   MState = HES_NotComplete;
 }
 
-event_impl::event_impl() : MState(HES_Complete) {}
+event_impl::event_impl(HostEventState State) : MState(State) {}
 
 event_impl::event_impl(RT::PiEvent Event, const context &SyclContext)
     : MEvent(Event), MContext(detail::getSyclObjImpl(SyclContext)),
@@ -187,6 +192,10 @@ void event_impl::instrumentationEpilog(void *TelemetryEvent,
 
 void event_impl::wait(
     std::shared_ptr<cl::sycl::detail::event_impl> Self) const {
+  if (MState == HES_Invalid)
+    throw sycl::exception(make_error_code(errc::invalid),
+                          "wait method cannot be used for an invalid event.");
+
 #ifdef XPTI_ENABLE_INSTRUMENTATION
   void *TelemetryEvent = nullptr;
   uint64_t IId;
@@ -303,6 +312,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_Invalid)
+    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());
@@ -332,6 +344,11 @@ pi_native_handle event_impl::getNative() const {
 }
 
 std::vector<EventImplPtr> event_impl::getWaitList() {
+  if (MState == HES_Invalid)
+    throw sycl::exception(
+        make_error_code(errc::invalid),
+        "get_wait_list() cannot be used for an invalid event.");
+
   std::lock_guard<std::mutex> Lock(MMutex);
 
   std::vector<EventImplPtr> Result;

sycl/source/detail/event_impl.hpp

@@ -32,10 +32,12 @@ using EventImplPtr = std::shared_ptr<cl::sycl::detail::event_impl>;
 
 class event_impl {
 public:
+  enum HostEventState : int { HES_NotComplete = 0, HES_Complete, HES_Invalid };
+
   /// 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
@@ -205,8 +207,6 @@ class event_impl {
   std::vector<EventImplPtr> MPreparedDepsEvents;
   std::vector<EventImplPtr> MPreparedHostDepsEvents;
 
-  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.

sycl/source/detail/memory_manager.cpp

@@ -647,14 +647,14 @@ void MemoryManager::unmap(SYCLMemObjI *, void *Mem, QueueImplPtr Queue,
 void MemoryManager::copy_usm(const void *SrcMem, QueueImplPtr SrcQueue,
                              size_t Len, void *DstMem,
                              std::vector<RT::PiEvent> DepEvents,
-                             RT::PiEvent &OutEvent) {
+                             RT::PiEvent *OutEvent) {
   sycl::context Context = SrcQueue->get_context();
 
   if (!Len) { // no-op, but ensure DepEvents will still be waited on
     if (!Context.is_host() && !DepEvents.empty()) {
       SrcQueue->getPlugin().call<PiApiKind::piEnqueueEventsWait>(
           SrcQueue->getHandleRef(), DepEvents.size(), DepEvents.data(),
-          &OutEvent);
+          OutEvent);
     }
     return;
   }
@@ -670,19 +670,19 @@ void MemoryManager::copy_usm(const void *SrcMem, QueueImplPtr SrcQueue,
     Plugin.call<PiApiKind::piextUSMEnqueueMemcpy>(SrcQueue->getHandleRef(),
                                                   /* blocking */ false, DstMem,
                                                   SrcMem, Len, DepEvents.size(),
-                                                  DepEvents.data(), &OutEvent);
+                                                  DepEvents.data(), OutEvent);
   }
 }
 
 void MemoryManager::fill_usm(void *Mem, QueueImplPtr Queue, size_t Length,
                              int Pattern, std::vector<RT::PiEvent> DepEvents,
-                             RT::PiEvent &OutEvent) {
+                             RT::PiEvent *OutEvent) {
   sycl::context Context = Queue->get_context();
 
   if (!Length) { // no-op, but ensure DepEvents will still be waited on
     if (!Context.is_host() && !DepEvents.empty()) {
       Queue->getPlugin().call<PiApiKind::piEnqueueEventsWait>(
-          Queue->getHandleRef(), DepEvents.size(), DepEvents.data(), &OutEvent);
+          Queue->getHandleRef(), DepEvents.size(), DepEvents.data(), OutEvent);
     }
     return;
   }
@@ -697,13 +697,13 @@ void MemoryManager::fill_usm(void *Mem, QueueImplPtr Queue, size_t Length,
     const detail::plugin &Plugin = Queue->getPlugin();
     Plugin.call<PiApiKind::piextUSMEnqueueMemset>(
         Queue->getHandleRef(), Mem, Pattern, Length, DepEvents.size(),
-        DepEvents.data(), &OutEvent);
+        DepEvents.data(), OutEvent);
   }
 }
 
 void MemoryManager::prefetch_usm(void *Mem, QueueImplPtr Queue, size_t Length,
                                  std::vector<RT::PiEvent> DepEvents,
-                                 RT::PiEvent &OutEvent) {
+                                 RT::PiEvent *OutEvent) {
   sycl::context Context = Queue->get_context();
 
   if (Context.is_host()) {
@@ -712,23 +712,52 @@ void MemoryManager::prefetch_usm(void *Mem, QueueImplPtr Queue, size_t Length,
     const detail::plugin &Plugin = Queue->getPlugin();
     Plugin.call<PiApiKind::piextUSMEnqueuePrefetch>(
         Queue->getHandleRef(), Mem, Length, _pi_usm_migration_flags(0),
-        DepEvents.size(), DepEvents.data(), &OutEvent);
+        DepEvents.size(), DepEvents.data(), OutEvent);
   }
 }
 
 void MemoryManager::advise_usm(const void *Mem, QueueImplPtr Queue,
                                size_t Length, pi_mem_advice Advice,
                                std::vector<RT::PiEvent> /*DepEvents*/,
-                               RT::PiEvent &OutEvent) {
+                               RT::PiEvent *OutEvent) {
   sycl::context Context = Queue->get_context();
 
   if (!Context.is_host()) {
     const detail::plugin &Plugin = Queue->getPlugin();
     Plugin.call<PiApiKind::piextUSMEnqueueMemAdvise>(Queue->getHandleRef(), Mem,
-                                                     Length, Advice, &OutEvent);
+                                                     Length, Advice, OutEvent);
   }
 }
 
+// TODO: Delete this function when ABI breaking changes are allowed.
+void MemoryManager::copy_usm(const void *SrcMem, QueueImplPtr Queue, size_t Len,
+                             void *DstMem, std::vector<RT::PiEvent> DepEvents,
+                             RT::PiEvent &OutEvent) {
+  copy_usm(SrcMem, Queue, Len, DstMem, DepEvents, &OutEvent);
+}
+
+// TODO: Delete this function when ABI breaking changes are allowed.
+void MemoryManager::fill_usm(void *DstMem, QueueImplPtr Queue, size_t Len,
+                             int Pattern, std::vector<RT::PiEvent> DepEvents,
+                             RT::PiEvent &OutEvent) {
+  fill_usm(DstMem, Queue, Len, Pattern, DepEvents, &OutEvent);
+}
+
+// TODO: Delete this function when ABI breaking changes are allowed.
+void MemoryManager::prefetch_usm(void *Ptr, QueueImplPtr Queue, size_t Len,
+                                 std::vector<RT::PiEvent> DepEvents,
+                                 RT::PiEvent &OutEvent) {
+  prefetch_usm(Ptr, Queue, Len, DepEvents, &OutEvent);
+}
+
+// TODO: Delete this function when ABI breaking changes are allowed.
+void MemoryManager::advise_usm(const void *Ptr, QueueImplPtr Queue, size_t Len,
+                               pi_mem_advice Advice,
+                               std::vector<RT::PiEvent> DepEvents,
+                               RT::PiEvent &OutEvent) {
+  advise_usm(Ptr, Queue, Len, Advice, DepEvents, &OutEvent);
+}
+
 } // namespace detail
 } // namespace sycl
 } // __SYCL_INLINE_NAMESPACE(cl)