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.

Sign up for GitHub

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

Jump to bottom

Updates to SYCL backend #127

Draft
wants to merge 2 commits into
base: main
Choose a base branch
from
Draft

Updates to SYCL backend #127

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
2 commits
Select commit Hold shift + click to select a range
7151c5f
initial updates to sycl backend
abagusetty Apr 12, 2023
c58ee2f
fix build error with cmake macro
abagusetty Apr 12, 2023
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
10 changes: 9 additions & 1 deletion CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -9,6 +9,7 @@ set(camp_VERSION_MINOR ${PROJECT_VERSION_MINOR})
set(camp_VERSION_PATCH ${PROJECT_VERSION_PATCH})

include(CheckCXXCompilerFlag)
include(FindPackageHandleStandardArgs)
if(NOT DEFINED BLT_CXX_STD)
set(CXX_VERSIONS 17 14)
foreach(cxxver ${CXX_VERSIONS})
Expand Down Expand Up @@ -113,6 +114,14 @@ if (ENABLE_HIP)
endif()
endif ()

if (ENABLE_SYCL)
check_cxx_compiler_flag("-fsycl" CXX_HAS_FSYCL)
find_package_handle_standard_args( SYCL
REQUIRED_VARS CXX_HAS_FSYCL
)
add_compile_options("-fsycl")
endif ()

# end backends

# Configure the config header file to allow config time options
Expand Down Expand Up @@ -189,4 +198,3 @@ if(CAMP_ENABLE_TESTS)
enable_testing()
add_subdirectory(test)
endif()

4 changes: 4 additions & 0 deletions include/camp/defines.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -25,6 +25,10 @@ For details about use and distribution, please read LICENSE and NOTICE from
#include <hip/hip_runtime.h>
#endif

#ifdef CAMP_ENABLE_SYCL
#include <sycl/sycl.hpp>
#endif

namespace camp
{

Expand Down
110 changes: 57 additions & 53 deletions include/camp/resource/sycl.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -16,10 +16,9 @@ For details about use and distribution, please read LICENSE and NOTICE from
#include "camp/resource/platform.hpp"

#ifdef CAMP_ENABLE_SYCL
#include <CL/sycl.hpp>
#include <sycl/sycl.hpp>
#include <map>
#include <array>
using namespace cl;

namespace camp
{
Expand All @@ -32,7 +31,9 @@ namespace resources
{
public:
SyclEvent(sycl::queue *qu) { m_event = sycl::event(); }
bool check() const { return true; }
bool check() const {
return (m_event.get_info<sycl::info::event::command_execution_status>() == sycl::info::event_command_status::complete);
}
void wait() const { getSyclEvent_t().wait(); }
sycl::event getSyclEvent_t() const { return m_event; }

Expand All @@ -42,15 +43,13 @@ namespace resources

class Sycl
{
static sycl::queue *get_a_queue(sycl::context &syclContext,
static sycl::queue *get_a_queue(sycl::context *syclContext,
int num,
bool useContext)
{
static sycl::gpu_selector gpuSelector;
static sycl::device gpuSelector { sycl::gpu_selector_v };
static sycl::property_list propertyList =
sycl::property_list(sycl::property::queue::in_order());
static sycl::context privateContext;
static sycl::context *contextInUse = NULL;
static std::map<sycl::context *, std::array<sycl::queue, 16>> queueMap;


Expand All @@ -59,47 +58,45 @@ namespace resources

// User passed a context, use it
if (useContext) {
contextInUse = &syclContext;

if (queueMap.find(contextInUse) == queueMap.end()) {
queueMap[contextInUse] = {
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList)};
queueMap[syclContext] = {
sycl::queue(*syclContext, gpuSelector, propertyList),
sycl::queue(*syclContext, gpuSelector, propertyList),
sycl::queue(*syclContext, gpuSelector, propertyList),
sycl::queue(*syclContext, gpuSelector, propertyList),
sycl::queue(*syclContext, gpuSelector, propertyList),
sycl::queue(*syclContext, gpuSelector, propertyList),
sycl::queue(*syclContext, gpuSelector, propertyList),
sycl::queue(*syclContext, gpuSelector, propertyList),
sycl::queue(*syclContext, gpuSelector, propertyList),
sycl::queue(*syclContext, gpuSelector, propertyList),
sycl::queue(*syclContext, gpuSelector, propertyList),
sycl::queue(*syclContext, gpuSelector, propertyList),
sycl::queue(*syclContext, gpuSelector, propertyList),
sycl::queue(*syclContext, gpuSelector, propertyList),
sycl::queue(*syclContext, gpuSelector, propertyList),
sycl::queue(*syclContext, gpuSelector, propertyList)};
}
} else { // User did not pass context, use last used or private one
if (contextInUse == NULL) {
contextInUse = &privateContext;
queueMap[contextInUse] = {
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList),
sycl::queue(*contextInUse, gpuSelector, propertyList)};
if (syclContext == nullptr) {
sycl::context* privateContext = new sycl::context(gpuSelector);
queueMap[privateContext] = {
sycl::queue(*privateContext, gpuSelector, propertyList),
sycl::queue(*privateContext, gpuSelector, propertyList),
sycl::queue(*privateContext, gpuSelector, propertyList),
sycl::queue(*privateContext, gpuSelector, propertyList),
sycl::queue(*privateContext, gpuSelector, propertyList),
sycl::queue(*privateContext, gpuSelector, propertyList),
sycl::queue(*privateContext, gpuSelector, propertyList),
sycl::queue(*privateContext, gpuSelector, propertyList),
sycl::queue(*privateContext, gpuSelector, propertyList),
sycl::queue(*privateContext, gpuSelector, propertyList),
sycl::queue(*privateContext, gpuSelector, propertyList),
sycl::queue(*privateContext, gpuSelector, propertyList),
sycl::queue(*privateContext, gpuSelector, propertyList),
sycl::queue(*privateContext, gpuSelector, propertyList),
sycl::queue(*privateContext, gpuSelector, propertyList),
sycl::queue(*privateContext, gpuSelector, propertyList)};
}
}
m_mtx.unlock();
Expand All @@ -116,16 +113,14 @@ namespace resources

return &queueMap[contextInUse][num % 16];
}

public:
Sycl(int group = -1)
{
sycl::context temp;
qu = get_a_queue(temp, group, false);
qu = get_a_queue(nullptr, group, false);
}

Sycl(sycl::context &syclContext, int group = -1)
: qu(get_a_queue(syclContext, group, true))
: qu(get_a_queue(&syclContext, group, true))
{
}

Expand All @@ -138,12 +133,22 @@ namespace resources
}
SyclEvent get_event() { return SyclEvent(get_queue()); }
Event get_event_erased() { return Event{SyclEvent(get_queue())}; }
void wait() { qu->wait(); }
void wait() {
#if defined(SYCL_EXT_ONEAPI_ENQUEUE_BARRIER)
qu->ext_oneapi_submit_barrier();
#else
qu->wait();
#endif
}
void wait_for(Event *e)
{
auto *sycl_event = e->try_get<SyclEvent>();
if (sycl_event) {
#if defined(SYCL_EXT_ONEAPI_ENQUEUE_BARRIER)
qu->ext_oneapi_submit_barrier( {sycl_event->getSyclEvent_t()} );
#else
(sycl_event->getSyclEvent_t()).wait();
#endif
} else {
e->wait();
}
Expand All @@ -155,7 +160,6 @@ namespace resources
{
T *ret = nullptr;
if (size > 0) {
ret = sycl::malloc_shared<T>(size, *qu);
switch (ma) {
case MemoryAccess::Unknown:
case MemoryAccess::Device:
Expand All @@ -181,13 +185,13 @@ namespace resources
void memcpy(void *dst, const void *src, size_t size)
{
if (size > 0) {
qu->memcpy(dst, src, size).wait();
qu->memcpy(dst, src, size);
}
}
void memset(void *p, int val, size_t size)
{
if (size > 0) {
qu->memset(p, val, size).wait();
qu->memset(p, val, size);
}
}

Expand Down
4 changes: 4 additions & 0 deletions test/CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -24,6 +24,10 @@ function(camp_add_test TESTNAME)
list(APPEND _depends blt::hip)
endif()

if(ENABLE_SYCL)
list(APPEND _depends sycl)
endif()

if(ABT_DEPENDS_ON)
list(APPEND _depends ${ABT_DEPENDS_ON})
endif()
Expand Down
Loading