[Bf-blender-cvs] [29a1cb5cc8e] cycles_oneapi: Cycles: Improve runtime error handling for oneAPI rendering
Nikita Sirgienko
noreply at git.blender.org
Fri Apr 1 01:05:27 CEST 2022
Commit: 29a1cb5cc8ea68d35fbb6556c3816d96fb14c6cf
Author: Nikita Sirgienko
Date: Thu Mar 31 23:15:41 2022 +0200
Branches: cycles_oneapi
https://developer.blender.org/rB29a1cb5cc8ea68d35fbb6556c3816d96fb14c6cf
Cycles: Improve runtime error handling for oneAPI rendering
Now errors in oneAPI kernel during initialization and pre-render
Cycles stages (like device-side memory allocaitons) will be properly
handled during oneAPI rendering and will be shown into UI.
===================================================================
M intern/cycles/device/oneapi/device_impl.cpp
M intern/cycles/device/oneapi/device_impl.h
M intern/cycles/device/oneapi/queue.cpp
M intern/cycles/device/oneapi/queue.h
M intern/cycles/kernel/device/oneapi/dll_interface_template.h
M intern/cycles/kernel/device/oneapi/kernel.cpp
===================================================================
diff --git a/intern/cycles/device/oneapi/device_impl.cpp b/intern/cycles/device/oneapi/device_impl.cpp
index c9dff23690b..99974f25405 100644
--- a/intern/cycles/device/oneapi/device_impl.cpp
+++ b/intern/cycles/device/oneapi/device_impl.cpp
@@ -12,6 +12,13 @@
CCL_NAMESPACE_BEGIN
+static void queue_error_cb(const char *message, void *user_ptr)
+{
+ if (user_ptr) {
+ *((std::string *)user_ptr) = message;
+ }
+}
+
OneapiDevice::OneapiDevice(const DeviceInfo &info,
oneAPIDLLInterface &oneapi_dll_object,
Stats &stats,
@@ -26,12 +33,14 @@ OneapiDevice::OneapiDevice(const DeviceInfo &info,
{
need_texture_info = false;
+ (oneapi_dll.oneapi_set_error_cb)(queue_error_cb, &oneapi_error_string);
+
// Oneapi calls should be initialised on this moment;
assert(oneapi_dll.oneapi_create_queue != nullptr);
bool is_finished_ok = (oneapi_dll.oneapi_create_queue)(device_queue, info.num);
if (is_finished_ok == false) {
- set_error("oneAPI queue initialization error: got runtime exception");
+ set_error("oneAPI queue initialization error: got runtime exception \"" + oneapi_error_string + "\"");
}
else {
VLOG(1) << "oneAPI queue has been successfully created for the device \"" << info.description
@@ -42,7 +51,7 @@ OneapiDevice::OneapiDevice(const DeviceInfo &info,
size_t globals_segment_size;
is_finished_ok = (oneapi_dll.oneapi_kernel_globals_size)(device_queue, globals_segment_size);
if (is_finished_ok == false) {
- set_error("oneAPI constant memory initialization got runtime exception");
+ set_error("oneAPI constant memory initialization got runtime exception \"" + oneapi_error_string + "\"");
}
else {
VLOG(1) << "Successfuly created global/constant memory segment (kernel globals object)";
@@ -94,7 +103,7 @@ bool OneapiDevice::load_kernels(const uint requested_features)
bool is_finished_ok = (oneapi_dll.oneapi_trigger_runtime_compilation)(device_queue);
if (is_finished_ok == false) {
- set_error("oneAPI kernel load: got runtime exception");
+ set_error("oneAPI kernel load: got runtime exception \"" + oneapi_error_string + "\"");
}
else {
VLOG(1) << "Runtime compilation done for \"" << info.description << "\"";
@@ -163,6 +172,11 @@ SyclQueue *OneapiDevice::sycl_queue()
return device_queue;
}
+string OneapiDevice::oneapi_error_message()
+{
+ return string(oneapi_error_string.c_str());
+}
+
oneAPIDLLInterface OneapiDevice::oneapi_dll_object()
{
return oneapi_dll;
@@ -248,7 +262,10 @@ void OneapiDevice::mem_copy_from(device_memory &mem, size_t y, size_t w, size_t
assert(mem.device_pointer);
char *shifted_host = (char *)mem.host_pointer + offset;
char *shifted_device = (char *)mem.device_pointer + offset;
- (oneapi_dll.oneapi_usm_memcpy)(device_queue, shifted_host, shifted_device, size);
+ bool is_finished_ok = (oneapi_dll.oneapi_usm_memcpy)(device_queue, shifted_host, shifted_device, size);
+ if (is_finished_ok == false) {
+ set_error("oneAPI memory operation error: got runtime exception \"" + oneapi_error_string + "\"");
+ }
}
}
@@ -268,7 +285,10 @@ void OneapiDevice::mem_zero(device_memory &mem)
}
assert(device_queue);
- (oneapi_dll.oneapi_usm_memset)(device_queue, (void *)mem.device_pointer, 0, mem.memory_size());
+ bool is_finished_ok = (oneapi_dll.oneapi_usm_memset)(device_queue, (void *)mem.device_pointer, 0, mem.memory_size());
+ if (is_finished_ok == false) {
+ set_error("oneAPI memory operation error: got runtime exception \"" + oneapi_error_string + "\"");
+ }
}
void OneapiDevice::mem_free(device_memory &mem)
diff --git a/intern/cycles/device/oneapi/device_impl.h b/intern/cycles/device/oneapi/device_impl.h
index 1176b85b40d..d18018bbc4c 100644
--- a/intern/cycles/device/oneapi/device_impl.h
+++ b/intern/cycles/device/oneapi/device_impl.h
@@ -26,6 +26,7 @@ class OneapiDevice : public Device {
void *kg_memory_device;
size_t kg_memory_size = (size_t)0;
oneAPIDLLInterface oneapi_dll;
+ std::string oneapi_error_string;
public:
virtual BVHLayoutMask get_bvh_layout_mask() const override;
@@ -56,6 +57,8 @@ class OneapiDevice : public Device {
SyclQueue *sycl_queue();
+ string oneapi_error_message();
+
oneAPIDLLInterface oneapi_dll_object();
void *kernel_globals_device_pointer();
diff --git a/intern/cycles/device/oneapi/queue.cpp b/intern/cycles/device/oneapi/queue.cpp
index 1ea8c0e69ae..8747796bb76 100644
--- a/intern/cycles/device/oneapi/queue.cpp
+++ b/intern/cycles/device/oneapi/queue.cpp
@@ -146,13 +146,6 @@ int OneapiDeviceQueue::num_concurrent_busy_states() const
}
}
-static void queue_error_cb(const char *message, void *user_ptr)
-{
- if (user_ptr) {
- *((std::string *)user_ptr) = message;
- }
-}
-
void OneapiDeviceQueue::init_execution()
{
oneapi_device->load_texture_info();
@@ -162,7 +155,6 @@ void OneapiDeviceQueue::init_execution()
assert(device_queue);
assert(kg_dptr);
kernel_context = new KernelContext{device_queue, kg_dptr, with_kernel_statistics};
- (oneapi_dll.oneapi_set_error_cb)(queue_error_cb, &kernel_error_string);
debug_init_execution();
}
@@ -268,7 +260,7 @@ bool OneapiDeviceQueue::enqueue(DeviceKernel kernel,
if (is_finished_ok == false) {
oneapi_device->set_error("oneAPI kernel \"" + std::string(device_kernel_as_string(kernel)) +
- "\" execution error: got runtime exception \"" + kernel_error_string +
+ "\" execution error: got runtime exception \"" + oneapi_device->oneapi_error_message() +
"\"");
}
@@ -281,7 +273,9 @@ bool OneapiDeviceQueue::synchronize()
return false;
}
- (oneapi_dll.oneapi_queue_synchronize)(oneapi_device->sycl_queue());
+ bool is_finished_ok = (oneapi_dll.oneapi_queue_synchronize)(oneapi_device->sycl_queue());
+ if (is_finished_ok == false)
+ oneapi_device->set_error("oneAPI unknown kernel execution error: got runtime exception \"" + oneapi_device->oneapi_error_message() + "\"");
debug_synchronize();
diff --git a/intern/cycles/device/oneapi/queue.h b/intern/cycles/device/oneapi/queue.h
index 1ca184c8cbe..7bf37382026 100644
--- a/intern/cycles/device/oneapi/queue.h
+++ b/intern/cycles/device/oneapi/queue.h
@@ -46,7 +46,6 @@ class OneapiDeviceQueue : public DeviceQueue {
OneapiDevice *oneapi_device;
oneAPIDLLInterface oneapi_dll;
KernelContext *kernel_context;
- std::string kernel_error_string;
static std::set<DeviceKernel> SUPPORTED_KERNELS;
bool with_kernel_statistics;
};
diff --git a/intern/cycles/kernel/device/oneapi/dll_interface_template.h b/intern/cycles/kernel/device/oneapi/dll_interface_template.h
index cf6facdb105..18d66407974 100644
--- a/intern/cycles/kernel/device/oneapi/dll_interface_template.h
+++ b/intern/cycles/kernel/device/oneapi/dll_interface_template.h
@@ -15,10 +15,10 @@ DLL_INTERFACE_CALL(oneapi_usm_alloc_device, void *, SyclQueue *queue, size_t mem
DLL_INTERFACE_CALL(oneapi_usm_free, void, SyclQueue *queue, void *usm_ptr)
DLL_INTERFACE_CALL(
- oneapi_usm_memcpy, void, SyclQueue *queue, void *dest, void *src, size_t num_bytes)
-DLL_INTERFACE_CALL(oneapi_queue_synchronize, void, SyclQueue *queue)
+ oneapi_usm_memcpy, bool, SyclQueue *queue, void *dest, void *src, size_t num_bytes)
+DLL_INTERFACE_CALL(oneapi_queue_synchronize, bool, SyclQueue *queue)
DLL_INTERFACE_CALL(oneapi_usm_memset,
- void,
+ bool,
SyclQueue *queue,
void *usm_ptr,
unsigned char value,
diff --git a/intern/cycles/kernel/device/oneapi/kernel.cpp b/intern/cycles/kernel/device/oneapi/kernel.cpp
index c0736e9b8b7..18ae5249646 100644
--- a/intern/cycles/kernel/device/oneapi/kernel.cpp
+++ b/intern/cycles/kernel/device/oneapi/kernel.cpp
@@ -116,34 +116,56 @@ void oneapi_usm_free(SyclQueue *queue_, void *usm_ptr)
sycl::free(usm_ptr, *queue);
}
-void oneapi_usm_memcpy(SyclQueue *queue_, void *dest, void *src, size_t num_bytes)
+bool oneapi_usm_memcpy(SyclQueue *queue_, void *dest, void *src, size_t num_bytes)
{
assert(queue_);
sycl::queue *queue = reinterpret_cast<sycl::queue *>(queue_);
check_usm(queue_, dest, true);
check_usm(queue_, src, true);
- queue->memcpy(dest, src, num_bytes);
+ try {
+ sycl::event mem_event = queue->memcpy(dest, src, num_bytes);
+ mem_event.wait_and_throw();
+ return true;
+ }
+ catch (sycl::exception const &e) {
+ if (s_error_cb) {
+ s_error_cb(e.what(), s_error_user_ptr);
+ }
+ return false;
+ }
}
-void oneapi_usm_memset(SyclQueue *queue_, void *usm_ptr, unsigned char value, size_t num_bytes)
+bool oneapi_usm_memset(SyclQueue *queue_, void *usm_ptr, unsigned char value, size_t num_bytes)
{
assert(queue_);
sycl::queue *queue = reinterpret_cast<sycl::queue *>(queue_);
check_usm(queue_, usm_ptr, true);
- queue->memset(usm_ptr, value, num_bytes);
+ try {
+ sycl::event mem_event = queue->memset(usm_ptr, value, num_bytes);
+ mem_event.wait_and_throw();
+ return true;
+ }
+ catch (sycl::exception const &e) {
+ if (s_error_cb) {
+ s_error_cb(e.what(), s_error_user_ptr);
+ }
+ return false;
+ }
}
-void oneapi_queue_synchronize(SyclQueue *queue_)
+bool oneapi_queue_synchronize(SyclQueue *queue_)
{
assert(queue_);
sycl::queue *queue = reinterpret_cast<sycl::queue *>(queue_);
try {
queue->wait_and_throw();
+ return true;
}
catch (sycl::exception const &e) {
if (s_error_cb) {
s_error_cb(e.what(), s_error_user_ptr);
}
+ return false;
}
}
More information about the Bf-blender-cvs
mailing list