DPCT1102

メッセージ

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

説明

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

修正方法の提案

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

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

 struct Foo { 
... float data[0]; 
 }; 
 
 __global__ void kernel(Foo *foo) { 
 Foo->data[DATA_NUM - 1] = 123.f; 
 } 
 
 int main() { 
 Foo* foo; 
 cudaMalloc(&foo, sizeof(Foo) + DATA_NUM * sizeof(float)); 
 kernel<<<1, 1>>>(foo); 
 ...  }

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

struct Foo { 
...  std::byte data[0]; 
}; 

void kernel(Foo *foo) { 
foo->data[DATA_NUM - 1] = 123.f; 
} 

int main() { 
dpct::device_ext &dev_ct1 = dpct::get_current_device(); 
sycl::queue &q_ct1 = dev_ct1.default_queue(); 
Foo* foo; 
foo = (Foo *)sycl::malloc_device(sizeof(Foo) + DATA_NUM * sizeof(float), q_ct1); 
q_ct1.parallel_for( 
sycl::nd_range<3>(sycl::range<3>(1, 1, 1), sycl::range<3>(1, 1, 1)), 
[=](sycl::nd_item<3> item_ct1) { 
kernel(foo); 
}); 
...  }

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

struct Foo { 
...  std::byte data[DATA_NUM]; 
}; 

void kernel(Foo *foo) { 
foo->data[DATA_NUM - 1] = 123.f; 
} 

int main() { 
dpct::device_ext &dev_ct1 = dpct::get_current_device(); 
sycl::queue &q_ct1 = dev_ct1.default_queue(); 
Foo* foo; 
foo = (Foo *)sycl::malloc_device(sizeof(Foo), q_ct1); 
q_ct1.parallel_for( 
sycl::nd_range<3>(sycl::range<3>(1, 1, 1), sycl::range<3>(1, 1, 1)), 
[=](sycl::nd_item<3> item_ct1) { 
kernel(foo); 
}); 
...  }