Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

[SYCL][COMPAT] Add set_default_queue functionality #12835

Merged
merged 3 commits into from
Feb 27, 2024
Merged
Show file tree
Hide file tree
Changes from all commits
Commits
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
20 changes: 20 additions & 0 deletions sycl/doc/syclcompat/README.md
Original file line number Diff line number Diff line change
Expand Up @@ -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);
PietroGhg marked this conversation as resolved.
Show resolved Hide resolved

// Util function to wait for the queued kernels.
void wait(sycl::queue q = get_default_queue());

Expand Down Expand Up @@ -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++
Expand All @@ -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);
Expand Down
20 changes: 20 additions & 0 deletions sycl/include/syclcompat/device.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -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);
PietroGhg marked this conversation as resolved.
Show resolved Hide resolved
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() {
Expand Down Expand Up @@ -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
Expand Down
11 changes: 11 additions & 0 deletions sycl/test-e2e/syclcompat/device/device.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -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
Expand Down
128 changes: 128 additions & 0 deletions sycl/test-e2e/syclcompat/memory/memory_management_diff_queues.cpp
Original file line number Diff line number Diff line change
@@ -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;
}
Loading