triSYCL throws non_cl_error, when tricycle::device::~device is called

Question

I'm trying to run a parallel for loop with triSYCL. This is my code:

#define TRISYCL_OPENCL
#define OMP_NUM_THREADS 8
#define BOOST_COMPUTE_USE_CPP11

//standart libraries
#include <iostream>
#include <functional>

//deps
#include "CL/sycl.hpp"

struct Color
{
    float r, g, b, a;

    friend std::ostream& operator<<(std::ostream& os, const Color& c)
    {
        os << "(" << c.r << ", " << c.g << ", " << c.b << ", " << c.a << ")";
        return os;
    }
};

struct Vertex
{
    float x, y;
    Color color;

    friend std::ostream& operator<<(std::ostream& os, const Vertex& v)
    {
        os << "x: " << v.x << ", y: " << v.y << ", color: " << v.color;
        return os;
    }
};

template<typename T>
T mapNumber(T x, T a, T b, T c, T d)
{
    return (x - a) / (b - a) * (d - c) + c;
}

int windowWidth = 640;
int windowHeight = 720;

int main()
{
    auto exception_handler = [](cl::sycl::exception_list exceptions) {
        for (std::exception_ptr const& e : exceptions)
        {
            try
            {
                std::rethrow_exception(e);
            } catch (cl::sycl::exception const& e)
            {
                std::cout << "Caught asynchronous SYCL exception: " << e.what() << std::endl;
            }
        }
    };

    cl::sycl::default_selector defaultSelector;
    cl::sycl::context context(defaultSelector, exception_handler);
    cl::sycl::queue queue(context, defaultSelector, exception_handler);

        auto* pixelColors = new Color[windowWidth * windowHeight];
        {
            cl::sycl::buffer<Color, 2> color_buffer(pixelColors, cl::sycl::range < 2 > {(unsigned long) windowWidth,
                                                                                        (unsigned long) windowHeight});

            cl::sycl::buffer<int, 1> b_windowWidth(&windowWidth, cl::sycl::range < 1 > {1});
            cl::sycl::buffer<int, 1> b_windowHeight(&windowHeight, cl::sycl::range < 1 > {1});

            queue.submit([&](cl::sycl::handler& cgh) {
                auto color_buffer_acc = color_buffer.get_access<cl::sycl::access::mode::write>(cgh);
                auto width_buffer_acc = b_windowWidth.get_access<cl::sycl::access::mode::read>(cgh);
                auto height_buffer_acc = b_windowHeight.get_access<cl::sycl::access::mode::read>(cgh);

                cgh.parallel_for<class init_pixelColors>(
                        cl::sycl::range<2>((unsigned long) width_buffer_acc[0], (unsigned long) height_buffer_acc[0]),
                        [=](cl::sycl::id<2> index) {
                            color_buffer_acc[index[0]][index[1]] = {
                                    mapNumber<float>(index[0], 0.f, width_buffer_acc[0], 0.f, 1.f),
                                    mapNumber<float>(index[1], 0.f, height_buffer_acc[0], 0.f, 1.f),
                                    0.f,
                                    1.f};
                        });
            });

            std::cout << "cl::sycl::queue check - selected device: "
                      << queue.get_device().get_info<cl::sycl::info::device::name>() << std::endl;
        }//here the error appears 

        delete[] pixelColors;
    return 0;
}

I'm building it with this CMakeLists.txt file:

cmake_minimum_required(VERSION 3.16.2)

project(acMandelbrotSet_stackoverflow)

set(CMAKE_CXX_STANDARD 17)

set(SRC_FILES
        path/to/main.cpp
        )
find_package(OpenCL REQUIRED)

set(Boost_INCLUDE_DIR path/to/boost)
include_directories(${Boost_INCLUDE_DIR})

include_directories(path/to/SYCL/include)

set(LIBS PRIVATE ${Boost_LIBRARIES} OpenCL::OpenCL)


add_executable(${PROJECT_NAME} ${SRC_FILES})

set_target_properties(${PROJECT_NAME} PROPERTIES DEBUG_POSTFIX _d)
target_link_libraries(${PROJECT_NAME} ${LIBS})

When I try to run it, I get this message: libc++abi.dylib: terminating with uncaught exception of type trisycl::non_cl_error from path/to/SYCL/include/triSYCL/command_group/detail/task.hpp line: 278 function: trisycl::detail::task::get_kernel, the message was: "Cannot use an OpenCL kernel in this context".

I've tried to create a lambda of mapNumber in the kernel but that didn't make any difference. I've also tried to use this before the end of the scope to catch errors:

try
{
    queue.wait_and_throw();
} catch (cl::sycl::exception const& e)
{
    std::cout << "Caught synchronous SYCL exception: " << e.what() << std::endl;
}

but nothing was printed to the console except the error from before. And I've also tried to make an event of the queue.submit call and then call event.wait() before the end of the scope but again the exact same output.

Does any body have an idea what else I could try?

Answer 1

The problem is that triSYCL is a research project looking deeper at some aspects of SYCL while not providing a global generic SYCL support for an end-user. I have just clarified this on the README of the project. :-( Probably the problem here is that the OpenCL SPIR kernel has not been generated. So you need to first compile the specific (old) Clang & LLVM from triSYCL https://github.com/triSYCL/triSYCL/blob/master/doc/architecture.rst#trisycl-architecture-for-accelerator. But unfortunately there is no simple Clang driver to use all the specific Clang & LLVM to generate the kernels from the SYCL source. Right know it is done with some ad-hoc awful Makefiles (look around https://github.com/triSYCL/triSYCL/blob/master/tests/Makefile#L360) and, even if you can survive to this, you might encounter some bugs...

The good news is now there are several other implementations of SYCL which are quite easier to use, quite more complete and quite less buggy! :-) Look at ComputeCpp, DPC++ and hipSYCL for example.