From b3efd3ab62cdb5ec0d25a96218f894a07f4eb6be Mon Sep 17 00:00:00 2001 From: Gilles Grospellier Date: Tue, 25 Feb 2025 18:40:08 +0100 Subject: [PATCH 1/2] =?UTF-8?q?[arcane,accelerator]=20Ajoute=20m=C3=A9thod?= =?UTF-8?q?e=20'RunQueueEvent::hasPendingWork()'=20pour=20savoir=20si=20le?= =?UTF-8?q?s=20RunQueue=20associ=C3=A9es=20=C3=A0=20un=20=C3=A9v=C3=A8neme?= =?UTF-8?q?nt=20ont=20du=20travail.?= MIME-Version: 1.0 Content-Type: text/plain; charset=UTF-8 Content-Transfer-Encoding: 8bit Cette méthode encapsule 'cudaQueryEvent()' ou 'hipQueryEvent()'. --- arcane/src/arcane/accelerator/core/RunQueueEvent.cc | 11 +++++++++++ arcane/src/arcane/accelerator/core/RunQueueEvent.h | 8 ++++++++ .../src/arcane/accelerator/core/RunQueueRuntime.cc | 5 +++-- .../accelerator/core/internal/IRunQueueEventImpl.h | 6 ++++-- .../cuda/runtime/CudaAcceleratorRuntime.cc | 9 +++++++++ .../hip/runtime/HipAcceleratorRuntime.cc | 13 +++++++++++-- .../sycl/runtime/SyclAcceleratorRuntime.cc | 10 ++++++++-- 7 files changed, 54 insertions(+), 8 deletions(-) diff --git a/arcane/src/arcane/accelerator/core/RunQueueEvent.cc b/arcane/src/arcane/accelerator/core/RunQueueEvent.cc index 9e3eb89284..85ec3a1512 100644 --- a/arcane/src/arcane/accelerator/core/RunQueueEvent.cc +++ b/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 { diff --git a/arcane/src/arcane/accelerator/core/RunQueueEvent.h b/arcane/src/arcane/accelerator/core/RunQueueEvent.h index ac9aba7b24..ae157b5c7f 100644 --- a/arcane/src/arcane/accelerator/core/RunQueueEvent.h +++ b/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; diff --git a/arcane/src/arcane/accelerator/core/RunQueueRuntime.cc b/arcane/src/arcane/accelerator/core/RunQueueRuntime.cc index 12cfb8297a..b2016832ac 100644 --- a/arcane/src/arcane/accelerator/core/RunQueueRuntime.cc +++ b/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); diff --git a/arcane/src/arcane/accelerator/core/internal/IRunQueueEventImpl.h b/arcane/src/arcane/accelerator/core/internal/IRunQueueEventImpl.h index 6c2b215033..0904806d34 100644 --- a/arcane/src/arcane/accelerator/core/internal/IRunQueueEventImpl.h +++ b/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; }; /*---------------------------------------------------------------------------*/ diff --git a/arcane/src/arcane/accelerator/cuda/runtime/CudaAcceleratorRuntime.cc b/arcane/src/arcane/accelerator/cuda/runtime/CudaAcceleratorRuntime.cc index 46cb22bd61..4d0d4a9f29 100644 --- a/arcane/src/arcane/accelerator/cuda/runtime/CudaAcceleratorRuntime.cc +++ b/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; diff --git a/arcane/src/arcane/accelerator/hip/runtime/HipAcceleratorRuntime.cc b/arcane/src/arcane/accelerator/hip/runtime/HipAcceleratorRuntime.cc index 530c452fdc..6c6de44d9c 100644 --- a/arcane/src/arcane/accelerator/hip/runtime/HipAcceleratorRuntime.cc +++ b/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; diff --git a/arcane/src/arcane/accelerator/sycl/runtime/SyclAcceleratorRuntime.cc b/arcane/src/arcane/accelerator/sycl/runtime/SyclAcceleratorRuntime.cc index 1d2655b191..667ceffcdf 100644 --- a/arcane/src/arcane/accelerator/sycl/runtime/SyclAcceleratorRuntime.cc +++ b/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; From e43e36a7013cde0da0e81ec444fa28bb3b57af3e Mon Sep 17 00:00:00 2001 From: Gilles Grospellier Date: Tue, 25 Feb 2025 18:43:38 +0100 Subject: [PATCH 2/2] [arcane,tests] Ajoute test pour 'RunQueueEvent::hasPendingWork()'. --- .../tests/accelerator/RunQueueUnitTest.cc | 62 +++++++++---------- 1 file changed, 30 insertions(+), 32 deletions(-) diff --git a/arcane/src/arcane/tests/accelerator/RunQueueUnitTest.cc b/arcane/src/arcane/tests/accelerator/RunQueueUnitTest.cc index 1998b0fd41..2f6889ad97 100644 --- a/arcane/src/arcane/tests/accelerator/RunQueueUnitTest.cc +++ b/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,7 +177,7 @@ _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)); } @@ -193,6 +185,7 @@ _executeTest1(bool use_priority) 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> 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));