DPCT1102
目次
DPCT1102#
メッセージ#
SYCL* デバイスコードでは、ゼロレングスの配列は許されていません。
詳細な説明#
ゼロレングスの配列は、C++ 標準または SYCL* 仕様ではサポートされません。一部のホスト・コンパイラーの実装ではサポートされる場合がありますが、デバイスコード向けのインテル® oneAPI DPC++/C++ コンパイラーではサポートされていません。
修正方法の提案#
ゼロより大きなレングスの配列を使用してください。
例えば、以下のオリジナル CUDA* コードについて考えてみます。
1 #define DATA_NUM 16
2
3 struct Foo {
4 ...
5 float data[0];
6 };
7
8 __global__ void kernel(Foo *foo) {
9 foo->data[DATA_NUM - 1] = 123.f;
10 }
11
12 void test() {
13 Foo* foo;
14 cudaMalloc(&foo, sizeof(Foo) + DATA_NUM * sizeof(float));
15 kernel<<<1, 1>>>(foo);
16 ...
17 }
このコードは、以下の SYCL* コードに移行されます。
1 #define DATA_NUM 16
2
3 struct Foo {
4 ...
5 /*
6 DPCT1102:0: SYCL* デバイスコードでは、ゼロレングスの配列は許されていません。
7 */
8 float data[0];
9 };
10
11 void kernel(Foo *foo) {
12 foo->data[DATA_NUM - 1] = 123.f;
13 }
14
15 void test() {
16 sycl::device dev_ct1;
17 sycl::queue q_ct1(dev_ct1,
18 sycl::property_list{sycl::property::queue::in_order()});
19 Foo* foo;
20 foo =
21 (Foo *)sycl::malloc_device(sizeof(Foo) + DATA_NUM * sizeof(float), q_ct1);
22 q_ct1.parallel_for(
23 sycl::nd_range<3>(sycl::range<3>(1, 1, 1), sycl::range<3>(1, 1, 1)),
24 [=](sycl::nd_item<3> item_ct1) {
25 kernel(foo);
26 });
27 ...
28 }
このコードは次のように書き換えられます。
1 #define DATA_NUM 16
2
3 struct Foo {
4 ...
5 std::byte data[DATA_NUM];
6 };
7
8 void kernel(Foo *foo) {
9 foo->data[DATA_NUM - 1] = 123.f;
10 }
11
12 void test() {
13 sycl::device dev_ct1;
14 sycl::queue q_ct1(dev_ct1,
15 sycl::property_list{sycl::property::queue::in_order()});
16 Foo* foo;
17 foo = (Foo *)sycl::malloc_device(sizeof(Foo), q_ct1);
18 q_ct1.parallel_for(
19 sycl::nd_range<3>(sycl::range<3>(1, 1, 1), sycl::range<3>(1, 1, 1)),
20 [=](sycl::nd_item<3> item_ct1) {
21 kernel(foo);
22 });
23 ...
24 }