From 44a77a1080ddce5df5410bd601cc10015ce23022 Mon Sep 17 00:00:00 2001 From: rbegam Date: Mon, 20 Jul 2020 16:17:51 -0700 Subject: [PATCH 1/7] [SYCL] Separate OpenCL specific constructs from lit tests. Adding 'opencl' requirement disabled several tests some of which have non-opencl constructs as well. This patch separates those tests to enable testing of SYCL core functionalities. Signed-off-by: rbegam --- sycl/test/inorder_queue/in_order_buffs.cpp | 22 -- .../test/inorder_queue/in_order_buffs_ocl.cpp | 51 +++++ sycl/test/inorder_queue/in_order_dmemll.cpp | 10 +- .../inorder_queue/in_order_dmemll_ocl.cpp | 46 ++++ .../kernel-and-program-interop.cpp | 205 ++++++++++++++++++ .../kernel-and-program/kernel-and-program.cpp | 165 -------------- 6 files changed, 307 insertions(+), 192 deletions(-) create mode 100644 sycl/test/inorder_queue/in_order_buffs_ocl.cpp create mode 100644 sycl/test/inorder_queue/in_order_dmemll_ocl.cpp create mode 100644 sycl/test/kernel-and-program/kernel-and-program-interop.cpp diff --git a/sycl/test/inorder_queue/in_order_buffs.cpp b/sycl/test/inorder_queue/in_order_buffs.cpp index ef317031291ce..c92868fdd742c 100644 --- a/sycl/test/inorder_queue/in_order_buffs.cpp +++ b/sycl/test/inorder_queue/in_order_buffs.cpp @@ -1,5 +1,3 @@ -// REQUIRES: opencl - // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out -L %opencl_libs_dir -lOpenCL // RUN: %ACC_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out @@ -18,14 +16,6 @@ using namespace cl::sycl; const int dataSize = 32; -bool isQueueInOrder(cl_command_queue cq) { - cl_command_queue_properties reportedProps; - cl_int iRet = clGetCommandQueueInfo( - cq, CL_QUEUE_PROPERTIES, sizeof(reportedProps), &reportedProps, nullptr); - assert(CL_SUCCESS == iRet && "Failed to obtain queue info from ocl device"); - return (!(reportedProps & CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE)); -} - int main() { int dataA[dataSize] = {0}; int dataB[dataSize] = {0}; @@ -63,18 +53,6 @@ int main() { cgh.parallel_for(myRange, myKernel); }); - bool result = true; - cl_command_queue cq = Queue.get(); - device dev = Queue.get_device(); - bool expected_result = dev.is_host() ? true : isQueueInOrder(cq); - - if (expected_result != result) { - std::cout << "Resulting queue order is OOO but expected order is inorder" - << std::endl; - - return -1; - } - auto readBufferB = bufB.get_access(); for (size_t i = 0; i != dataSize; ++i) { if (readBufferB[i] != i) { diff --git a/sycl/test/inorder_queue/in_order_buffs_ocl.cpp b/sycl/test/inorder_queue/in_order_buffs_ocl.cpp new file mode 100644 index 0000000000000..46dc6fed4fe06 --- /dev/null +++ b/sycl/test/inorder_queue/in_order_buffs_ocl.cpp @@ -0,0 +1,51 @@ +// REQUIRES: opencl + +// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out -L %opencl_libs_dir -lOpenCL +// RUN: %ACC_RUN_PLACEHOLDER %t.out +// RUN: %CPU_RUN_PLACEHOLDER %t.out +// RUN: %GPU_RUN_PLACEHOLDER %t.out +//==-------- ordered_buffs.cpp - SYCL buffers in ordered queues test--------==// +// +// 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 +// +//===----------------------------------------------------------------------===// +#include +#include + +using namespace cl::sycl; + +const int dataSize = 32; + +bool isQueueInOrder(cl_command_queue cq) { + cl_command_queue_properties reportedProps; + cl_int iRet = clGetCommandQueueInfo( + cq, CL_QUEUE_PROPERTIES, sizeof(reportedProps), &reportedProps, nullptr); + assert(CL_SUCCESS == iRet && "Failed to obtain queue info from ocl device"); + return (!(reportedProps & CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE)); +} + +int main() { + int dataA[dataSize] = {0}; + int dataB[dataSize] = {0}; + + { + queue Queue{property::queue::in_order()}; + + bool result = true; + cl_command_queue cq = Queue.get(); + device dev = Queue.get_device(); + bool expected_result = dev.is_host() ? true : isQueueInOrder(cq); + + if (expected_result != result) { + std::cout << "Resulting queue order is OOO but expected order is inorder" + << std::endl; + + return -1; + } + } + + return 0; +} + diff --git a/sycl/test/inorder_queue/in_order_dmemll.cpp b/sycl/test/inorder_queue/in_order_dmemll.cpp index defb64115d560..e856f828f9e99 100644 --- a/sycl/test/inorder_queue/in_order_dmemll.cpp +++ b/sycl/test/inorder_queue/in_order_dmemll.cpp @@ -1,4 +1,4 @@ -// REQUIRES: opencl +// R-EQUIRES: opencl // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t1.out -L %opencl_libs_dir -lOpenCL // RUN: %CPU_RUN_PLACEHOLDER %t1.out @@ -19,14 +19,14 @@ using namespace cl::sycl; constexpr int numNodes = 4; -bool getQueueOrder(cl_command_queue cq) { +/*bool getQueueOrder(cl_command_queue cq) { cl_command_queue_properties reportedProps; cl_int iRet = clGetCommandQueueInfo( cq, CL_QUEUE_PROPERTIES, sizeof(reportedProps), &reportedProps, nullptr); assert(CL_SUCCESS == iRet && "Failed to obtain queue info from ocl device"); return (reportedProps & CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE) ? false : true; -} +}*/ struct Node { Node() : pNext(nullptr), Num(0xDEADBEEF) {} @@ -103,7 +103,7 @@ int main() { d_cur = h_cur.pNext; } - bool result = true; + /*bool result = true; cl_command_queue cq = q.get(); bool expected_result = dev.is_host() ? true : getQueueOrder(cq); if (expected_result != result) { @@ -111,7 +111,7 @@ int main() { << std::endl; return -1; - } + }*/ return 0; } diff --git a/sycl/test/inorder_queue/in_order_dmemll_ocl.cpp b/sycl/test/inorder_queue/in_order_dmemll_ocl.cpp new file mode 100644 index 0000000000000..424bf9c64419d --- /dev/null +++ b/sycl/test/inorder_queue/in_order_dmemll_ocl.cpp @@ -0,0 +1,46 @@ +// REQUIRES: opencl + +// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t1.out -L %opencl_libs_dir -lOpenCL +// RUN: %CPU_RUN_PLACEHOLDER %t1.out +// RUN: %GPU_RUN_PLACEHOLDER %t1.out +//==----------- ordered_dmemll.cpp - Device Memory Linked List test --------==// +// It uses an ordered queue where explicit waiting is not necessary between +// kernels +// +// 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 +// +//===----------------------------------------------------------------------===// + +#include + +using namespace cl::sycl; + +constexpr int numNodes = 4; + +bool getQueueOrder(cl_command_queue cq) { + cl_command_queue_properties reportedProps; + cl_int iRet = clGetCommandQueueInfo( + cq, CL_QUEUE_PROPERTIES, sizeof(reportedProps), &reportedProps, nullptr); + assert(CL_SUCCESS == iRet && "Failed to obtain queue info from ocl device"); + return (reportedProps & CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE) ? false + : true; +} + +int main() { + queue q{property::queue::in_order()}; + auto dev = q.get_device(); + + bool result = true; + cl_command_queue cq = q.get(); + bool expected_result = dev.is_host() ? true : getQueueOrder(cq); + if (expected_result != result) { + std::cout << "Resulting queue order is OOO but expected order is inorder" + << std::endl; + + return -1; + } + + return 0; +} diff --git a/sycl/test/kernel-and-program/kernel-and-program-interop.cpp b/sycl/test/kernel-and-program/kernel-and-program-interop.cpp new file mode 100644 index 0000000000000..991a313712fae --- /dev/null +++ b/sycl/test/kernel-and-program/kernel-and-program-interop.cpp @@ -0,0 +1,205 @@ +// REQUIRES: opencl + +// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out -L %opencl_libs_dir -lOpenCL +// RUN: env SYCL_DEVICE_TYPE=HOST %t.out +// RUN: %CPU_RUN_PLACEHOLDER %t.out +// RUNx: %GPU_RUN_PLACEHOLDER %t.out +// RUNx: %ACC_RUN_PLACEHOLDER %t.out + +//==--- kernel-and-program.cpp - SYCL kernel/program test ------------------==// +// +// 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 +// +//===----------------------------------------------------------------------===// + +#include + +#include +#include +#include +#include + +int main() { + // Single task invocation methods + { + cl::sycl::queue q; + int data = 0; + + // OpenCL interoperability kernel invocation + if (!q.is_host()) { + { + cl_int err; + cl::sycl::context ctx = q.get_context(); + cl_context clCtx = ctx.get(); + cl_command_queue clQ = q.get(); + cl_mem clBuffer = + clCreateBuffer(clCtx, CL_MEM_WRITE_ONLY, sizeof(int), NULL, NULL); + err = clEnqueueWriteBuffer(clQ, clBuffer, CL_TRUE, 0, sizeof(int), + &data, 0, NULL, NULL); + assert(err == CL_SUCCESS); + clFinish(clQ); + cl::sycl::program prog(ctx); + prog.build_with_source( + "kernel void SingleTask(global int* a) {*a+=1; }\n"); + q.submit([&](cl::sycl::handler &cgh) { + cgh.set_args(clBuffer); + cgh.single_task(prog.get_kernel("SingleTask")); + }); + q.wait(); + err = clEnqueueReadBuffer(clQ, clBuffer, CL_TRUE, 0, sizeof(int), &data, + 0, NULL, NULL); + clReleaseCommandQueue(clQ); + clReleaseContext(clCtx); + assert(err == CL_SUCCESS); + assert(data == 1); + } + { + cl::sycl::queue sycl_queue; + cl::sycl::program prog(sycl_queue.get_context()); + prog.build_with_source("kernel void foo(global int* a, global int* b, " + "global int* c) {*a=*b+*c; }\n"); + int a = 13, b = 14, c = 15; + { + cl::sycl::buffer bufa(&a, cl::sycl::range<1>(1)); + cl::sycl::buffer bufb(&b, cl::sycl::range<1>(1)); + cl::sycl::buffer bufc(&c, cl::sycl::range<1>(1)); + sycl_queue.submit([&](cl::sycl::handler &cgh) { + auto A = bufa.get_access(cgh); + auto B = bufb.get_access(cgh); + auto C = bufc.get_access(cgh); + cgh.set_args(A, B, C); + cgh.single_task(prog.get_kernel("foo")); + }); + } + assert(a == b + c); + } + } + { + cl::sycl::queue Queue; + if (!Queue.is_host()) { + cl::sycl::sampler first( + cl::sycl::coordinate_normalization_mode::normalized, + cl::sycl::addressing_mode::clamp, cl::sycl::filtering_mode::linear); + cl::sycl::sampler second( + cl::sycl::coordinate_normalization_mode::unnormalized, + cl::sycl::addressing_mode::clamp_to_edge, + cl::sycl::filtering_mode::nearest); + cl::sycl::program prog(Queue.get_context()); + prog.build_with_source( + "kernel void sampler_args(int a, sampler_t first, " + "int b, sampler_t second, int c) {}\n"); + cl::sycl::kernel krn = prog.get_kernel("sampler_args"); + + Queue.submit([&](cl::sycl::handler &cgh) { + cgh.set_args(0, first, 2, second, 3); + cgh.single_task(krn); + }); + } + } + } + // Parallel for with range + { + cl::sycl::queue q; + std::vector dataVec(10); + std::iota(dataVec.begin(), dataVec.end(), 0); + + if (!q.is_host()) { + cl_int err; + { + cl::sycl::context ctx = q.get_context(); + cl_context clCtx = ctx.get(); + cl_command_queue clQ = q.get(); + cl_mem clBuffer = clCreateBuffer( + clCtx, CL_MEM_WRITE_ONLY, sizeof(int) * dataVec.size(), NULL, NULL); + err = clEnqueueWriteBuffer(clQ, clBuffer, CL_TRUE, 0, + sizeof(int) * dataVec.size(), dataVec.data(), + 0, NULL, NULL); + assert(err == CL_SUCCESS); + + cl::sycl::program prog(ctx); + prog.build_with_source( + "kernel void ParallelFor(__global int* a, int v, __local int *l) " + "{ size_t index = get_global_id(0); l[index] = a[index];" + " l[index] += v; a[index] = l[index]; }\n"); + + q.submit([&](cl::sycl::handler &cgh) { + const int value = 1; + auto local_acc = + cl::sycl::accessor( + cl::sycl::range<1>(10), cgh); + cgh.set_args(clBuffer, value, local_acc); + cgh.parallel_for(cl::sycl::range<1>(10), + prog.get_kernel("ParallelFor")); + }); + + q.wait(); + err = clEnqueueReadBuffer(clQ, clBuffer, CL_TRUE, 0, + sizeof(int) * dataVec.size(), dataVec.data(), + 0, NULL, NULL); + clReleaseCommandQueue(clQ); + clReleaseContext(clCtx); + assert(err == CL_SUCCESS); + for (size_t i = 0; i < dataVec.size(); ++i) { + assert(dataVec[i] == i + 1); + } + } + } + } + + // Parallel for with nd_range + { + cl::sycl::queue q; + std::vector dataVec(10); + std::iota(dataVec.begin(), dataVec.end(), 0); + + if (!q.is_host()) { + cl_int err; + { + cl::sycl::context ctx = q.get_context(); + cl_context clCtx = ctx.get(); + cl_command_queue clQ = q.get(); + cl_mem clBuffer = clCreateBuffer( + clCtx, CL_MEM_WRITE_ONLY, sizeof(int) * dataVec.size(), NULL, NULL); + err = clEnqueueWriteBuffer(clQ, clBuffer, CL_TRUE, 0, + sizeof(int) * dataVec.size(), dataVec.data(), + 0, NULL, NULL); + assert(err == CL_SUCCESS); + + cl::sycl::program prog(ctx); + prog.build_with_source( + "kernel void ParallelForND( local int* l,global int* a)" + "{ size_t idx = get_global_id(0);" + " int pos = idx & 1;" + " int opp = pos ^ 1;" + " l[pos] = a[get_global_id(0)];" + " barrier(CLK_LOCAL_MEM_FENCE);" + " a[idx]=l[opp]; }"); + + // TODO is there no way to set local memory size via interoperability? + cl::sycl::kernel krn = prog.get_kernel("ParallelForND"); + clSetKernelArg(krn.get(), 0, sizeof(int) * 2, NULL); + + q.submit([&](cl::sycl::handler &cgh) { + cgh.set_arg(1, clBuffer); + cgh.parallel_for(cl::sycl::nd_range<1>(cl::sycl::range<1>(10), + cl::sycl::range<1>(2)), + krn); + }); + + q.wait(); + err = clEnqueueReadBuffer(clQ, clBuffer, CL_TRUE, 0, + sizeof(int) * dataVec.size(), dataVec.data(), + 0, NULL, NULL); + clReleaseCommandQueue(clQ); + clReleaseContext(clCtx); + assert(err == CL_SUCCESS); + } + for (size_t i = 0; i < dataVec.size(); ++i) { + assert(dataVec[i] == (i ^ 1)); + } + } + } +} diff --git a/sycl/test/kernel-and-program/kernel-and-program.cpp b/sycl/test/kernel-and-program/kernel-and-program.cpp index 8d3f57eb5a671..ece94d9d4b38c 100644 --- a/sycl/test/kernel-and-program/kernel-and-program.cpp +++ b/sycl/test/kernel-and-program/kernel-and-program.cpp @@ -1,5 +1,3 @@ -// REQUIRES: opencl - // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out -L %opencl_libs_dir -lOpenCL // RUN: env SYCL_DEVICE_TYPE=HOST %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out @@ -54,77 +52,6 @@ int main() { } assert(data == 1); - // OpenCL interoperability kernel invocation - if (!q.is_host()) { - { - cl_int err; - cl::sycl::context ctx = q.get_context(); - cl_context clCtx = ctx.get(); - cl_command_queue clQ = q.get(); - cl_mem clBuffer = - clCreateBuffer(clCtx, CL_MEM_WRITE_ONLY, sizeof(int), NULL, NULL); - err = clEnqueueWriteBuffer(clQ, clBuffer, CL_TRUE, 0, sizeof(int), - &data, 0, NULL, NULL); - assert(err == CL_SUCCESS); - clFinish(clQ); - cl::sycl::program prog(ctx); - prog.build_with_source( - "kernel void SingleTask(global int* a) {*a+=1; }\n"); - q.submit([&](cl::sycl::handler &cgh) { - cgh.set_args(clBuffer); - cgh.single_task(prog.get_kernel("SingleTask")); - }); - q.wait(); - err = clEnqueueReadBuffer(clQ, clBuffer, CL_TRUE, 0, sizeof(int), &data, - 0, NULL, NULL); - clReleaseCommandQueue(clQ); - clReleaseContext(clCtx); - assert(err == CL_SUCCESS); - assert(data == 2); - } - { - cl::sycl::queue sycl_queue; - cl::sycl::program prog(sycl_queue.get_context()); - prog.build_with_source("kernel void foo(global int* a, global int* b, " - "global int* c) {*a=*b+*c; }\n"); - int a = 13, b = 14, c = 15; - { - cl::sycl::buffer bufa(&a, cl::sycl::range<1>(1)); - cl::sycl::buffer bufb(&b, cl::sycl::range<1>(1)); - cl::sycl::buffer bufc(&c, cl::sycl::range<1>(1)); - sycl_queue.submit([&](cl::sycl::handler &cgh) { - auto A = bufa.get_access(cgh); - auto B = bufb.get_access(cgh); - auto C = bufc.get_access(cgh); - cgh.set_args(A, B, C); - cgh.single_task(prog.get_kernel("foo")); - }); - } - assert(a == b + c); - } - } - { - cl::sycl::queue Queue; - if (!Queue.is_host()) { - cl::sycl::sampler first( - cl::sycl::coordinate_normalization_mode::normalized, - cl::sycl::addressing_mode::clamp, cl::sycl::filtering_mode::linear); - cl::sycl::sampler second( - cl::sycl::coordinate_normalization_mode::unnormalized, - cl::sycl::addressing_mode::clamp_to_edge, - cl::sycl::filtering_mode::nearest); - cl::sycl::program prog(Queue.get_context()); - prog.build_with_source( - "kernel void sampler_args(int a, sampler_t first, " - "int b, sampler_t second, int c) {}\n"); - cl::sycl::kernel krn = prog.get_kernel("sampler_args"); - - Queue.submit([&](cl::sycl::handler &cgh) { - cgh.set_args(0, first, 2, second, 3); - cgh.single_task(krn); - }); - } - } } // Parallel for with range { @@ -157,50 +84,6 @@ int main() { for (size_t i = 0; i < dataVec.size(); ++i) { assert(dataVec[i] == i + 1); } - - // OpenCL interoperability kernel invocation - if (!q.is_host()) { - cl_int err; - { - cl::sycl::context ctx = q.get_context(); - cl_context clCtx = ctx.get(); - cl_command_queue clQ = q.get(); - cl_mem clBuffer = clCreateBuffer( - clCtx, CL_MEM_WRITE_ONLY, sizeof(int) * dataVec.size(), NULL, NULL); - err = clEnqueueWriteBuffer(clQ, clBuffer, CL_TRUE, 0, - sizeof(int) * dataVec.size(), dataVec.data(), - 0, NULL, NULL); - assert(err == CL_SUCCESS); - - cl::sycl::program prog(ctx); - prog.build_with_source( - "kernel void ParallelFor(__global int* a, int v, __local int *l) " - "{ size_t index = get_global_id(0); l[index] = a[index];" - " l[index] += v; a[index] = l[index]; }\n"); - - q.submit([&](cl::sycl::handler &cgh) { - const int value = 1; - auto local_acc = - cl::sycl::accessor( - cl::sycl::range<1>(10), cgh); - cgh.set_args(clBuffer, value, local_acc); - cgh.parallel_for(cl::sycl::range<1>(10), - prog.get_kernel("ParallelFor")); - }); - - q.wait(); - err = clEnqueueReadBuffer(clQ, clBuffer, CL_TRUE, 0, - sizeof(int) * dataVec.size(), dataVec.data(), - 0, NULL, NULL); - clReleaseCommandQueue(clQ); - clReleaseContext(clCtx); - assert(err == CL_SUCCESS); - for (size_t i = 0; i < dataVec.size(); ++i) { - assert(dataVec[i] == i + 2); - } - } - } } // Parallel for with nd_range @@ -250,53 +133,5 @@ int main() { assert(dataVec[i] == (i ^ 1)); } } - - // OpenCL interoperability kernel invocation - if (!q.is_host()) { - cl_int err; - { - cl::sycl::context ctx = q.get_context(); - cl_context clCtx = ctx.get(); - cl_command_queue clQ = q.get(); - cl_mem clBuffer = clCreateBuffer( - clCtx, CL_MEM_WRITE_ONLY, sizeof(int) * dataVec.size(), NULL, NULL); - err = clEnqueueWriteBuffer(clQ, clBuffer, CL_TRUE, 0, - sizeof(int) * dataVec.size(), dataVec.data(), - 0, NULL, NULL); - assert(err == CL_SUCCESS); - - cl::sycl::program prog(ctx); - prog.build_with_source( - "kernel void ParallelForND( local int* l,global int* a)" - "{ size_t idx = get_global_id(0);" - " int pos = idx & 1;" - " int opp = pos ^ 1;" - " l[pos] = a[get_global_id(0)];" - " barrier(CLK_LOCAL_MEM_FENCE);" - " a[idx]=l[opp]; }"); - - // TODO is there no way to set local memory size via interoperability? - cl::sycl::kernel krn = prog.get_kernel("ParallelForND"); - clSetKernelArg(krn.get(), 0, sizeof(int) * 2, NULL); - - q.submit([&](cl::sycl::handler &cgh) { - cgh.set_arg(1, clBuffer); - cgh.parallel_for(cl::sycl::nd_range<1>(cl::sycl::range<1>(10), - cl::sycl::range<1>(2)), - krn); - }); - - q.wait(); - err = clEnqueueReadBuffer(clQ, clBuffer, CL_TRUE, 0, - sizeof(int) * dataVec.size(), dataVec.data(), - 0, NULL, NULL); - clReleaseCommandQueue(clQ); - clReleaseContext(clCtx); - assert(err == CL_SUCCESS); - } - for (size_t i = 0; i < dataVec.size(); ++i) { - assert(dataVec[i] == i); - } - } } } From 0966142a1f13046f1a8b7609debe961748b9c14c Mon Sep 17 00:00:00 2001 From: rbegam Date: Mon, 20 Jul 2020 16:34:00 -0700 Subject: [PATCH 2/7] removes redundant opencl requirement and comments. Signed-off-by: rbegam --- sycl/test/inorder_queue/in_order_dmemll.cpp | 21 --------------------- 1 file changed, 21 deletions(-) diff --git a/sycl/test/inorder_queue/in_order_dmemll.cpp b/sycl/test/inorder_queue/in_order_dmemll.cpp index e856f828f9e99..d51316fd08641 100644 --- a/sycl/test/inorder_queue/in_order_dmemll.cpp +++ b/sycl/test/inorder_queue/in_order_dmemll.cpp @@ -1,5 +1,3 @@ -// R-EQUIRES: opencl - // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t1.out -L %opencl_libs_dir -lOpenCL // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out @@ -19,15 +17,6 @@ using namespace cl::sycl; constexpr int numNodes = 4; -/*bool getQueueOrder(cl_command_queue cq) { - cl_command_queue_properties reportedProps; - cl_int iRet = clGetCommandQueueInfo( - cq, CL_QUEUE_PROPERTIES, sizeof(reportedProps), &reportedProps, nullptr); - assert(CL_SUCCESS == iRet && "Failed to obtain queue info from ocl device"); - return (reportedProps & CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE) ? false - : true; -}*/ - struct Node { Node() : pNext(nullptr), Num(0xDEADBEEF) {} @@ -103,15 +92,5 @@ int main() { d_cur = h_cur.pNext; } - /*bool result = true; - cl_command_queue cq = q.get(); - bool expected_result = dev.is_host() ? true : getQueueOrder(cq); - if (expected_result != result) { - std::cout << "Resulting queue order is OOO but expected order is inorder" - << std::endl; - - return -1; - }*/ - return 0; } From 1cb78c1a3e7d88e1c6ba1679a4bd9af5b8966111 Mon Sep 17 00:00:00 2001 From: rbegam Date: Mon, 20 Jul 2020 16:41:49 -0700 Subject: [PATCH 3/7] adds a new file. Signed-off-by: rbegam --- sycl/test/basic_tests/event.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/test/basic_tests/event.cpp b/sycl/test/basic_tests/event.cpp index e8e30015079f7..e4884562f9b82 100644 --- a/sycl/test/basic_tests/event.cpp +++ b/sycl/test/basic_tests/event.cpp @@ -1,4 +1,4 @@ -// REQUIRES: opencl +// REQUIRES: opencl || level0 // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out -L %opencl_libs_dir -lOpenCL // RUN: env SYCL_DEVICE_TYPE=HOST %t.out From adfb330b7ae934893934f38378f9317045533238 Mon Sep 17 00:00:00 2001 From: rbegam Date: Tue, 21 Jul 2020 16:25:08 -0700 Subject: [PATCH 4/7] clang formatted. Signed-off-by: rbegam --- sycl/test/inorder_queue/in_order_buffs_ocl.cpp | 1 - sycl/test/inorder_queue/in_order_dmemll_ocl.cpp | 2 +- 2 files changed, 1 insertion(+), 2 deletions(-) diff --git a/sycl/test/inorder_queue/in_order_buffs_ocl.cpp b/sycl/test/inorder_queue/in_order_buffs_ocl.cpp index 46dc6fed4fe06..d42bdd2ec4e02 100644 --- a/sycl/test/inorder_queue/in_order_buffs_ocl.cpp +++ b/sycl/test/inorder_queue/in_order_buffs_ocl.cpp @@ -48,4 +48,3 @@ int main() { return 0; } - diff --git a/sycl/test/inorder_queue/in_order_dmemll_ocl.cpp b/sycl/test/inorder_queue/in_order_dmemll_ocl.cpp index 424bf9c64419d..e5fb04e2b1e96 100644 --- a/sycl/test/inorder_queue/in_order_dmemll_ocl.cpp +++ b/sycl/test/inorder_queue/in_order_dmemll_ocl.cpp @@ -33,7 +33,7 @@ int main() { auto dev = q.get_device(); bool result = true; - cl_command_queue cq = q.get(); + cl_command_queue cq = q.get(); bool expected_result = dev.is_host() ? true : getQueueOrder(cq); if (expected_result != result) { std::cout << "Resulting queue order is OOO but expected order is inorder" From 9cc541f69df6a2bab8ef6b2d4549147d3fdef591 Mon Sep 17 00:00:00 2001 From: rbegam Date: Tue, 21 Jul 2020 16:37:39 -0700 Subject: [PATCH 5/7] corrects RUN commands. Signed-off-by: rbegam --- sycl/test/inorder_queue/in_order_buffs.cpp | 2 +- sycl/test/inorder_queue/in_order_dmemll.cpp | 2 +- sycl/test/kernel-and-program/kernel-and-program.cpp | 2 +- 3 files changed, 3 insertions(+), 3 deletions(-) diff --git a/sycl/test/inorder_queue/in_order_buffs.cpp b/sycl/test/inorder_queue/in_order_buffs.cpp index c92868fdd742c..be96e3b849f2a 100644 --- a/sycl/test/inorder_queue/in_order_buffs.cpp +++ b/sycl/test/inorder_queue/in_order_buffs.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out -L %opencl_libs_dir -lOpenCL +// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out diff --git a/sycl/test/inorder_queue/in_order_dmemll.cpp b/sycl/test/inorder_queue/in_order_dmemll.cpp index d51316fd08641..4ea98ba4ba419 100644 --- a/sycl/test/inorder_queue/in_order_dmemll.cpp +++ b/sycl/test/inorder_queue/in_order_dmemll.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t1.out -L %opencl_libs_dir -lOpenCL +// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t1.out // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out //==----------- ordered_dmemll.cpp - Device Memory Linked List test --------==// diff --git a/sycl/test/kernel-and-program/kernel-and-program.cpp b/sycl/test/kernel-and-program/kernel-and-program.cpp index ece94d9d4b38c..1233ec3664ee6 100644 --- a/sycl/test/kernel-and-program/kernel-and-program.cpp +++ b/sycl/test/kernel-and-program/kernel-and-program.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out -L %opencl_libs_dir -lOpenCL +// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out // RUN: env SYCL_DEVICE_TYPE=HOST %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // RUNx: %GPU_RUN_PLACEHOLDER %t.out From 2a079d9f745156554f8b433ac951192a45139a21 Mon Sep 17 00:00:00 2001 From: rbegam Date: Wed, 22 Jul 2020 12:01:45 -0700 Subject: [PATCH 6/7] handles cuda fail. Signed-off-by: rbegam --- sycl/test/inorder_queue/in_order_dmemll.cpp | 2 ++ 1 file changed, 2 insertions(+) diff --git a/sycl/test/inorder_queue/in_order_dmemll.cpp b/sycl/test/inorder_queue/in_order_dmemll.cpp index 4ea98ba4ba419..3eb6ecc39a34c 100644 --- a/sycl/test/inorder_queue/in_order_dmemll.cpp +++ b/sycl/test/inorder_queue/in_order_dmemll.cpp @@ -1,6 +1,8 @@ // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t1.out // RUN: %CPU_RUN_PLACEHOLDER %t1.out // RUN: %GPU_RUN_PLACEHOLDER %t1.out +// +// XFAIL: cuda //==----------- ordered_dmemll.cpp - Device Memory Linked List test --------==// // It uses an ordered queue where explicit waiting is not necessary between // kernels From 1bc78a51ba4bde809d82fead8c784b692d733af4 Mon Sep 17 00:00:00 2001 From: rbegam Date: Thu, 30 Jul 2020 11:10:12 -0700 Subject: [PATCH 7/7] [SYCL] update test. Signed-off-by: rbegam --- sycl/test/basic_tests/event.cpp | 3 +++ 1 file changed, 3 insertions(+) diff --git a/sycl/test/basic_tests/event.cpp b/sycl/test/basic_tests/event.cpp index e4884562f9b82..be6ba2f3aa43a 100644 --- a/sycl/test/basic_tests/event.cpp +++ b/sycl/test/basic_tests/event.cpp @@ -2,6 +2,9 @@ // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out -L %opencl_libs_dir -lOpenCL // RUN: env SYCL_DEVICE_TYPE=HOST %t.out +// RUN: %CPU_RUN_PLACEHOLDER %t.out +// RUN: %GPU_RUN_PLACEHOLDER %t.out + //==--------------- event.cpp - SYCL event test ----------------------------==// // // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.