12
12
#include < detail/global_handler.hpp>
13
13
#include < detail/graph_impl.hpp>
14
14
#include < detail/handler_impl.hpp>
15
+ #include < detail/host_task.hpp>
15
16
#include < detail/image_impl.hpp>
16
17
#include < detail/kernel_bundle_impl.hpp>
17
18
#include < detail/kernel_impl.hpp>
18
19
#include < detail/queue_impl.hpp>
19
20
#include < detail/scheduler/commands.hpp>
20
21
#include < detail/scheduler/scheduler.hpp>
21
- #include < detail/host_task.hpp>
22
22
#include < detail/usm/usm_impl.hpp>
23
23
#include < sycl/detail/common.hpp>
24
24
#include < sycl/detail/helpers.hpp>
@@ -90,8 +90,8 @@ handler::handler(std::shared_ptr<detail::queue_impl> Queue,
90
90
std::shared_ptr<detail::queue_impl> SecondaryQueue,
91
91
bool CallerNeedsEvent)
92
92
: impl(std::make_shared<detail::handler_impl>(std::move(PrimaryQueue),
93
- std::move (SecondaryQueue),
94
- CallerNeedsEvent)),
93
+ std::move (SecondaryQueue),
94
+ CallerNeedsEvent)),
95
95
MQueue(std::move(Queue)) {}
96
96
97
97
handler::handler (
@@ -123,7 +123,8 @@ bool handler::isStateExplicitKernelBundle() const {
123
123
std::shared_ptr<detail::kernel_bundle_impl>
124
124
handler::getOrInsertHandlerKernelBundle (bool Insert) const {
125
125
if (!impl->MKernelBundle && Insert) {
126
- auto Ctx = impl->MGraph ? impl->MGraph ->getContext () : MQueue->get_context ();
126
+ auto Ctx =
127
+ impl->MGraph ? impl->MGraph ->getContext () : MQueue->get_context ();
127
128
auto Dev = impl->MGraph ? impl->MGraph ->getDevice () : MQueue->get_device ();
128
129
impl->MKernelBundle = detail::getSyclObjImpl (
129
130
get_kernel_bundle<bundle_state::input>(Ctx, {Dev}, {}));
@@ -199,7 +200,8 @@ event handler::finalize() {
199
200
// Make sure implicit non-interop kernel bundles have the kernel
200
201
if (!KernelBundleImpPtr->isInterop () &&
201
202
!impl->isStateExplicitKernelBundle ()) {
202
- auto Dev = impl->MGraph ? impl->MGraph ->getDevice () : MQueue->get_device ();
203
+ auto Dev =
204
+ impl->MGraph ? impl->MGraph ->getDevice () : MQueue->get_device ();
203
205
kernel_id KernelID =
204
206
detail::ProgramManager::getInstance ().getSYCLKernelID (
205
207
MKernelName.c_str ());
@@ -356,21 +358,22 @@ event handler::finalize() {
356
358
new detail::CGUpdateHost (MDstPtr, std::move (impl->CGData ), MCodeLoc));
357
359
break ;
358
360
case detail::CGType::CopyUSM:
359
- CommandGroup.reset (new detail::CGCopyUSM (MSrcPtr, MDstPtr, MLength,
360
- std::move (impl->CGData ), MCodeLoc));
361
+ CommandGroup.reset (new detail::CGCopyUSM (
362
+ MSrcPtr, MDstPtr, MLength, std::move (impl->CGData ), MCodeLoc));
361
363
break ;
362
364
case detail::CGType::FillUSM:
363
- CommandGroup.reset (new detail::CGFillUSM (
364
- std::move (MPattern), MDstPtr, MLength, std::move (impl->CGData ), MCodeLoc));
365
+ CommandGroup.reset (new detail::CGFillUSM (std::move (MPattern), MDstPtr,
366
+ MLength, std::move (impl->CGData ),
367
+ MCodeLoc));
365
368
break ;
366
369
case detail::CGType::PrefetchUSM:
367
- CommandGroup.reset (new detail::CGPrefetchUSM (MDstPtr, MLength,
368
- std::move (impl->CGData ), MCodeLoc));
370
+ CommandGroup.reset (new detail::CGPrefetchUSM (
371
+ MDstPtr, MLength, std::move (impl->CGData ), MCodeLoc));
369
372
break ;
370
373
case detail::CGType::AdviseUSM:
371
374
CommandGroup.reset (new detail::CGAdviseUSM (MDstPtr, MLength, impl->MAdvice ,
372
- std::move (impl->CGData ), getType (),
373
- MCodeLoc));
375
+ std::move (impl->CGData ),
376
+ getType (), MCodeLoc));
374
377
break ;
375
378
case detail::CGType::Copy2DUSM:
376
379
CommandGroup.reset (new detail::CGCopy2DUSM (
@@ -388,8 +391,9 @@ event handler::finalize() {
388
391
std::move (impl->CGData ), MCodeLoc));
389
392
break ;
390
393
case detail::CGType::CodeplayHostTask: {
391
- auto context = impl->MGraph ? detail::getSyclObjImpl (impl->MGraph ->getContext ())
392
- : MQueue->getContextImplPtr ();
394
+ auto context = impl->MGraph
395
+ ? detail::getSyclObjImpl (impl->MGraph ->getContext ())
396
+ : MQueue->getContextImplPtr ();
393
397
CommandGroup.reset (new detail::CGHostTask (
394
398
std::move (impl->MHostTask ), MQueue, context, std::move (impl->MArgs ),
395
399
std::move (impl->CGData ), getType (), MCodeLoc));
@@ -399,13 +403,13 @@ event handler::finalize() {
399
403
case detail::CGType::BarrierWaitlist: {
400
404
if (auto GraphImpl = getCommandGraph (); GraphImpl != nullptr ) {
401
405
impl->CGData .MEvents .insert (std::end (impl->CGData .MEvents ),
402
- std::begin (impl->MEventsWaitWithBarrier ),
403
- std::end (impl->MEventsWaitWithBarrier ));
406
+ std::begin (impl->MEventsWaitWithBarrier ),
407
+ std::end (impl->MEventsWaitWithBarrier ));
404
408
// Barrier node is implemented as an empty node in Graph
405
409
// but keep the barrier type to help managing dependencies
406
410
setType (detail::CGType::Barrier);
407
- CommandGroup.reset (
408
- new detail::CG (detail::CGType::Barrier, std::move (impl->CGData ), MCodeLoc));
411
+ CommandGroup.reset (new detail::CG (detail::CGType::Barrier,
412
+ std::move (impl->CGData ), MCodeLoc));
409
413
} else {
410
414
CommandGroup.reset (
411
415
new detail::CGBarrier (std::move (impl->MEventsWaitWithBarrier ),
@@ -414,7 +418,8 @@ event handler::finalize() {
414
418
break ;
415
419
}
416
420
case detail::CGType::ProfilingTag: {
417
- CommandGroup.reset (new detail::CGProfilingTag (std::move (impl->CGData ), MCodeLoc));
421
+ CommandGroup.reset (
422
+ new detail::CGProfilingTag (std::move (impl->CGData ), MCodeLoc));
418
423
break ;
419
424
}
420
425
case detail::CGType::CopyToDeviceGlobal: {
@@ -466,17 +471,18 @@ event handler::finalize() {
466
471
CommandGroup.reset (new detail::CGCopyImage (
467
472
MSrcPtr, MDstPtr, impl->MImageDesc , impl->MImageFormat ,
468
473
impl->MImageCopyFlags , impl->MSrcOffset , impl->MDestOffset ,
469
- impl->MHostExtent , impl->MCopyExtent , std::move (impl->CGData ), MCodeLoc));
474
+ impl->MHostExtent , impl->MCopyExtent , std::move (impl->CGData ),
475
+ MCodeLoc));
470
476
break ;
471
477
case detail::CGType::SemaphoreWait:
472
478
CommandGroup.reset (new detail::CGSemaphoreWait (
473
- impl->MInteropSemaphoreHandle , impl->MWaitValue , std::move (impl-> CGData ),
474
- MCodeLoc));
479
+ impl->MInteropSemaphoreHandle , impl->MWaitValue ,
480
+ std::move (impl-> CGData ), MCodeLoc));
475
481
break ;
476
482
case detail::CGType::SemaphoreSignal:
477
483
CommandGroup.reset (new detail::CGSemaphoreSignal (
478
- impl->MInteropSemaphoreHandle , impl->MSignalValue , std::move (impl-> CGData ),
479
- MCodeLoc));
484
+ impl->MInteropSemaphoreHandle , impl->MSignalValue ,
485
+ std::move (impl-> CGData ), MCodeLoc));
480
486
break ;
481
487
case detail::CGType::None:
482
488
if (detail::pi::trace (detail::pi::TraceLevel::PI_TRACE_ALL)) {
@@ -487,8 +493,8 @@ event handler::finalize() {
487
493
// For Standard mode (non-graph),
488
494
// empty nodes are not sent to the scheduler to save time
489
495
if (impl->MGraph || (MQueue && MQueue->getCommandGraph ())) {
490
- CommandGroup.reset (
491
- new detail::CG (detail::CGType::None, std::move (impl->CGData ), MCodeLoc));
496
+ CommandGroup.reset (new detail::CG (detail::CGType::None,
497
+ std::move (impl->CGData ), MCodeLoc));
492
498
} else {
493
499
detail::EventImplPtr Event = std::make_shared<sycl::detail::event_impl>();
494
500
MLastEvent = detail::createSyclObjFromImpl<event>(Event);
@@ -524,8 +530,7 @@ event handler::finalize() {
524
530
GraphImpl->MMutex );
525
531
526
532
ext::oneapi::experimental::node_type NodeType =
527
- impl->MUserFacingNodeType !=
528
- ext::oneapi::experimental::node_type::empty
533
+ impl->MUserFacingNodeType != ext::oneapi::experimental::node_type::empty
529
534
? impl->MUserFacingNodeType
530
535
: ext::oneapi::experimental::detail::getNodeTypeFromCG (getType ());
531
536
@@ -675,17 +680,17 @@ void handler::processArg(void *Ptr, const detail::kernel_param_kind_t &Kind,
675
680
static_cast <detail::AccessorBaseHost *>(&S->GlobalBuf );
676
681
detail::AccessorImplPtr GBufImpl = detail::getSyclObjImpl (*GBufBase);
677
682
detail::Requirement *GBufReq = GBufImpl.get ();
678
- addArgsForGlobalAccessor (GBufReq, Index, IndexShift, Size,
679
- IsKernelCreatedFromSource,
680
- impl->MNDRDesc .GlobalSize .size (), impl->MArgs , IsESIMD);
683
+ addArgsForGlobalAccessor (
684
+ GBufReq, Index, IndexShift, Size, IsKernelCreatedFromSource,
685
+ impl->MNDRDesc .GlobalSize .size (), impl->MArgs , IsESIMD);
681
686
++IndexShift;
682
687
detail::AccessorBaseHost *GOffsetBase =
683
688
static_cast <detail::AccessorBaseHost *>(&S->GlobalOffset );
684
689
detail::AccessorImplPtr GOfssetImpl = detail::getSyclObjImpl (*GOffsetBase);
685
690
detail::Requirement *GOffsetReq = GOfssetImpl.get ();
686
- addArgsForGlobalAccessor (GOffsetReq, Index, IndexShift, Size,
687
- IsKernelCreatedFromSource,
688
- impl->MNDRDesc .GlobalSize .size (), impl->MArgs , IsESIMD);
691
+ addArgsForGlobalAccessor (
692
+ GOffsetReq, Index, IndexShift, Size, IsKernelCreatedFromSource,
693
+ impl->MNDRDesc .GlobalSize .size (), impl->MArgs , IsESIMD);
689
694
++IndexShift;
690
695
detail::AccessorBaseHost *GFlushBase =
691
696
static_cast <detail::AccessorBaseHost *>(&S->GlobalFlushBuf );
@@ -738,7 +743,7 @@ void handler::processArg(void *Ptr, const detail::kernel_param_kind_t &Kind,
738
743
// make it a minimum allocation of 1 byte.
739
744
SizeInBytes = std::max (SizeInBytes, 1 );
740
745
impl->MArgs .emplace_back (kernel_param_kind_t ::kind_std_layout, nullptr ,
741
- SizeInBytes, Index + IndexShift);
746
+ SizeInBytes, Index + IndexShift);
742
747
// TODO ESIMD currently does not suport MSize field passing yet
743
748
// accessor::init for ESIMD-mode accessor has a single field, translated
744
749
// to a single kernel argument set above.
@@ -880,8 +885,8 @@ void handler::verifyUsedKernelBundleInternal(detail::string_view KernelName) {
880
885
return ;
881
886
882
887
kernel_id KernelID = detail::get_kernel_id_impl (KernelName);
883
- device Dev =
884
- impl-> MGraph ? impl-> MGraph -> getDevice () : detail::getDeviceFromHandler (*this );
888
+ device Dev = impl-> MGraph ? impl-> MGraph -> getDevice ()
889
+ : detail::getDeviceFromHandler (*this );
885
890
if (!UsedKernelBundleImplPtr->has_kernel (KernelID, Dev))
886
891
throw sycl::exception (
887
892
make_error_code (errc::kernel_not_supported),
@@ -1459,8 +1464,10 @@ void handler::use_kernel_bundle(
1459
1464
const kernel_bundle<bundle_state::executable> &ExecBundle) {
1460
1465
std::shared_ptr<detail::queue_impl> PrimaryQueue =
1461
1466
impl->MSubmissionPrimaryQueue ;
1462
- if ((!impl->MGraph && (PrimaryQueue->get_context () != ExecBundle.get_context ())) ||
1463
- (impl->MGraph && (impl->MGraph ->getContext () != ExecBundle.get_context ())))
1467
+ if ((!impl->MGraph &&
1468
+ (PrimaryQueue->get_context () != ExecBundle.get_context ())) ||
1469
+ (impl->MGraph &&
1470
+ (impl->MGraph ->getContext () != ExecBundle.get_context ())))
1464
1471
throw sycl::exception (
1465
1472
make_error_code (errc::invalid),
1466
1473
" Context associated with the primary queue is different from the "
@@ -1834,16 +1841,14 @@ void handler::addArg(detail::kernel_param_kind_t ArgKind, void *Req,
1834
1841
impl->MArgs .emplace_back (ArgKind, Req, AccessTarget, ArgIndex);
1835
1842
}
1836
1843
1837
- void handler::clearArgs () {
1838
- impl->MArgs .clear ();
1839
- }
1844
+ void handler::clearArgs () { impl->MArgs .clear (); }
1840
1845
1841
1846
void handler::setArgsToAssociatedAccessors () {
1842
1847
impl->MArgs = impl->MAssociatedAccesors ;
1843
1848
}
1844
1849
1845
1850
bool handler::HasAssociatedAccessor (detail::AccessorImplHost *Req,
1846
- access::target AccessTarget) const {
1851
+ access::target AccessTarget) const {
1847
1852
return std::find_if (
1848
1853
impl->MAssociatedAccesors .cbegin (),
1849
1854
impl->MAssociatedAccesors .cend (), [&](const detail::ArgDesc &AD) {
@@ -1865,8 +1870,8 @@ void handler::setNDRangeDescriptorPadded(sycl::range<3> NumWorkItems,
1865
1870
impl->MNDRDesc = NDRDescT{NumWorkItems, Offset, Dims};
1866
1871
}
1867
1872
void handler::setNDRangeDescriptorPadded (sycl::range<3 > NumWorkItems,
1868
- sycl::range< 3 > LocalSize, sycl::id <3 > Offset ,
1869
- int Dims) {
1873
+ sycl::range <3 > LocalSize ,
1874
+ sycl::id< 3 > Offset, int Dims) {
1870
1875
impl->MNDRDesc = NDRDescT{NumWorkItems, LocalSize, Offset, Dims};
1871
1876
}
1872
1877
0 commit comments