DPCT1059
目次
DPCT1059#
メッセージ#
SYCL* は、4 チャネルのイメージ形式のみをサポートしています。コードを調整します。
詳細な説明#
SYCL* は、4 チャネルのイメージ形式のみをサポートしています。この警告は、ツールがオリジナルコードに対応するサポートされていないイメージ形式のコードを生成すると出力されます。移行したコードでイメージ形式を変更できます。この回避策は、コードのパフォーマンスに影響する可能性があります。
例えば、以下のオリジナル CUDA* コードについて考えてみます。
1 static texture<uint2, 2> tex;
2
3 __global__ void test_image() {
4 uint2 tex_data = tex2D(tex, 0, 0);
5 }
6 void foo(uint2 *image_data) {
7 ...
8 test_image<<<1, 1>>>();
9 }
このコードは、以下の SYCL* コードに移行されます。
1 /*
2 DPCT1059:0: SYCL only supports 4-channel image format. Adjust the code.
3 */
4 dpct::image_wrapper<sycl::uint2, 2> tex;
5
6 void test_image(dpct::image_accessor_ext<sycl::uint2, 2> tex) {
7 sycl::uint2 tex_data = tex.read(0, 0);
8 }
9 void foo(sycl::uint2 *image_data) {
10 ...
11 dpct::get_default_queue().submit([&](sycl::handler &cgh) {
12 auto tex_acc = tex.get_access(cgh);
13
14 auto tex_smpl = tex.get_sampler();
15
16 cgh.parallel_for(
17 sycl::nd_range<3>(sycl::range<3>(1, 1, 1), sycl::range<3>(1, 1, 1)),
18 [=](sycl::nd_item<3> item_ct1) {
19 test_image(
20 dpct::image_accessor_ext<sycl::uint2, 2>(tex_smpl, tex_acc));
21 });
22 });
23 }
このコードは次のように書き換えられます。
1 dpct::image_wrapper<sycl::uint4, 2> tex;
2
3 void test_image(dpct::image_accessor_ext<sycl::uint4, 2> tex) {
4 sycl::uint4 tex_data = tex.read(0, 0);
5 }
6 void foo(sycl::uint2 *image_data) {
7 // (1) Please allocate sycl::uint4 type new_image_data and initialize the value with image_data.
8 // (2) Bind new_image_data to the tex object instead of the image_data.
9 // (3) Free the memory of new_image_data after usage.
10 ...
11 dpct::get_default_queue().submit([&](sycl::handler &cgh) {
12 auto tex_acc = tex.get_access(cgh);
13
14 auto tex_smpl = tex.get_sampler();
15
16 cgh.parallel_for(
17 sycl::nd_range<3>(sycl::range<3>(1, 1, 1), sycl::range<3>(1, 1, 1)),
18 [=](sycl::nd_item<3> item_ct1) {
19 test_image(
20 dpct::image_accessor_ext<sycl::uint4, 2>(tex_smpl, tex_acc));
21 });
22 });
23 }
修正方法の提案
コードを修正する必要があります。