Ensure thread safety

Fiber scheduling is now thread local, and truly thread-shared data structures are protected by a global lock.

include/simsycl/detail/check.hh

@@ -14,6 +14,7 @@ void check(
     bool condition, const char *cond_string, std::source_location location, int default_mode, const char *message, ...);
 
 struct override_check_mode {
+    // effect is thread-local
     override_check_mode(int mode);
     ~override_check_mode();
 };

include/simsycl/detail/lock.hh

@@ -0,0 +1,44 @@
+#pragma once
+
+#include <cassert>
+#include <mutex>
+#include <utility>
+
+
+namespace simsycl::detail {
+
+class system_lock {
+  public:
+    system_lock();
+
+  private:
+    template<typename F>
+    friend decltype(auto) with_system_lock(F &&f);
+
+    std::lock_guard<std::recursive_mutex> m_lock;
+};
+
+template<typename T>
+class shared_value {
+  public:
+    shared_value() : m_value() {}
+    shared_value(const T &v) : m_value(v) {}
+    shared_value(T &&v) : m_value(std::move(v)) {}
+
+    template<typename... Params>
+    shared_value(const std::in_place_t /* tag */, Params &&...params) : m_value(std::forward<Params>(params)...) {}
+
+    shared_value(const shared_value &) = delete;
+    shared_value(shared_value &&) = delete;
+    shared_value &operator=(const shared_value &) = delete;
+    shared_value &operator=(shared_value &&) = delete;
+
+    ~shared_value() = default;
+
+    T &with(system_lock & /* lock */) { return m_value; }
+
+  private:
+    T m_value;
+};
+
+} // namespace simsycl::detail

include/simsycl/detail/reference_type.hh

@@ -45,7 +45,7 @@ class reference_type {
         static_assert(std::is_base_of_v<reference_type, Derived>);
     }
 
-    state_type &state() const {
+    const state_type &state() const {
         SIMSYCL_CHECK(m_state != nullptr);
         return *m_state;
     }

include/simsycl/schedule.hh

@@ -40,7 +40,8 @@ class shuffle_schedule final : public cooperative_schedule {
     uint64_t m_seed = 1234567890;
 };
 
+// effect is thread-local
 const cooperative_schedule &get_cooperative_schedule();
-void set_cooperative_schedule(std::unique_ptr<cooperative_schedule> schedule);
+void set_cooperative_schedule(std::shared_ptr<const cooperative_schedule> schedule);
 
 } // namespace simsycl

include/simsycl/sycl/accessor.hh

@@ -148,7 +148,7 @@ class host_access_guard {
     template<typename T, typename AllocatorT>
     explicit host_access_guard(
         const sycl::buffer<T, Dimensions, AllocatorT> &buf, const accessed_range<Dimensions> &range)
-        : m_validator(&detail::get_buffer_access_validator(buf)), m_range(range) //
+        : m_validator(&detail::get_buffer_access_validator(buf, m_lock)), m_range(range) //
     {
         m_validator->begin_host_access(m_range);
     }
@@ -161,10 +161,27 @@ class host_access_guard {
     ~host_access_guard() { m_validator->end_host_access(m_range); }
 
   private:
+    system_lock m_lock;
     buffer_access_validator<Dimensions> *m_validator;
     accessed_range<Dimensions> m_range;
 };
 
+template<int Dimensions>
+class command_group_access_guard {
+  public:
+    template<typename T, typename AllocatorT>
+    explicit command_group_access_guard(const sycl::buffer<T, Dimensions, AllocatorT> &buf)
+        : m_validator(&detail::get_buffer_access_validator(buf, m_lock)) {}
+
+    void check_access_from_command_group(const accessed_range<Dimensions> &range) {
+        m_validator->check_access_from_command_group(range);
+    }
+
+  private:
+    system_lock m_lock;
+    buffer_access_validator<Dimensions> *m_validator;
+};
+
 } // namespace simsycl::detail
 
 namespace simsycl::sycl {
@@ -382,7 +399,7 @@ class accessor : public simsycl::detail::property_interface {
     } constexpr inline static internal{};
 
     DataT *m_buffer = nullptr;
-    detail::buffer_access_validator<Dimensions> *m_access_validator = nullptr;
+    std::shared_ptr<detail::command_group_access_guard<Dimensions>> m_guard; // shared_ptr: accessors must be copyable
     range<Dimensions> m_buffer_range;
     id<Dimensions> m_access_offset;
     range<Dimensions> m_access_range;
@@ -392,9 +409,9 @@ class accessor : public simsycl::detail::property_interface {
     template<typename AllocatorT>
     void init(buffer<DataT, Dimensions, AllocatorT> &buffer_ref) {
         m_buffer = detail::get_buffer_data(buffer_ref);
+        m_guard = std::make_shared<detail::command_group_access_guard<Dimensions>>(buffer_ref);
         m_buffer_range = buffer_ref.get_range();
         m_access_range = m_buffer_range;
-        m_access_validator = &detail::get_buffer_access_validator(buffer_ref);
     }
     void init(const id<Dimensions> &access_offset) { m_access_offset = access_offset; }
 
@@ -416,8 +433,8 @@ class accessor : public simsycl::detail::property_interface {
 
     void require() {
         SIMSYCL_CHECK(m_buffer != nullptr);
-        SIMSYCL_CHECK(m_access_validator != nullptr);
-        m_access_validator->check_access_from_command_group({m_access_offset, m_access_range, AccessMode});
+        SIMSYCL_CHECK(m_guard != nullptr);
+        m_guard->check_access_from_command_group({m_access_offset, m_access_range, AccessMode});
         *m_required = true;
     }
 
@@ -429,8 +446,8 @@ accessor(buffer<DataT, Dimensions, AllocatorT> &, detail::accessor_tag<AccessMod
     -> accessor<DataT, Dimensions, AccessMode, AccessTarget, access::placeholder::false_t>;
 
 template<typename DataT, int Dimensions, typename AllocatorT, access_mode AccessMode, target AccessTarget>
-accessor(buffer<DataT, Dimensions, AllocatorT> &, detail::accessor_tag<AccessMode, AccessTarget>, const property_list &)
-    -> accessor<DataT, Dimensions, AccessMode, AccessTarget, access::placeholder::false_t>;
+accessor(buffer<DataT, Dimensions, AllocatorT> &, detail::accessor_tag<AccessMode, AccessTarget>,
+    const property_list &) -> accessor<DataT, Dimensions, AccessMode, AccessTarget, access::placeholder::false_t>;
 
 template<typename DataT, int Dimensions, typename AllocatorT, access_mode AccessMode, target AccessTarget>
 accessor(buffer<DataT, Dimensions, AllocatorT> &, handler &, detail::accessor_tag<AccessMode, AccessTarget>)
@@ -455,8 +472,8 @@ accessor(buffer<DataT, Dimensions, AllocatorT> &, handler &, range<Dimensions>,
 
 template<typename DataT, int Dimensions, typename AllocatorT, access_mode AccessMode, target AccessTarget>
 accessor(buffer<DataT, Dimensions, AllocatorT> &, handler &, range<Dimensions>,
-    detail::accessor_tag<AccessMode, AccessTarget>, const property_list &)
-    -> accessor<DataT, Dimensions, AccessMode, AccessTarget, access::placeholder::false_t>;
+    detail::accessor_tag<AccessMode, AccessTarget>,
+    const property_list &) -> accessor<DataT, Dimensions, AccessMode, AccessTarget, access::placeholder::false_t>;
 
 template<typename DataT, int Dimensions, typename AllocatorT, access_mode AccessMode, target AccessTarget>
 accessor(buffer<DataT, Dimensions, AllocatorT> &, range<Dimensions>, id<Dimensions>,
@@ -465,8 +482,8 @@ accessor(buffer<DataT, Dimensions, AllocatorT> &, range<Dimensions>, id<Dimensio
 
 template<typename DataT, int Dimensions, typename AllocatorT, access_mode AccessMode, target AccessTarget>
 accessor(buffer<DataT, Dimensions, AllocatorT> &, range<Dimensions>, id<Dimensions>,
-    detail::accessor_tag<AccessMode, AccessTarget>, const property_list &)
-    -> accessor<DataT, Dimensions, AccessMode, AccessTarget, access::placeholder::false_t>;
+    detail::accessor_tag<AccessMode, AccessTarget>,
+    const property_list &) -> accessor<DataT, Dimensions, AccessMode, AccessTarget, access::placeholder::false_t>;
 
 template<typename DataT, int Dimensions, typename AllocatorT, access_mode AccessMode, target AccessTarget>
 accessor(buffer<DataT, Dimensions, AllocatorT> &, handler &, range<Dimensions>, id<Dimensions>,
@@ -475,8 +492,8 @@ accessor(buffer<DataT, Dimensions, AllocatorT> &, handler &, range<Dimensions>,
 
 template<typename DataT, int Dimensions, typename AllocatorT, access_mode AccessMode, target AccessTarget>
 accessor(buffer<DataT, Dimensions, AllocatorT> &, handler &, range<Dimensions>, id<Dimensions>,
-    detail::accessor_tag<AccessMode, AccessTarget>, const property_list &)
-    -> accessor<DataT, Dimensions, AccessMode, AccessTarget, access::placeholder::false_t>;
+    detail::accessor_tag<AccessMode, AccessTarget>,
+    const property_list &) -> accessor<DataT, Dimensions, AccessMode, AccessTarget, access::placeholder::false_t>;
 
 
 template<typename DataT, access_mode AccessMode, target AccessTarget, access::placeholder IsPlaceholder>
@@ -509,7 +526,7 @@ class accessor<DataT, 0, AccessMode, AccessTarget, IsPlaceholder> : public simsy
     accessor(buffer<DataT, 1, AllocatorT> &buffer_ref, const property_list &prop_list = {})
         : simsycl::detail::property_interface(prop_list, property_compatibility()),
           m_buffer(detail::get_buffer_data(buffer_ref)),
-          m_access_validator(&detail::get_buffer_access_validator(buffer_ref)) {}
+          m_guard(std::make_shared<detail::command_group_access_guard<1>>(buffer_ref)) {}
 
     template<typename AllocatorT>
     accessor(buffer<DataT, 1, AllocatorT> &buffer_ref, handler &command_group_handler_ref,
@@ -615,14 +632,14 @@ class accessor<DataT, 0, AccessMode, AccessTarget, IsPlaceholder> : public simsy
     friend struct std::hash;
 
     DataT *m_buffer = nullptr;
-    detail::buffer_access_validator<1> *m_access_validator = nullptr;
+    std::shared_ptr<detail::command_group_access_guard<1>> m_guard; // shared_ptr: accessors must be copyable
     // shared: require() on a copy is equivalent to require() on the original instance
     std::shared_ptr<bool> m_required = std::make_shared<bool>(false);
 
     void require() {
         SIMSYCL_CHECK(m_buffer != nullptr);
-        SIMSYCL_CHECK(m_access_validator != nullptr);
-        m_access_validator->check_access_from_command_group({0, 1, AccessMode});
+        SIMSYCL_CHECK(m_guard != nullptr);
+        m_guard->check_access_from_command_group({0, 1, AccessMode});
         *m_required = true;
     }
 };
@@ -972,8 +989,8 @@ class host_accessor : public simsycl::detail::property_interface {
 
 template<typename DataT, int Dimensions, typename AllocatorT, access_mode AccessMode>
 host_accessor(buffer<DataT, Dimensions, AllocatorT> &buffer_ref, range<Dimensions> access_range,
-    detail::accessor_tag<AccessMode, target::device> tag, const property_list &prop_list = {})
-    -> host_accessor<DataT, Dimensions, AccessMode>;
+    detail::accessor_tag<AccessMode, target::device> tag,
+    const property_list &prop_list = {}) -> host_accessor<DataT, Dimensions, AccessMode>;
 
 template<typename DataT, int Dimensions, typename AllocatorT, access_mode AccessMode>
 host_accessor(buffer<DataT, Dimensions, AllocatorT> &buffer_ref, range<Dimensions> access_range,
@@ -1173,7 +1190,7 @@ class accessor<DataT, Dimensions, AccessMode, target::constant_buffer, IsPlaceho
     } constexpr inline static internal{};
 
     DataT *m_buffer = nullptr;
-    detail::buffer_access_validator<Dimensions> *m_access_validator = nullptr;
+    std::shared_ptr<detail::command_group_access_guard<1>> m_guard; // shared_ptr: accessors must be copyable
     range<Dimensions> m_buffer_range;
     id<Dimensions> m_access_offset;
     range<Dimensions> m_access_range;
@@ -1183,9 +1200,9 @@ class accessor<DataT, Dimensions, AccessMode, target::constant_buffer, IsPlaceho
     template<typename AllocatorT>
     void init(buffer<DataT, Dimensions, AllocatorT> &buffer_ref) {
         m_buffer = detail::get_buffer_data(buffer_ref);
+        m_guard = std::make_shared<detail::command_group_access_guard<Dimensions>>(buffer_ref);
         m_buffer_range = buffer_ref.get_range();
         m_access_range = m_buffer_range;
-        m_access_validator = &detail::get_buffer_access_validator(buffer_ref);
     }
     void init(const id<Dimensions> &access_offset) { m_access_offset = access_offset; }
 
@@ -1205,8 +1222,8 @@ class accessor<DataT, Dimensions, AccessMode, target::constant_buffer, IsPlaceho
 
     void require() {
         SIMSYCL_CHECK(m_buffer != nullptr);
-        SIMSYCL_CHECK(m_access_validator != nullptr);
-        m_access_validator->check_access_from_command_group({m_access_offset, m_access_range, AccessMode});
+        SIMSYCL_CHECK(m_guard != nullptr);
+        m_guard->check_access_from_command_group({m_access_offset, m_access_range, AccessMode});
         *m_required = true;
     }
 
@@ -1231,7 +1248,7 @@ class accessor<DataT, 0, AccessMode, target::constant_buffer, IsPlaceholder> fin
     accessor(buffer<DataT, 1, AllocatorT> &buffer_ref, const property_list &prop_list = {})
         : simsycl::detail::property_interface(prop_list, property_compatibility()),
           m_buffer(detail::get_buffer_data(buffer_ref)),
-          m_access_validator(&detail::get_buffer_access_validator(buffer_ref)) {}
+          m_guard(std::make_shared<detail::command_group_access_guard<1>>(buffer_ref)) {}
 
     template<typename AllocatorT>
     accessor(buffer<DataT, 1, AllocatorT> &buffer_ref, handler &command_group_handler_ref,
@@ -1268,14 +1285,14 @@ class accessor<DataT, 0, AccessMode, target::constant_buffer, IsPlaceholder> fin
     friend struct std::hash;
 
     DataT *m_buffer = nullptr;
-    detail::buffer_access_validator<1> *m_access_validator = nullptr;
+    std::shared_ptr<detail::command_group_access_guard<1>> m_guard;
     // shared: require() on a copy is equivalent to require() on the original instance
     std::shared_ptr<bool> m_required = std::make_shared<bool>(false);
 
     void require() {
         SIMSYCL_CHECK(m_buffer != nullptr);
-        SIMSYCL_CHECK(m_access_validator != nullptr);
-        m_access_validator->check_access_from_command_group({0, 1, AccessMode});
+        SIMSYCL_CHECK(m_guard != nullptr);
+        m_guard->check_access_from_command_group({0, 1, AccessMode});
         *m_required = true;
     }
 };
@@ -1313,6 +1330,8 @@ class accessor<DataT, Dimensions, AccessMode, target::host_buffer, IsPlaceholder
           m_access_guard(std::make_shared<detail::host_access_guard<Dimensions>>(
               buffer_ref, detail::accessed_range<Dimensions>(m_access_offset, m_access_range, AccessMode))) {}
 
+    // non-copyable and immovable, because it holds a system_lock
+
     bool is_placeholder() const { return false; }
 
     size_t get_size() const { return get_count() * sizeof(DataT); }
@@ -1371,6 +1390,8 @@ class accessor<DataT, 0, AccessMode, target::host_buffer, IsPlaceholder> : publi
                                                              buffer_ref, detail::accessed_range<1>(0, 1, AccessMode))) {
     }
 
+    // non-copyable and immovable, because it holds a system_lock
+
     bool is_placeholder() const { return false; }
 
     size_t get_size() const { return sizeof(DataT); }
@@ -1390,6 +1411,7 @@ class accessor<DataT, 0, AccessMode, target::host_buffer, IsPlaceholder> : publi
     template<typename>
     friend struct std::hash;
 
+    detail::system_lock m_lock; // active host accessors must block command-group submission on other threads
     DataT *m_buffer = nullptr;
     std::shared_ptr<detail::host_access_guard<1>> m_access_guard;
 };

include/simsycl/sycl/buffer.hh

@@ -5,6 +5,7 @@
 #include "property.hh"
 
 #include "../detail/allocation.hh"
+#include "../detail/lock.hh"
 #include "../detail/reference_type.hh"
 
 #include <cstring>
@@ -128,7 +129,7 @@ struct buffer_access_validator {
 };
 
 template<typename T, int Dimensions, typename AllocatorT>
-struct buffer_state : buffer_access_validator<Dimensions> {
+struct buffer_state {
     using write_back_fn = std::function<void(const T *, size_t)>;
 
     sycl::range<Dimensions> range;
@@ -137,6 +138,7 @@ struct buffer_state : buffer_access_validator<Dimensions> {
     write_back_fn write_back;
     bool write_back_enabled;
     std::shared_ptr<const void> shared_host_ptr; // keep the std::shared_ptr host pointer alive
+    mutable shared_value<buffer_access_validator<Dimensions>> validator;
 
     buffer_state(sycl::range<Dimensions> range, const AllocatorT &allocator = {}, const T *init_from = nullptr,
         write_back_fn write_back = {}, const std::shared_ptr<const void> &shared_host_ptr = nullptr)
@@ -162,7 +164,10 @@ struct buffer_state : buffer_access_validator<Dimensions> {
     buffer_state &operator=(buffer_state &&) = delete;
 
     ~buffer_state() {
-        if(write_back_enabled) { write_back(buffer, range.size()); }
+        if(write_back_enabled) {
+            system_lock lock; // writeback must not overlap with command groups in other threads
+            write_back(buffer, range.size());
+        }
         allocator.deallocate(buffer, range.size());
     }
 };
@@ -343,7 +348,8 @@ class buffer final : public detail::reference_type<buffer<T, Dimensions, Allocat
     friend U *simsycl::detail::get_buffer_data(sycl::buffer<U, D, A> &buf);
 
     template<typename U, int D, typename A>
-    friend detail::buffer_access_validator<D> &detail::get_buffer_access_validator(const sycl::buffer<U, D, A> &buf);
+    friend detail::buffer_access_validator<D> &detail::get_buffer_access_validator(
+        const sycl::buffer<U, D, A> &buf, detail::system_lock &lock);
 
     using reference_type::state;
 
@@ -412,8 +418,9 @@ T *get_buffer_data(sycl::buffer<T, Dimensions, AllocatorT> &buf) {
 }
 
 template<typename T, int Dimensions, typename AllocatorT>
-buffer_access_validator<Dimensions> &get_buffer_access_validator(const sycl::buffer<T, Dimensions, AllocatorT> &buf) {
-    return buf.state();
+buffer_access_validator<Dimensions> &get_buffer_access_validator(
+    const sycl::buffer<T, Dimensions, AllocatorT> &buf, system_lock &lock) {
+    return buf.state().validator.with(lock);
 }
 
 } // namespace simsycl::detail

include/simsycl/sycl/device.hh

@@ -40,8 +40,9 @@ struct accelerator_selector {
 };
 
 struct device_state;
+class system_lock;
 
-size_t *device_bytes_free(const sycl::device &device);
+size_t &device_bytes_free(const sycl::device &device, system_lock &lock);
 
 } // namespace simsycl::detail
 
@@ -109,7 +110,7 @@ class device final : public detail::reference_type<device, detail::device_state>
 
     friend device simsycl::make_device(sycl::platform &platform, const device_config &config);
     friend void simsycl::set_parent_device(sycl::device &device, const sycl::device &parent);
-    friend size_t *detail::device_bytes_free(const sycl::device &device);
+    friend size_t &detail::device_bytes_free(const sycl::device &device, detail::system_lock &lock);
 
     device(const detail::device_selector &selector);
     device(std::shared_ptr<detail::device_state> &&state) : reference_type(std::move(state)) {}