diff --git a/test_conformance/extensions/cl_khr_command_buffer/cl_khr_command_buffer_mutable_dispatch/mutable_command_basic.h b/test_conformance/extensions/cl_khr_command_buffer/cl_khr_command_buffer_mutable_dispatch/mutable_command_basic.h index 51938dce02..c1697accbd 100644 --- a/test_conformance/extensions/cl_khr_command_buffer/cl_khr_command_buffer_mutable_dispatch/mutable_command_basic.h +++ b/test_conformance/extensions/cl_khr_command_buffer/cl_khr_command_buffer_mutable_dispatch/mutable_command_basic.h @@ -126,7 +126,7 @@ struct BasicMutableCommandBufferTest : BasicCommandBufferTest clUpdateMutableCommandsKHR_fn clUpdateMutableCommandsKHR = nullptr; const char* kernelString = "__kernel void empty() {}"; - const size_t global_work_size = 4 * 16; + const size_t global_work_size = 256 * 1024; }; struct InfoMutableCommandBufferTest : BasicMutableCommandBufferTest diff --git a/test_conformance/extensions/cl_khr_command_buffer/cl_khr_command_buffer_mutable_dispatch/mutable_command_global_size.cpp b/test_conformance/extensions/cl_khr_command_buffer/cl_khr_command_buffer_mutable_dispatch/mutable_command_global_size.cpp index 946fa995b3..22a58ff593 100644 --- a/test_conformance/extensions/cl_khr_command_buffer/cl_khr_command_buffer_mutable_dispatch/mutable_command_global_size.cpp +++ b/test_conformance/extensions/cl_khr_command_buffer/cl_khr_command_buffer_mutable_dispatch/mutable_command_global_size.cpp @@ -153,9 +153,9 @@ struct MutableDispatchGlobalSize : public InfoMutableCommandBufferTest } size_t info_global_size = 0; - const size_t update_global_size = 3; - const size_t sizeToAllocate = global_work_size; - const size_t num_elements = sizeToAllocate / sizeof(cl_int); + const size_t update_global_size = 16 * 1024; + const size_t sizeToAllocate = global_work_size * sizeof(cl_int); + const size_t num_elements = global_work_size; cl_mutable_command_khr command = nullptr; }; diff --git a/test_conformance/extensions/cl_khr_command_buffer/cl_khr_command_buffer_mutable_dispatch/mutable_command_info.cpp b/test_conformance/extensions/cl_khr_command_buffer/cl_khr_command_buffer_mutable_dispatch/mutable_command_info.cpp index 2b5b3b42e8..4dd7d66efb 100644 --- a/test_conformance/extensions/cl_khr_command_buffer/cl_khr_command_buffer_mutable_dispatch/mutable_command_info.cpp +++ b/test_conformance/extensions/cl_khr_command_buffer/cl_khr_command_buffer_mutable_dispatch/mutable_command_info.cpp @@ -120,10 +120,10 @@ struct PropertiesArray : public InfoMutableCommandBufferTest virtual bool Skip() override { Version device_version = get_device_cl_version(device); + if (InfoMutableCommandBufferTest::Skip()) return true; if ((device_version >= Version(3, 0)) || is_extension_available(device, "cl_khr_extended_versioning")) { - cl_version extension_version = get_extension_version( device, "cl_khr_command_buffer_mutable_dispatch"); @@ -132,7 +132,17 @@ struct PropertiesArray : public InfoMutableCommandBufferTest return true; } } - return InfoMutableCommandBufferTest::Skip(); + + cl_mutable_dispatch_fields_khr mutable_capabilities; + cl_int error = clGetDeviceInfo( + device, CL_DEVICE_MUTABLE_DISPATCH_CAPABILITIES_KHR, + sizeof(mutable_capabilities), &mutable_capabilities, nullptr); + test_error(error, "clGetDeviceInfo failed"); + + if ((mutable_capabilities & CL_MUTABLE_DISPATCH_ARGUMENTS_KHR) == 0) + return true; + + return false; } cl_int Run() override diff --git a/test_conformance/extensions/cl_khr_command_buffer/cl_khr_command_buffer_mutable_dispatch/mutable_command_work_groups.cpp b/test_conformance/extensions/cl_khr_command_buffer/cl_khr_command_buffer_mutable_dispatch/mutable_command_work_groups.cpp index ad20fbe3b2..03b42e0df2 100644 --- a/test_conformance/extensions/cl_khr_command_buffer/cl_khr_command_buffer_mutable_dispatch/mutable_command_work_groups.cpp +++ b/test_conformance/extensions/cl_khr_command_buffer/cl_khr_command_buffer_mutable_dispatch/mutable_command_work_groups.cpp @@ -278,10 +278,9 @@ struct MutableDispatchWorkGroups : public BasicMutableCommandBufferTest Configuration config; size_t info_global_size = 0; - static constexpr size_t test_global_work_size = 64; - static constexpr size_t update_global_size = 16; + static constexpr size_t update_global_size = 16 * 1024; const size_t local_work_size = 8; - const size_t sizeToAllocate = 64 * sizeof(cl_int); + const size_t sizeToAllocate = global_work_size * sizeof(cl_int); }; int test_command_buffer_with_no_additional_work_groups(cl_device_id device,