[SYCL][COMPAT] Add set_default_queue functionality (#12835)

This PR adds mechanisms to change the default queue of the current
device both via the device extension and a free function.

---------

Signed-off-by: Alberto Cabrera <alberto.cabrera@codeplay.com>

sycl/doc/syclcompat/README.md

@@ -742,6 +742,14 @@ sycl::queue create_queue(bool print_on_async_exceptions = false,
 // device manager.
 sycl::queue get_default_queue();
 
+// Util function to set the default queue of the current device in the
+// device manager.
+// If the device extension saved queue is the default queue, 
+// the previous saved queue will be overwritten as well.
+// This function will be blocking if there are submitted kernels in the
+// previous default queue.
+void set_default_queue(const sycl::queue &q);
+
 // Util function to wait for the queued kernels.
 void wait(sycl::queue q = get_default_queue());
 
@@ -785,6 +793,17 @@ destructor waits on a set of `sycl::event` which can be added to via
 schedule release of memory after a kernel or `mempcy`. SYCL device properties
 can be queried through `device_ext` as well.
 
+Users can manage queues through the `syclcompat::set_default_queue(sycl::queue q)`
+free function, and the `device_ext` `set_saved_queue`, `set_default_queue`,
+and `get_saved_queue` member functions.
+`set_default_queue` is blocking, and
+overwrites the previous default queue with a user defined one, waiting for any
+submitted kernels to finish.
+The `device_ext` automatically sets the saved queue to the default queue.
+Therefore, it's important to note that if the previous default queue was the
+device's saved queue, setting a new default queue will update the reference of
+the saved queue to the new default one to keep the state of the class consistent.
+
 The class is exposed as follows:
 
 ```c++
@@ -807,6 +826,7 @@ class device_ext : public sycl::device {
   void reset();
 
   sycl::queue *default_queue();
+  void set_default_queue(const sycl::queue &q);
   void queues_wait_and_throw();
   sycl::queue *create_queue(bool print_on_async_exceptions = false,
                             bool in_order = true);

sycl/include/syclcompat/device.hpp

@@ -309,6 +309,15 @@ class device_ext : public sycl::device {
     _saved_queue = _default_queue = _queues.front().get();
   }
 
+  void set_default_queue(const sycl::queue &q) {
+    std::lock_guard<std::mutex> lock(m_mutex);
+    _queues.front().get()->wait_and_throw();
+    _queues[0] = std::make_shared<sycl::queue>(q);
+    if (_saved_queue == _default_queue)
+      _saved_queue = _queues.front().get();
+    _default_queue = _queues.front().get();
+  }
+
   sycl::queue *default_queue() { return _default_queue; }
 
   void queues_wait_and_throw() {
@@ -517,6 +526,17 @@ static inline sycl::queue get_default_queue() {
   return *detail::dev_mgr::instance().current_device().default_queue();
 }
 
+/// Util function to change the default queue of the current device in the
+/// device manager
+/// If the device extension saved queue is the default queue,
+/// the previous saved queue will be overwritten as well.
+/// This function will be blocking if there are submitted kernels in the
+/// previous default queue.
+/// @param q New user-defined queue
+static inline void set_default_queue(const sycl::queue &q) {
+  detail::dev_mgr::instance().current_device().set_default_queue(q);
+}
+
 inline void wait(sycl::queue q = get_default_queue()) { q.wait(); }
 
 /// Util function to get the id of current device in

sycl/test-e2e/syclcompat/device/device.cpp

@@ -36,6 +36,17 @@
 
 #include "device_fixt.hpp"
 
+void test_set_default_queue() {
+  std::cout << __PRETTY_FUNCTION__ << std::endl;
+
+  DeviceExtFixt dev_ext;
+  auto &dev_ = dev_ext.get_dev_ext();
+  sycl::queue old_default_queue = syclcompat::get_default_queue();
+  dev_.set_default_queue(syclcompat::create_queue());
+  assert(*dev_.default_queue() == *dev_.get_saved_queue());
+  assert(*dev_.default_queue() != old_default_queue);
+}
+
 int main() {
   /*
     Device Tests

sycl/test-e2e/syclcompat/memory/memory_management_diff_queues.cpp

@@ -0,0 +1,128 @@
+/***************************************************************************
+ *
+ *  Copyright (C) Codeplay Software Ltd.
+ *
+ *  Part of the LLVM Project, under the Apache License v2.0 with LLVM
+ *  Exceptions. See https://llvm.org/LICENSE.txt for license information.
+ *  SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+ *
+ *  Unless required by applicable law or agreed to in writing, software
+ *  distributed under the License is distributed on an "AS IS" BASIS,
+ *  WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+ *  See the License for the specific language governing permissions and
+ *  limitations under the License.
+ *
+ *  SYCLcompat API
+ *
+ *  memory_management_diff_queue.cpp
+ *
+ *  Description:
+ *    memory operations tests for operations when changing the default queue
+ **************************************************************************/
+
+// RUN: %clangxx -std=c++20 -fsycl -fsycl-targets=%{sycl_triple} %s -o %t.out
+// RUN: %{run} %t.out
+
+#include <sycl/sycl.hpp>
+
+#include <syclcompat/memory.hpp>
+
+#include "../common.hpp"
+#include "memory_common.hpp"
+#include "memory_fixt.hpp"
+
+void test_memcpy() {
+  std::cout << __PRETTY_FUNCTION__ << std::endl;
+  sycl::queue q{{sycl::property::queue::in_order()}};
+
+  constexpr int ELEMENTS = 5000;
+  constexpr int N1 = 1000;
+  float *h_A = (float *)malloc(ELEMENTS * sizeof(float));
+  float *h_B = (float *)malloc(ELEMENTS * sizeof(float));
+  float *h_C = (float *)malloc(ELEMENTS * sizeof(float));
+
+  for (int i = 0; i < ELEMENTS; i++) {
+    h_A[i] = 1.0f;
+    h_B[i] = 2.0f;
+  }
+
+  float *d_A = nullptr;
+  // hostA[0..999] -> deviceA[0..999]
+  // hostB[0..3999] -> deviceA[1000..4999]
+  // deviceA[0..4999] -> hostC[0..4999]
+  d_A = (float *)syclcompat::malloc(ELEMENTS * sizeof(float), q);
+  syclcompat::memcpy((void *)d_A, (void *)h_A, N1 * sizeof(float), q);
+
+  syclcompat::set_default_queue(q);
+  syclcompat::memcpy((void *)(d_A + N1), (void *)h_B,
+                     (ELEMENTS - N1) * sizeof(float));
+
+  syclcompat::memcpy((void *)h_C, (void *)d_A, ELEMENTS * sizeof(float));
+
+  // verify
+  for (int i = 0; i < N1; i++) {
+    assert(h_A[i] == h_C[i]);
+  }
+
+  for (int i = N1; i < ELEMENTS; i++) {
+    assert(h_B[i] == h_C[i]);
+  }
+
+  free(h_A);
+  free(h_B);
+  free(h_C);
+  syclcompat::free((void *)d_A);
+}
+
+void test_memset() {
+  std::cout << __PRETTY_FUNCTION__ << std::endl;
+  sycl::queue q{{sycl::property::queue::in_order()}};
+
+  constexpr int PORTION = 5;
+  constexpr int ELEMENTS = PORTION * 3;
+
+  int *h_A = (int *)malloc(ELEMENTS * sizeof(int));
+  for (int i = 0; i < ELEMENTS; i++) {
+    h_A[i] = 4;
+  }
+
+  int *d_A = nullptr;
+
+  d_A = (int *)syclcompat::malloc(ELEMENTS * sizeof(int));
+  // hostA -> deviceA
+  syclcompat::memcpy((void *)d_A, (void *)h_A, ELEMENTS * sizeof(int), q);
+
+  // set d_A[0,..., PORTION - 1] = 0
+  syclcompat::memset((void *)d_A, 0, PORTION * sizeof(int), q);
+
+  syclcompat::set_default_queue(q);
+  // set d_A[PORTION,..., 2 * PORTION - 1] = 0x01010101
+  syclcompat::memset((void *)(d_A + PORTION), 1, PORTION * sizeof(int));
+  // deviceA -> hostA
+  syclcompat::memcpy((void *)h_A, (void *)d_A, ELEMENTS * sizeof(int));
+
+  // check d_A[0,..., PORTION - 1] = 0
+  for (int i = 0; i < PORTION; i++) {
+    assert(h_A[i] == 0);
+  }
+
+  // check d_A[PORTION,..., 2 * PORTION - 1] = 0x01010101
+  for (int i = PORTION; i < (2 * PORTION - 1); i++) {
+    assert(h_A[i] == 0x01010101);
+  }
+
+  // check d_A[2 * PORTION,..., ELEMENTS] = 4
+  for (int i = 2 * PORTION; i < ELEMENTS; i++) {
+    assert(h_A[i] == 4);
+  }
+
+  free(h_A);
+  syclcompat::free((void *)d_A);
+}
+
+int main() {
+  test_memcpy();
+  test_memset();
+
+  return 0;
+}