DPCT1042

メッセージ

SYCL* カーネルに渡される引数のサイズが、カスタムではない SYCL* デバイスの最小サイズ制限 (1024) を超えています。ハードウェア引数のサイズの上限は、info::device::max_parameter_size で照会できます。引数のサイズがハードウェアの制限を超える場合、このコードを書き換える必要があります。

説明

カスタムでない SYCL* デバイスの SYCL* カーネルに渡される引数のサイズには制限があります (SYCL* 1.2.1 仕様の「4.6.4.2 デバイス情報記述子」を参照)。

この警告が出力された場合は、SYCL* カーネルのラムダでキャプチャーされるアクセサーやその他の引数の数を減らすように、コードを手動で調整する必要があります。

以下の例では、同じ型の 2 つのバッファーをマージすることで、1 つのアクセサーを削除しています。

以下に例を示します。この例では、SYCL* カーネルのラムダでキャプチャーされる引数がサイズ制限を超えています。

...// device0 から device30 の宣言 
 dpct::constant_memory<int, 1> device31(ARRAY_SIZE); 
 dpct::constant_memory<int, 1> device32(ARRAY_SIZE); 
 
 // カーネル関数の宣言 
 void kernel(sycl::nd_item<3> item_ct1, int *device0, ..., int *device32); 
 
 void test_function() { 
 ... // device0 から device30 の memcpy 操作 
 dpct::dpct_memcpy(device31.get_ptr(), host, ARRAY_SIZE * sizeof(int)); 
 dpct::dpct_memcpy(device32.get_ptr(), host, ARRAY_SIZE * sizeof(int)); 
 { 
 dpct::get_default_queue().submit([&](sycl::handler &cgh) { 
 ... // device0 から device30 のアクセサー、SYCL* カーネルのラムダで取得 
 auto device31_acc_ct1 = device31.get_access(cgh); 
 auto device32_acc_ct1 = device32.get_access(cgh); 
 cgh.parallel_for(sycl::nd_range<3>(sycl::range<3>(1), sycl::range<3>(1)), 
 [=](sycl::nd_item<3> item_ct1) { 
 kernel(item_ct1, ... /*device0 から device30 の引数*/ 
 device31_acc_ct1.get_pointer(), 
 device32_acc_ct1.get_pointer()); 
 }); 
 }); 
 } 
 }

次のようにコードを更新することで、同じ型の 2 つのバッファー (device31 と device32) を 1 つ (device31) に統合し、アクセサー数を 1 つ減らします。

... // device0 から device30 の宣言 
dpct::constant_memory<int, 1> device32(ARRAY_SIZE + ARRAY_SIZE); // ２つではなく 1 つのバッファー  

// カーネル関数の宣言 
void kernel(sycl::nd_item<3> item_ct1, int *device0, ..., int *device32); 

void test_function() { 
... // device0 から device30 の memcpy 操作 
dpct::dpct_memcpy(device31.get_ptr(), host, ARRAY_SIZE * sizeof(int)); 
// オフセットでメモリーコピー:  dpct::dpct_memcpy(device31.get_ptr() + ARRAY_SIZE, host, 
ARRAY_SIZE * sizeof(int)); 
{ 
dpct::get_default_queue().submit([&](sycl::handler &cgh) { 
... // device0 から device30 のアクセサー、SYCL* カーネルのラムダで取得 
auto device31_acc_ct1 = 
device31.get_access(cgh); // ２つではなく 1 つのアクセサー 
cgh.parallel_for(sycl::nd_range<3>(sycl::range<3>(1), sycl::range<3>(1)), 
[=](sycl::nd_item<3> item_ct1) { 
// 最後のパラメーターは、オフセットと前のパラメーター 
// と同じアクセサーを使用するように変更されています 
kernel(item_ct1, ... /* device0 から device30 の引数 */ 
device31_acc_ct1.get_pointer(), 
device31_acc_ct1.get_pointer() + ARRAY_SIZE); 
}); 
}); 
} 
}

修正方法の提案

コードを確認して、調整してください。