[SYCL] Don't return last event in ext_oneapi_submit_barrier

devops/compat_ci_exclude.sycl-rel-6_3

@@ -9,6 +9,13 @@ Matrix/joint_matrix_bf16_fill_k_cache_runtime_dim.cpp
 # See GSD-11097.
 Assert/assert_in_kernels.cpp
 
+# https://github.com/intel/llvm/pull/20159 prevents returning last event as an
+# optimization for submitting barrier to an empty IOQ. However, the test
+# actually checks whether last event is returned or not, so it needs to be
+# updated to match the new behavior. ext_oneapi_submit_barrier spec doesn't
+# require last event to be returned, so this is not an ABI break.
+InorderQueue/in_order_ext_oneapi_submit_barrier.cpp
+
 # Likely OK, but need author to provide justification, get approval/confirmation
 # from someone:
 

sycl/source/queue.cpp

@@ -334,24 +334,6 @@ void queue::wait_and_throw_proxy(const detail::code_location &CodeLoc) {
   impl->wait_and_throw(CodeLoc);
 }
 
-static event
-getBarrierEventForInorderQueueHelper(detail::queue_impl &QueueImpl) {
-  // This function should not be called when a queue is recording to a graph,
-  // as a graph can record from multiple queues and we cannot guarantee the
-  // last node added by an in-order queue will be the last node added to the
-  // graph.
-  assert(!QueueImpl.hasCommandGraph() &&
-         "Should not be called in on graph recording.");
-
-  sycl::detail::optional<event> LastEvent = QueueImpl.getLastEvent();
-  if (LastEvent)
-    return *LastEvent;
-
-  // If there was no last event, we create an empty one.
-  return detail::createSyclObjFromImpl<event>(
-      detail::event_impl::create_default_event());
-}
-
 /// Prevents any commands submitted afterward to this queue from executing
 /// until all commands previously submitted to this queue have entered the
 /// complete state.
@@ -374,18 +356,17 @@ event queue::ext_oneapi_submit_barrier(const detail::code_location &CodeLoc) {
 /// group is being enqueued on.
 event queue::ext_oneapi_submit_barrier(const std::vector<event> &WaitList,
                                        const detail::code_location &CodeLoc) {
+
+  // If waitlist contains only empty, default constructed events, ignore
+  // them.
   bool AllEventsEmptyOrNop = std::all_of(
       begin(WaitList), end(WaitList), [&](const event &Event) -> bool {
         detail::event_impl &EventImpl = *detail::getSyclObjImpl(Event);
         return (EventImpl.isDefaultConstructed() || EventImpl.isNOP()) &&
                !EventImpl.hasCommandGraph();
       });
-  if (is_in_order() && !impl->hasCommandGraph() && !impl->MIsProfilingEnabled &&
-      AllEventsEmptyOrNop) {
-    return getBarrierEventForInorderQueueHelper(*impl);
-  }
 
-  if (WaitList.empty())
+  if (WaitList.empty() || AllEventsEmptyOrNop)
     return submit([=](handler &CGH) { CGH.ext_oneapi_barrier(); }, CodeLoc);
   else
     return submit([=](handler &CGH) { CGH.ext_oneapi_barrier(WaitList); },

sycl/test-e2e/InorderQueue/in_order_ext_oneapi_submit_barrier.cpp

@@ -53,28 +53,13 @@ int main() {
     std::cout << "Test 2" << std::endl;
     *Res = 0;
 
-    auto Event1 = Q.submit(
-        [&](sycl::handler &CGH) { CGH.host_task([&] { *Res += 1; }); });
-    auto BarrierEvent1 = Q.ext_oneapi_submit_barrier();
-    assert(checkBarrierEvent(Q.get_backend(), Event1, BarrierEvent1,
-                             false /* host tasks used */));
-    auto Event2 = Q.submit([&](sycl::handler &CGH) { CGH.fill(Res, 10, 1); });
+    Q.submit([&](sycl::handler &CGH) { CGH.host_task([&] { *Res += 1; }); });
+    Q.ext_oneapi_submit_barrier();
+    Q.submit([&](sycl::handler &CGH) { CGH.fill(Res, 10, 1); });
 
     Q.wait();
     assert(*Res == 10);
   }
-
-  {
-    // Test cast 3 - empty queue.
-    std::cout << "Test 3" << std::endl;
-    sycl::queue EmptyQ({sycl::property::queue::in_order{}});
-    auto BarrierEvent = EmptyQ.ext_oneapi_submit_barrier();
-    assert(
-        BarrierEvent.get_info<sycl::info::event::command_execution_status>() ==
-        sycl::info::event_command_status::complete);
-    BarrierEvent.wait();
-  }
-
   {
     // Test cast 4 - graph.
     sycl::queue GQueue{sycl::property::queue::in_order{}};

sycl/unittests/Extensions/CMakeLists.txt

@@ -13,6 +13,7 @@ add_sycl_unittest(ExtensionsTests OBJECT
   OneAPIProd.cpp
   EnqueueFunctionsEvents.cpp
   EnqueueFunctionsPrefetch.cpp
+  ExtOneapiBarrierOpt.cpp
   ProfilingTag.cpp
   KernelProperties.cpp
   NoDeviceIPVersion.cpp

sycl/unittests/Extensions/ExtOneapiBarrierOpt.cpp

@@ -0,0 +1,60 @@
+//==------------------- ExtOneapiBarrierOpt.cpp ----------------------------==//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===----------------------------------------------------------------------===//
+
+#include <gtest/gtest.h>
+#include <helpers/ScopedEnvVar.hpp>
+#include <helpers/UrMock.hpp>
+#include <sycl/sycl.hpp>
+
+using namespace sycl;
+
+inline thread_local uint32_t NumEventsInWaitList;
+
+static ur_result_t redefinedEnqueueEventsWaitWithBarrierExt(void *pParams) {
+  auto params =
+      *static_cast<ur_enqueue_events_wait_with_barrier_ext_params_t *>(pParams);
+  NumEventsInWaitList = *(params.pnumEventsInWaitList);
+  return UR_RESULT_SUCCESS;
+}
+
+class ExtOneapiBarrierOptTest : public ::testing::Test {
+public:
+  ExtOneapiBarrierOptTest() : Mock{} {}
+
+protected:
+  void SetUp() override { NumEventsInWaitList = 0; }
+
+protected:
+  sycl::unittest::UrMock<> Mock;
+};
+
+// Check that ext_oneapi_submit_barrier works fine in the scenarios
+// when provided waitlist consists of only empty events.
+// Tets for https://github.com/intel/llvm/pull/12951
+TEST(ExtOneapiBarrierOptTest, EmptyEventTest) {
+  sycl::queue q1{{sycl::property::queue::in_order()}};
+
+  mock::getCallbacks().set_after_callback(
+      "urEnqueueEventsWaitWithBarrierExt",
+      &redefinedEnqueueEventsWaitWithBarrierExt);
+
+  NumEventsInWaitList = 100;
+  q1.ext_oneapi_submit_barrier();
+  ASSERT_EQ(0u, NumEventsInWaitList);
+
+  // ext_oneapi_submit_barrier should ignore empty, default constructed events.
+  sycl::event E1{};
+  NumEventsInWaitList = 100;
+  q1.ext_oneapi_submit_barrier({E1});
+  ASSERT_EQ(0u, NumEventsInWaitList);
+
+  sycl::event E2{};
+  NumEventsInWaitList = 100;
+  q1.ext_oneapi_submit_barrier({E1, E2});
+  ASSERT_EQ(0u, NumEventsInWaitList);
+}