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: Zero-length arrays are not permitted in SYCL device code. 
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  }