Arcane  v4.1.1.0
Documentation utilisateur
Chargement...
Recherche...
Aucune correspondance
RunCommandImpl.cc
1// -*- tab-width: 2; indent-tabs-mode: nil; coding: utf-8-with-signature -*-
2//-----------------------------------------------------------------------------
3// Copyright 2000-2025 CEA (www.cea.fr) IFPEN (www.ifpenergiesnouvelles.com)
4// See the top-level COPYRIGHT file for details.
5// SPDX-License-Identifier: Apache-2.0
6//-----------------------------------------------------------------------------
7/*---------------------------------------------------------------------------*/
8/* RunCommandImpl.cc (C) 2000-2025 */
9/* */
10/* Implémentation de la gestion d'une commande sur accélérateur. */
11/*---------------------------------------------------------------------------*/
12/*---------------------------------------------------------------------------*/
13
14#include "arccore/common/accelerator/internal/RunCommandImpl.h"
15#include "arccore/common/accelerator/internal/AcceleratorCoreGlobalInternal.h"
16
17#include "arccore/base/FatalErrorException.h"
18#include "arccore/base/ForLoopTraceInfo.h"
19#include "arccore/base/PlatformUtils.h"
20#include "arccore/base/Convert.h"
21#include "arccore/base/ConcurrencyBase.h"
22
23#include "arccore/common/accelerator/Runner.h"
24#include "arccore/common/accelerator/internal/IRunQueueEventImpl.h"
25#include "arccore/common/accelerator/internal/IRunQueueStream.h"
26#include "arccore/common/accelerator/internal/IRunnerRuntime.h"
27#include "arccore/common/accelerator/internal/RunQueueImpl.h"
28#include "arccore/common/accelerator/internal/ReduceMemoryImpl.h"
29#include "arccore/common/accelerator/internal/RunnerImpl.h"
30
31/*---------------------------------------------------------------------------*/
32/*---------------------------------------------------------------------------*/
33
34namespace Arcane::Accelerator::impl
35{
36
37/*---------------------------------------------------------------------------*/
38/*---------------------------------------------------------------------------*/
39
40RunCommandImpl::
41RunCommandImpl(RunQueueImpl* queue)
42: m_queue(queue)
43, m_use_accelerator(impl::isAcceleratorPolicy(queue->runner()->executionPolicy()))
44{
45 _init();
46}
47
48/*---------------------------------------------------------------------------*/
49/*---------------------------------------------------------------------------*/
50
51RunCommandImpl::
52~RunCommandImpl()
53{
54 _freePools();
55 delete m_start_event;
56 delete m_stop_event;
57}
58
59/*---------------------------------------------------------------------------*/
60/*---------------------------------------------------------------------------*/
61
62void RunCommandImpl::
63_freePools()
64{
65 while (!m_reduce_memory_pool.empty()) {
66 delete m_reduce_memory_pool.top();
67 m_reduce_memory_pool.pop();
68 }
69}
70
71/*---------------------------------------------------------------------------*/
72/*---------------------------------------------------------------------------*/
73
74IRunQueueEventImpl* RunCommandImpl::
75_createEvent()
76{
77 if (m_use_sequential_timer_event)
78 return getSequentialRunQueueRuntime()->createEventImplWithTimer();
79 return runner()->_createEventWithTimer();
80}
81
82/*---------------------------------------------------------------------------*/
83/*---------------------------------------------------------------------------*/
84
85void RunCommandImpl::
86_init()
87{
88 // N'utilise les timers accélérateur que si le profiling est activé.
89 // On fait cela pour éviter d'appeler les évènements accélérateurs car on
90 // ne connait pas encore leur influence sur les performances. Si elle est
91 // négligeable alors on pourra l'activer par défaut.
92
93 // TODO: il faudrait éventuellement avoir une instance séquentielle et
94 // une associée à runner() pour gérer le cas ou ProfilingRegistry::hasProfiling()
95 // change en cours d'exécution.
96 if (m_use_accelerator && !ProfilingRegistry::hasProfiling())
97 m_use_sequential_timer_event = true;
98
99 m_start_event = _createEvent();
100 m_stop_event = _createEvent();
101
102 if (auto v = Convert::Type<Int32>::tryParseFromEnvironment("ARCANE_ACCELERATOR_ALLOW_REUSE_COMMAND", true))
103 m_is_allow_reuse_command = (v.value() != 0);
104}
105
106/*---------------------------------------------------------------------------*/
107/*---------------------------------------------------------------------------*/
108
109RunCommandImpl* RunCommandImpl::
110create(RunQueueImpl* r)
111{
112 RunCommandImpl* c = r->_internalCreateOrGetRunCommandImpl();
113 c->_reset();
114 return c;
115}
116
117/*---------------------------------------------------------------------------*/
118/*---------------------------------------------------------------------------*/
119/*!
120 * \brief Notification du début d'exécution de la commande.
121 */
122void RunCommandImpl::
123notifyBeginLaunchKernel()
124{
125 if (m_has_been_launched) {
126 if (!m_is_allow_reuse_command)
127 ARCCORE_FATAL("Command has already been launched. You can not re-use the same command.\n"
128 " You can temporarily allow it if you set environment variable\n"
129 " ARCANE_ACCELERATOR_ALLOW_REUSE_COMMAND to 1\n");
130 }
131 IRunQueueStream* stream = internalStream();
132 stream->notifyBeginLaunchKernel(*this);
133 // TODO: utiliser la bonne stream en séquentiel
134 m_has_been_launched = true;
135 if (m_use_profiling) {
136 m_start_event->recordQueue(stream);
137 m_begin_time = platform::getRealTimeNS();
138 m_loop_one_exec_stat_ptr = &m_loop_one_exec_stat;
139 m_loop_one_exec_stat.setBeginTime(m_begin_time);
140 }
141}
142
143/*---------------------------------------------------------------------------*/
144/*---------------------------------------------------------------------------*/
145/*!
146 * \brief Notification de la fin de lancement de la commande.
147 *
148 * La commande continue à s'exécuter en tâche de fond.
149 */
150void RunCommandImpl::
151notifyEndLaunchKernel()
152{
153 IRunQueueStream* stream = internalStream();
154 // TODO: utiliser la bonne stream en séquentiel
155 if (m_use_profiling)
156 m_stop_event->recordQueue(stream);
157 stream->notifyEndLaunchKernel(*this);
158 m_queue->_addRunningCommand(this);
159}
160
161/*---------------------------------------------------------------------------*/
162/*---------------------------------------------------------------------------*/
163/*!
164 * \brief Notification du lancement d'un kernel SYCL.
165 *
166 * \a sycl_event_ptr est de type sycl::event* et contient
167 * l'évènement associé à la commande qui est retourné lors
168 * des appels à sycl::queue::submit().
169 */
170void RunCommandImpl::
171notifyLaunchKernelSyclEvent(void* sycl_event_ptr)
172{
173 IRunQueueStream* stream = internalStream();
174 stream->_setSyclLastCommandEvent(sycl_event_ptr);
175 // Il faut enregistrer à nouveau la file associée à l'évènement
176 // car lors de l'appel à notifyBeginLaunchKernel() il n'y avait pas
177 // encore l'évènement associé à cette file.
178 m_start_event->recordQueue(stream);
179}
180
181/*---------------------------------------------------------------------------*/
182/*---------------------------------------------------------------------------*/
183/*!
184 * \brief Notification de la fin d'exécution du noyau.
185 *
186 * Après cet appel, on est sur que la commande a fini de s'exécuter et on
187 * peut la recycler. En asynchrone, cette méthode est appelée lors de la
188 * synchronisation d'une file.
189 */
190void RunCommandImpl::
191notifyEndExecuteKernel()
192{
193 // Ne fait rien si la commande n'a pas été lancée.
194 if (!m_has_been_launched)
195 return;
196
197 Int64 diff_time_ns = 0;
198 if (m_use_profiling){
199 diff_time_ns = m_stop_event->elapsedTime(m_start_event);
200 runner()->addTime((double)diff_time_ns / 1.0e9);
201 }
202
203 ForLoopOneExecStat* exec_info = m_loop_one_exec_stat_ptr;
204 if (exec_info) {
205 exec_info->setEndTime(m_begin_time + diff_time_ns);
206 //std::cout << "END_EXEC exec_info=" << m_loop_run_info.traceInfo().traceInfo() << "\n";
207 ForLoopTraceInfo flti(traceInfo(), kernelName());
208 ProfilingRegistry::_threadLocalForLoopInstance()->merge(*exec_info, flti);
209 }
210}
211
212/*---------------------------------------------------------------------------*/
213/*---------------------------------------------------------------------------*/
214
215void RunCommandImpl::
216_reset()
217{
218 m_kernel_name = String();
219 m_trace_info = TraceInfo();
220 m_nb_thread_per_block = 0;
221 m_use_profiling = ProfilingRegistry::hasProfiling();
222 m_parallel_loop_options = ConcurrencyBase::defaultParallelLoopOptions();
223 m_begin_time = 0;
224 m_loop_one_exec_stat.reset();
225 m_loop_one_exec_stat_ptr = nullptr;
226 m_has_been_launched = false;
227 m_has_living_run_command = false;
228 m_may_be_put_in_pool = false;
229 m_shared_memory_size = 0;
230}
231
232/*---------------------------------------------------------------------------*/
233/*---------------------------------------------------------------------------*/
234
235IReduceMemoryImpl* RunCommandImpl::
236getOrCreateReduceMemoryImpl()
237{
238 ReduceMemoryImpl* p = _getOrCreateReduceMemoryImpl();
239 if (p) {
240 m_active_reduce_memory_list.insert(p);
241 }
242 return p;
243}
244
245/*---------------------------------------------------------------------------*/
246/*---------------------------------------------------------------------------*/
247
248void RunCommandImpl::
249releaseReduceMemoryImpl(ReduceMemoryImpl* p)
250{
251 auto x = m_active_reduce_memory_list.find(p);
252 if (x == m_active_reduce_memory_list.end())
253 ARCCORE_FATAL("ReduceMemoryImpl in not in active list");
254 m_active_reduce_memory_list.erase(x);
255 m_reduce_memory_pool.push(p);
256}
257
258/*---------------------------------------------------------------------------*/
259/*---------------------------------------------------------------------------*/
260
261IRunQueueStream* RunCommandImpl::
262internalStream() const
263{
264 return m_queue->_internalStream();
265}
266
267/*---------------------------------------------------------------------------*/
268/*---------------------------------------------------------------------------*/
269
270RunnerImpl* RunCommandImpl::
271runner() const
272{
273 return m_queue->runner();
274}
275
276/*---------------------------------------------------------------------------*/
277/*---------------------------------------------------------------------------*/
278
279ReduceMemoryImpl* RunCommandImpl::
280_getOrCreateReduceMemoryImpl()
281{
282 // Pas besoin d'allouer de la mémoire spécifique si on n'est pas
283 // sur un accélérateur
284 if (!m_use_accelerator)
285 return nullptr;
286
287 auto& pool = m_reduce_memory_pool;
288
289 if (!pool.empty()) {
290 ReduceMemoryImpl* p = pool.top();
291 pool.pop();
292 return p;
293 }
294 return new ReduceMemoryImpl(this);
295}
296
297/*---------------------------------------------------------------------------*/
298/*---------------------------------------------------------------------------*/
299/*!
300 * \brief Méthode appelée quand l'instance RunCommand associée est détruite.
301 */
302void RunCommandImpl::
303_notifyDestroyRunCommand()
304{
305 // Si la commande n'a pas été lancé, il faut la remettre dans le pool
306 // des commandes de la file (sinon on aura une fuite mémoire)
307 if (!m_has_been_launched || m_may_be_put_in_pool)
308 m_queue->_putInCommandPool(this);
309}
310
311/*---------------------------------------------------------------------------*/
312/*---------------------------------------------------------------------------*/
313
314} // namespace Arcane::Accelerator::impl
315
316/*---------------------------------------------------------------------------*/
317/*---------------------------------------------------------------------------*/