DPCT1083
目次
DPCT1083#
メッセージ#
移行したコードの <placeholder> サイズは、オリジナルコードとは異なる場合があります。移行したコードの割り当てメモリーサイズが正しいことを確認してください。
詳細な説明#
一部の型は、移行後のコードと移行前のコードではサイズが異なります (例: sycl::float3
と float3
)。移行したコードで、メモリーの割り当てサイズが正しいかどうかを確認します。
以下の例では、オリジナルコードの float3
を表すため 3*sizeof(float)
を使用しています。移行したコードでは、sycl::float3
のサイズが異なるため、割り当てサイズの調整が必要です。
例えば、以下のオリジナル CUDA* コードについて考えてみます。
1 __global__ void kernel() {
2 extern __shared__ float3 shared_memory[];
3 ...
4 }
5
6 void foo() {
7 size_t shared_size = 3 * sizeof(float);
8 kernel<<<1, 1, shared_size>>>();
9 }
このコードは、以下の SYCL* コードに移行されます。
1 void kernel(uint8_t *dpct_local) {
2 auto shared_memory = (sycl::float3 *)dpct_local;
3 ...
4 }
5
6 void foo() {
7 /*
8 DPCT1083:0: The size of local memory in the migrated code may be different
9 from the original code.Check that the allocated memory size in the migrated
10 code is correct.
11 */
12 size_t shared_size = 3 * sizeof(float);
13 dpct::get_in_order_queue().submit([&](sycl::handler &cgh) {
14 sycl::local_accessor<uint8_t, 1> dpct_local_acc_ct1(
15 sycl::range<1>(shared_size), cgh);
16
17 cgh.parallel_for(
18 sycl::nd_range<3>(sycl::range<3>(1, 1, 1), sycl::range<3>(1, 1, 1)),
19 [=](sycl::nd_item<3> item_ct1) {
20 kernel(dpct_local_acc_ct1.get_pointer());
21 });
22 });
23 }
このコードを以下のように手動で調整します。
1 void kernel(uint8_t *dpct_local) {
2 auto shared_memory = (sycl::float3 *)dpct_local;
3 ...
4 }
5
6 void foo() {
7 size_t shared_size = 1 * sizeof(sycl::float3);
8 dpct::get_in_order_queue().submit([&](sycl::handler &cgh) {
9 sycl::local_accessor<uint8_t, 1> dpct_local_acc_ct1(
10 sycl::range<1>(shared_size), cgh);
11
12 cgh.parallel_for(
13 sycl::nd_range<3>(sycl::range<3>(1, 1, 1), sycl::range<3>(1, 1, 1)),
14 [=](sycl::nd_item<3> item_ct1) {
15 kernel(dpct_local_acc_ct1.get_pointer());
16 });
17 });
18 }
修正方法の提案
メモリーの割り当てサイズを確認し、必要に応じて正しいサイズに置き換えます。