Skip to content

Commit 9f8ae50

Browse files
romanovvladbader
romanovvlad
authored and
bader
committed
[SYCL] Add diagnostic on command group with multiple actions (#917)
Signed-off-by: Vlad Romanov <[email protected]>
1 parent 3acdb07 commit 9f8ae50

File tree

2 files changed

+72
-0
lines changed

2 files changed

+72
-0
lines changed

‎sycl/include/CL/sycl/handler.hpp

Lines changed: 35 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -190,6 +190,14 @@ class handler {
190190
return Storage;
191191
}
192192

193+
void throwIfActionIsCreated() {
194+
if (detail::CG::NONE != MCGType)
195+
throw sycl::runtime_error("Attempt to set multiple actions for the "
196+
"command group. Command group must consist of "
197+
"a single kernel or explicit memory operation.",
198+
CL_INVALID_OPERATION);
199+
}
200+
193201
// The method extracts and prepares kernel arguments from the lambda using
194202
// integration header.
195203
void
@@ -676,6 +684,7 @@ class handler {
676684
// single_task version with a kernel represented as a lambda.
677685
template <typename KernelName = detail::auto_name, typename KernelType>
678686
void single_task(KernelType KernelFunc) {
687+
throwIfActionIsCreated();
679688
using NameT =
680689
typename detail::get_kernel_name_t<KernelName, KernelType>::name;
681690
#ifdef __SYCL_DEVICE_ONLY__
@@ -693,6 +702,7 @@ class handler {
693702
template <typename KernelName = detail::auto_name, typename KernelType,
694703
int Dims>
695704
void parallel_for(range<Dims> NumWorkItems, KernelType KernelFunc) {
705+
throwIfActionIsCreated();
696706
using NameT =
697707
typename detail::get_kernel_name_t<KernelName, KernelType>::name;
698708
#ifdef __SYCL_DEVICE_ONLY__
@@ -706,6 +716,7 @@ class handler {
706716

707717
// Similar to single_task, but passed lambda will be executed on host.
708718
template <typename FuncT> void run_on_host_intel(FuncT Func) {
719+
throwIfActionIsCreated();
709720
MNDRDesc.set(range<1>{1});
710721

711722
MArgs = std::move(MAssociatedAccesors);
@@ -719,6 +730,7 @@ class handler {
719730
int Dims>
720731
void parallel_for(range<Dims> NumWorkItems, id<Dims> WorkItemOffset,
721732
KernelType KernelFunc) {
733+
throwIfActionIsCreated();
722734
using NameT =
723735
typename detail::get_kernel_name_t<KernelName, KernelType>::name;
724736
#ifdef __SYCL_DEVICE_ONLY__
@@ -735,6 +747,7 @@ class handler {
735747
template <typename KernelName = detail::auto_name, typename KernelType,
736748
int Dims>
737749
void parallel_for(nd_range<Dims> ExecutionRange, KernelType KernelFunc) {
750+
throwIfActionIsCreated();
738751
using NameT =
739752
typename detail::get_kernel_name_t<KernelName, KernelType>::name;
740753
#ifdef __SYCL_DEVICE_ONLY__
@@ -750,6 +763,7 @@ class handler {
750763
int Dims>
751764
void parallel_for_work_group(range<Dims> NumWorkGroups,
752765
KernelType KernelFunc) {
766+
throwIfActionIsCreated();
753767
using NameT =
754768
typename detail::get_kernel_name_t<KernelName, KernelType>::name;
755769
#ifdef __SYCL_DEVICE_ONLY__
@@ -766,6 +780,7 @@ class handler {
766780
void parallel_for_work_group(range<Dims> NumWorkGroups,
767781
range<Dims> WorkGroupSize,
768782
KernelType KernelFunc) {
783+
throwIfActionIsCreated();
769784
using NameT =
770785
typename detail::get_kernel_name_t<KernelName, KernelType>::name;
771786
#ifdef __SYCL_DEVICE_ONLY__
@@ -780,6 +795,7 @@ class handler {
780795
// single_task version with a kernel represented as a sycl::kernel.
781796
// The kernel invocation method has no functors and cannot be called on host.
782797
void single_task(kernel SyclKernel) {
798+
throwIfActionIsCreated();
783799
verifySyclKernelInvoc(SyclKernel);
784800
MNDRDesc.set(range<1>{1});
785801
MSyclKernel = detail::getSyclObjImpl(std::move(SyclKernel));
@@ -792,6 +808,7 @@ class handler {
792808
// functors and cannot be called on host.
793809
template <int Dims>
794810
void parallel_for(range<Dims> NumWorkItems, kernel SyclKernel) {
811+
throwIfActionIsCreated();
795812
verifySyclKernelInvoc(SyclKernel);
796813
MSyclKernel = detail::getSyclObjImpl(std::move(SyclKernel));
797814
MNDRDesc.set(std::move(NumWorkItems));
@@ -805,6 +822,7 @@ class handler {
805822
template <int Dims>
806823
void parallel_for(range<Dims> NumWorkItems, id<Dims> workItemOffset,
807824
kernel SyclKernel) {
825+
throwIfActionIsCreated();
808826
verifySyclKernelInvoc(SyclKernel);
809827
MSyclKernel = detail::getSyclObjImpl(std::move(SyclKernel));
810828
MNDRDesc.set(std::move(NumWorkItems), std::move(workItemOffset));
@@ -817,6 +835,7 @@ class handler {
817835
// method has no functors and cannot be called on host.
818836
template <int Dims>
819837
void parallel_for(nd_range<Dims> NDRange, kernel SyclKernel) {
838+
throwIfActionIsCreated();
820839
verifySyclKernelInvoc(SyclKernel);
821840
MSyclKernel = detail::getSyclObjImpl(std::move(SyclKernel));
822841
MNDRDesc.set(std::move(NDRange));
@@ -833,6 +852,7 @@ class handler {
833852
// which is used otherwise.
834853
template <typename KernelName = detail::auto_name, typename KernelType>
835854
void single_task(kernel SyclKernel, KernelType KernelFunc) {
855+
throwIfActionIsCreated();
836856
using NameT =
837857
typename detail::get_kernel_name_t<KernelName, KernelType>::name;
838858
#ifdef __SYCL_DEVICE_ONLY__
@@ -855,6 +875,7 @@ class handler {
855875
int Dims>
856876
void parallel_for(kernel SyclKernel, range<Dims> NumWorkItems,
857877
KernelType KernelFunc) {
878+
throwIfActionIsCreated();
858879
using NameT =
859880
typename detail::get_kernel_name_t<KernelName, KernelType>::name;
860881
#ifdef __SYCL_DEVICE_ONLY__
@@ -877,6 +898,7 @@ class handler {
877898
int Dims>
878899
void parallel_for(kernel SyclKernel, range<Dims> NumWorkItems,
879900
id<Dims> WorkItemOffset, KernelType KernelFunc) {
901+
throwIfActionIsCreated();
880902
using NameT =
881903
typename detail::get_kernel_name_t<KernelName, KernelType>::name;
882904
#ifdef __SYCL_DEVICE_ONLY__
@@ -899,6 +921,7 @@ class handler {
899921
int Dims>
900922
void parallel_for(kernel SyclKernel, nd_range<Dims> NDRange,
901923
KernelType KernelFunc) {
924+
throwIfActionIsCreated();
902925
using NameT =
903926
typename detail::get_kernel_name_t<KernelName, KernelType>::name;
904927
#ifdef __SYCL_DEVICE_ONLY__
@@ -924,6 +947,7 @@ class handler {
924947
int Dims>
925948
void parallel_for_work_group(kernel SyclKernel, range<Dims> NumWorkGroups,
926949
KernelType KernelFunc) {
950+
throwIfActionIsCreated();
927951
using NameT =
928952
typename detail::get_kernel_name_t<KernelName, KernelType>::name;
929953
#ifdef __SYCL_DEVICE_ONLY__
@@ -943,6 +967,7 @@ class handler {
943967
void parallel_for_work_group(kernel SyclKernel, range<Dims> NumWorkGroups,
944968
range<Dims> WorkGroupSize,
945969
KernelType KernelFunc) {
970+
throwIfActionIsCreated();
946971
using NameT =
947972
typename detail::get_kernel_name_t<KernelName, KernelType>::name;
948973
#ifdef __SYCL_DEVICE_ONLY__
@@ -977,6 +1002,7 @@ class handler {
9771002
access::placeholder IsPlaceholder = access::placeholder::false_t>
9781003
void copy(accessor<T_Src, Dims, AccessMode, AccessTarget, IsPlaceholder> Src,
9791004
shared_ptr_class<T_Dst> Dst) {
1005+
throwIfActionIsCreated();
9801006
static_assert(isValidTargetForExplicitOp(AccessTarget),
9811007
"Invalid accessor target for the copy method.");
9821008
// Make sure data shared_ptr points to is not released until we finish
@@ -993,6 +1019,7 @@ class handler {
9931019
void
9941020
copy(shared_ptr_class<T_Src> Src,
9951021
accessor<T_Dst, Dims, AccessMode, AccessTarget, IsPlaceholder> Dst) {
1022+
throwIfActionIsCreated();
9961023
static_assert(isValidTargetForExplicitOp(AccessTarget),
9971024
"Invalid accessor target for the copy method.");
9981025
// Make sure data shared_ptr points to is not released until we finish
@@ -1008,6 +1035,7 @@ class handler {
10081035
access::placeholder IsPlaceholder = access::placeholder::false_t>
10091036
void copy(accessor<T_Src, Dims, AccessMode, AccessTarget, IsPlaceholder> Src,
10101037
T_Dst *Dst) {
1038+
throwIfActionIsCreated();
10111039
static_assert(isValidTargetForExplicitOp(AccessTarget),
10121040
"Invalid accessor target for the copy method.");
10131041
#ifndef __SYCL_DEVICE_ONLY__
@@ -1047,6 +1075,7 @@ class handler {
10471075
void
10481076
copy(const T_Src *Src,
10491077
accessor<T_Dst, Dims, AccessMode, AccessTarget, IsPlaceholder> Dst) {
1078+
throwIfActionIsCreated();
10501079
static_assert(isValidTargetForExplicitOp(AccessTarget),
10511080
"Invalid accessor target for the copy method.");
10521081
#ifndef __SYCL_DEVICE_ONLY__
@@ -1124,6 +1153,7 @@ class handler {
11241153
accessor<T_Dst, Dims_Dst, AccessMode_Dst, AccessTarget_Dst,
11251154
IsPlaceholder_Dst>
11261155
Dst) {
1156+
throwIfActionIsCreated();
11271157
static_assert(isValidTargetForExplicitOp(AccessTarget_Src),
11281158
"Invalid source accessor target for the copy method.");
11291159
static_assert(isValidTargetForExplicitOp(AccessTarget_Dst),
@@ -1177,6 +1207,7 @@ class handler {
11771207
access::placeholder IsPlaceholder = access::placeholder::false_t>
11781208
void
11791209
update_host(accessor<T, Dims, AccessMode, AccessTarget, IsPlaceholder> Acc) {
1210+
throwIfActionIsCreated();
11801211
static_assert(isValidTargetForExplicitOp(AccessTarget),
11811212
"Invalid accessor target for the update_host method.");
11821213
MCGType = detail::CG::UPDATE_HOST;
@@ -1198,6 +1229,7 @@ class handler {
11981229
access::placeholder IsPlaceholder = access::placeholder::false_t>
11991230
void fill(accessor<T, Dims, AccessMode, AccessTarget, IsPlaceholder> Dst,
12001231
const T &Pattern) {
1232+
throwIfActionIsCreated();
12011233
// TODO add check:T must be an integral scalar value or a SYCL vector type
12021234
static_assert(isValidTargetForExplicitOp(AccessTarget),
12031235
"Invalid accessor target for the fill method.");
@@ -1229,6 +1261,7 @@ class handler {
12291261

12301262
// Copy memory from the source to the destination.
12311263
void memcpy(void *Dest, const void *Src, size_t Count) {
1264+
throwIfActionIsCreated();
12321265
MSrcPtr = const_cast<void *>(Src);
12331266
MDstPtr = Dest;
12341267
MLength = Count;
@@ -1237,6 +1270,7 @@ class handler {
12371270

12381271
// Fill the memory pointed to by the destination with the given bytes.
12391272
void memset(void *Dest, int Value, size_t Count) {
1273+
throwIfActionIsCreated();
12401274
MDstPtr = Dest;
12411275
MPattern.push_back((char)Value);
12421276
MLength = Count;
@@ -1245,6 +1279,7 @@ class handler {
12451279

12461280
// Prefetch the memory pointed to by the pointer.
12471281
void prefetch(const void *Ptr, size_t Count) {
1282+
throwIfActionIsCreated();
12481283
MDstPtr = const_cast<void *>(Ptr);
12491284
MLength = Count;
12501285
MCGType = detail::CG::PREFETCH_USM;

‎sycl/test/basic_tests/diagnostics/handler.cpp

Lines changed: 37 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,37 @@
1+
// RUN: %clangxx -fsycl %s -o %t.out
2+
// RUN: %t.out | FileCheck %s
3+
//==------------------- handler.cpp ----------------------------------------==//
4+
//
5+
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
6+
// See https://llvm.org/LICENSE.txt for license information.
7+
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
8+
//
9+
//===----------------------------------------------------------------------===//
10+
11+
#include <CL/sycl.hpp>
12+
13+
using namespace cl;
14+
15+
int main() {
16+
17+
bool Failed = false;
18+
19+
sycl::queue Queue([](sycl::exception_list ExceptionList) {
20+
if (ExceptionList.size() != 1) {
21+
std::cerr << "Should be one exception in exception list" << std::endl;
22+
std::abort();
23+
}
24+
std::rethrow_exception(*ExceptionList.begin());
25+
});
26+
27+
try {
28+
Queue.submit([&](sycl::handler &CGH) {
29+
CGH.single_task<class Dummy1>([]() {});
30+
CGH.single_task<class Dummy2>([]() {});
31+
});
32+
Queue.throw_asynchronous();
33+
} catch (sycl::exception &E) {
34+
// CHECK: Attempt to set multiple actions for the command group
35+
std::cout << E.what() << std::endl;
36+
}
37+
}

0 commit comments

Comments
 (0)