Core dump with writable 2d image accessor

[Ouzounoudis__George]

Hello,

When running the following code i get core dumps.
It fails with either

• a cpu selector: Intel(R) Core(TM) i7-6700K CPU @ 4.00GHz
• or gpu selector: Intel(R) Gen9 HD Graphics NEO

int main()
{
    const int nx = 1024;
    const int ny = 1024;

    sycl::default_selector device_selector;

    sycl::queue queue(device_selector);
    {
        const auto &device = queue.get_device();
        std::cout << "Running on " << device.get_info<sycl::info::device::name>() << std::endl;
    }

    sycl::image<2> img(sycl::image_channel_order::rgba, sycl::image_channel_type::unorm_int8, sycl::range<2>(nx, ny));

    queue.submit([&](sycl::handler &cgh) {
        auto img_acc = img.get_access<sycl::float4, sycl::access::mode::write>(cgh);
        sycl::nd_range<2> nd_range(sycl::range<2>(nx, ny), sycl::range<2>(16, 16));
        cgh.parallel_for<class sample>(nd_range, [=](sycl::nd_item<2> i) {
            sycl::id<2> index = i.get_global_id();
            sycl::float4 color (0.5f, 0.5f, 0.2f, 1.0f);
            img_acc.write(sycl::int2(index[0], index[1]), color);
        });
    });
    return 0;
}

I would appreciate any insight on this. Thanks

I run this on:
OS: Ubuntu 19.10
CPU: Intel core i7 6700K
GPU: Nvidia RTX 2070
package: intel-basekit/all,now 2021.1-85.beta03

[Jie_L_Intel]

i can think of two issues in your code. 

1. the "sycl::image_channel_order::rgba" and "sycl::image_channel_type::unorm_int8" is not compatible, use fp16, fp32 or int32 or change channel_order to be "argb". 

2. it is SYCL spec. bug that no query API for image format defined. So your provided image format is neither supported by Gen9 nor i7 CPU. And current Beta03 does not support Nvidia GPU.

So you have to convert your image data first to those ordinary format that supported by Intel devices, and compute on that image format, and then convert them back, if needed.

[RahulV_intel]

Hi George,

Kindly let us know if the solution provided above has worked for you.

[Ouzounoudis__George]

Hello!
 

I tried to change from unorm_int8 format to fp32, fp16 or unsigned_int32 and i still get crashes on the provided piece of code.
Also changing to argb channel ordering with unorm_int8 or fp32 format did crash also:

Running on Intel(R) Core(TM) i7-6700K CPU @ 4.00GHz
terminate called after throwing an instance of 'cl::sycl::runtime_error'
  what():  OpenCL API failed. OpenCL API returns: -50 (CL_INVALID_ARG_VALUE) -50 (CL_INVALID_ARG_VALUE)
Aborted (core dumped)

Running on Intel(R) Gen9 HD Graphics NEO
terminate called after throwing an instance of 'cl::sycl::runtime_error'
  what():  OpenCL API failed. OpenCL API returns: -50 (CL_INVALID_ARG_VALUE) -50 (CL_INVALID_ARG_VALUE)
Aborted (core dumped)

EDIT: it seems that it crashes on the destructor of the sycl::image

[Jie_L_Intel]

current Beta03 DPC++ runtime looks like not able to construct an image without a host pointer initialization. To work-around, add a "set_final_data()" explicitly after the image constructor. i tried with below code of rgba and fp32, no OpenCL error reported.

#include <CL/sycl.hpp>
using namespace cl;

int main() {

  int nx = 1024;
  int ny = 1024;

  sycl::image<2> img(sycl::image_channel_order::rgba, sycl::image_channel_type::fp32, sycl::range<2>(nx, ny));
  img.set_final_data();

  sycl::queue myQueue;

    myQueue.submit([&](sycl::handler &cgh) {

        auto img_acc = img.get_access<sycl::float4, sycl::access::mode::write>(cgh);

        sycl::nd_range<2> nd_range(sycl::range<2>(nx, ny), sycl::range<2>(16, 16));

        cgh.parallel_for<class sample>(nd_range, [=](sycl::nd_item<2> i) {

            sycl::id<2> index = i.get_global_id();

            sycl::float4 color (0.5f, 0.5f, 0.2f, 1.0f);

            img_acc.write(sycl::int2(index[0], index[1]), color);

        });

    });

  return 0;
}

[Ouzounoudis__George]

Thanks, it now seems to work for unorm_int8 and rgba.

[RahulV_intel]

George,

Glad to know that your issue got resolved. We are closing this thread.