ahmed/OpenCL-CTS
1
0
Fork 0
mirror of https://github.com/KhronosGroup/OpenCL-CTS.git synced 2026-03-25 08:19:02 +00:00
Code Packages Projects Releases Wiki Activity

Semaphore types bug fixes revised (#1822)

Browse Source 
* Added support for SYNC_FD and other handle types

* Fix consistency test

Deleted test cases that are no longer testable
according to the spec.

* Fix multi-import tests

-Delete obsolete code relating to offsets
-Propagate dedicated memory change

* Fix error handling

Some subtests did not fail on incorrect result.
Changes to macros to fail, so this does not occur
again.

* Delete invalid test cases

Test cases are not related to this extension.

* External memory test

Add support for any handle type supported by
the platform.

Change-Id: I6765fde5e7929988f49bfbf2df2f41d5263b6abc

* Update multi-import tests to use new semaphore types

* Fix formatting

* Addressed review comments. Deleted VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_NT_KMT as it appears to be redundant.
This commit is contained in:
joshqti
2023-11-29 02:32:59 -08:00
committed by GitHub
parent 5815e2ce33
commit f5bd92b83e
16 changed files with 1542 additions and 1451 deletions
Download Patch File Download Diff File Expand all files Collapse all files

15
test_common/harness/errorHelpers.h
View File

@@ -61,6 +61,21 @@ static int vlog_win32(const char *format, ...);
log_error(msg, ##__VA_ARGS__); \ log_error(msg, ##__VA_ARGS__); \
return TEST_FAIL; \ return TEST_FAIL; \
} }
#define test_fail_and_cleanup(errRet, cleanup, msg, ...) \
{ \
log_error(msg, ##__VA_ARGS__); \
errRet = TEST_FAIL; \
goto cleanup; \
}
#define test_error_and_cleanup(errCode, cleanup, msg, ...) \
{ \
auto errCodeResult = errCode; \
if (errCodeResult != CL_SUCCESS) \
{ \
print_error(errCodeResult, msg); \
goto cleanup; \
} \
}
#define test_error(errCode, msg) test_error_ret(errCode, msg, errCode) #define test_error(errCode, msg) test_error_ret(errCode, msg, errCode)
#define test_error_fail(errCode, msg) test_error_ret(errCode, msg, TEST_FAIL) #define test_error_fail(errCode, msg) test_error_ret(errCode, msg, TEST_FAIL)
#define test_error_ret(errCode, msg, retValue) \ #define test_error_ret(errCode, msg, retValue) \

156
test_conformance/common/vulkan_wrapper/opencl_vulkan_wrapper.cpp
View File

@@ -33,6 +33,7 @@ pfnclEnqueueAcquireExternalMemObjectsKHR
pfnclEnqueueReleaseExternalMemObjectsKHR pfnclEnqueueReleaseExternalMemObjectsKHR
clEnqueueReleaseExternalMemObjectsKHRptr; clEnqueueReleaseExternalMemObjectsKHRptr;
pfnclReleaseSemaphoreKHR clReleaseSemaphoreKHRptr; pfnclReleaseSemaphoreKHR clReleaseSemaphoreKHRptr;
pfnclGetSemaphoreHandleForTypeKHR clGetSemaphoreHandleForTypeKHRptr;
void init_cl_vk_ext(cl_platform_id opencl_platform) void init_cl_vk_ext(cl_platform_id opencl_platform)
{ {
@@ -69,6 +70,15 @@ void init_cl_vk_ext(cl_platform_id opencl_platform)
throw std::runtime_error("Failed to get the function pointer of " throw std::runtime_error("Failed to get the function pointer of "
"clCreateSemaphoreWithPropertiesKHRptr!"); "clCreateSemaphoreWithPropertiesKHRptr!");
} }
clGetSemaphoreHandleForTypeKHRptr = (pfnclGetSemaphoreHandleForTypeKHR)
clGetExtensionFunctionAddressForPlatform(
opencl_platform, "clGetSemaphoreHandleForTypeKHR");
if (NULL == clGetSemaphoreHandleForTypeKHRptr)
{
throw std::runtime_error("Failed to get the function pointer of "
"clGetSemaphoreHandleForTypeKHRptr!");
}
} }
cl_int setMaxImageDimensions(cl_device_id deviceID, size_t &max_width, cl_int setMaxImageDimensions(cl_device_id deviceID, size_t &max_width,
@@ -522,8 +532,8 @@ clExternalMemory::clExternalMemory(const clExternalMemory &externalMemory)
clExternalMemory::clExternalMemory( clExternalMemory::clExternalMemory(
const VulkanDeviceMemory *deviceMemory, const VulkanDeviceMemory *deviceMemory,
VulkanExternalMemoryHandleType externalMemoryHandleType, uint64_t offset, VulkanExternalMemoryHandleType externalMemoryHandleType, uint64_t size,
uint64_t size, cl_context context, cl_device_id deviceId) cl_context context, cl_device_id deviceId)
{ {
int err = 0; int err = 0;
m_externalMemory = NULL; m_externalMemory = NULL;
@@ -548,9 +558,9 @@ clExternalMemory::clExternalMemory(
{ {
case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_FD: case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_FD:
#ifdef _WIN32 #ifdef _WIN32
log_info("Opaque file descriptors are not supported on Windows\n");
ASSERT(0); ASSERT(0);
#endif #endif
log_info("Opaque file descriptors are not supported on Windows\n");
fd = (int)deviceMemory->getHandle(externalMemoryHandleType); fd = (int)deviceMemory->getHandle(externalMemoryHandleType);
err = check_external_memory_handle_type( err = check_external_memory_handle_type(
devList[0], CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR); devList[0], CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR);
@@ -728,56 +738,19 @@ clExternalMemoryImage::clExternalMemoryImage() {}
// clExternalSemaphore implementation // // clExternalSemaphore implementation //
////////////////////////////////////////// //////////////////////////////////////////
clExternalSemaphore::clExternalSemaphore(
const clExternalSemaphore &externalSemaphore)
: m_externalSemaphore(externalSemaphore.m_externalSemaphore)
{}
clExternalSemaphore::clExternalSemaphore( clExternalSemaphore::clExternalSemaphore(
const VulkanSemaphore &semaphore, cl_context context, const VulkanSemaphore &semaphore, cl_context context,
VulkanExternalSemaphoreHandleType externalSemaphoreHandleType, VulkanExternalSemaphoreHandleType externalSemaphoreHandleType,
cl_device_id deviceId) cl_device_id deviceId)
: m_deviceSemaphore(semaphore)
{ {
cl_int err = 0; cl_int err = 0;
cl_device_id devList[] = { deviceId, NULL }; cl_device_id devList[] = { deviceId, NULL };
m_externalHandleType = externalSemaphoreHandleType;
switch (externalSemaphoreHandleType) m_externalSemaphore = nullptr;
{ m_device = deviceId;
case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_FD: m_context = context;
if (!is_extension_available(devList[0],
"cl_khr_external_semaphore_opaque_fd"))
{
throw std::runtime_error("Device does not support "
"cl_khr_external_semaphore_opaque_fd "
"extension \n");
}
break;
case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_NT:
case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_KMT:
case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_NT_KMT:
if (!is_extension_available(devList[0],
"cl_khr_external_semaphore_win32"))
{
throw std::runtime_error(
"Device does not support "
"cl_khr_external_semaphore_win32 extension\n");
}
break;
case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_SYNC_FD:
if (!is_extension_available(devList[0],
"cl_khr_external_semaphore_sync_fd"))
{
throw std::runtime_error(
"Device does not support cl_khr_external_semaphore_sync_fd "
"extension \n");
}
break;
default:
throw std::runtime_error(
"Unsupported external semaphore handle type\n");
break;
}
std::vector<cl_semaphore_properties_khr> sema_props{ std::vector<cl_semaphore_properties_khr> sema_props{
(cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_KHR, (cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_KHR,
@@ -873,16 +846,97 @@ clExternalSemaphore::~clExternalSemaphore() noexcept(false)
} }
} }
void clExternalSemaphore::signal(cl_command_queue cmd_queue) int clExternalSemaphore::signal(cl_command_queue cmd_queue)
{ {
clEnqueueSignalSemaphoresKHRptr(cmd_queue, 1, &m_externalSemaphore, NULL, 0, int err = clEnqueueSignalSemaphoresKHRptr(
NULL, NULL); cmd_queue, 1, &m_externalSemaphore, NULL, 0, NULL, nullptr);
if (err != CL_SUCCESS)
{
return err;
}
if (m_externalHandleType == VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_SYNC_FD)
{
err = clGetSemaphoreHandleForTypeKHRptr(m_externalSemaphore, m_device,
CL_SEMAPHORE_HANDLE_SYNC_FD_KHR,
sizeof(int), &fd, nullptr);
if (err != CL_SUCCESS)
{
log_error("Failed to export fd from semaphore\n");
return err;
}
VkImportSemaphoreFdInfoKHR import = {};
import.sType = VK_STRUCTURE_TYPE_IMPORT_SEMAPHORE_FD_INFO_KHR;
import.semaphore = m_deviceSemaphore;
import.fd = fd;
import.pNext = nullptr;
import.handleType = VK_EXTERNAL_SEMAPHORE_HANDLE_TYPE_SYNC_FD_BIT_KHR;
import.flags = 0;
VkResult res =
vkImportSemaphoreFdKHR(m_deviceSemaphore.getDevice(), &import);
ASSERT(res == VK_SUCCESS);
if (res != VK_SUCCESS)
{
err = CL_INVALID_OPERATION;
}
}
return err;
} }
void clExternalSemaphore::wait(cl_command_queue cmd_queue) int clExternalSemaphore::wait(cl_command_queue cmd_queue)
{ {
clEnqueueWaitSemaphoresKHRptr(cmd_queue, 1, &m_externalSemaphore, NULL, 0, int err = CL_SUCCESS;
NULL, NULL); if (m_externalHandleType == VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_SYNC_FD)
{
cl_int err = 0;
cl_device_id devList[] = { m_device, NULL };
std::vector<cl_semaphore_properties_khr> sema_props{
(cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_KHR,
(cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_BINARY_KHR,
};
fd = (int)m_deviceSemaphore.getHandle(m_externalHandleType);
err = check_external_semaphore_handle_type(
devList[0], CL_SEMAPHORE_HANDLE_SYNC_FD_KHR);
if (CL_SUCCESS != err)
{
log_error("CL_SEMAPHORE_HANDLE_SYNC_FD_KHR not supported\n");
return err;
}
sema_props.push_back(
(cl_semaphore_properties_khr)CL_SEMAPHORE_HANDLE_SYNC_FD_KHR);
sema_props.push_back((cl_semaphore_properties_khr)fd);
sema_props.push_back(0);
if (m_externalSemaphore)
{
err = clReleaseSemaphoreKHRptr(m_externalSemaphore);
if (err != CL_SUCCESS)
{
log_error("Failed to release CL external semaphore\n");
return err;
}
m_externalSemaphore = nullptr;
}
m_externalSemaphore = clCreateSemaphoreWithPropertiesKHRptr(
m_context, sema_props.data(), &err);
if (CL_SUCCESS != err)
{
log_error("clCreateSemaphoreWithPropertiesKHRptr failed with %d\n",
err);
return err;
}
}
err = clEnqueueWaitSemaphoresKHRptr(cmd_queue, 1, &m_externalSemaphore,
NULL, 0, NULL, NULL);
return err;
} }
cl_semaphore_khr &clExternalSemaphore::getCLSemaphore() cl_semaphore_khr &clExternalSemaphore::getCLSemaphore()

16
test_conformance/common/vulkan_wrapper/opencl_vulkan_wrapper.hpp
View File

@@ -50,6 +50,10 @@ typedef cl_int (*pfnclEnqueueReleaseExternalMemObjectsKHR)(
const cl_mem *mem_objects, cl_uint num_events_in_wait_list, const cl_mem *mem_objects, cl_uint num_events_in_wait_list,
const cl_event *event_wait_list, cl_event *event); const cl_event *event_wait_list, cl_event *event);
typedef cl_int (*pfnclReleaseSemaphoreKHR)(cl_semaphore_khr sema_object); typedef cl_int (*pfnclReleaseSemaphoreKHR)(cl_semaphore_khr sema_object);
typedef cl_int (*pfnclGetSemaphoreHandleForTypeKHR)(
cl_semaphore_khr sema_object, cl_device_id device,
cl_external_semaphore_handle_type_khr handleType, size_t handle_size,
void *handle, size_t *handleSize);
extern pfnclCreateSemaphoreWithPropertiesKHR extern pfnclCreateSemaphoreWithPropertiesKHR
clCreateSemaphoreWithPropertiesKHRptr; clCreateSemaphoreWithPropertiesKHRptr;
@@ -83,8 +87,7 @@ public:
clExternalMemory(); clExternalMemory();
clExternalMemory(const VulkanDeviceMemory *deviceMemory, clExternalMemory(const VulkanDeviceMemory *deviceMemory,
VulkanExternalMemoryHandleType externalMemoryHandleType, VulkanExternalMemoryHandleType externalMemoryHandleType,
uint64_t offset, uint64_t size, cl_context context, uint64_t size, cl_context context, cl_device_id deviceId);
cl_device_id deviceId);
virtual ~clExternalMemory(); virtual ~clExternalMemory();
cl_mem getExternalMemoryBuffer(); cl_mem getExternalMemoryBuffer();
@@ -111,9 +114,12 @@ public:
class clExternalSemaphore { class clExternalSemaphore {
protected: protected:
cl_semaphore_khr m_externalSemaphore; cl_semaphore_khr m_externalSemaphore;
VulkanExternalSemaphoreHandleType m_externalHandleType;
cl_device_id m_device;
cl_context m_context;
const VulkanSemaphore &m_deviceSemaphore;
int fd; int fd;
void *handle; void *handle;
clExternalSemaphore(const clExternalSemaphore &externalSemaphore);
public: public:
clExternalSemaphore( clExternalSemaphore(
@@ -121,8 +127,8 @@ public:
VulkanExternalSemaphoreHandleType externalSemaphoreHandleType, VulkanExternalSemaphoreHandleType externalSemaphoreHandleType,
cl_device_id deviceId); cl_device_id deviceId);
virtual ~clExternalSemaphore() noexcept(false); virtual ~clExternalSemaphore() noexcept(false);
void signal(cl_command_queue command_queue); int signal(cl_command_queue command_queue);
void wait(cl_command_queue command_queue); int wait(cl_command_queue command_queue);
cl_semaphore_khr &getCLSemaphore(); cl_semaphore_khr &getCLSemaphore();
// operator openclExternalSemaphore_t() const; // operator openclExternalSemaphore_t() const;
}; };

9
test_conformance/common/vulkan_wrapper/vulkan_api_list.hpp
View File

@@ -98,8 +98,9 @@
VK_FUNC_DECL(vkGetPhysicalDeviceSurfaceFormatsKHR) \ VK_FUNC_DECL(vkGetPhysicalDeviceSurfaceFormatsKHR) \
VK_FUNC_DECL(vkGetPhysicalDeviceSurfacePresentModesKHR) \ VK_FUNC_DECL(vkGetPhysicalDeviceSurfacePresentModesKHR) \
VK_FUNC_DECL(vkEnumerateDeviceExtensionProperties) \ VK_FUNC_DECL(vkEnumerateDeviceExtensionProperties) \
VK_FUNC_DECL(vkGetPhysicalDeviceSurfaceSupportKHR) VK_FUNC_DECL(vkGetPhysicalDeviceSurfaceSupportKHR) \
VK_FUNC_DECL(vkImportSemaphoreFdKHR) \
VK_FUNC_DECL(vkGetPhysicalDeviceExternalSemaphorePropertiesKHR)
#define VK_WINDOWS_FUNC_LIST \ #define VK_WINDOWS_FUNC_LIST \
VK_FUNC_DECL(vkGetMemoryWin32HandleKHR) \ VK_FUNC_DECL(vkGetMemoryWin32HandleKHR) \
VK_FUNC_DECL(vkGetSemaphoreWin32HandleKHR) VK_FUNC_DECL(vkGetSemaphoreWin32HandleKHR)
@@ -192,7 +193,9 @@
_vkEnumerateDeviceExtensionProperties _vkEnumerateDeviceExtensionProperties
#define vkGetPhysicalDeviceSurfaceSupportKHR \ #define vkGetPhysicalDeviceSurfaceSupportKHR \
_vkGetPhysicalDeviceSurfaceSupportKHR _vkGetPhysicalDeviceSurfaceSupportKHR
#define vkImportSemaphoreFdKHR _vkImportSemaphoreFdKHR
#define vkGetPhysicalDeviceExternalSemaphorePropertiesKHR \
_vkGetPhysicalDeviceExternalSemaphorePropertiesKHR
#define vkGetMemoryWin32HandleKHR _vkGetMemoryWin32HandleKHR #define vkGetMemoryWin32HandleKHR _vkGetMemoryWin32HandleKHR
#define vkGetSemaphoreWin32HandleKHR _vkGetSemaphoreWin32HandleKHR #define vkGetSemaphoreWin32HandleKHR _vkGetSemaphoreWin32HandleKHR

2
test_conformance/common/vulkan_wrapper/vulkan_list_map.hpp
View File

@@ -336,6 +336,8 @@ const VulkanWrapper &
return (m_wrapperList.size() > 0) ? m_wrapperList[idx].get() return (m_wrapperList.size() > 0) ? m_wrapperList[idx].get()
: m_constWrapperList[idx].get(); : m_constWrapperList[idx].get();
} }
throw std::runtime_error("Out of bounds operator access");
} }
template <class VulkanWrapper, class VulkanNative> template <class VulkanWrapper, class VulkanNative>

118
test_conformance/common/vulkan_wrapper/vulkan_utility.cpp
View File

@@ -24,6 +24,7 @@
#include <algorithm> #include <algorithm>
#include <CL/cl.h> #include <CL/cl.h>
#include <CL/cl_ext.h> #include <CL/cl_ext.h>
#include "deviceInfo.h"
#if defined(_WIN32) || defined(_WIN64) #if defined(_WIN32) || defined(_WIN64)
#include <versionhelpers.h> #include <versionhelpers.h>
#endif #endif
@@ -174,7 +175,7 @@ getVulkanMemoryType(const VulkanDevice &device,
} }
} }
// CHECK_LT(mtIdx, memoryTypeList.size()); ASSERT(mtIdx < memoryTypeList.size());
return memoryTypeList[mtIdx]; return memoryTypeList[mtIdx];
} }
@@ -236,30 +237,112 @@ getSupportedVulkanExternalMemoryHandleTypeList()
} }
const std::vector<VulkanExternalSemaphoreHandleType> const std::vector<VulkanExternalSemaphoreHandleType>
getSupportedVulkanExternalSemaphoreHandleTypeList() getSupportedVulkanExternalSemaphoreHandleTypeList(const VulkanDevice &vkDevice)
{ {
typedef struct
{
const char *extension_name;
VkExternalSemaphoreHandleTypeFlagBits vk_type;
VulkanExternalSemaphoreHandleType enum_type;
} VkSemaphoreHandleMap;
// Add all known handle types, use Vulkan queries to determine what is
// supported.
std::vector<VkSemaphoreHandleMap> all_known_handle_types;
all_known_handle_types.push_back(
{ "VK_KHR_external_semaphore_fd",
VK_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_FD_BIT_KHR,
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_FD });
all_known_handle_types.push_back(
{ "VK_KHR_external_semaphore_fd",
VK_EXTERNAL_SEMAPHORE_HANDLE_TYPE_SYNC_FD_BIT_KHR,
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_SYNC_FD });
all_known_handle_types.push_back(
{ "VK_KHR_external_semaphore_win32",
VK_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_KMT_BIT_KHR,
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_KMT });
all_known_handle_types.push_back(
{ "VK_KHR_external_semaphore_win32",
VK_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_BIT_KHR,
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_NT });
std::vector<VulkanExternalSemaphoreHandleType> std::vector<VulkanExternalSemaphoreHandleType>
externalSemaphoreHandleTypeList; externalSemaphoreHandleTypeList;
#if _WIN32 for (auto handle_type : all_known_handle_types)
if (IsWindows8OrGreater())
{ {
externalSemaphoreHandleTypeList.push_back( if (!vkDevice.getPhysicalDevice().hasExtension(
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_NT); handle_type.extension_name))
{
continue;
}
VkPhysicalDeviceExternalSemaphoreInfo handle_query = {
VK_STRUCTURE_TYPE_PHYSICAL_DEVICE_EXTERNAL_SEMAPHORE_INFO, nullptr,
handle_type.vk_type
};
VkExternalSemaphoreProperties query_result = {};
vkGetPhysicalDeviceExternalSemaphorePropertiesKHR(
vkDevice.getPhysicalDevice(), &handle_query, &query_result);
if (query_result.externalSemaphoreFeatures
& (VK_EXTERNAL_SEMAPHORE_FEATURE_EXPORTABLE_BIT_KHR
| VK_EXTERNAL_SEMAPHORE_FEATURE_IMPORTABLE_BIT_KHR))
{
externalSemaphoreHandleTypeList.push_back(handle_type.enum_type);
}
} }
externalSemaphoreHandleTypeList.push_back(
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_KMT);
#elif defined(__ANDROID__)
externalSemaphoreHandleTypeList.push_back(
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_SYNC_FD);
#else
externalSemaphoreHandleTypeList.push_back(
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_FD);
#endif
return externalSemaphoreHandleTypeList; return externalSemaphoreHandleTypeList;
} }
std::vector<VulkanExternalSemaphoreHandleType>
getSupportedInteropExternalSemaphoreHandleTypes(cl_device_id device,
VulkanDevice &vkDevice)
{
const std::vector<VulkanExternalSemaphoreHandleType>
supportedVkSemaphoreTypes =
getSupportedVulkanExternalSemaphoreHandleTypeList(vkDevice);
std::vector<VulkanExternalSemaphoreHandleType> supportedSemaphoreTypes;
if (is_extension_available(device, "cl_khr_external_semaphore_opaque_fd")
&& std::count(supportedVkSemaphoreTypes.begin(),
supportedVkSemaphoreTypes.end(),
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_FD))
{
supportedSemaphoreTypes.push_back(
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_FD);
}
if (is_extension_available(device, "cl_khr_external_semaphore_sync_fd")
&& std::count(supportedVkSemaphoreTypes.begin(),
supportedVkSemaphoreTypes.end(),
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_SYNC_FD))
{
supportedSemaphoreTypes.push_back(
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_SYNC_FD);
}
if (is_extension_available(device, "cl_khr_external_semaphore_win32")
&& std::count(supportedVkSemaphoreTypes.begin(),
supportedVkSemaphoreTypes.end(),
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_NT))
{
supportedSemaphoreTypes.push_back(
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_NT);
}
if (is_extension_available(device, "cl_khr_external_semaphore_win32")
&& std::count(supportedVkSemaphoreTypes.begin(),
supportedVkSemaphoreTypes.end(),
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_KMT))
{
supportedSemaphoreTypes.push_back(
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_KMT);
}
return supportedSemaphoreTypes;
}
const std::vector<VulkanFormat> getSupportedVulkanFormatList() const std::vector<VulkanFormat> getSupportedVulkanFormatList()
{ {
std::vector<VulkanFormat> formatList; std::vector<VulkanFormat> formatList;
@@ -498,7 +581,6 @@ cl_external_semaphore_handle_type_khr getCLSemaphoreTypeFromVulkanType(
clExternalSemaphoreHandleTypeKhr = clExternalSemaphoreHandleTypeKhr =
CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KHR; CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KHR;
break; break;
case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_NT_KMT:
case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_KMT: case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_KMT:
clExternalSemaphoreHandleTypeKhr = clExternalSemaphoreHandleTypeKhr =
CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KMT_KHR; CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KMT_KHR;
@@ -631,8 +713,8 @@ operator<<(std::ostream &os,
return os << "Opaque NT handle"; return os << "Opaque NT handle";
case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_KMT: case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_KMT:
return os << "Opaque D3DKMT handle"; return os << "Opaque D3DKMT handle";
case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_NT_KMT: case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_SYNC_FD:
return os << "Opaque NT and D3DKMT handle"; return os << "Sync fd semaphore handle";
} }
return os; return os;

5
test_conformance/common/vulkan_wrapper/vulkan_utility.hpp
View File

@@ -46,7 +46,10 @@ getDefaultVulkanQueueFamilyToQueueCountMap();
const std::vector<VulkanExternalMemoryHandleType> const std::vector<VulkanExternalMemoryHandleType>
getSupportedVulkanExternalMemoryHandleTypeList(); getSupportedVulkanExternalMemoryHandleTypeList();
const std::vector<VulkanExternalSemaphoreHandleType> const std::vector<VulkanExternalSemaphoreHandleType>
getSupportedVulkanExternalSemaphoreHandleTypeList(); getSupportedVulkanExternalSemaphoreHandleTypeList(const VulkanDevice& vkDevice);
std::vector<VulkanExternalSemaphoreHandleType>
getSupportedInteropExternalSemaphoreHandleTypes(cl_device_id device,
VulkanDevice& vkDevice);
const std::vector<VulkanFormat> getSupportedVulkanFormatList(); const std::vector<VulkanFormat> getSupportedVulkanFormatList();
uint32_t getVulkanFormatElementSize(VulkanFormat format); uint32_t getVulkanFormatElementSize(VulkanFormat format);

41
test_conformance/common/vulkan_wrapper/vulkan_wrapper.cpp
View File

@@ -335,6 +335,16 @@ VulkanPhysicalDevice::VulkanPhysicalDevice(VkPhysicalDevice vkPhysicalDevice)
memoryHeap); memoryHeap);
m_memoryTypeList.add(*memoryType); m_memoryTypeList.add(*memoryType);
} }
uint32_t num_extensions = 0;
vkEnumerateDeviceExtensionProperties(m_vkPhysicalDevice, nullptr,
&num_extensions, nullptr);
if (num_extensions)
{
m_extensions.resize(num_extensions);
vkEnumerateDeviceExtensionProperties(
m_vkPhysicalDevice, nullptr, &num_extensions, m_extensions.data());
}
} }
VulkanPhysicalDevice::~VulkanPhysicalDevice() VulkanPhysicalDevice::~VulkanPhysicalDevice()
@@ -388,6 +398,18 @@ VulkanPhysicalDevice::operator VkPhysicalDevice() const
return m_vkPhysicalDevice; return m_vkPhysicalDevice;
} }
bool VulkanPhysicalDevice::hasExtension(const char *extension_name) const
{
for (const auto &m_extension : m_extensions)
{
if (!strcmp(m_extension.extensionName, extension_name))
{
return true;
}
}
return false;
}
bool operator<(const VulkanQueueFamily &queueFamilyA, bool operator<(const VulkanQueueFamily &queueFamilyA,
const VulkanQueueFamily &queueFamilyB) const VulkanQueueFamily &queueFamilyB)
{ {
@@ -2256,6 +2278,8 @@ VulkanSemaphore::VulkanSemaphore(
vkCreateSemaphore(m_device, &vkSemaphoreCreateInfo, NULL, &m_vkSemaphore); vkCreateSemaphore(m_device, &vkSemaphoreCreateInfo, NULL, &m_vkSemaphore);
} }
const VulkanDevice &VulkanSemaphore::getDevice() const { return m_device; }
VulkanSemaphore::~VulkanSemaphore() VulkanSemaphore::~VulkanSemaphore()
{ {
vkDestroySemaphore(m_device, m_vkSemaphore, NULL); vkDestroySemaphore(m_device, m_vkSemaphore, NULL);
@@ -2301,6 +2325,23 @@ int VulkanSemaphore::getHandle(
return fd; return fd;
} }
else if (externalSemaphoreHandleType
== VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_SYNC_FD)
{
int fd;
VkSemaphoreGetFdInfoKHR vkSemaphoreGetFdInfoKHR = {};
vkSemaphoreGetFdInfoKHR.sType =
VK_STRUCTURE_TYPE_SEMAPHORE_GET_FD_INFO_KHR;
vkSemaphoreGetFdInfoKHR.pNext = NULL;
vkSemaphoreGetFdInfoKHR.semaphore = m_vkSemaphore;
vkSemaphoreGetFdInfoKHR.handleType =
VK_EXTERNAL_SEMAPHORE_HANDLE_TYPE_SYNC_FD_BIT_KHR;
vkGetSemaphoreFdKHR(m_device, &vkSemaphoreGetFdInfoKHR, &fd);
return fd;
}
return HANDLE_ERROR; return HANDLE_ERROR;
} }
#endif #endif

5
test_conformance/common/vulkan_wrapper/vulkan_wrapper.hpp
View File

@@ -53,12 +53,15 @@ protected:
VulkanQueueFamilyList m_queueFamilyList; VulkanQueueFamilyList m_queueFamilyList;
VulkanMemoryHeapList m_memoryHeapList; VulkanMemoryHeapList m_memoryHeapList;
VulkanMemoryTypeList m_memoryTypeList; VulkanMemoryTypeList m_memoryTypeList;
std::vector<VkExtensionProperties> m_extensions;
VulkanPhysicalDevice(const VulkanPhysicalDevice &physicalDevice); VulkanPhysicalDevice(const VulkanPhysicalDevice &physicalDevice);
VulkanPhysicalDevice(VkPhysicalDevice vkPhysicalDevice); VulkanPhysicalDevice(VkPhysicalDevice vkPhysicalDevice);
virtual ~VulkanPhysicalDevice(); virtual ~VulkanPhysicalDevice();
public: public:
bool hasExtension(const char *extension_name) const;
const VulkanQueueFamilyList &getQueueFamilyList() const; const VulkanQueueFamilyList &getQueueFamilyList() const;
const VulkanMemoryHeapList &getMemoryHeapList() const; const VulkanMemoryHeapList &getMemoryHeapList() const;
const VulkanMemoryTypeList &getMemoryTypeList() const; const VulkanMemoryTypeList &getMemoryTypeList() const;
@@ -537,6 +540,7 @@ protected:
uint64_t m_size; uint64_t m_size;
bool m_isDedicated; bool m_isDedicated;
VulkanDeviceMemory(const VulkanDeviceMemory &deviceMemory); VulkanDeviceMemory(const VulkanDeviceMemory &deviceMemory);
public: public:
@@ -588,6 +592,7 @@ public:
VulkanExternalSemaphoreHandleType externalSemaphoreHandleType = VulkanExternalSemaphoreHandleType externalSemaphoreHandleType =
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_NONE, VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_NONE,
const std::wstring name = L""); const std::wstring name = L"");
const VulkanDevice &getDevice() const;
virtual ~VulkanSemaphore(); virtual ~VulkanSemaphore();
#ifdef _WIN32 #ifdef _WIN32
HANDLE getHandle( HANDLE getHandle(

3
test_conformance/common/vulkan_wrapper/vulkan_wrapper_types.hpp
View File

@@ -167,9 +167,6 @@ enum VulkanExternalSemaphoreHandleType
VK_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_BIT_KHR, VK_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_BIT_KHR,
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_KMT = VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_KMT =
VK_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_KMT_BIT_KHR, VK_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_KMT_BIT_KHR,
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_NT_KMT =
VK_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_BIT_KHR
| VK_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_KMT_BIT_KHR,
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_SYNC_FD = VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_SYNC_FD =
VK_EXTERNAL_SEMAPHORE_HANDLE_TYPE_SYNC_FD_BIT_KHR VK_EXTERNAL_SEMAPHORE_HANDLE_TYPE_SYNC_FD_BIT_KHR
}; };

957
test_conformance/extensions/cl_khr_external_semaphore/test_external_semaphore.cpp
View File

File diff suppressed because it is too large Load Diff

6
test_conformance/vulkan/main.cpp
View File

@@ -53,7 +53,7 @@ static void params_reset()
extern int test_buffer_common(cl_device_id device_, cl_context context_, extern int test_buffer_common(cl_device_id device_, cl_context context_,
cl_command_queue queue_, int numElements_, cl_command_queue queue_, int numElements_,
float use_fence); bool use_fence);
extern int test_image_common(cl_device_id device_, cl_context context_, extern int test_image_common(cl_device_id device_, cl_context context_,
cl_command_queue queue_, int numElements_); cl_command_queue queue_, int numElements_);
@@ -227,10 +227,6 @@ size_t parseParams(int argc, const char *argv[], const char **argList)
{ {
disableNTHandleType = true; disableNTHandleType = true;
} }
if (!strcmp(argv[i], "--enableOffset"))
{
enableOffset = true;
}
if (strcmp(argv[i], "-h") == 0) if (strcmp(argv[i], "-h") == 0)
{ {
printUsage(argv[0]); printUsage(argv[0]);

207
test_conformance/vulkan/test_vulkan_api_consistency.cpp
View File

@@ -219,9 +219,8 @@ int test_consistency_external_image(cl_device_id deviceID, cl_context _context,
#else #else
if (!is_extension_available(devList[0], "cl_khr_external_memory_opaque_fd")) if (!is_extension_available(devList[0], "cl_khr_external_memory_opaque_fd"))
{ {
throw std::runtime_error( test_fail("Device does not support cl_khr_external_memory_opaque_fd "
"Device does not support cl_khr_external_memory_opaque_fd " "extension \n");
"extension \n");
} }
#endif #endif
uint32_t width = 256; uint32_t width = 256;
@@ -324,26 +323,6 @@ int test_consistency_external_image(cl_device_id deviceID, cl_context _context,
test_error(errNum, "Unable to create Image with Properties"); test_error(errNum, "Unable to create Image with Properties");
image.reset(); image.reset();
// Passing properties, image_desc and image_format all as NULL
image = clCreateImageWithProperties(context, NULL, CL_MEM_READ_WRITE, NULL,
NULL, NULL, &errNum);
test_failure_error(
errNum, CL_INVALID_IMAGE_DESCRIPTOR,
"Image creation must fail with CL_INVALID_IMAGE_DESCRIPTOR "
"when all are passed as NULL");
image.reset();
// Passing NULL properties and a valid image_format and image_desc
image =
clCreateImageWithProperties(context, NULL, CL_MEM_READ_WRITE,
&img_format, &image_desc, NULL, &errNum);
test_error(errNum,
"Unable to create image with NULL properties "
"with valid image format and image desc");
image.reset();
// Passing image_format as NULL // Passing image_format as NULL
image = clCreateImageWithProperties(context, extMemProperties.data(), image = clCreateImageWithProperties(context, extMemProperties.data(),
CL_MEM_READ_WRITE, NULL, &image_desc, CL_MEM_READ_WRITE, NULL, &image_desc,
@@ -396,88 +375,92 @@ int test_consistency_external_semaphore(cl_device_id deviceID,
cl_device_id devList[] = { deviceID, NULL }; cl_device_id devList[] = { deviceID, NULL };
#ifdef _WIN32 std::vector<VulkanExternalSemaphoreHandleType> supportedExternalSemaphores =
if (!is_extension_available(devList[0], "cl_khr_external_semaphore_win32")) getSupportedInteropExternalSemaphoreHandleTypes(devList[0], vkDevice);
{
throw std::runtime_error(
"Device does not support cl_khr_external_semaphore_win32 "
"extension \n");
}
#else
if (!is_extension_available(devList[0],
"cl_khr_external_semaphore_opaque_fd"))
{
throw std::runtime_error(
"Device does not support "
"cl_khr_external_semaphore_opaque_fd extension \n");
}
#endif
VulkanExternalSemaphoreHandleType vkExternalSemaphoreHandleType =
getSupportedVulkanExternalSemaphoreHandleTypeList()[0];
VulkanSemaphore vkVk2Clsemaphore(vkDevice, vkExternalSemaphoreHandleType);
VulkanSemaphore vkCl2Vksemaphore(vkDevice, vkExternalSemaphoreHandleType);
cl_semaphore_khr clCl2Vksemaphore;
cl_semaphore_khr clVk2Clsemaphore;
void* handle1 = NULL; if (supportedExternalSemaphores.empty())
void* handle2 = NULL;
int fd1, fd2;
std::vector<cl_semaphore_properties_khr> sema_props1{
(cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_KHR,
(cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_BINARY_KHR,
};
std::vector<cl_semaphore_properties_khr> sema_props2{
(cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_KHR,
(cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_BINARY_KHR,
};
switch (vkExternalSemaphoreHandleType)
{ {
test_fail("No supported external semaphore types found\n");
}
for (VulkanExternalSemaphoreHandleType semaphoreHandleType :
supportedExternalSemaphores)
{
VulkanSemaphore vkVk2Clsemaphore(vkDevice, semaphoreHandleType);
VulkanSemaphore vkCl2Vksemaphore(vkDevice, semaphoreHandleType);
cl_semaphore_khr clCl2Vksemaphore;
cl_semaphore_khr clVk2Clsemaphore;
void* handle1 = NULL;
void* handle2 = NULL;
int fd1, fd2;
std::vector<cl_semaphore_properties_khr> sema_props1{
(cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_KHR,
(cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_BINARY_KHR,
};
std::vector<cl_semaphore_properties_khr> sema_props2{
(cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_KHR,
(cl_semaphore_properties_khr)CL_SEMAPHORE_TYPE_BINARY_KHR,
};
switch (semaphoreHandleType)
{
#ifdef _WIN32 #ifdef _WIN32
case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_NT: case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_NT:
log_info(" Opaque NT handles are only supported on Windows\n"); log_info(" Opaque NT handles are only supported on Windows\n");
handle1 = vkVk2Clsemaphore.getHandle(vkExternalSemaphoreHandleType); handle1 = vkVk2Clsemaphore.getHandle(semaphoreHandleType);
handle2 = vkCl2Vksemaphore.getHandle(vkExternalSemaphoreHandleType); handle2 = vkCl2Vksemaphore.getHandle(semaphoreHandleType);
errNum = check_external_semaphore_handle_type( errNum = check_external_semaphore_handle_type(
devList[0], CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KHR); devList[0], CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KHR);
sema_props1.push_back((cl_semaphore_properties_khr) sema_props1.push_back((cl_semaphore_properties_khr)
CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KHR); CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KHR);
sema_props1.push_back((cl_semaphore_properties_khr)handle1); sema_props1.push_back((cl_semaphore_properties_khr)handle1);
sema_props2.push_back((cl_semaphore_properties_khr) sema_props2.push_back((cl_semaphore_properties_khr)
CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KHR); CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KHR);
sema_props2.push_back((cl_semaphore_properties_khr)handle2); sema_props2.push_back((cl_semaphore_properties_khr)handle2);
break; break;
case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_KMT: case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_KMT:
log_info(" Opaque D3DKMT handles are only supported on Windows\n"); log_info(
handle1 = vkVk2Clsemaphore.getHandle(vkExternalSemaphoreHandleType); " Opaque D3DKMT handles are only supported on Windows\n");
handle2 = vkCl2Vksemaphore.getHandle(vkExternalSemaphoreHandleType); handle1 = vkVk2Clsemaphore.getHandle(semaphoreHandleType);
errNum = check_external_semaphore_handle_type( handle2 = vkCl2Vksemaphore.getHandle(semaphoreHandleType);
devList[0], CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KMT_KHR); errNum = check_external_semaphore_handle_type(
sema_props1.push_back((cl_semaphore_properties_khr) devList[0], CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KMT_KHR);
CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KMT_KHR); sema_props1.push_back(
sema_props1.push_back((cl_semaphore_properties_khr)handle1); (cl_semaphore_properties_khr)
sema_props2.push_back((cl_semaphore_properties_khr) CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KMT_KHR);
CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KMT_KHR); sema_props1.push_back((cl_semaphore_properties_khr)handle1);
sema_props2.push_back((cl_semaphore_properties_khr)handle2); sema_props2.push_back(
break; (cl_semaphore_properties_khr)
CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KMT_KHR);
sema_props2.push_back((cl_semaphore_properties_khr)handle2);
break;
#else #else
case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_FD: case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_FD:
log_info(" Opaque file descriptors are not supported on Windows\n"); fd1 = (int)vkVk2Clsemaphore.getHandle(semaphoreHandleType);
fd1 = fd2 = (int)vkCl2Vksemaphore.getHandle(semaphoreHandleType);
(int)vkVk2Clsemaphore.getHandle(vkExternalSemaphoreHandleType); errNum = check_external_semaphore_handle_type(
fd2 = devList[0], CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR);
(int)vkCl2Vksemaphore.getHandle(vkExternalSemaphoreHandleType); sema_props1.push_back((cl_semaphore_properties_khr)
errNum = check_external_semaphore_handle_type( CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR);
devList[0], CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR); sema_props1.push_back((cl_semaphore_properties_khr)fd1);
sema_props1.push_back( sema_props2.push_back((cl_semaphore_properties_khr)
(cl_semaphore_properties_khr)CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR); CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR);
sema_props1.push_back((cl_semaphore_properties_khr)fd1); sema_props2.push_back((cl_semaphore_properties_khr)fd2);
sema_props2.push_back( break;
(cl_semaphore_properties_khr)CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR); case VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_SYNC_FD:
sema_props2.push_back((cl_semaphore_properties_khr)fd2); fd1 = -1;
break; fd2 = -1;
errNum = check_external_semaphore_handle_type(
devList[0], CL_SEMAPHORE_HANDLE_SYNC_FD_KHR);
sema_props1.push_back((cl_semaphore_properties_khr)
CL_SEMAPHORE_HANDLE_SYNC_FD_KHR);
sema_props1.push_back((cl_semaphore_properties_khr)fd1);
sema_props2.push_back((cl_semaphore_properties_khr)
CL_SEMAPHORE_HANDLE_SYNC_FD_KHR);
sema_props2.push_back((cl_semaphore_properties_khr)fd2);
break;
#endif #endif
default: log_error("Unsupported external memory handle type\n"); break; default: log_error("Unsupported external memory handle type\n"); break;
} }
if (CL_SUCCESS != errNum) if (CL_SUCCESS != errNum)
{ {
throw std::runtime_error( throw std::runtime_error(
@@ -532,31 +515,6 @@ int test_consistency_external_semaphore(cl_device_id deviceID,
test_error(errNum, test_error(errNum,
"Unable to create semaphore with valid semaphore properties"); "Unable to create semaphore with valid semaphore properties");
// Call Signal twice consecutively
errNum = clEnqueueSignalSemaphoresKHRptr(cmd_queue, 1, &clVk2Clsemaphore,
NULL, 0, NULL, NULL);
test_error(errNum, "clEnqueueSignalSemaphoresKHRptr failed");
errNum = clEnqueueSignalSemaphoresKHRptr(cmd_queue, 1, &clCl2Vksemaphore,
NULL, 0, NULL, NULL);
test_error(errNum,
"clEnqueueSignalSemaphoresKHRptr failed for two "
"consecutive wait events");
// Call Wait twice consecutively
errNum = clEnqueueWaitSemaphoresKHRptr(cmd_queue, 1, &clVk2Clsemaphore,
NULL, 0, NULL, NULL);
test_error(errNum, "clEnqueueWaitSemaphoresKHRptr failed");
errNum = clEnqueueWaitSemaphoresKHRptr(cmd_queue, 1, &clCl2Vksemaphore,
NULL, 0, NULL, NULL);
test_error(errNum,
"clEnqueueWaitSemaphoresKHRptr failed for two "
" consecutive wait events");
// Pass invalid object to release call // Pass invalid object to release call
errNum = clReleaseSemaphoreKHRptr(NULL); errNum = clReleaseSemaphoreKHRptr(NULL);
test_failure_error(errNum, CL_INVALID_VALUE, test_failure_error(errNum, CL_INVALID_VALUE,
@@ -569,6 +527,7 @@ int test_consistency_external_semaphore(cl_device_id deviceID,
errNum = clReleaseSemaphoreKHRptr(clCl2Vksemaphore); errNum = clReleaseSemaphoreKHRptr(clCl2Vksemaphore);
test_error(errNum, "clReleaseSemaphoreKHRptr failed"); test_error(errNum, "clReleaseSemaphoreKHRptr failed");
}
return TEST_PASS; return TEST_PASS;
} }

1146
test_conformance/vulkan/test_vulkan_interop_buffer.cpp
View File

File diff suppressed because it is too large Load Diff

305
test_conformance/vulkan/test_vulkan_interop_image.cpp
View File

@@ -18,6 +18,7 @@
#include <string> #include <string>
#include "harness/errorHelpers.h" #include "harness/errorHelpers.h"
#include <algorithm> #include <algorithm>
#include "deviceInfo.h"
#define MAX_2D_IMAGES 5 #define MAX_2D_IMAGES 5
#define MAX_2D_IMAGE_WIDTH 1024 #define MAX_2D_IMAGE_WIDTH 1024
@@ -189,11 +190,11 @@ const cl_kernel getKernelType(VulkanFormat format, cl_kernel kernel_float,
return kernel; return kernel;
} }
int run_test_with_two_queue(cl_context &context, cl_command_queue &cmd_queue1, int run_test_with_two_queue(
cl_command_queue &cmd_queue2, cl_context &context, cl_command_queue &cmd_queue1,
cl_kernel *kernel_unsigned, cl_command_queue &cmd_queue2, cl_kernel *kernel_unsigned,
cl_kernel *kernel_signed, cl_kernel *kernel_float, cl_kernel *kernel_signed, cl_kernel *kernel_float, VulkanDevice &vkDevice,
VulkanDevice &vkDevice) VulkanExternalSemaphoreHandleType vkExternalSemaphoreHandleType)
{ {
cl_int err = CL_SUCCESS; cl_int err = CL_SUCCESS;
size_t origin[3] = { 0, 0, 0 }; size_t origin[3] = { 0, 0, 0 };
@@ -245,8 +246,6 @@ int run_test_with_two_queue(cl_context &context, cl_command_queue &cmd_queue1,
VulkanCommandBuffer vkShaderCommandBuffer(vkDevice, vkCommandPool); VulkanCommandBuffer vkShaderCommandBuffer(vkDevice, vkCommandPool);
VulkanQueue &vkQueue = vkDevice.getQueue(); VulkanQueue &vkQueue = vkDevice.getQueue();
VulkanExternalSemaphoreHandleType vkExternalSemaphoreHandleType =
getSupportedVulkanExternalSemaphoreHandleTypeList()[0];
VulkanSemaphore vkVk2CLSemaphore(vkDevice, vkExternalSemaphoreHandleType); VulkanSemaphore vkVk2CLSemaphore(vkDevice, vkExternalSemaphoreHandleType);
VulkanSemaphore vkCl2VkSemaphore(vkDevice, vkExternalSemaphoreHandleType); VulkanSemaphore vkCl2VkSemaphore(vkDevice, vkExternalSemaphoreHandleType);
clExternalSemaphore *clVk2CLExternalSemaphore = NULL; clExternalSemaphore *clVk2CLExternalSemaphore = NULL;
@@ -462,7 +461,11 @@ int run_test_with_two_queue(cl_context &context, cl_command_queue &cmd_queue1,
->getExternalMemoryImage(); ->getExternalMemoryImage();
} }
clCl2VkExternalSemaphore->signal(cmd_queue1); err = clCl2VkExternalSemaphore->signal(cmd_queue1);
test_error_and_cleanup(
err, CLEANUP,
"Failed to signal CL semaphore\n");
if (!useSingleImageKernel) if (!useSingleImageKernel)
{ {
vkDescriptorSet.updateArray(1, vkDescriptorSet.updateArray(1,
@@ -499,6 +502,7 @@ int run_test_with_two_queue(cl_context &context, cl_command_queue &cmd_queue1,
1); 1);
vkShaderCommandBuffer.end(); vkShaderCommandBuffer.end();
} }
for (uint32_t iter = 0; iter < innerIterations; for (uint32_t iter = 0; iter < innerIterations;
iter++) iter++)
{ {
@@ -552,7 +556,17 @@ int run_test_with_two_queue(cl_context &context, cl_command_queue &cmd_queue1,
vkQueue.submit(vkCl2VkSemaphore, vkQueue.submit(vkCl2VkSemaphore,
vkShaderCommandBuffer, vkShaderCommandBuffer,
vkVk2CLSemaphore); vkVk2CLSemaphore);
clVk2CLExternalSemaphore->wait(cmd_queue1);
err =
clVk2CLExternalSemaphore->wait(cmd_queue1);
if (err != CL_SUCCESS)
{
print_error(err,
"Error: failed to wait on CL "
"external semaphore\n");
goto CLEANUP;
}
switch (num2DImages) switch (num2DImages)
{ {
case 2: case 2:
@@ -626,14 +640,10 @@ int run_test_with_two_queue(cl_context &context, cl_command_queue &cmd_queue1,
err |= clSetKernelArg(updateKernelCQ1, ++j, err |= clSetKernelArg(updateKernelCQ1, ++j,
sizeof(unsigned int), sizeof(unsigned int),
&numMipLevels); &numMipLevels);
test_error_and_cleanup(
err, CLEANUP,
"Error: Failed to set arg values \n");
if (err != CL_SUCCESS)
{
print_error(
err,
"Error: Failed to set arg values \n");
goto CLEANUP;
}
// clVk2CLExternalSemaphore->wait(cmd_queue1); // clVk2CLExternalSemaphore->wait(cmd_queue1);
size_t global_work_size[3] = { width, height, size_t global_work_size[3] = { width, height,
1 }; 1 };
@@ -642,21 +652,24 @@ int run_test_with_two_queue(cl_context &context, cl_command_queue &cmd_queue1,
cmd_queue1, updateKernelCQ1, 2, NULL, cmd_queue1, updateKernelCQ1, 2, NULL,
global_work_size, NULL, 0, NULL, global_work_size, NULL, 0, NULL,
&first_launch); &first_launch);
if (err != CL_SUCCESS) test_error_and_cleanup(
{ err, CLEANUP,
goto CLEANUP; "Failed to enqueue updateKernelCQ1\n");
}
err = clEnqueueNDRangeKernel( err = clEnqueueNDRangeKernel(
cmd_queue2, updateKernelCQ2, 2, NULL, cmd_queue2, updateKernelCQ2, 2, NULL,
global_work_size, NULL, 1, &first_launch, global_work_size, NULL, 1, &first_launch,
NULL); NULL);
if (err != CL_SUCCESS) test_error_and_cleanup(
{ err, CLEANUP,
goto CLEANUP; "Failed to enqueue updateKernelCQ2\n");
}
clFinish(cmd_queue2); clFinish(cmd_queue2);
clCl2VkExternalSemaphore->signal(cmd_queue2); err = clCl2VkExternalSemaphore->signal(
cmd_queue2);
test_error_and_cleanup(
err, CLEANUP,
"Failed to signal CL semaphore\n");
} }
unsigned int flags = 0; unsigned int flags = 0;
@@ -668,14 +681,11 @@ int run_test_with_two_queue(cl_context &context, cl_command_queue &cmd_queue1,
err = clEnqueueReadImage( err = clEnqueueReadImage(
cmd_queue1, external_mem_image2[i], CL_TRUE, cmd_queue1, external_mem_image2[i], CL_TRUE,
origin, region, 0, 0, dstBufferPtr, 0, NULL, origin, region, 0, 0, dstBufferPtr, 0, NULL,
&eventReadImage); NULL);
test_error_and_cleanup(
if (err != CL_SUCCESS) err, CLEANUP,
{ "clEnqueueReadImage failed with"
print_error(err, "error\n");
"clEnqueueReadImage failed with"
"error\n");
}
if (memcmp(srcBufferPtr, dstBufferPtr, if (memcmp(srcBufferPtr, dstBufferPtr,
srcBufSize)) srcBufSize))
@@ -727,10 +737,8 @@ int run_test_with_two_queue(cl_context &context, cl_command_queue &cmd_queue1,
externalMemory2.erase(externalMemory2.begin(), externalMemory2.erase(externalMemory2.begin(),
externalMemory2.begin() externalMemory2.begin()
+ num2DImages); + num2DImages);
if (CL_SUCCESS != err) test_error_and_cleanup(err, CLEANUP,
{ "Test error detected\n");
goto CLEANUP;
}
} }
} }
} }
@@ -748,10 +756,11 @@ CLEANUP:
return err; return err;
} }
int run_test_with_one_queue(cl_context &context, cl_command_queue &cmd_queue1, int run_test_with_one_queue(
cl_kernel *kernel_unsigned, cl_context &context, cl_command_queue &cmd_queue1,
cl_kernel *kernel_signed, cl_kernel *kernel_float, cl_kernel *kernel_unsigned, cl_kernel *kernel_signed,
VulkanDevice &vkDevice) cl_kernel *kernel_float, VulkanDevice &vkDevice,
VulkanExternalSemaphoreHandleType vkExternalSemaphoreHandleType)
{ {
cl_int err = CL_SUCCESS; cl_int err = CL_SUCCESS;
size_t origin[3] = { 0, 0, 0 }; size_t origin[3] = { 0, 0, 0 };
@@ -802,8 +811,6 @@ int run_test_with_one_queue(cl_context &context, cl_command_queue &cmd_queue1,
VulkanCommandBuffer vkShaderCommandBuffer(vkDevice, vkCommandPool); VulkanCommandBuffer vkShaderCommandBuffer(vkDevice, vkCommandPool);
VulkanQueue &vkQueue = vkDevice.getQueue(); VulkanQueue &vkQueue = vkDevice.getQueue();
VulkanExternalSemaphoreHandleType vkExternalSemaphoreHandleType =
getSupportedVulkanExternalSemaphoreHandleTypeList()[0];
VulkanSemaphore vkVk2CLSemaphore(vkDevice, vkExternalSemaphoreHandleType); VulkanSemaphore vkVk2CLSemaphore(vkDevice, vkExternalSemaphoreHandleType);
VulkanSemaphore vkCl2VkSemaphore(vkDevice, vkExternalSemaphoreHandleType); VulkanSemaphore vkCl2VkSemaphore(vkDevice, vkExternalSemaphoreHandleType);
clExternalSemaphore *clVk2CLExternalSemaphore = NULL; clExternalSemaphore *clVk2CLExternalSemaphore = NULL;
@@ -925,8 +932,8 @@ int run_test_with_one_queue(cl_context &context, cl_command_queue &cmd_queue1,
vkClExternalMemoryHandleTilingAssumption( vkClExternalMemoryHandleTilingAssumption(
deviceId, deviceId,
vkExternalMemoryHandleTypeList[emhtIdx], &err); vkExternalMemoryHandleTypeList[emhtIdx], &err);
ASSERT_SUCCESS(err, test_error_and_cleanup(
"Failed to query OpenCL tiling mode"); err, CLEANUP, "Failed to query OpenCL tiling mode");
VulkanImage2D vkDummyImage2D( VulkanImage2D vkDummyImage2D(
vkDevice, vkFormatList[0], widthList[0], vkDevice, vkFormatList[0], widthList[0],
@@ -1024,7 +1031,11 @@ int run_test_with_one_queue(cl_context &context, cl_command_queue &cmd_queue1,
->getExternalMemoryImage(); ->getExternalMemoryImage();
} }
clCl2VkExternalSemaphore->signal(cmd_queue1); err = clCl2VkExternalSemaphore->signal(cmd_queue1);
test_error_and_cleanup(
err, CLEANUP,
"Failed to signal CL semaphore\n");
if (!useSingleImageKernel) if (!useSingleImageKernel)
{ {
vkDescriptorSet.updateArray(1, vkDescriptorSet.updateArray(1,
@@ -1061,6 +1072,7 @@ int run_test_with_one_queue(cl_context &context, cl_command_queue &cmd_queue1,
1); 1);
vkShaderCommandBuffer.end(); vkShaderCommandBuffer.end();
} }
for (uint32_t iter = 0; iter < innerIterations; for (uint32_t iter = 0; iter < innerIterations;
iter++) iter++)
{ {
@@ -1114,7 +1126,14 @@ int run_test_with_one_queue(cl_context &context, cl_command_queue &cmd_queue1,
vkQueue.submit(vkCl2VkSemaphore, vkQueue.submit(vkCl2VkSemaphore,
vkShaderCommandBuffer, vkShaderCommandBuffer,
vkVk2CLSemaphore); vkVk2CLSemaphore);
clVk2CLExternalSemaphore->wait(cmd_queue1);
err =
clVk2CLExternalSemaphore->wait(cmd_queue1);
test_error_and_cleanup(
err, CLEANUP,
"Error: failed to wait on CL external "
"semaphore\n");
switch (num2DImages) switch (num2DImages)
{ {
case 1: case 1:
@@ -1158,25 +1177,25 @@ int run_test_with_one_queue(cl_context &context, cl_command_queue &cmd_queue1,
err |= clSetKernelArg(updateKernelCQ1, ++j, err |= clSetKernelArg(updateKernelCQ1, ++j,
sizeof(unsigned int), sizeof(unsigned int),
&numMipLevels); &numMipLevels);
test_error_and_cleanup(
if (err != CL_SUCCESS) err, CLEANUP,
{ "Error: Failed to set arg "
print_error(err, "values for kernel-1\n");
"Error: Failed to set arg "
"values for kernel-1\n");
goto CLEANUP;
}
size_t global_work_size[3] = { width, height, size_t global_work_size[3] = { width, height,
1 }; 1 };
err = clEnqueueNDRangeKernel( err = clEnqueueNDRangeKernel(
cmd_queue1, updateKernelCQ1, 2, NULL, cmd_queue1, updateKernelCQ1, 2, NULL,
global_work_size, NULL, 0, NULL, NULL); global_work_size, NULL, 0, NULL, NULL);
if (err != CL_SUCCESS) test_error_and_cleanup(
{ err, CLEANUP,
goto CLEANUP; "Failed to enqueue updateKernelCQ1\n");
}
clCl2VkExternalSemaphore->signal(cmd_queue1); err = clCl2VkExternalSemaphore->signal(
cmd_queue1);
test_error_and_cleanup(
err, CLEANUP,
"Failed to signal CL semaphore\n");
} }
unsigned int flags = 0; unsigned int flags = 0;
@@ -1187,14 +1206,11 @@ int run_test_with_one_queue(cl_context &context, cl_command_queue &cmd_queue1,
err = clEnqueueReadImage( err = clEnqueueReadImage(
cmd_queue1, external_mem_image2[i], CL_TRUE, cmd_queue1, external_mem_image2[i], CL_TRUE,
origin, region, 0, 0, dstBufferPtr, 0, NULL, origin, region, 0, 0, dstBufferPtr, 0, NULL,
&eventReadImage); NULL);
test_error_and_cleanup(
if (err != CL_SUCCESS) err, CLEANUP,
{ "clEnqueueReadImage failed with"
print_error(err, "error\n");
"clEnqueueReadImage failed with"
"error\n");
}
if (memcmp(srcBufferPtr, dstBufferPtr, if (memcmp(srcBufferPtr, dstBufferPtr,
srcBufSize)) srcBufSize))
@@ -1246,10 +1262,8 @@ int run_test_with_one_queue(cl_context &context, cl_command_queue &cmd_queue1,
externalMemory2.erase(externalMemory2.begin(), externalMemory2.erase(externalMemory2.begin(),
externalMemory2.begin() externalMemory2.begin()
+ num2DImages); + num2DImages);
if (CL_SUCCESS != err) test_error_and_cleanup(err, CLEANUP,
{ "Test detected error\n");
goto CLEANUP;
}
} }
} }
} }
@@ -1293,44 +1307,35 @@ int test_image_common(cl_device_id device_, cl_context context_,
char source_2[4096]; char source_2[4096];
char source_3[4096]; char source_3[4096];
size_t program_source_length; size_t program_source_length;
cl_program program[num_kernel_types]; cl_program program[num_kernel_types] = { NULL };
cl_kernel kernel_float[num_kernels] = { NULL, NULL, NULL, NULL }; cl_kernel kernel_float[num_kernels] = { NULL };
cl_kernel kernel_signed[num_kernels] = { NULL, NULL, NULL, NULL }; cl_kernel kernel_signed[num_kernels] = { NULL };
cl_kernel kernel_unsigned[num_kernels] = { NULL, NULL, NULL, NULL }; cl_kernel kernel_unsigned[num_kernels] = { NULL };
cl_mem external_mem_image1; cl_mem external_mem_image1;
cl_mem external_mem_image2; cl_mem external_mem_image2;
std::vector<VulkanExternalSemaphoreHandleType> supportedSemaphoreTypes;
VulkanDevice vkDevice; VulkanDevice vkDevice;
cl_context_properties contextProperties[] = { CL_CONTEXT_PLATFORM, 0, 0 }; cl_context_properties contextProperties[] = { CL_CONTEXT_PLATFORM, 0, 0 };
// get the platform ID // get the platform ID
err = clGetPlatformIDs(1, &platform, NULL); err = clGetPlatformIDs(1, &platform, NULL);
if (err != CL_SUCCESS) test_error_and_cleanup(err, CLEANUP, "Error: Failed to get platform\n");
{
print_error(err, "Error: Failed to get platform\n");
goto CLEANUP;
}
err = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, 0, NULL, &num_devices); err = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, 0, NULL, &num_devices);
if (CL_SUCCESS != err) test_error_and_cleanup(
{ err, CLEANUP, "clGetDeviceIDs failed in returning no. of devices\n");
print_error(err, "clGetDeviceIDs failed in returning no. of devices\n");
goto CLEANUP;
}
devices = (cl_device_id *)malloc(num_devices * sizeof(cl_device_id)); devices = (cl_device_id *)malloc(num_devices * sizeof(cl_device_id));
if (NULL == devices) if (NULL == devices)
{ {
err = CL_OUT_OF_HOST_MEMORY; test_fail_and_cleanup(err, CLEANUP,
print_error(err, "Unable to allocate memory for devices\n"); "Unable to allocate memory for devices\n");
goto CLEANUP;
} }
err = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, num_devices, devices, err = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, num_devices, devices,
NULL); NULL);
if (CL_SUCCESS != err) test_error_and_cleanup(err, CLEANUP, "Failed to get deviceID.\n");
{
print_error(err, "Failed to get deviceID.\n");
goto CLEANUP;
}
contextProperties[1] = (cl_context_properties)platform; contextProperties[1] = (cl_context_properties)platform;
log_info("Assigned contextproperties for platform\n"); log_info("Assigned contextproperties for platform\n");
for (device_no = 0; device_no < num_devices; device_no++) for (device_no = 0; device_no < num_devices; device_no++)
@@ -1360,12 +1365,20 @@ int test_image_common(cl_device_id device_, cl_context context_,
goto CLEANUP; goto CLEANUP;
} }
err = clGetDeviceInfo(devices[device_no], CL_DEVICE_UUID_KHR, err = clGetDeviceInfo(devices[device_no], CL_DEVICE_UUID_KHR,
CL_UUID_SIZE_KHR, uuid, &extensionSize); CL_UUID_SIZE_KHR, uuid, NULL);
if (CL_SUCCESS != err) test_error_and_cleanup(err, CLEANUP,
"clGetDeviceInfo failed with error");
supportedSemaphoreTypes =
getSupportedInteropExternalSemaphoreHandleTypes(devices[device_no],
vkDevice);
// If device does not support any semaphores, try the next one
if (supportedSemaphoreTypes.empty())
{ {
print_error(err, "clGetDeviceInfo failed with error"); continue;
goto CLEANUP;
} }
err = err =
memcmp(uuid, vkDevice.getPhysicalDevice().getUUID(), VK_UUID_SIZE); memcmp(uuid, vkDevice.getPhysicalDevice().getUUID(), VK_UUID_SIZE);
if (err == 0) if (err == 0)
@@ -1373,48 +1386,41 @@ int test_image_common(cl_device_id device_, cl_context context_,
break; break;
} }
} }
if (supportedSemaphoreTypes.empty())
{
test_fail_and_cleanup(
err, CLEANUP, "No devices found that support OpenCL semaphores\n");
}
if (device_no >= num_devices) if (device_no >= num_devices)
{ {
err = EXIT_FAILURE; test_fail_and_cleanup(err, CLEANUP,
print_error(err, "OpenCL error:"
"OpenCL error:" "No Vulkan-OpenCL Interop capable GPU found.\n");
"No Vulkan-OpenCL Interop capable GPU found.\n");
goto CLEANUP;
} }
deviceId = devices[device_no]; deviceId = devices[device_no];
err = setMaxImageDimensions(deviceId, max_width, max_height); err = setMaxImageDimensions(deviceId, max_width, max_height);
if (CL_SUCCESS != err) test_error_and_cleanup(err, CLEANUP, "error setting max image dimensions");
{
print_error(err, "error setting max image dimensions");
goto CLEANUP;
}
log_info("Set max_width to %lu and max_height to %lu\n", max_width, log_info("Set max_width to %lu and max_height to %lu\n", max_width,
max_height); max_height);
context = clCreateContextFromType(contextProperties, CL_DEVICE_TYPE_GPU, context = clCreateContextFromType(contextProperties, CL_DEVICE_TYPE_GPU,
NULL, NULL, &err); NULL, NULL, &err);
if (CL_SUCCESS != err) test_error_and_cleanup(err, CLEANUP, "error creating context");
{
print_error(err, "error creating context");
goto CLEANUP;
}
log_info("Successfully created context !!!\n"); log_info("Successfully created context !!!\n");
cmd_queue1 = clCreateCommandQueue(context, devices[device_no], 0, &err); cmd_queue1 = clCreateCommandQueue(context, devices[device_no], 0, &err);
if (CL_SUCCESS != err) test_error_and_cleanup(err, CLEANUP,
{ "Error: Failed to create command queue!\n");
err = CL_INVALID_COMMAND_QUEUE;
print_error(err, "Error: Failed to create command queue!\n");
goto CLEANUP;
}
log_info("clCreateCommandQueue successfull \n"); log_info("clCreateCommandQueue successfull \n");
cmd_queue2 = clCreateCommandQueue(context, devices[device_no], 0, &err); cmd_queue2 = clCreateCommandQueue(context, devices[device_no], 0, &err);
if (CL_SUCCESS != err) test_error_and_cleanup(err, CLEANUP,
{ "Error: Failed to create command queue!\n");
err = CL_INVALID_COMMAND_QUEUE;
print_error(err, "Error: Failed to create command queue!\n");
goto CLEANUP;
}
log_info("clCreateCommandQueue2 successful \n"); log_info("clCreateCommandQueue2 successful \n");
for (int i = 0; i < num_kernels; i++) for (int i = 0; i < num_kernels; i++)
@@ -1473,42 +1479,33 @@ int test_image_common(cl_device_id device_, cl_context context_,
context, 1, &sourceTexts[k], &program_source_length, &err); context, 1, &sourceTexts[k], &program_source_length, &err);
err |= clBuildProgram(program[k], 0, NULL, NULL, NULL, NULL); err |= clBuildProgram(program[k], 0, NULL, NULL, NULL, NULL);
} }
test_error_and_cleanup(err, CLEANUP, "Error: Failed to build program");
if (err != CL_SUCCESS)
{
print_error(err, "Error: Failed to build program");
goto CLEANUP;
}
// create the kernel // create the kernel
kernel_float[i] = clCreateKernel(program[0], "image2DKernel", &err); kernel_float[i] = clCreateKernel(program[0], "image2DKernel", &err);
if (err != CL_SUCCESS) test_error_and_cleanup(err, CLEANUP, "clCreateKernel failed");
{
print_error(err, "clCreateKernel failed");
goto CLEANUP;
}
kernel_signed[i] = clCreateKernel(program[1], "image2DKernel", &err); kernel_signed[i] = clCreateKernel(program[1], "image2DKernel", &err);
if (err != CL_SUCCESS) test_error_and_cleanup(err, CLEANUP, "clCreateKernel failed");
{
print_error(err, "clCreateKernel failed");
goto CLEANUP;
}
kernel_unsigned[i] = clCreateKernel(program[2], "image2DKernel", &err); kernel_unsigned[i] = clCreateKernel(program[2], "image2DKernel", &err);
if (err != CL_SUCCESS) test_error_and_cleanup(err, CLEANUP, "clCreateKernel failed ");
}
for (VulkanExternalSemaphoreHandleType externalSemaphoreType :
supportedSemaphoreTypes)
{
if (numCQ == 2)
{ {
print_error(err, "clCreateKernel failed "); err = run_test_with_two_queue(
goto CLEANUP; context, cmd_queue1, cmd_queue2, kernel_unsigned, kernel_signed,
kernel_float, vkDevice, externalSemaphoreType);
}
else
{
err = run_test_with_one_queue(context, cmd_queue1, kernel_unsigned,
kernel_signed, kernel_float, vkDevice,
externalSemaphoreType);
} }
}
if (numCQ == 2)
{
err = run_test_with_two_queue(context, cmd_queue1, cmd_queue2,
kernel_unsigned, kernel_signed,
kernel_float, vkDevice);
}
else
{
err = run_test_with_one_queue(context, cmd_queue1, kernel_unsigned,
kernel_signed, kernel_float, vkDevice);
} }
CLEANUP: CLEANUP:
for (int i = 0; i < num_kernels; i++) for (int i = 0; i < num_kernels; i++)

2
test_conformance/vulkan/vulkan_interop_common.hpp
View File

@@ -43,7 +43,5 @@ extern bool debug_trace;
extern bool useSingleImageKernel; extern bool useSingleImageKernel;
extern bool useDeviceLocal; extern bool useDeviceLocal;
extern bool disableNTHandleType; extern bool disableNTHandleType;
// Enable offset for multiImport of vulkan device memory
extern bool enableOffset;
#endif // _vulkan_interop_common_hpp_ #endif // _vulkan_interop_common_hpp_