From 3d07790d5425b68990acf9b2a0548936efec1e33 Mon Sep 17 00:00:00 2001
From: Sergey Semenov <sergey.semenov@intel.com>
Date: Wed, 4 Sep 2024 08:39:27 -0700
Subject: [PATCH] [SYCL] Guard access to MCreateShadowCopy (#15273)

Fixes a flaky failure when getting write access to a buffer from
multiple threads.

---------

Co-authored-by: Artur Gainullin <artur.gainullin@intel.com>
---
 sycl/source/detail/sycl_mem_obj_t.cpp         |  7 ++--
 sycl/source/detail/sycl_mem_obj_t.hpp         |  4 +++
 .../Regression/multithread_write_accessor.cpp | 35 +++++++++++++++++++
 3 files changed, 44 insertions(+), 2 deletions(-)
 create mode 100644 sycl/test-e2e/Regression/multithread_write_accessor.cpp

diff --git a/sycl/source/detail/sycl_mem_obj_t.cpp b/sycl/source/detail/sycl_mem_obj_t.cpp
index 457c1c6faada5..42e4cfc5a1860 100644
--- a/sycl/source/detail/sycl_mem_obj_t.cpp
+++ b/sycl/source/detail/sycl_mem_obj_t.cpp
@@ -228,8 +228,11 @@ void SYCLMemObjT::detachMemoryObject(
 
 void SYCLMemObjT::handleWriteAccessorCreation() {
   const auto InitialUserPtr = MUserPtr;
-  MCreateShadowCopy();
-  MCreateShadowCopy = []() -> void {};
+  {
+    std::lock_guard<std::mutex> Lock(MCreateShadowCopyMtx);
+    MCreateShadowCopy();
+    MCreateShadowCopy = []() -> void {};
+  }
   if (MRecord != nullptr && MUserPtr != InitialUserPtr) {
     for (auto &it : MRecord->MAllocaCommands) {
       if (it->MMemAllocation == InitialUserPtr) {
diff --git a/sycl/source/detail/sycl_mem_obj_t.hpp b/sycl/source/detail/sycl_mem_obj_t.hpp
index 9a3e3e23af6c8..de400e85267b0 100644
--- a/sycl/source/detail/sycl_mem_obj_t.hpp
+++ b/sycl/source/detail/sycl_mem_obj_t.hpp
@@ -23,6 +23,7 @@
 #include <atomic>
 #include <cstring>
 #include <memory>
+#include <mutex>
 #include <type_traits>
 
 namespace sycl {
@@ -196,6 +197,7 @@ class SYCLMemObjT : public SYCLMemObjI {
         MUserPtr = HostPtr;
       } else if (canReadHostPtr(HostPtr, RequiredAlign)) {
         MUserPtr = HostPtr;
+        std::lock_guard<std::mutex> Lock(MCreateShadowCopyMtx);
         MCreateShadowCopy = [this, RequiredAlign, HostPtr]() -> void {
           setAlign(RequiredAlign);
           MShadowCopy = allocateHostMem();
@@ -229,6 +231,7 @@ class SYCLMemObjT : public SYCLMemObjI {
         MUserPtr = HostPtr.get();
       } else if (canReadHostPtr(HostPtr.get(), RequiredAlign)) {
         MUserPtr = HostPtr.get();
+        std::lock_guard<std::mutex> Lock(MCreateShadowCopyMtx);
         MCreateShadowCopy = [this, RequiredAlign, HostPtr]() -> void {
           setAlign(RequiredAlign);
           MShadowCopy = allocateHostMem();
@@ -375,6 +378,7 @@ class SYCLMemObjT : public SYCLMemObjI {
   // defer the memory allocation and copying to the point where a writable
   // accessor is created.
   std::function<void(void)> MCreateShadowCopy = []() -> void {};
+  std::mutex MCreateShadowCopyMtx;
   bool MOwnNativeHandle = true;
 };
 } // namespace detail
diff --git a/sycl/test-e2e/Regression/multithread_write_accessor.cpp b/sycl/test-e2e/Regression/multithread_write_accessor.cpp
new file mode 100644
index 0000000000000..d3228a2d2f49a
--- /dev/null
+++ b/sycl/test-e2e/Regression/multithread_write_accessor.cpp
@@ -0,0 +1,35 @@
+// RUN: %{build} -o %t.out
+// RUN: %{run} %t.out
+#include <sycl/detail/core.hpp>
+
+#include <cassert>
+#include <thread>
+#include <vector>
+
+constexpr int NThreads = 8;
+
+class KernelA;
+
+void threadFunction(sycl::buffer<int, 1> &Buf) {
+  sycl::queue Q;
+  Q.submit([&](sycl::handler &Cgh) {
+    auto Acc = Buf.get_access<sycl::access::mode::read_write>(Cgh);
+    Cgh.single_task<class KernelA>([=]() { Acc[0] += 1; });
+  });
+}
+int main() {
+  std::vector<std::thread> Threads;
+  Threads.reserve(NThreads);
+
+  int Val = 0;
+  {
+    sycl::buffer<int, 1> Buf(&Val, sycl::range<1>(1));
+    sycl::queue Q;
+
+    for (int I = 0; I < NThreads; ++I)
+      Threads.emplace_back(threadFunction, std::ref(Buf));
+    for (auto &t : Threads)
+      t.join();
+  }
+  assert(Val == NThreads);
+}