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

Remove deprecated SYCL usage for newer compilers #1772

Merged
merged 3 commits into from
Dec 11, 2024
Merged

Remove deprecated SYCL usage for newer compilers #1772

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
3 commits
Select commit Hold shift + click to select a range
24e3453
Remove deprecated SYCL usage for newer compilers
rhornung67 Nov 25, 2024
856a098
Merge branch 'develop' into task/rhornung67/squash_sycl_warnings
rhornung67 Dec 10, 2024
8397649
Global scope for consistency
rhornung67 Dec 10, 2024
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
2 changes: 1 addition & 1 deletion examples/memoryManager.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -76,7 +76,7 @@ void deallocate(T *&ptr)
hipErrchk(hipMalloc((void **)&ptr, sizeof(T) * size));
#elif defined(RAJA_ENABLE_SYCL)
auto qu = sycl_res->get<camp::resources::Sycl>().get_queue();
ptr = cl::sycl::malloc_device<T>(size, *qu);
ptr = ::sycl::malloc_device<T>(size, *qu);
#endif
return ptr;
}
Expand Down
2 changes: 1 addition & 1 deletion exercises/memoryManager.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -76,7 +76,7 @@ void deallocate(T *&ptr)
hipErrchk(hipMalloc((void **)&ptr, sizeof(T) * size));
#elif defined(RAJA_ENABLE_SYCL)
auto qu = sycl_res->get<camp::resources::Sycl>().get_queue();
ptr = cl::sycl::malloc_device<T>(size, *qu);
ptr = ::sycl::malloc_device<T>(size, *qu);
#endif
return ptr;
}
Expand Down
2 changes: 1 addition & 1 deletion include/RAJA/pattern/launch/launch_core.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -162,7 +162,7 @@ class LaunchContext
void *shared_mem_ptr;

#if defined(RAJA_ENABLE_SYCL)
mutable cl::sycl::nd_item<3> *itm;
mutable ::sycl::nd_item<3> *itm;
#endif

RAJA_HOST_DEVICE LaunchContext()
Expand Down
4 changes: 2 additions & 2 deletions include/RAJA/policy/sycl/MemUtils_SYCL.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -50,7 +50,7 @@ namespace detail
struct syclInfo {
sycl_dim_t gridDim{0};
sycl_dim_t blockDim{0};
cl::sycl::queue qu = cl::sycl::queue();
::sycl::queue qu = ::sycl::queue();
bool setup_reducers = false;
#if defined(RAJA_ENABLE_OPENMP)
syclInfo* thread_states = nullptr;
Expand All @@ -62,7 +62,7 @@ extern syclInfo g_status;

extern syclInfo tl_status;

extern std::unordered_map<cl::sycl::queue, bool> g_queue_info_map;
extern std::unordered_map<::sycl::queue, bool> g_queue_info_map;

} // namespace detail

Expand Down
4 changes: 2 additions & 2 deletions include/RAJA/policy/sycl/forall.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -206,8 +206,8 @@ resources::EventProxy<resources::Sycl> forall_impl(resources::Sycl &sycl_res,
}).wait(); // Need to wait for completion to free memory

// Free our device memory
cl::sycl::free(lbody, *q);
cl::sycl::free(beg, *q);
::sycl::free(lbody, *q);
::sycl::free(beg, *q);

RAJA_FT_END;
}
Expand Down
2 changes: 1 addition & 1 deletion include/RAJA/policy/sycl/kernel/Conditional.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -51,7 +51,7 @@ struct SyclStatementExecutor<Data,
static
inline
RAJA_DEVICE
void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
if (Conditional::eval(data)) {

Expand Down
14 changes: 7 additions & 7 deletions include/RAJA/policy/sycl/kernel/For.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -59,7 +59,7 @@ struct SyclStatementExecutor<
using diff_t = segment_diff_type<ArgumentId, Data>;

static
inline RAJA_DEVICE void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
inline RAJA_DEVICE void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
auto len = segment_length<ArgumentId>(data);
auto i = item.get_global_id(Dim);
Expand Down Expand Up @@ -124,7 +124,7 @@ struct SyclStatementExecutor<
using diff_t = segment_diff_type<ArgumentId, Data>;

static
inline RAJA_DEVICE void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
inline RAJA_DEVICE void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
auto len = segment_length<ArgumentId>(data);
auto i = item.get_group(Dim);
Expand Down Expand Up @@ -187,7 +187,7 @@ struct SyclStatementExecutor<
using diff_t = segment_diff_type<ArgumentId, Data>;

static
inline RAJA_DEVICE void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
inline RAJA_DEVICE void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
auto len = segment_length<ArgumentId>(data);
auto i0 = item.get_group(Dim);
Expand Down Expand Up @@ -253,7 +253,7 @@ struct SyclStatementExecutor<
using diff_t = segment_diff_type<ArgumentId, Data>;

static
inline RAJA_DEVICE void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
inline RAJA_DEVICE void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
auto len = segment_length<ArgumentId>(data);
auto i = item.get_local_id(Dim);
Expand Down Expand Up @@ -317,7 +317,7 @@ struct SyclStatementExecutor<
using diff_t = segment_diff_type<ArgumentId, Data>;

static
inline RAJA_DEVICE void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
inline RAJA_DEVICE void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
auto len = segment_length<ArgumentId>(data);
auto i0 = item.get_local_id(Dim);
Expand Down Expand Up @@ -393,7 +393,7 @@ struct SyclStatementExecutor<
using diff_t = segment_diff_type<ArgumentId, Data>;

static
inline RAJA_DEVICE void exec(Data &data, cl::sycl::nd_item<3> item)
inline RAJA_DEVICE void exec(Data &data, ::sycl::nd_item<3> item)
{
auto len = segment_length<ArgumentId>(data);
auto i = item.get_global_id(0);
Expand Down Expand Up @@ -454,7 +454,7 @@ struct SyclStatementExecutor<
static
inline
RAJA_DEVICE
void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{

using idx_type = camp::decay<decltype(camp::get<ArgumentId>(data.offset_tuple))>;
Expand Down
14 changes: 7 additions & 7 deletions include/RAJA/policy/sycl/kernel/ForICount.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -63,7 +63,7 @@ struct SyclStatementExecutor<
static
inline
RAJA_DEVICE
void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
diff_t len = segment_length<ArgumentId>(data);
auto i = item.get_local_id(ThreadDim);
Expand Down Expand Up @@ -121,7 +121,7 @@ struct SyclStatementExecutor<
static
inline
RAJA_DEVICE
void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
diff_t len = segment_length<ArgumentId>(data);
auto i0 = item.get_local_id(0);
Expand Down Expand Up @@ -181,7 +181,7 @@ struct SyclStatementExecutor<
static
inline
RAJA_DEVICE
void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
// masked size strided loop
diff_t len = segment_length<ArgumentId>(data);
Expand Down Expand Up @@ -243,7 +243,7 @@ struct SyclStatementExecutor<
using typename Base::diff_t;

static
inline RAJA_DEVICE void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
inline RAJA_DEVICE void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
// block stride loop
diff_t len = segment_length<ArgumentId>(data);
Expand Down Expand Up @@ -300,7 +300,7 @@ struct SyclStatementExecutor<
using typename Base::diff_t;

static
inline RAJA_DEVICE void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
inline RAJA_DEVICE void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
// grid stride loop
diff_t len = segment_length<ArgumentId>(data);
Expand Down Expand Up @@ -349,7 +349,7 @@ struct SyclStatementExecutor<
using typename Base::diff_t;

static
inline RAJA_DEVICE void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
inline RAJA_DEVICE void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
// grid stride loop
diff_t len = segment_length<ArgumentId>(data);
Expand Down Expand Up @@ -399,7 +399,7 @@ struct SyclStatementExecutor<
static
inline
RAJA_DEVICE
void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
diff_t len = segment_length<ArgumentId>(data);

Expand Down
2 changes: 1 addition & 1 deletion include/RAJA/policy/sycl/kernel/Lambda.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -46,7 +46,7 @@ template <typename Data, camp::idx_t LambdaIndex, typename... Args, typename Typ
struct SyclStatementExecutor<Data, statement::Lambda<LambdaIndex, Args...>, Types> {

static
inline RAJA_DEVICE void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
inline RAJA_DEVICE void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
// Only execute the lambda if it hasn't been masked off
if(thread_active){
Expand Down
18 changes: 9 additions & 9 deletions include/RAJA/policy/sycl/kernel/SyclKernel.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -93,7 +93,7 @@ namespace internal
* SYCL global function for launching SyclKernel policies.
*/
template <typename Data, typename Exec>
void SyclKernelLauncher(Data data, cl::sycl::nd_item<3> item)
void SyclKernelLauncher(Data data, ::sycl::nd_item<3> item)
{

using data_t = camp::decay<Data>;
Expand Down Expand Up @@ -128,28 +128,28 @@ struct SyclLaunchHelper<false,sycl_launch<async0>,StmtList,Data,Types>
static void launch(Data &&data,
internal::LaunchDims launch_dims,
size_t shmem,
cl::sycl::queue* qu)
::sycl::queue* qu)
{

//
// Setup shared memory buffers
// Kernel body is nontrivially copyable, create space on device and copy to
// Workaround until "is_device_copyable" is supported
//
data_t* m_data = (data_t*) cl::sycl::malloc_device(sizeof(data_t), *qu);
data_t* m_data = (data_t*) ::sycl::malloc_device(sizeof(data_t), *qu);
qu->memcpy(m_data, &data, sizeof(data_t)).wait();

qu->submit([&](cl::sycl::handler& h) {
qu->submit([&](::sycl::handler& h) {

h.parallel_for(launch_dims.fit_nd_range(qu),
[=] (cl::sycl::nd_item<3> item) {
[=] (::sycl::nd_item<3> item) {

SyclKernelLauncher<Data, executor_t>(*m_data, item);

});
}).wait(); // Need to wait to free memory

cl::sycl::free(m_data, *qu);
::sycl::free(m_data, *qu);

}
};
Expand All @@ -172,13 +172,13 @@ struct SyclLaunchHelper<true,sycl_launch<async0>,StmtList,Data,Types>
static void launch(Data &&data,
internal::LaunchDims launch_dims,
size_t shmem,
cl::sycl::queue* qu)
::sycl::queue* qu)
{

qu->submit([&](cl::sycl::handler& h) {
qu->submit([&](::sycl::handler& h) {

h.parallel_for(launch_dims.fit_nd_range(qu),
[=] (cl::sycl::nd_item<3> item) {
[=] (::sycl::nd_item<3> item) {

SyclKernelLauncher<Data, executor_t>(data, item);

Expand Down
10 changes: 5 additions & 5 deletions include/RAJA/policy/sycl/kernel/Tile.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -61,7 +61,7 @@ struct SyclStatementExecutor<
using enclosed_stmts_t = SyclStatementListExecutor<Data, stmt_list_t, Types>;
using diff_t = segment_diff_type<ArgumentId, Data>;

static inline RAJA_DEVICE void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active){
static inline RAJA_DEVICE void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active){
// Get the segment referenced by this Tile statement
auto &segment = camp::get<ArgumentId>(data.segment_tuple);

Expand Down Expand Up @@ -139,7 +139,7 @@ struct SyclStatementExecutor<

using diff_t = segment_diff_type<ArgumentId, Data>;

static inline RAJA_DEVICE void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
static inline RAJA_DEVICE void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
// Get the segment referenced by this Tile statement
auto &segment = camp::get<ArgumentId>(data.segment_tuple);
Expand Down Expand Up @@ -231,7 +231,7 @@ struct SyclStatementExecutor<

using diff_t = segment_diff_type<ArgumentId, Data>;

static inline RAJA_DEVICE void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
static inline RAJA_DEVICE void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
// Get the segment referenced by this Tile statement
auto &segment = camp::get<ArgumentId>(data.segment_tuple);
Expand Down Expand Up @@ -321,7 +321,7 @@ struct SyclStatementExecutor<

using diff_t = segment_diff_type<ArgumentId, Data>;

static inline RAJA_DEVICE void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
static inline RAJA_DEVICE void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
// Get the segment referenced by this Tile statement
auto &segment = camp::get<ArgumentId>(data.segment_tuple);
Expand Down Expand Up @@ -409,7 +409,7 @@ struct SyclStatementExecutor<

using diff_t = segment_diff_type<ArgumentId, Data>;

static inline RAJA_DEVICE void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
static inline RAJA_DEVICE void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
// Get the segment referenced by this Tile statement
auto &segment = camp::get<ArgumentId>(data.segment_tuple);
Expand Down
10 changes: 5 additions & 5 deletions include/RAJA/policy/sycl/kernel/TileTCount.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -70,7 +70,7 @@ struct SyclStatementExecutor<
static
inline
RAJA_DEVICE
void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active){
void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active){
// Get the segment referenced by this Tile statement
auto &segment = camp::get<ArgumentId>(data.segment_tuple);

Expand Down Expand Up @@ -141,7 +141,7 @@ struct SyclStatementExecutor<
static
inline
RAJA_DEVICE
void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
// Get the segment referenced by this Tile statement
auto &segment = camp::get<ArgumentId>(data.segment_tuple);
Expand Down Expand Up @@ -214,7 +214,7 @@ struct SyclStatementExecutor<
static
inline
RAJA_DEVICE
void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
// Get the segment referenced by this Tile statement
auto &segment = camp::get<ArgumentId>(data.segment_tuple);
Expand Down Expand Up @@ -289,7 +289,7 @@ struct SyclStatementExecutor<
static
inline
RAJA_DEVICE
void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
// Get the segment referenced by this Tile statement
auto &segment = camp::get<ArgumentId>(data.segment_tuple);
Expand Down Expand Up @@ -363,7 +363,7 @@ struct SyclStatementExecutor<
static
inline
RAJA_DEVICE
void exec(Data &data, cl::sycl::nd_item<3> item, bool thread_active)
void exec(Data &data, ::sycl::nd_item<3> item, bool thread_active)
{
// Get the segment referenced by this Tile statement
auto &segment = camp::get<ArgumentId>(data.segment_tuple);
Expand Down
Loading
Loading