DPCT1102

メッセージ

SYCL* デバイスコードでは、ゼロレングスの配列は許されていません。

説明

ゼロレングスの配列は、C++ 標準または SYCL* 仕様ではサポートされません。一部のホスト・コンパイラーの実装ではサポートされる場合がありますが、デバイスコード向けのインテル® oneAPI DPC++/C++ コンパイラー (英語) ではサポートされていません。

修正方法の提案

ゼロより大きなレングスの配列を使用してください。

例えば、以下のオリジナル CUDA* コードについて考えてみます。


1 struct Foo { 
2   ...
3 float data[0]; 
4 }; 
5 
6 __global__ void kernel(Foo *foo) { 
7 Foo->data[DATA_NUM - 1] = 123.f; 
8 } 
9 
10 int main() { 
11    Foo* foo; 
12    cudaMalloc(&foo, sizeof(Foo) + DATA_NUM * sizeof(float)); 
13    kernel<<<1, 1>>>(foo); 
14       ... 
15 }

このコードは、以下の SYCL* コードに移行されます。


1 struct Foo { 
2    ... 
3 std::byte data[0]; 
4 }; 
5 
6 void kernel(Foo *foo) { 
7    foo->data[DATA_NUM - 1] = 123.f; 
8 } 
9 
10 int main() { 
11    dpct::device_ext &dev_ct1 = dpct::get_current_device(); 
12    sycl::queue &q_ct1 = dev_ct1.default_queue(); 
13    Foo* foo; 
14    foo = (Foo *)sycl::malloc_device(sizeof(Foo) + DATA_NUM * sizeof(float), q_ct1); 
15    q_ct1.parallel_for( 
16    sycl::nd_range<3>(sycl::range<3>(1, 1, 1), sycl::range<3>(1, 1, 1)), 
17              [=](sycl::nd_item<3> item_ct1) { 
18       kernel(foo); 
19    }); 
20       ... 
21 }

このコードは次のように書き換えられます。


1 struct Foo { 
2    ... 
3    std::byte data[DATA_NUM]; 
4 }; 
5 
6 void kernel(Foo *foo) { 
7 foo->data[DATA_NUM - 1] = 123.f; 
8 } 
9 
10 int main() { 
11    dpct::device_ext &dev_ct1 = dpct::get_current_device(); 
12    sycl::queue &q_ct1 = dev_ct1.default_queue(); 
13    Foo* foo; 
14    foo = (Foo *)sycl::malloc_device(sizeof(Foo), q_ct1); 
15    q_ct1.parallel_for( 
16    sycl::nd_range<3>(sycl::range<3>(1, 1, 1), sycl::range<3>(1, 1, 1)), 
17              [=](sycl::nd_item<3> item_ct1) { 
18       kernel(foo); 
19    }); 
20       ... 
21 }