Merge pull request #2006 from arcaneframework/dev/gg-add-accelerator-query-event

Add support to query work on `RunQueueEvent`

Author: grospelliergilles

arcane/src/arcane/accelerator/core/RunQueueEvent.cc

@@ -142,6 +142,17 @@ wait()
 /*---------------------------------------------------------------------------*/
 /*---------------------------------------------------------------------------*/
 
+bool RunQueueEvent::
+hasPendingWork() const
+{
+  if (m_p)
+    return m_p->m_impl->hasPendingWork();
+  return false;
+}
+
+/*---------------------------------------------------------------------------*/
+/*---------------------------------------------------------------------------*/
+
 impl::IRunQueueEventImpl* RunQueueEvent::
 _internalEventImpl() const
 {

arcane/src/arcane/accelerator/core/RunQueueEvent.h

@@ -74,6 +74,14 @@ class ARCANE_ACCELERATOR_CORE_EXPORT RunQueueEvent
   //! Bloque tant que les files associées à cet évènement n'ont pas fini leur travail.
   void wait();
 
+  /*!
+   * \brief Indique si les RunQueue associées à cet évènement ont fini leur travail.
+   *
+   * Retourne \a false si les RunQueue enregistrées via RunQueue::recordEvent() ont
+   * fini leur travail. Retourn \a true sinon.
+   */
+  bool hasPendingWork() const;
+
  private:
 
   impl::IRunQueueEventImpl* _internalEventImpl() const;

arcane/src/arcane/accelerator/core/RunQueueRuntime.cc

@@ -1,11 +1,11 @@
 // -*- tab-width: 2; indent-tabs-mode: nil; coding: utf-8-with-signature -*-
 //-----------------------------------------------------------------------------
-// Copyright 2000-2024 CEA (www.cea.fr) IFPEN (www.ifpenergiesnouvelles.com)
+// Copyright 2000-2025 CEA (www.cea.fr) IFPEN (www.ifpenergiesnouvelles.com)
 // See the top-level COPYRIGHT file for details.
 // SPDX-License-Identifier: Apache-2.0
 //-----------------------------------------------------------------------------
 /*---------------------------------------------------------------------------*/
-/* RunQueueRuntime.cc                                          (C) 2000-2024 */
+/* RunQueueRuntime.cc                                          (C) 2000-2025 */
 /*                                                                           */
 /* Implémentation d'un RunQueue pour une cible donnée.                       */
 /*---------------------------------------------------------------------------*/
@@ -83,6 +83,7 @@ class ARCANE_ACCELERATOR_CORE_EXPORT HostRunQueueEvent
   }
   void wait() final {}
   void waitForEvent(IRunQueueStream*) final {}
+  bool hasPendingWork() final { return false; }
   Int64 elapsedTime(IRunQueueEventImpl* start_event) final
   {
     ARCANE_CHECK_POINTER(start_event);

arcane/src/arcane/accelerator/core/internal/IRunQueueEventImpl.h

@@ -1,11 +1,11 @@
 // -*- tab-width: 2; indent-tabs-mode: nil; coding: utf-8-with-signature -*-
 //-----------------------------------------------------------------------------
-// Copyright 2000-2024 CEA (www.cea.fr) IFPEN (www.ifpenergiesnouvelles.com)
+// Copyright 2000-2025 CEA (www.cea.fr) IFPEN (www.ifpenergiesnouvelles.com)
 // See the top-level COPYRIGHT file for details.
 // SPDX-License-Identifier: Apache-2.0
 //-----------------------------------------------------------------------------
 /*---------------------------------------------------------------------------*/
-/* IRunQueueEventImpl.h                                        (C) 2000-2024 */
+/* IRunQueueEventImpl.h                                        (C) 2000-2025 */
 /*                                                                           */
 /* Interface de l'implémentation d'un évènement.                             */
 /*---------------------------------------------------------------------------*/
@@ -42,6 +42,8 @@ class ARCANE_ACCELERATOR_CORE_EXPORT IRunQueueEventImpl
 
   //! Temps écoulé (en nanoseconde) entre l'évènement \a from_event et cet évènement.
   virtual Int64 elapsedTime(IRunQueueEventImpl* from_event) = 0;
+
+  virtual bool hasPendingWork() =0;
 };
 
 /*---------------------------------------------------------------------------*/

arcane/src/arcane/accelerator/cuda/runtime/CudaAcceleratorRuntime.cc

@@ -232,6 +232,15 @@ class CudaRunQueueEvent
     return nano_time;
   }
 
+  bool hasPendingWork() final
+  {
+    cudaError_t v = cudaEventQuery(m_cuda_event);
+    if (v == cudaErrorNotReady)
+      return true;
+    ARCANE_CHECK_CUDA(v);
+    return false;
+  }
+
  private:
 
   cudaEvent_t m_cuda_event;

arcane/src/arcane/accelerator/hip/runtime/HipAcceleratorRuntime.cc

@@ -1,11 +1,11 @@
 // -*- tab-width: 2; indent-tabs-mode: nil; coding: utf-8-with-signature -*-
 //-----------------------------------------------------------------------------
-// Copyright 2000-2024 CEA (www.cea.fr) IFPEN (www.ifpenergiesnouvelles.com)
+// Copyright 2000-2025 CEA (www.cea.fr) IFPEN (www.ifpenergiesnouvelles.com)
 // See the top-level COPYRIGHT file for details.
 // SPDX-License-Identifier: Apache-2.0
 //-----------------------------------------------------------------------------
 /*---------------------------------------------------------------------------*/
-/* HipAcceleratorRuntime.cc                                    (C) 2000-2024 */
+/* HipAcceleratorRuntime.cc                                    (C) 2000-2025 */
 /*                                                                           */
 /* Runtime pour 'HIP'.                                                       */
 /*---------------------------------------------------------------------------*/
@@ -187,6 +187,15 @@ class HipRunQueueEvent
     return nano_time;
   }
 
+  bool hasPendingWork() final
+  {
+    hipError_t v = hipEventQuery(m_hip_event);
+    if (v == hipErrorNotReady)
+      return true;
+    ARCANE_CHECK_HIP(v);
+    return false;
+  }
+
  private:
 
   hipEvent_t m_hip_event;

arcane/src/arcane/accelerator/sycl/runtime/SyclAcceleratorRuntime.cc

@@ -1,11 +1,11 @@
 // -*- tab-width: 2; indent-tabs-mode: nil; coding: utf-8-with-signature -*-
 //-----------------------------------------------------------------------------
-// Copyright 2000-2024 CEA (www.cea.fr) IFPEN (www.ifpenergiesnouvelles.com)
+// Copyright 2000-2025 CEA (www.cea.fr) IFPEN (www.ifpenergiesnouvelles.com)
 // See the top-level COPYRIGHT file for details.
 // SPDX-License-Identifier: Apache-2.0
 //-----------------------------------------------------------------------------
 /*---------------------------------------------------------------------------*/
-/* SyclAcceleratorRuntime.cc                                   (C) 2000-2024 */
+/* SyclAcceleratorRuntime.cc                                   (C) 2000-2025 */
 /*                                                                           */
 /* Runtime pour 'SYCL'.                                                      */
 /*---------------------------------------------------------------------------*/
@@ -16,6 +16,7 @@
 
 #include "arcane/utils/PlatformUtils.h"
 #include "arcane/utils/NotSupportedException.h"
+#include "arcane/utils/NotImplementedException.h"
 #include "arcane/utils/FatalErrorException.h"
 #include "arcane/utils/IMemoryRessourceMng.h"
 #include "arcane/utils/internal/IMemoryRessourceMngInternal.h"
@@ -216,6 +217,11 @@ class SyclRunQueueEvent
     return (end - start);
   }
 
+  bool hasPendingWork() final
+  {
+    ARCANE_THROW(NotImplementedException,"hasPendingWork()");
+  }
+
  private:
 
   sycl::event m_sycl_event;

arcane/src/arcane/tests/accelerator/RunQueueUnitTest.cc

@@ -14,6 +14,7 @@
 #include "arcane/utils/NumArray.h"
 #include "arcane/utils/ValueChecker.h"
 #include "arcane/utils/MemoryUtils.h"
+#include "arcane/utils/PlatformUtils.h"
 
 #include "arcane/core/BasicUnitTest.h"
 #include "arcane/core/ServiceFactory.h"
@@ -22,6 +23,7 @@
 #include "arcane/accelerator/core/Runner.h"
 #include "arcane/accelerator/core/RunQueueEvent.h"
 #include "arcane/accelerator/core/IAcceleratorMng.h"
+#include "arcane/accelerator/core/internal/RunQueueImpl.h"
 
 #include "arcane/accelerator/NumArrayViews.h"
 #include "arcane/accelerator/SpanViews.h"
@@ -49,7 +51,6 @@ class RunQueueUnitTest
  public:
 
   explicit RunQueueUnitTest(const ServiceBuildInfo& cb);
-  ~RunQueueUnitTest();
 
  public:
 
@@ -58,14 +59,14 @@ class RunQueueUnitTest
 
  private:
 
-  ax::Runner* m_runner = nullptr;
+  ax::Runner m_runner;
 
  public:
 
   void _executeTestNullQueue();
   void _executeTest1(bool use_priority);
   void _executeTest2();
-  void _executeTest3();
+  void _executeTest3(bool use_pooling);
   void _executeTest4();
   void _executeTest5();
 };
@@ -87,21 +88,13 @@ RunQueueUnitTest(const ServiceBuildInfo& sb)
 /*---------------------------------------------------------------------------*/
 /*---------------------------------------------------------------------------*/
 
-RunQueueUnitTest::
-~RunQueueUnitTest()
-{
-}
-
-/*---------------------------------------------------------------------------*/
-/*---------------------------------------------------------------------------*/
-
 /*---------------------------------------------------------------------------*/
 /*---------------------------------------------------------------------------*/
 
 void RunQueueUnitTest::
 initializeTest()
 {
-  m_runner = subDomain()->acceleratorMng()->defaultRunner();
+  m_runner = subDomain()->acceleratorMng()->runner();
 }
 
 /*---------------------------------------------------------------------------*/
@@ -112,14 +105,13 @@ executeTest()
 {
   _executeTestNullQueue();
   _executeTest2();
-  bool old_v = m_runner->isConcurrentQueueCreation();
-  m_runner->setConcurrentQueueCreation(true);
   _executeTest1(false);
   _executeTest1(true);
-  _executeTest3();
+  _executeTest3(false);
+  if (m_runner.executionPolicy() != ax::eExecutionPolicy::SYCL)
+    _executeTest3(true);
   _executeTest4();
   _executeTest5();
-  m_runner->setConcurrentQueueCreation(old_v);
 }
 
 /*---------------------------------------------------------------------------*/
@@ -140,13 +132,13 @@ _executeTestNullQueue()
   if (queue.allocationOptions() != default_mem_opt)
     ARCANE_FATAL("Bad null allocationOptions()");
 
-  queue = makeQueue(*m_runner);
+  queue = makeQueue(m_runner);
   vc.areEqual(queue.isNull(), false, "not null");
 
   queue = RunQueue();
   vc.areEqual(queue.isNull(), true, "is null (2)");
 
-  queue = makeQueue(*m_runner);
+  queue = makeQueue(m_runner);
   if (queue.executionPolicy() == eExecutionPolicy::None)
     ARCANE_FATAL("Bad execution policy");
 }
@@ -185,14 +177,15 @@ _executeTest1(bool use_priority)
     ax::RunQueueBuildInfo bi;
     if (use_priority && (i > 3))
       bi.setPriority(-8);
-    auto queue_ref = makeQueueRef(*m_runner, bi);
+    auto queue_ref = makeQueueRef(m_runner, bi);
     queue_ref->setAsync(true);
     allthreads.add(new std::thread(task_func, queue_ref, i));
   }
   for (auto thr : allthreads) {
     thr->join();
     delete thr;
   }
+  info() << "End of wait";
 
   Int64 true_total = 0;
   Int64 expected_true_total = 0;
@@ -217,10 +210,10 @@ _executeTest2()
   info() << "Test2: use events";
   ValueChecker vc(A_FUNCINFO);
 
-  auto event{ makeEvent(*m_runner) };
-  auto queue1{ makeQueue(*m_runner) };
+  auto event{ makeEvent(m_runner) };
+  auto queue1{ makeQueue(m_runner) };
   queue1.setAsync(true);
-  auto queue2{ makeQueue(*m_runner) };
+  auto queue2{ makeQueue(m_runner) };
   queue2.setAsync(true);
 
   Integer nb_value = 100000;
@@ -244,7 +237,7 @@ _executeTest2()
       v(iter) = v(iter) * 2;
     };
   }
-  queue1.barrier();
+
   queue2.barrier();
 
   // Vérifie les valeurs
@@ -257,19 +250,19 @@ _executeTest2()
 
 /*---------------------------------------------------------------------------*/
 /*---------------------------------------------------------------------------*/
-// Test la synchronisation de avec un évènement.
+// Teste la synchronisation avec un évènement.
 void RunQueueUnitTest::
-_executeTest3()
+_executeTest3(bool use_pooling)
 {
-  info() << "Test3: use events with wait()";
+  info() << "Test3: use events with wait() or pooling is_pooling?=" << use_pooling;
   ValueChecker vc(A_FUNCINFO);
 
   UniqueArray<Ref<ax::RunQueueEvent>> event_array;
-  event_array.add(makeEventRef(*m_runner));
+  event_array.add(makeEventRef(m_runner));
 
-  auto queue1{ makeQueue(*m_runner) };
+  auto queue1{ makeQueue(m_runner) };
   queue1.setAsync(true);
-  auto queue2{ makeQueue(*m_runner) };
+  auto queue2{ makeQueue(m_runner) };
   queue2.setAsync(true);
 
   Integer nb_value = 100000;
@@ -284,7 +277,12 @@ _executeTest3()
     };
     queue1.recordEvent(event_array[0]);
   }
-  event_array[0]->wait();
+  if (use_pooling)
+    while (event_array[0]->hasPendingWork()) {
+      // Do something ...
+    }
+  else
+    event_array[0]->wait();
   {
     auto command2 = makeCommand(queue2);
     auto v = viewInOut(command2, values);
@@ -316,7 +314,7 @@ _executeTest4()
     Arcane::Accelerator::RunQueueEvent event0;
     if (!event0.isNull())
       ARCANE_FATAL("Event is not null");
-    event0 = makeEvent(*m_runner);
+    event0 = makeEvent(m_runner);
     if (event0.isNull())
       ARCANE_FATAL("Event is null");
     Arcane::Accelerator::RunQueueEvent event1(event0);
@@ -326,7 +324,7 @@ _executeTest4()
 
   ValueChecker vc(A_FUNCINFO);
   //![SampleRunQueueEventSample1]
-  Arcane::Accelerator::Runner runner = *m_runner;
+  Arcane::Accelerator::Runner runner = m_runner;
 
   Arcane::Accelerator::RunQueueEvent event(makeEvent(runner));