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

Corrected test_vulkan to use specific platform/device from harness (#2154)

Browse Source 
Fixes #1926 according to task description
This commit is contained in:
Marcin Hajder
2025-01-07 19:09:38 +01:00
committed by GitHub
parent 4c70fecad7
commit d058dfdeef
14 changed files with 2090 additions and 2085 deletions
Download Patch File Download Diff File Expand all files Collapse all files

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

@@ -464,7 +464,7 @@ getCLImageInfoFromVkImageInfo(const VkImageCreateInfo *VulkanImageCreateInfo,
memcpy(img_fmt, &clImgFormat, sizeof(cl_image_format)); memcpy(img_fmt, &clImgFormat, sizeof(cl_image_format));
img_desc->image_type = getImageTypeFromVk(VulkanImageCreateInfo->imageType); img_desc->image_type = getImageTypeFromVk(VulkanImageCreateInfo->imageType);
if (CL_INVALID_VALUE == img_desc->image_type) if (CL_INVALID_VALUE == static_cast<cl_int>(img_desc->image_type))
{ {
return CL_INVALID_VALUE; return CL_INVALID_VALUE;
} }
@@ -503,6 +503,8 @@ cl_int check_external_memory_handle_type(
errNum = clGetDeviceInfo(deviceID, errNum = clGetDeviceInfo(deviceID,
CL_DEVICE_EXTERNAL_MEMORY_IMPORT_HANDLE_TYPES_KHR, CL_DEVICE_EXTERNAL_MEMORY_IMPORT_HANDLE_TYPES_KHR,
0, NULL, &handle_type_size); 0, NULL, &handle_type_size);
test_error(errNum, "clGetDeviceInfo failed");
handle_type = handle_type =
(cl_external_memory_handle_type_khr *)malloc(handle_type_size); (cl_external_memory_handle_type_khr *)malloc(handle_type_size);
@@ -539,6 +541,7 @@ cl_int check_external_semaphore_handle_type(
errNum = errNum =
clGetDeviceInfo(deviceID, queryParamName, 0, NULL, &handle_type_size); clGetDeviceInfo(deviceID, queryParamName, 0, NULL, &handle_type_size);
test_error(errNum, "clGetDeviceInfo failed");
if (handle_type_size == 0) if (handle_type_size == 0)
{ {

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

@@ -1,5 +1,5 @@
// //
// Copyright (c) 2022 The Khronos Group Inc. // Copyright (c) 2024 The Khronos Group Inc.
// //
// Licensed under the Apache License, Version 2.0 (the "License"); // Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License. // you may not use this file except in compliance with the License.
@@ -40,13 +40,10 @@ const VulkanInstance &getVulkanInstance()
const VulkanPhysicalDevice &getVulkanPhysicalDevice() const VulkanPhysicalDevice &getVulkanPhysicalDevice()
{ {
size_t pdIdx; size_t pdIdx = 0;
cl_int errNum = 0; cl_int errNum = 0;
cl_platform_id platform = NULL; cl_platform_id platform = nullptr;
cl_uchar uuid[CL_UUID_SIZE_KHR]; cl_uchar uuid[CL_UUID_SIZE_KHR];
cl_device_id *devices;
char *extensions = NULL;
size_t extensionSize = 0;
cl_uint num_devices = 0; cl_uint num_devices = 0;
cl_uint device_no = 0; cl_uint device_no = 0;
const size_t bufsize = BUFFERSIZE; const size_t bufsize = BUFFERSIZE;
@@ -69,14 +66,9 @@ const VulkanPhysicalDevice &getVulkanPhysicalDevice()
throw std::runtime_error( throw std::runtime_error(
"Error: clGetDeviceIDs failed in returning of devices\n"); "Error: clGetDeviceIDs failed in returning of devices\n");
} }
devices = (cl_device_id *)malloc(num_devices * sizeof(cl_device_id)); std::vector<cl_device_id> devices(num_devices);
if (NULL == devices) errNum = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, num_devices,
{ devices.data(), NULL);
throw std::runtime_error(
"Error: Unable to allocate memory for devices\n");
}
errNum = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, num_devices, devices,
NULL);
if (CL_SUCCESS != errNum) if (CL_SUCCESS != errNum)
{ {
throw std::runtime_error("Error: Failed to get deviceID.\n"); throw std::runtime_error("Error: Failed to get deviceID.\n");
@@ -84,34 +76,14 @@ const VulkanPhysicalDevice &getVulkanPhysicalDevice()
bool is_selected = false; bool is_selected = false;
for (device_no = 0; device_no < num_devices; device_no++) for (device_no = 0; device_no < num_devices; device_no++)
{ {
errNum = clGetDeviceInfo(devices[device_no], CL_DEVICE_EXTENSIONS, 0,
NULL, &extensionSize);
if (CL_SUCCESS != errNum)
{
throw std::runtime_error("Error in clGetDeviceInfo for getting "
"device_extension size....\n");
}
extensions = (char *)malloc(extensionSize);
if (NULL == extensions)
{
throw std::runtime_error(
"Unable to allocate memory for extensions\n");
}
errNum = clGetDeviceInfo(devices[device_no], CL_DEVICE_EXTENSIONS,
extensionSize, extensions, NULL);
if (CL_SUCCESS != errNum)
{
throw std::runtime_error("Error: Error in clGetDeviceInfo for "
"getting device_extension\n");
}
errNum = clGetDeviceInfo(devices[device_no], CL_DEVICE_UUID_KHR, errNum = clGetDeviceInfo(devices[device_no], CL_DEVICE_UUID_KHR,
CL_UUID_SIZE_KHR, uuid, &extensionSize); CL_UUID_SIZE_KHR, uuid, nullptr);
if (CL_SUCCESS != errNum) if (CL_SUCCESS != errNum)
{ {
throw std::runtime_error( throw std::runtime_error(
"Error: clGetDeviceInfo failed with error\n"); "Error: clGetDeviceInfo failed with error\n");
} }
free(extensions);
for (pdIdx = 0; pdIdx < physicalDeviceList.size(); pdIdx++) for (pdIdx = 0; pdIdx < physicalDeviceList.size(); pdIdx++)
{ {
if (!memcmp(&uuid, physicalDeviceList[pdIdx].getUUID(), if (!memcmp(&uuid, physicalDeviceList[pdIdx].getUUID(),
@@ -139,10 +111,48 @@ const VulkanPhysicalDevice &getVulkanPhysicalDevice()
return physicalDeviceList[pdIdx]; return physicalDeviceList[pdIdx];
} }
const VulkanQueueFamily &getVulkanQueueFamily(uint32_t queueFlags) const VulkanPhysicalDevice &
getAssociatedVulkanPhysicalDevice(cl_device_id deviceId)
{
size_t pdIdx;
cl_int errNum = 0;
cl_uchar uuid[CL_UUID_SIZE_KHR];
const VulkanInstance &instance = getVulkanInstance();
const VulkanPhysicalDeviceList &physicalDeviceList =
instance.getPhysicalDeviceList();
errNum = clGetDeviceInfo(deviceId, CL_DEVICE_UUID_KHR, CL_UUID_SIZE_KHR,
uuid, nullptr);
if (CL_SUCCESS != errNum)
{
throw std::runtime_error("Error: clGetDeviceInfo failed with error\n");
}
for (pdIdx = 0; pdIdx < physicalDeviceList.size(); pdIdx++)
{
if (!memcmp(&uuid, physicalDeviceList[pdIdx].getUUID(), VK_UUID_SIZE))
{
std::cout << "Selected physical device = "
<< physicalDeviceList[pdIdx] << std::endl;
break;
}
}
if ((pdIdx >= physicalDeviceList.size())
|| (physicalDeviceList[pdIdx] == (VkPhysicalDevice)VK_NULL_HANDLE))
{
throw std::runtime_error("failed to find a suitable GPU!");
}
std::cout << "Selected physical device is: " << physicalDeviceList[pdIdx]
<< std::endl;
return physicalDeviceList[pdIdx];
}
const VulkanQueueFamily &
getVulkanQueueFamily(const VulkanPhysicalDevice &physicalDevice,
uint32_t queueFlags)
{ {
size_t qfIdx; size_t qfIdx;
const VulkanPhysicalDevice &physicalDevice = getVulkanPhysicalDevice();
const VulkanQueueFamilyList &queueFamilyList = const VulkanQueueFamilyList &queueFamilyList =
physicalDevice.getQueueFamilyList(); physicalDevice.getQueueFamilyList();

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

@@ -1,5 +1,5 @@
// //
// Copyright (c) 2022 The Khronos Group Inc. // Copyright (c) 2024 The Khronos Group Inc.
// //
// Licensed under the Apache License, Version 2.0 (the "License"); // Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License. // you may not use this file except in compliance with the License.
@@ -32,8 +32,11 @@
const VulkanInstance& getVulkanInstance(); const VulkanInstance& getVulkanInstance();
const VulkanPhysicalDevice& getVulkanPhysicalDevice(); const VulkanPhysicalDevice& getVulkanPhysicalDevice();
const VulkanQueueFamily& const VulkanPhysicalDevice&
getVulkanQueueFamily(uint32_t queueFlags = VULKAN_QUEUE_FLAG_GRAPHICS getAssociatedVulkanPhysicalDevice(cl_device_id deviceId);
const VulkanQueueFamily& getVulkanQueueFamily(
const VulkanPhysicalDevice& physicalDevice = getVulkanPhysicalDevice(),
uint32_t queueFlags = VULKAN_QUEUE_FLAG_GRAPHICS
| VULKAN_QUEUE_FLAG_COMPUTE); | VULKAN_QUEUE_FLAG_COMPUTE);
const VulkanMemoryType& const VulkanMemoryType&
getVulkanMemoryType(const VulkanDevice& device, getVulkanMemoryType(const VulkanDevice& device,

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

@@ -145,7 +145,7 @@ public:
virtual ~VulkanDevice(); virtual ~VulkanDevice();
const VulkanPhysicalDevice &getPhysicalDevice() const; const VulkanPhysicalDevice &getPhysicalDevice() const;
VulkanQueue & VulkanQueue &
getQueue(const VulkanQueueFamily &queueFamily = getVulkanQueueFamily(), getQueue(const VulkanQueueFamily &queueFamily /* = getVulkanQueueFamily()*/,
uint32_t queueIndex = 0); uint32_t queueIndex = 0);
operator VkDevice() const; operator VkDevice() const;
}; };

1
test_conformance/vulkan/CMakeLists.txt
View File

@@ -25,6 +25,7 @@ set (${MODULE_NAME}_SOURCES
test_vulkan_api_consistency_for_1dimages.cpp test_vulkan_api_consistency_for_1dimages.cpp
test_vulkan_platform_device_info.cpp test_vulkan_platform_device_info.cpp
vulkan_interop_common.cpp vulkan_interop_common.cpp
vulkan_test_base.h
) )
include_directories("../common/vulkan_wrapper") include_directories("../common/vulkan_wrapper")

230
test_conformance/vulkan/main.cpp
View File

@@ -1,5 +1,5 @@
// //
// Copyright (c) 2022 The Khronos Group Inc. // Copyright (c) 2024 The Khronos Group Inc.
// //
// Licensed under the Apache License, Version 2.0 (the "License"); // Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License. // you may not use this file except in compliance with the License.
@@ -30,121 +30,15 @@
#include <OpenCL/cl.h> #include <OpenCL/cl.h>
#endif #endif
#include "procs.h" #include "procs.h"
#include "harness/testHarness.h" #include "harness/testHarness.h"
#include "harness/parseParameters.h"
#include "harness/deviceInfo.h"
#if !defined(_WIN32) #if !defined(_WIN32)
#include <unistd.h> #include <unistd.h>
#endif #endif
#include <vulkan_interop_common.hpp>
#include <vulkan_wrapper.hpp>
#define BUFFERSIZE 3000 #define BUFFERSIZE 3000
static void params_reset()
{
numCQ = 1;
multiImport = false;
multiCtx = false;
}
extern int test_buffer_common(cl_device_id device_, cl_context context_,
cl_command_queue queue_, int numElements_,
bool use_fence);
extern int test_image_common(cl_device_id device_, cl_context context_,
cl_command_queue queue_, int numElements_);
int test_buffer_single_queue(cl_device_id device_, cl_context context_,
cl_command_queue queue_, int numElements_)
{
params_reset();
log_info("RUNNING TEST WITH ONE QUEUE...... \n\n");
return test_buffer_common(device_, context_, queue_, numElements_, false);
}
int test_buffer_multiple_queue(cl_device_id device_, cl_context context_,
cl_command_queue queue_, int numElements_)
{
params_reset();
numCQ = 2;
log_info("RUNNING TEST WITH TWO QUEUE...... \n\n");
return test_buffer_common(device_, context_, queue_, numElements_, false);
}
int test_buffer_multiImport_sameCtx(cl_device_id device_, cl_context context_,
cl_command_queue queue_, int numElements_)
{
params_reset();
multiImport = true;
log_info("RUNNING TEST WITH MULTIPLE DEVICE MEMORY IMPORT "
"IN SAME CONTEXT...... \n\n");
return test_buffer_common(device_, context_, queue_, numElements_, false);
}
int test_buffer_multiImport_diffCtx(cl_device_id device_, cl_context context_,
cl_command_queue queue_, int numElements_)
{
params_reset();
multiImport = true;
multiCtx = true;
log_info("RUNNING TEST WITH MULTIPLE DEVICE MEMORY IMPORT "
"IN DIFFERENT CONTEXT...... \n\n");
return test_buffer_common(device_, context_, queue_, numElements_, false);
}
int test_buffer_single_queue_fence(cl_device_id device_, cl_context context_,
cl_command_queue queue_, int numElements_)
{
params_reset();
log_info("RUNNING TEST WITH ONE QUEUE...... \n\n");
return test_buffer_common(device_, context_, queue_, numElements_, true);
}
int test_buffer_multiple_queue_fence(cl_device_id device_, cl_context context_,
cl_command_queue queue_, int numElements_)
{
params_reset();
numCQ = 2;
log_info("RUNNING TEST WITH TWO QUEUE...... \n\n");
return test_buffer_common(device_, context_, queue_, numElements_, true);
}
int test_buffer_multiImport_sameCtx_fence(cl_device_id device_,
cl_context context_,
cl_command_queue queue_,
int numElements_)
{
params_reset();
multiImport = true;
log_info("RUNNING TEST WITH MULTIPLE DEVICE MEMORY IMPORT "
"IN SAME CONTEXT...... \n\n");
return test_buffer_common(device_, context_, queue_, numElements_, true);
}
int test_buffer_multiImport_diffCtx_fence(cl_device_id device_,
cl_context context_,
cl_command_queue queue_,
int numElements_)
{
params_reset();
multiImport = true;
multiCtx = true;
log_info("RUNNING TEST WITH MULTIPLE DEVICE MEMORY IMPORT "
"IN DIFFERENT CONTEXT...... \n\n");
return test_buffer_common(device_, context_, queue_, numElements_, true);
}
int test_image_single_queue(cl_device_id device_, cl_context context_,
cl_command_queue queue_, int numElements_)
{
params_reset();
log_info("RUNNING TEST WITH ONE QUEUE...... \n\n");
return test_image_common(device_, context_, queue_, numElements_);
}
int test_image_multiple_queue(cl_device_id device_, cl_context context_,
cl_command_queue queue_, int numElements_)
{
params_reset();
numCQ = 2;
log_info("RUNNING TEST WITH TWO QUEUE...... \n\n");
return test_image_common(device_, context_, queue_, numElements_);
}
test_definition test_list[] = { ADD_TEST(buffer_single_queue), test_definition test_list[] = { ADD_TEST(buffer_single_queue),
ADD_TEST(buffer_multiple_queue), ADD_TEST(buffer_multiple_queue),
ADD_TEST(buffer_multiImport_sameCtx), ADD_TEST(buffer_multiImport_sameCtx),
@@ -165,20 +59,6 @@ test_definition test_list[] = { ADD_TEST(buffer_single_queue),
const int test_num = ARRAY_SIZE(test_list); const int test_num = ARRAY_SIZE(test_list);
cl_device_type gDeviceType = CL_DEVICE_TYPE_DEFAULT;
char *choosen_platform_name = NULL;
cl_platform_id platform = NULL;
cl_int choosen_platform_index = -1;
char platform_name[1024] = "";
cl_platform_id select_platform = NULL;
char *extensions = NULL;
size_t extensionSize = 0;
cl_uint num_devices = 0;
cl_uint device_no = 0;
cl_device_id *devices;
const size_t bufsize = BUFFERSIZE;
char buf[BUFFERSIZE];
cl_uchar uuid[CL_UUID_SIZE_KHR];
unsigned int numCQ; unsigned int numCQ;
bool multiImport; bool multiImport;
bool multiCtx; bool multiCtx;
@@ -269,19 +149,7 @@ size_t parseParams(int argc, const char *argv[], const char **argList)
int main(int argc, const char *argv[]) int main(int argc, const char *argv[])
{ {
int errNum = 0;
test_start(); test_start();
params_reset();
if (!checkVkSupport())
{
log_info("Vulkan supported GPU not found \n");
log_info("TEST SKIPPED \n");
return 0;
}
VulkanDevice vkDevice;
cl_device_type requestedDeviceType = CL_DEVICE_TYPE_GPU; cl_device_type requestedDeviceType = CL_DEVICE_TYPE_GPU;
char *force_cpu = getenv("CL_DEVICE_TYPE"); char *force_cpu = getenv("CL_DEVICE_TYPE");
@@ -305,104 +173,10 @@ int main(int argc, const char *argv[])
log_info("Vulkan tests can only run on a GPU device.\n"); log_info("Vulkan tests can only run on a GPU device.\n");
return 0; return 0;
} }
gDeviceType = CL_DEVICE_TYPE_GPU;
const char **argList = (const char **)calloc(argc, sizeof(char *)); const char **argList = (const char **)calloc(argc, sizeof(char *));
size_t argCount = parseParams(argc, argv, argList); size_t argCount = parseParams(argc, argv, argList);
if (argCount == 0) return 0; if (argCount == 0) return 0;
// get the platform ID
errNum = clGetPlatformIDs(1, &platform, NULL);
if (errNum != CL_SUCCESS)
{
print_error(errNum, "Error: Failed to get platform\n");
return errNum;
}
errNum = return runTestHarness(argc, argv, test_num, test_list, false, 0);
clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, 0, NULL, &num_devices);
if (CL_SUCCESS != errNum)
{
print_error(errNum, "clGetDeviceIDs failed in returning of devices\n");
return errNum;
}
devices = (cl_device_id *)malloc(num_devices * sizeof(cl_device_id));
if (NULL == devices)
{
print_error(errNum, "Unable to allocate memory for devices\n");
return CL_OUT_OF_HOST_MEMORY;
}
errNum = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, num_devices, devices,
NULL);
if (CL_SUCCESS != errNum)
{
print_error(errNum, "Failed to get deviceID.\n");
return errNum;
}
for (device_no = 0; device_no < num_devices; device_no++)
{
errNum = clGetDeviceInfo(devices[device_no], CL_DEVICE_EXTENSIONS, 0,
NULL, &extensionSize);
if (CL_SUCCESS != errNum)
{
log_error("Error in clGetDeviceInfo for getting "
"device_extension size....\n");
return errNum;
}
extensions = (char *)malloc(extensionSize);
if (NULL == extensions)
{
log_error("Unable to allocate memory for extensions\n");
return CL_OUT_OF_HOST_MEMORY;
}
errNum =
clGetDeviceInfo(devices[device_no], CL_DEVICE_EXTENSIONS,
extensionSize, extensions, NULL /*&extensionSize*/);
if (CL_SUCCESS != errNum)
{
print_error(errNum,
"Error in clGetDeviceInfo for getting "
"device_extension\n");
return errNum;
}
errNum = clGetDeviceInfo(devices[device_no], CL_DEVICE_UUID_KHR,
CL_UUID_SIZE_KHR, uuid, &extensionSize);
if (CL_SUCCESS != errNum)
{
print_error(errNum, "clGetDeviceInfo failed with error\n ");
return errNum;
}
errNum =
memcmp(uuid, vkDevice.getPhysicalDevice().getUUID(), VK_UUID_SIZE);
if (errNum == 0)
{
break;
}
}
if (device_no >= num_devices)
{
fprintf(stderr,
"OpenCL error: "
"No Vulkan-OpenCL Interop capable GPU found.\n");
}
if (!(is_extension_available(devices[device_no], "cl_khr_external_memory")
&& is_extension_available(devices[device_no],
"cl_khr_external_semaphore")))
{
log_info("Device does not support cl_khr_external_memory "
"or cl_khr_external_semaphore\n");
log_info(" TEST SKIPPED\n");
return CL_SUCCESS;
}
init_cl_vk_ext(platform, num_devices, devices);
// Execute tests.
// Note: don't use the entire harness, because we have a different way of
// obtaining the device (via the context)
test_harness_config config{};
config.forceNoContextCreation = true;
config.numElementsToUse = 1024;
config.queueProps = 0;
errNum = parseAndCallCommandLineTests(argCount, argList, devices[device_no],
test_num, test_list, config);
return errNum;
} }

33
test_conformance/vulkan/procs.h
View File

@@ -44,3 +44,36 @@ extern int test_platform_info(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements); cl_command_queue queue, int num_elements);
extern int test_device_info(cl_device_id device, cl_context context, extern int test_device_info(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements); cl_command_queue queue, int num_elements);
extern int test_buffer_single_queue(cl_device_id device_, cl_context context_,
cl_command_queue queue_, int numElements_);
extern int test_buffer_multiple_queue(cl_device_id device_, cl_context context_,
cl_command_queue queue_,
int numElements_);
extern int test_buffer_multiImport_sameCtx(cl_device_id device_,
cl_context context_,
cl_command_queue queue_,
int numElements_);
extern int test_buffer_multiImport_diffCtx(cl_device_id device_,
cl_context context_,
cl_command_queue queue_,
int numElements_);
extern int test_buffer_single_queue_fence(cl_device_id device_,
cl_context context_,
cl_command_queue queue_,
int numElements_);
extern int test_buffer_multiple_queue_fence(cl_device_id device_,
cl_context context_,
cl_command_queue queue_,
int numElements_);
extern int test_buffer_multiImport_sameCtx_fence(cl_device_id device_,
cl_context context_,
cl_command_queue queue_,
int numElements_);
extern int test_buffer_multiImport_diffCtx_fence(cl_device_id device_,
cl_context context_,
cl_command_queue queue_,
int numElements_);
extern int test_image_single_queue(cl_device_id device_, cl_context context_,
cl_command_queue queue_, int numElements_);
extern int test_image_multiple_queue(cl_device_id device_, cl_context context_,
cl_command_queue queue_, int numElements_);

333
test_conformance/vulkan/test_vulkan_api_consistency.cpp
View File

@@ -1,5 +1,5 @@
// //
// Copyright (c) 2022 The Khronos Group Inc. // Copyright (c) 2024 The Khronos Group Inc.
// //
// Licensed under the Apache License, Version 2.0 (the "License"); // Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License. // you may not use this file except in compliance with the License.
@@ -33,40 +33,33 @@
#include "harness/typeWrappers.h" #include "harness/typeWrappers.h"
#include "harness/deviceInfo.h" #include "harness/deviceInfo.h"
int test_consistency_external_buffer(cl_device_id deviceID, cl_context _context, #include "vulkan_test_base.h"
cl_command_queue _queue, int num_elements) #include "opencl_vulkan_wrapper.hpp"
namespace {
struct ConsistencyExternalBufferTest : public VulkanTestBase
{ {
cl_int errNum; ConsistencyExternalBufferTest(cl_device_id device, cl_context context,
VulkanDevice vkDevice; cl_command_queue queue, cl_int nelems)
// Context and command queue creation : VulkanTestBase(device, context, queue, nelems)
cl_platform_id platform = NULL; {}
cl_context context = NULL;
cl_command_queue cmd_queue = NULL;
cl_context_properties contextProperties[] = { CL_CONTEXT_PLATFORM, 0, 0 }; cl_int Run() override
errNum = clGetPlatformIDs(1, &platform, NULL); {
test_error(errNum, "Failed to get platform Id");
contextProperties[1] = (cl_context_properties)platform;
context = clCreateContextFromType(contextProperties, CL_DEVICE_TYPE_GPU,
NULL, NULL, &errNum);
test_error(errNum, "Unable to create context with properties");
cmd_queue = clCreateCommandQueue(context, deviceID, 0, &errNum);
test_error(errNum, "Unable to create command queue");
cl_int errNum = CL_SUCCESS;
uint32_t bufferSize = 32; uint32_t bufferSize = 32;
cl_device_id devList[] = { deviceID, NULL };
#ifdef _WIN32 #ifdef _WIN32
if (!is_extension_available(devList[0], "cl_khr_external_memory_win32")) if (!is_extension_available(device, "cl_khr_external_memory_win32"))
{ {
throw std::runtime_error("Device does not support " throw std::runtime_error(
"Device does not support "
"cl_khr_external_memory_win32 extension \n"); "cl_khr_external_memory_win32 extension \n");
} }
#else #else
if (!is_extension_available(devList[0], "cl_khr_external_memory_opaque_fd")) if (!is_extension_available(device, "cl_khr_external_memory_opaque_fd"))
{ {
throw std::runtime_error( throw std::runtime_error(
"Device does not support " "Device does not support "
@@ -77,14 +70,15 @@ int test_consistency_external_buffer(cl_device_id deviceID, cl_context _context,
VulkanExternalMemoryHandleType vkExternalMemoryHandleType = VulkanExternalMemoryHandleType vkExternalMemoryHandleType =
getSupportedVulkanExternalMemoryHandleTypeList()[0]; getSupportedVulkanExternalMemoryHandleTypeList()[0];
VulkanBuffer vkDummyBuffer(vkDevice, 4 * 1024, vkExternalMemoryHandleType); VulkanBuffer vkDummyBuffer(*vkDevice, 4 * 1024,
vkExternalMemoryHandleType);
const VulkanMemoryTypeList& memoryTypeList = const VulkanMemoryTypeList& memoryTypeList =
vkDummyBuffer.getMemoryTypeList(); vkDummyBuffer.getMemoryTypeList();
VulkanBufferList vkBufferList(1, vkDevice, bufferSize, VulkanBufferList vkBufferList(1, *vkDevice, bufferSize,
vkExternalMemoryHandleType); vkExternalMemoryHandleType);
VulkanDeviceMemory* vkDeviceMem = VulkanDeviceMemory* vkDeviceMem = new VulkanDeviceMemory(
new VulkanDeviceMemory(vkDevice, vkBufferList[0], memoryTypeList[0], *vkDevice, vkBufferList[0], memoryTypeList[0],
vkExternalMemoryHandleType); vkExternalMemoryHandleType);
vkDeviceMem->bindBuffer(vkBufferList[0], 0); vkDeviceMem->bindBuffer(vkBufferList[0], 0);
@@ -94,7 +88,7 @@ int test_consistency_external_buffer(cl_device_id deviceID, cl_context _context,
std::vector<cl_mem_properties> extMemProperties{ std::vector<cl_mem_properties> extMemProperties{
(cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_KHR, (cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_KHR,
(cl_mem_properties)devList[0], (cl_mem_properties)device,
(cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_END_KHR, (cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_END_KHR,
}; };
cl_external_memory_handle_type_khr type; cl_external_memory_handle_type_khr type;
@@ -104,14 +98,14 @@ int test_consistency_external_buffer(cl_device_id deviceID, cl_context _context,
case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32_NT: case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32_NT:
handle = vkDeviceMem->getHandle(vkExternalMemoryHandleType); handle = vkDeviceMem->getHandle(vkExternalMemoryHandleType);
type = CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KHR; type = CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KHR;
errNum = check_external_memory_handle_type(devList[0], type); errNum = check_external_memory_handle_type(device, type);
extMemProperties.push_back((cl_mem_properties)type); extMemProperties.push_back((cl_mem_properties)type);
extMemProperties.push_back((cl_mem_properties)handle); extMemProperties.push_back((cl_mem_properties)handle);
break; break;
case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32_KMT: case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32_KMT:
handle = vkDeviceMem->getHandle(vkExternalMemoryHandleType); handle = vkDeviceMem->getHandle(vkExternalMemoryHandleType);
type = CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KMT_KHR; type = CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KMT_KHR;
errNum = check_external_memory_handle_type(devList[0], type); errNum = check_external_memory_handle_type(device, type);
extMemProperties.push_back((cl_mem_properties)type); extMemProperties.push_back((cl_mem_properties)type);
extMemProperties.push_back((cl_mem_properties)handle); extMemProperties.push_back((cl_mem_properties)handle);
break; break;
@@ -119,7 +113,7 @@ int test_consistency_external_buffer(cl_device_id deviceID, cl_context _context,
case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_FD: case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_FD:
fd = (int)vkDeviceMem->getHandle(vkExternalMemoryHandleType); fd = (int)vkDeviceMem->getHandle(vkExternalMemoryHandleType);
type = CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR; type = CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR;
errNum = check_external_memory_handle_type(devList[0], type); errNum = check_external_memory_handle_type(device, type);
extMemProperties.push_back((cl_mem_properties)type); extMemProperties.push_back((cl_mem_properties)type);
extMemProperties.push_back((cl_mem_properties)fd); extMemProperties.push_back((cl_mem_properties)fd);
break; break;
@@ -140,15 +134,15 @@ int test_consistency_external_buffer(cl_device_id deviceID, cl_context _context,
clMemWrapper buffer; clMemWrapper buffer;
// Passing NULL properties and a valid extMem_desc size // Passing NULL properties and a valid extMem_desc size
buffer = clCreateBufferWithProperties(context, NULL, 1, bufferSize, NULL, buffer = clCreateBufferWithProperties(context, NULL, 1, bufferSize,
&errNum); NULL, &errNum);
test_error(errNum, "Unable to create buffer with NULL properties"); test_error(errNum, "Unable to create buffer with NULL properties");
buffer.reset(); buffer.reset();
// Passing valid extMemProperties and buffersize // Passing valid extMemProperties and buffersize
buffer = clCreateBufferWithProperties(context, extMemProperties.data(), 1, buffer = clCreateBufferWithProperties(context, extMemProperties.data(),
bufferSize, NULL, &errNum); 1, bufferSize, NULL, &errNum);
test_error(errNum, "Unable to create buffer with Properties"); test_error(errNum, "Unable to create buffer with Properties");
buffer.reset(); buffer.reset();
@@ -163,12 +157,12 @@ int test_consistency_external_buffer(cl_device_id deviceID, cl_context _context,
(cl_mem_properties)-64, // Passing random invalid fd (cl_mem_properties)-64, // Passing random invalid fd
#endif #endif
(cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_KHR, (cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_KHR,
(cl_mem_properties)devList[0], (cl_mem_properties)device,
(cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_END_KHR, (cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_END_KHR,
0 0
}; };
buffer = clCreateBufferWithProperties(context, extMemProperties2.data(), 1, buffer = clCreateBufferWithProperties(context, extMemProperties2.data(),
bufferSize, NULL, &errNum); 1, bufferSize, NULL, &errNum);
test_failure_error(errNum, CL_INVALID_VALUE, test_failure_error(errNum, CL_INVALID_VALUE,
"Should return CL_INVALID_VALUE "); "Should return CL_INVALID_VALUE ");
@@ -176,50 +170,38 @@ int test_consistency_external_buffer(cl_device_id deviceID, cl_context _context,
// Passing extMem_desc size = 0 but valid memProperties, CL_INVALID_SIZE // Passing extMem_desc size = 0 but valid memProperties, CL_INVALID_SIZE
// should be returned. // should be returned.
buffer = clCreateBufferWithProperties(context, extMemProperties.data(), 1, buffer = clCreateBufferWithProperties(context, extMemProperties.data(),
0, NULL, &errNum); 1, 0, NULL, &errNum);
test_failure_error(errNum, CL_INVALID_BUFFER_SIZE, test_failure_error(errNum, CL_INVALID_BUFFER_SIZE,
"Should return CL_INVALID_BUFFER_SIZE"); "Should return CL_INVALID_BUFFER_SIZE");
return TEST_PASS; return TEST_PASS;
} }
};
int test_consistency_external_image(cl_device_id deviceID, cl_context _context, struct ConsistencyExternalImageTest : public VulkanTestBase
cl_command_queue _queue, int num_elements)
{ {
cl_int errNum; ConsistencyExternalImageTest(cl_device_id device, cl_context context,
VulkanDevice vkDevice; cl_command_queue queue, cl_int nelems)
: VulkanTestBase(device, context, queue, nelems)
{}
// Context and command queue creation cl_int Run() override
cl_platform_id platform = NULL; {
cl_context context = NULL; cl_int errNum = CL_SUCCESS;
cl_command_queue cmd_queue = NULL;
cl_context_properties contextProperties[] = { CL_CONTEXT_PLATFORM, 0, 0 };
errNum = clGetPlatformIDs(1, &platform, NULL);
test_error(errNum, "Failed to get platform id");
contextProperties[1] = (cl_context_properties)platform;
context = clCreateContextFromType(contextProperties, CL_DEVICE_TYPE_GPU,
NULL, NULL, &errNum);
test_error(errNum, "Unable to create context with properties");
cmd_queue = clCreateCommandQueue(context, deviceID, 0, &errNum);
test_error(errNum, "Unable to create command queue");
cl_device_id devList[] = { deviceID, NULL };
#ifdef _WIN32 #ifdef _WIN32
if (!is_extension_available(devList[0], "cl_khr_external_memory_win32")) if (!is_extension_available(device, "cl_khr_external_memory_win32"))
{ {
throw std::runtime_error("Device does not support" throw std::runtime_error(
"Device does not support"
"cl_khr_external_memory_win32 extension \n"); "cl_khr_external_memory_win32 extension \n");
} }
#else #else
if (!is_extension_available(devList[0], "cl_khr_external_memory_opaque_fd")) if (!is_extension_available(device, "cl_khr_external_memory_opaque_fd"))
{ {
test_fail("Device does not support cl_khr_external_memory_opaque_fd " test_fail(
"Device does not support cl_khr_external_memory_opaque_fd "
"extension \n"); "extension \n");
} }
#endif #endif
@@ -234,30 +216,32 @@ int test_consistency_external_image(cl_device_id deviceID, cl_context _context,
VulkanImageTiling vulkanImageTiling = VulkanImageTiling vulkanImageTiling =
vkClExternalMemoryHandleTilingAssumption( vkClExternalMemoryHandleTilingAssumption(
deviceID, vkExternalMemoryHandleType, &errNum); device, vkExternalMemoryHandleType, &errNum);
ASSERT_SUCCESS(errNum, "Failed to query OpenCL tiling mode"); ASSERT_SUCCESS(errNum, "Failed to query OpenCL tiling mode");
VulkanImage2D vkImage2D = VulkanImage2D vkImage2D = VulkanImage2D(
VulkanImage2D(vkDevice, VULKAN_FORMAT_R8G8B8A8_UNORM, width, height, *vkDevice, VULKAN_FORMAT_R8G8B8A8_UNORM, width, height,
vulkanImageTiling, 1, vkExternalMemoryHandleType); vulkanImageTiling, 1, vkExternalMemoryHandleType);
const VulkanMemoryTypeList& memoryTypeList = vkImage2D.getMemoryTypeList(); const VulkanMemoryTypeList& memoryTypeList =
vkImage2D.getMemoryTypeList();
uint64_t totalImageMemSize = vkImage2D.getSize(); uint64_t totalImageMemSize = vkImage2D.getSize();
log_info("Memory type index: %lu\n", (uint32_t)memoryTypeList[0]); log_info("Memory type index: %u\n", (uint32_t)memoryTypeList[0]);
log_info("Memory type property: %d\n", log_info("Memory type property: %d\n",
memoryTypeList[0].getMemoryTypeProperty()); memoryTypeList[0].getMemoryTypeProperty());
log_info("Image size : %d\n", totalImageMemSize); log_info("Image size : %ld\n", totalImageMemSize);
VulkanDeviceMemory* vkDeviceMem = new VulkanDeviceMemory( VulkanDeviceMemory* vkDeviceMem =
vkDevice, vkImage2D, memoryTypeList[0], vkExternalMemoryHandleType); new VulkanDeviceMemory(*vkDevice, vkImage2D, memoryTypeList[0],
vkExternalMemoryHandleType);
vkDeviceMem->bindImage(vkImage2D, 0); vkDeviceMem->bindImage(vkImage2D, 0);
void* handle = NULL; void* handle = NULL;
int fd; int fd;
std::vector<cl_mem_properties> extMemProperties{ std::vector<cl_mem_properties> extMemProperties{
(cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_KHR, (cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_KHR,
(cl_mem_properties)devList[0], (cl_mem_properties)device,
(cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_END_KHR, (cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_END_KHR,
}; };
switch (vkExternalMemoryHandleType) switch (vkExternalMemoryHandleType)
@@ -266,15 +250,16 @@ int test_consistency_external_image(cl_device_id deviceID, cl_context _context,
case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32_NT: case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32_NT:
handle = vkDeviceMem->getHandle(vkExternalMemoryHandleType); handle = vkDeviceMem->getHandle(vkExternalMemoryHandleType);
errNum = check_external_memory_handle_type( errNum = check_external_memory_handle_type(
devList[0], CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KHR); device, CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KHR);
extMemProperties.push_back( extMemProperties.push_back(
(cl_mem_properties)CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KHR); (cl_mem_properties)
CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KHR);
extMemProperties.push_back((cl_mem_properties)handle); extMemProperties.push_back((cl_mem_properties)handle);
break; break;
case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32_KMT: case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32_KMT:
handle = vkDeviceMem->getHandle(vkExternalMemoryHandleType); handle = vkDeviceMem->getHandle(vkExternalMemoryHandleType);
errNum = check_external_memory_handle_type( errNum = check_external_memory_handle_type(
devList[0], CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KMT_KHR); device, CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KMT_KHR);
extMemProperties.push_back( extMemProperties.push_back(
(cl_mem_properties) (cl_mem_properties)
CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KMT_KHR); CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KMT_KHR);
@@ -284,7 +269,7 @@ int test_consistency_external_image(cl_device_id deviceID, cl_context _context,
case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_FD: case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_FD:
fd = (int)vkDeviceMem->getHandle(vkExternalMemoryHandleType); fd = (int)vkDeviceMem->getHandle(vkExternalMemoryHandleType);
errNum = check_external_memory_handle_type( errNum = check_external_memory_handle_type(
devList[0], CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR); device, CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR);
extMemProperties.push_back( extMemProperties.push_back(
(cl_mem_properties)CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR); (cl_mem_properties)CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR);
extMemProperties.push_back((cl_mem_properties)fd); extMemProperties.push_back((cl_mem_properties)fd);
@@ -306,8 +291,9 @@ int test_consistency_external_image(cl_device_id deviceID, cl_context _context,
const VkImageCreateInfo VulkanImageCreateInfo = const VkImageCreateInfo VulkanImageCreateInfo =
vkImage2D.getVkImageCreateInfo(); vkImage2D.getVkImageCreateInfo();
errNum = getCLImageInfoFromVkImageInfo( errNum = getCLImageInfoFromVkImageInfo(&VulkanImageCreateInfo,
&VulkanImageCreateInfo, totalImageMemSize, &img_format, &image_desc); totalImageMemSize, &img_format,
&image_desc);
if (errNum != CL_SUCCESS) if (errNum != CL_SUCCESS)
{ {
log_error("getCLImageInfoFromVkImageInfo failed!!!"); log_error("getCLImageInfoFromVkImageInfo failed!!!");
@@ -325,8 +311,8 @@ int test_consistency_external_image(cl_device_id deviceID, cl_context _context,
// 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,
NULL, &errNum); &image_desc, NULL, &errNum);
test_failure_error(errNum, CL_INVALID_IMAGE_FORMAT_DESCRIPTOR, test_failure_error(errNum, CL_INVALID_IMAGE_FORMAT_DESCRIPTOR,
"Image creation must fail with " "Image creation must fail with "
"CL_INVALID_IMAGE_FORMAT_DESCRIPTOR" "CL_INVALID_IMAGE_FORMAT_DESCRIPTOR"
@@ -336,8 +322,8 @@ int test_consistency_external_image(cl_device_id deviceID, cl_context _context,
// Passing image_desc as NULL // Passing image_desc as NULL
image = clCreateImageWithProperties(context, extMemProperties.data(), image = clCreateImageWithProperties(context, extMemProperties.data(),
CL_MEM_READ_WRITE, &img_format, NULL, CL_MEM_READ_WRITE, &img_format,
NULL, &errNum); NULL, NULL, &errNum);
test_failure_error(errNum, CL_INVALID_IMAGE_DESCRIPTOR, test_failure_error(errNum, CL_INVALID_IMAGE_DESCRIPTOR,
"Image creation must fail with " "Image creation must fail with "
"CL_INVALID_IMAGE_DESCRIPTOR " "CL_INVALID_IMAGE_DESCRIPTOR "
@@ -345,38 +331,40 @@ int test_consistency_external_image(cl_device_id deviceID, cl_context _context,
image.reset(); image.reset();
return TEST_PASS; return TEST_PASS;
} }
};
int test_consistency_external_semaphore(cl_device_id deviceID, struct ConsistencyExternalSemaphoreTest : public VulkanTestBase
cl_context _context,
cl_command_queue _queue,
int num_elements)
{ {
cl_int errNum; ConsistencyExternalSemaphoreTest(cl_device_id device, cl_context context,
VulkanDevice vkDevice; cl_command_queue queue, cl_int nelems)
// Context and command queue creation : VulkanTestBase(device, context, queue, nelems)
cl_platform_id platform = NULL; {}
cl_context context = NULL;
cl_command_queue cmd_queue = NULL;
errNum = clGetPlatformIDs(1, &platform, NULL); cl_int Run() override
test_error(errNum, "Failed to get platform Id"); {
cl_int errNum = CL_SUCCESS;
cl_context_properties contextProperties[] = { CL_CONTEXT_PLATFORM, 0, 0 }; #ifdef _WIN32
if (!is_extension_available(device, "cl_khr_external_memory_win32"))
{
throw std::runtime_error(
"Device does not support"
"cl_khr_external_memory_win32 extension \n");
}
#else
if (!is_extension_available(device, "cl_khr_external_memory_opaque_fd"))
{
test_fail(
"Device does not support cl_khr_external_memory_opaque_fd "
"extension \n");
}
#endif
contextProperties[1] = (cl_context_properties)platform; std::vector<VulkanExternalSemaphoreHandleType>
supportedExternalSemaphores =
context = clCreateContextFromType(contextProperties, CL_DEVICE_TYPE_GPU, getSupportedInteropExternalSemaphoreHandleTypes(device,
NULL, NULL, &errNum); *vkDevice);
test_error(errNum, "Unable to create context with properties");
cmd_queue = clCreateCommandQueue(context, deviceID, 0, &errNum);
test_error(errNum, "Unable to create command queue");
cl_device_id devList[] = { deviceID, NULL };
std::vector<VulkanExternalSemaphoreHandleType> supportedExternalSemaphores =
getSupportedInteropExternalSemaphoreHandleTypes(devList[0], vkDevice);
if (supportedExternalSemaphores.empty()) if (supportedExternalSemaphores.empty())
{ {
@@ -386,8 +374,8 @@ int test_consistency_external_semaphore(cl_device_id deviceID,
for (VulkanExternalSemaphoreHandleType semaphoreHandleType : for (VulkanExternalSemaphoreHandleType semaphoreHandleType :
supportedExternalSemaphores) supportedExternalSemaphores)
{ {
VulkanSemaphore vkVk2Clsemaphore(vkDevice, semaphoreHandleType); VulkanSemaphore vkVk2Clsemaphore(*vkDevice, semaphoreHandleType);
VulkanSemaphore vkCl2Vksemaphore(vkDevice, semaphoreHandleType); VulkanSemaphore vkCl2Vksemaphore(*vkDevice, semaphoreHandleType);
cl_semaphore_khr clCl2Vksemaphore; cl_semaphore_khr clCl2Vksemaphore;
cl_semaphore_khr clVk2Clsemaphore; cl_semaphore_khr clVk2Clsemaphore;
void* handle1 = NULL; void* handle1 = NULL;
@@ -405,25 +393,28 @@ int test_consistency_external_semaphore(cl_device_id deviceID,
{ {
#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(semaphoreHandleType); handle1 = vkVk2Clsemaphore.getHandle(semaphoreHandleType);
handle2 = vkCl2Vksemaphore.getHandle(semaphoreHandleType); handle2 = vkCl2Vksemaphore.getHandle(semaphoreHandleType);
errNum = check_external_semaphore_handle_type( errNum = check_external_semaphore_handle_type(
devList[0], CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KHR); device, 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( log_info(" Opaque D3DKMT handles are only supported on "
" Opaque D3DKMT handles are only supported on Windows\n"); "Windows\n");
handle1 = vkVk2Clsemaphore.getHandle(semaphoreHandleType); handle1 = vkVk2Clsemaphore.getHandle(semaphoreHandleType);
handle2 = vkCl2Vksemaphore.getHandle(semaphoreHandleType); handle2 = vkCl2Vksemaphore.getHandle(semaphoreHandleType);
errNum = check_external_semaphore_handle_type( errNum = check_external_semaphore_handle_type(
devList[0], CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KMT_KHR); device, CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KMT_KHR);
sema_props1.push_back( sema_props1.push_back(
(cl_semaphore_properties_khr) (cl_semaphore_properties_khr)
CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KMT_KHR); CL_SEMAPHORE_HANDLE_OPAQUE_WIN32_KMT_KHR);
@@ -438,11 +429,13 @@ int test_consistency_external_semaphore(cl_device_id deviceID,
fd1 = (int)vkVk2Clsemaphore.getHandle(semaphoreHandleType); fd1 = (int)vkVk2Clsemaphore.getHandle(semaphoreHandleType);
fd2 = (int)vkCl2Vksemaphore.getHandle(semaphoreHandleType); fd2 = (int)vkCl2Vksemaphore.getHandle(semaphoreHandleType);
errNum = check_external_semaphore_handle_type( errNum = check_external_semaphore_handle_type(
devList[0], CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR); device, CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR);
sema_props1.push_back((cl_semaphore_properties_khr) sema_props1.push_back(
(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_props1.push_back((cl_semaphore_properties_khr)fd1);
sema_props2.push_back((cl_semaphore_properties_khr) sema_props2.push_back(
(cl_semaphore_properties_khr)
CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR); CL_SEMAPHORE_HANDLE_OPAQUE_FD_KHR);
sema_props2.push_back((cl_semaphore_properties_khr)fd2); sema_props2.push_back((cl_semaphore_properties_khr)fd2);
break; break;
@@ -450,7 +443,7 @@ int test_consistency_external_semaphore(cl_device_id deviceID,
fd1 = -1; fd1 = -1;
fd2 = -1; fd2 = -1;
errNum = check_external_semaphore_handle_type( errNum = check_external_semaphore_handle_type(
devList[0], CL_SEMAPHORE_HANDLE_SYNC_FD_KHR); device, CL_SEMAPHORE_HANDLE_SYNC_FD_KHR);
sema_props1.push_back((cl_semaphore_properties_khr) sema_props1.push_back((cl_semaphore_properties_khr)
CL_SEMAPHORE_HANDLE_SYNC_FD_KHR); CL_SEMAPHORE_HANDLE_SYNC_FD_KHR);
sema_props1.push_back((cl_semaphore_properties_khr)fd1); sema_props1.push_back((cl_semaphore_properties_khr)fd1);
@@ -459,68 +452,70 @@ int test_consistency_external_semaphore(cl_device_id deviceID,
sema_props2.push_back((cl_semaphore_properties_khr)fd2); sema_props2.push_back((cl_semaphore_properties_khr)fd2);
break; 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(
"Unsupported external sempahore handle type\n "); "Unsupported external sempahore handle type\n ");
} }
sema_props1.push_back( sema_props1.push_back((cl_semaphore_properties_khr)
(cl_semaphore_properties_khr)CL_SEMAPHORE_DEVICE_HANDLE_LIST_KHR); CL_SEMAPHORE_DEVICE_HANDLE_LIST_KHR);
sema_props1.push_back((cl_semaphore_properties_khr)devList[0]); sema_props1.push_back((cl_semaphore_properties_khr)device);
sema_props1.push_back( sema_props1.push_back((cl_semaphore_properties_khr)
(cl_semaphore_properties_khr)CL_SEMAPHORE_DEVICE_HANDLE_LIST_END_KHR); CL_SEMAPHORE_DEVICE_HANDLE_LIST_END_KHR);
sema_props2.push_back( sema_props2.push_back((cl_semaphore_properties_khr)
(cl_semaphore_properties_khr)CL_SEMAPHORE_DEVICE_HANDLE_LIST_KHR); CL_SEMAPHORE_DEVICE_HANDLE_LIST_KHR);
sema_props2.push_back((cl_semaphore_properties_khr)devList[0]); sema_props2.push_back((cl_semaphore_properties_khr)device);
sema_props2.push_back( sema_props2.push_back((cl_semaphore_properties_khr)
(cl_semaphore_properties_khr)CL_SEMAPHORE_DEVICE_HANDLE_LIST_END_KHR); CL_SEMAPHORE_DEVICE_HANDLE_LIST_END_KHR);
sema_props1.push_back(0); sema_props1.push_back(0);
sema_props2.push_back(0); sema_props2.push_back(0);
// Pass NULL properties // Pass NULL properties
cl_semaphore_khr cl_ext_semaphore =
clCreateSemaphoreWithPropertiesKHRptr(context, NULL, &errNum); clCreateSemaphoreWithPropertiesKHRptr(context, NULL, &errNum);
test_failure_error(errNum, CL_INVALID_VALUE, test_failure_error(
errNum, CL_INVALID_VALUE,
"Semaphore creation must fail with CL_INVALID_VALUE " "Semaphore creation must fail with CL_INVALID_VALUE "
" when properties are passed as NULL"); " when properties are passed as NULL");
// Pass invalid semaphore object to wait // Pass invalid semaphore object to wait
errNum = errNum = clEnqueueWaitSemaphoresKHRptr(queue, 1, NULL, NULL, 0,
clEnqueueWaitSemaphoresKHRptr(cmd_queue, 1, NULL, NULL, 0, NULL, NULL); NULL, NULL);
test_failure_error(errNum, CL_INVALID_VALUE, test_failure_error(
errNum, CL_INVALID_VALUE,
"clEnqueueWaitSemaphoresKHR fails with CL_INVALID_VALUE " "clEnqueueWaitSemaphoresKHR fails with CL_INVALID_VALUE "
"when invalid semaphore object is passed"); "when invalid semaphore object is passed");
// Pass invalid semaphore object to signal // Pass invalid semaphore object to signal
errNum = clEnqueueSignalSemaphoresKHRptr(cmd_queue, 1, NULL, NULL, 0, NULL, errNum = clEnqueueSignalSemaphoresKHRptr(queue, 1, NULL, NULL, 0,
NULL); NULL, NULL);
test_failure_error( test_failure_error(
errNum, CL_INVALID_VALUE, errNum, CL_INVALID_VALUE,
"clEnqueueSignalSemaphoresKHR fails with CL_INVALID_VALUE" "clEnqueueSignalSemaphoresKHR fails with CL_INVALID_VALUE"
"when invalid semaphore object is passed"); "when invalid semaphore object is passed");
// Create two semaphore objects // Create two semaphore objects
clVk2Clsemaphore = clCreateSemaphoreWithPropertiesKHRptr( clVk2Clsemaphore = clCreateSemaphoreWithPropertiesKHRptr(
context, sema_props1.data(), &errNum); context, sema_props1.data(), &errNum);
test_error(errNum, test_error(
errNum,
"Unable to create semaphore with valid semaphore properties"); "Unable to create semaphore with valid semaphore properties");
clCl2Vksemaphore = clCreateSemaphoreWithPropertiesKHRptr( clCl2Vksemaphore = clCreateSemaphoreWithPropertiesKHRptr(
context, sema_props2.data(), &errNum); context, sema_props2.data(), &errNum);
test_error(errNum, test_error(
errNum,
"Unable to create semaphore with valid semaphore properties"); "Unable to create semaphore with valid semaphore properties");
// Pass invalid object to release call // Pass invalid object to release call
errNum = clReleaseSemaphoreKHRptr(NULL); errNum = clReleaseSemaphoreKHRptr(NULL);
test_failure_error( test_failure_error(errNum, CL_INVALID_SEMAPHORE_KHR,
errNum, CL_INVALID_SEMAPHORE_KHR,
"clReleaseSemaphoreKHRptr fails with " "clReleaseSemaphoreKHRptr fails with "
"CL_INVALID_SEMAPHORE_KHR when NULL semaphore object is passed"); "CL_INVALID_SEMAPHORE_KHR when NULL semaphore "
"object is passed");
// Release both semaphore objects // Release both semaphore objects
errNum = clReleaseSemaphoreKHRptr(clVk2Clsemaphore); errNum = clReleaseSemaphoreKHRptr(clVk2Clsemaphore);
@@ -531,4 +526,32 @@ int test_consistency_external_semaphore(cl_device_id deviceID,
} }
return TEST_PASS; return TEST_PASS;
}
};
} // anonymous namespace
int test_consistency_external_buffer(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
return MakeAndRunTest<ConsistencyExternalBufferTest>(
deviceID, context, defaultQueue, num_elements);
}
int test_consistency_external_image(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
return MakeAndRunTest<ConsistencyExternalImageTest>(
deviceID, context, defaultQueue, num_elements);
}
int test_consistency_external_semaphore(cl_device_id deviceID,
cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
return MakeAndRunTest<ConsistencyExternalSemaphoreTest>(
deviceID, context, defaultQueue, num_elements);
} }

117
test_conformance/vulkan/test_vulkan_api_consistency_for_1dimages.cpp
View File

@@ -1,3 +1,19 @@
//
// Copyright (c) 2024 The Khronos Group Inc.
//
// Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License.
// You may obtain a copy of the License at
//
// http://www.apache.org/licenses/LICENSE-2.0
//
// Unless required by applicable law or agreed to in writing, software
// distributed under the License is distributed on an "AS IS" BASIS,
// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
// See the License for the specific language governing permissions and
// limitations under the License.
//
#include <vulkan_interop_common.hpp> #include <vulkan_interop_common.hpp>
#include <opencl_vulkan_wrapper.hpp> #include <opencl_vulkan_wrapper.hpp>
#include <vulkan_wrapper.hpp> #include <vulkan_wrapper.hpp>
@@ -17,42 +33,31 @@
#include "harness/typeWrappers.h" #include "harness/typeWrappers.h"
#include "harness/deviceInfo.h" #include "harness/deviceInfo.h"
int test_consistency_external_for_1dimage(cl_device_id deviceID, #include "vulkan_test_base.h"
cl_context _context, #include "opencl_vulkan_wrapper.hpp"
cl_command_queue _queue,
int num_elements) namespace {
struct ConsistencyExternalImage1DTest : public VulkanTestBase
{ {
cl_int errNum; ConsistencyExternalImage1DTest(cl_device_id device, cl_context context,
VulkanDevice vkDevice; cl_command_queue queue, cl_int nelems)
: VulkanTestBase(device, context, queue, nelems)
{}
// Context and command queue creation cl_int Run() override
cl_platform_id platform = NULL; {
cl_context context = NULL; cl_int errNum = CL_SUCCESS;
cl_command_queue cmd_queue = NULL;
cl_context_properties contextProperties[] = { CL_CONTEXT_PLATFORM, 0, 0 };
errNum = clGetPlatformIDs(1, &platform, NULL);
test_error(errNum, "Failed to get platform id");
contextProperties[1] = (cl_context_properties)platform;
context = clCreateContextFromType(contextProperties, CL_DEVICE_TYPE_GPU,
NULL, NULL, &errNum);
test_error(errNum, "Unable to create context with properties");
cmd_queue = clCreateCommandQueue(context, deviceID, 0, &errNum);
test_error(errNum, "Unable to create command queue");
cl_device_id devList[] = { deviceID, NULL };
#ifdef _WIN32 #ifdef _WIN32
if (!is_extension_available(devList[0], "cl_khr_external_memory_win32")) if (!is_extension_available(device, "cl_khr_external_memory_win32"))
{ {
throw std::runtime_error("Device does not support" throw std::runtime_error(
"Device does not support"
"cl_khr_external_memory_win32 extension \n"); "cl_khr_external_memory_win32 extension \n");
} }
#else #else
if (!is_extension_available(devList[0], "cl_khr_external_memory_opaque_fd")) if (!is_extension_available(device, "cl_khr_external_memory_opaque_fd"))
{ {
throw std::runtime_error( throw std::runtime_error(
"Device does not support cl_khr_external_memory_opaque_fd " "Device does not support cl_khr_external_memory_opaque_fd "
@@ -69,14 +74,15 @@ int test_consistency_external_for_1dimage(cl_device_id deviceID,
VulkanImageTiling vulkanImageTiling = VulkanImageTiling vulkanImageTiling =
vkClExternalMemoryHandleTilingAssumption( vkClExternalMemoryHandleTilingAssumption(
deviceID, vkExternalMemoryHandleType, &errNum); device, vkExternalMemoryHandleType, &errNum);
ASSERT_SUCCESS(errNum, "Failed to query OpenCL tiling mode"); ASSERT_SUCCESS(errNum, "Failed to query OpenCL tiling mode");
VulkanImage1D vkImage1D = VulkanImage1D vkImage1D =
VulkanImage1D(vkDevice, VULKAN_FORMAT_R8G8B8A8_UNORM, width, VulkanImage1D(*vkDevice, VULKAN_FORMAT_R8G8B8A8_UNORM, width,
vulkanImageTiling, 1, vkExternalMemoryHandleType); vulkanImageTiling, 1, vkExternalMemoryHandleType);
const VulkanMemoryTypeList& memoryTypeList = vkImage1D.getMemoryTypeList(); const VulkanMemoryTypeList& memoryTypeList =
vkImage1D.getMemoryTypeList();
uint64_t totalImageMemSize = vkImage1D.getSize(); uint64_t totalImageMemSize = vkImage1D.getSize();
log_info("Memory type index: %u\n", (uint32_t)memoryTypeList[0]); log_info("Memory type index: %u\n", (uint32_t)memoryTypeList[0]);
@@ -84,15 +90,16 @@ int test_consistency_external_for_1dimage(cl_device_id deviceID,
memoryTypeList[0].getMemoryTypeProperty()); memoryTypeList[0].getMemoryTypeProperty());
log_info("Image size : %lu\n", totalImageMemSize); log_info("Image size : %lu\n", totalImageMemSize);
VulkanDeviceMemory* vkDeviceMem = new VulkanDeviceMemory( VulkanDeviceMemory* vkDeviceMem =
vkDevice, vkImage1D, memoryTypeList[0], vkExternalMemoryHandleType); new VulkanDeviceMemory(*vkDevice, vkImage1D, memoryTypeList[0],
vkExternalMemoryHandleType);
vkDeviceMem->bindImage(vkImage1D, 0); vkDeviceMem->bindImage(vkImage1D, 0);
void* handle = NULL; void* handle = NULL;
int fd; int fd;
std::vector<cl_mem_properties> extMemProperties{ std::vector<cl_mem_properties> extMemProperties{
(cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_KHR, (cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_KHR,
(cl_mem_properties)devList[0], (cl_mem_properties)device,
(cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_END_KHR, (cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_END_KHR,
}; };
switch (vkExternalMemoryHandleType) switch (vkExternalMemoryHandleType)
@@ -101,15 +108,16 @@ int test_consistency_external_for_1dimage(cl_device_id deviceID,
case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32_NT: case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32_NT:
handle = vkDeviceMem->getHandle(vkExternalMemoryHandleType); handle = vkDeviceMem->getHandle(vkExternalMemoryHandleType);
errNum = check_external_memory_handle_type( errNum = check_external_memory_handle_type(
devList[0], CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KHR); device, CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KHR);
extMemProperties.push_back( extMemProperties.push_back(
(cl_mem_properties)CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KHR); (cl_mem_properties)
CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KHR);
extMemProperties.push_back((cl_mem_properties)handle); extMemProperties.push_back((cl_mem_properties)handle);
break; break;
case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32_KMT: case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32_KMT:
handle = vkDeviceMem->getHandle(vkExternalMemoryHandleType); handle = vkDeviceMem->getHandle(vkExternalMemoryHandleType);
errNum = check_external_memory_handle_type( errNum = check_external_memory_handle_type(
devList[0], CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KMT_KHR); device, CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KMT_KHR);
extMemProperties.push_back( extMemProperties.push_back(
(cl_mem_properties) (cl_mem_properties)
CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KMT_KHR); CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KMT_KHR);
@@ -119,7 +127,7 @@ int test_consistency_external_for_1dimage(cl_device_id deviceID,
case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_FD: case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_FD:
fd = (int)vkDeviceMem->getHandle(vkExternalMemoryHandleType); fd = (int)vkDeviceMem->getHandle(vkExternalMemoryHandleType);
errNum = check_external_memory_handle_type( errNum = check_external_memory_handle_type(
devList[0], CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR); device, CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR);
extMemProperties.push_back( extMemProperties.push_back(
(cl_mem_properties)CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR); (cl_mem_properties)CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR);
extMemProperties.push_back((cl_mem_properties)fd); extMemProperties.push_back((cl_mem_properties)fd);
@@ -141,8 +149,9 @@ int test_consistency_external_for_1dimage(cl_device_id deviceID,
const VkImageCreateInfo VulkanImageCreateInfo = const VkImageCreateInfo VulkanImageCreateInfo =
vkImage1D.getVkImageCreateInfo(); vkImage1D.getVkImageCreateInfo();
errNum = getCLImageInfoFromVkImageInfo( errNum = getCLImageInfoFromVkImageInfo(&VulkanImageCreateInfo,
&VulkanImageCreateInfo, totalImageMemSize, &img_format, &image_desc); totalImageMemSize, &img_format,
&image_desc);
if (errNum != CL_SUCCESS) if (errNum != CL_SUCCESS)
{ {
log_error("getCLImageInfoFromVkImageInfo failed!!!"); log_error("getCLImageInfoFromVkImageInfo failed!!!");
@@ -159,9 +168,9 @@ int test_consistency_external_for_1dimage(cl_device_id deviceID,
image.reset(); image.reset();
// Passing NULL properties and a valid image_format and image_desc // Passing NULL properties and a valid image_format and image_desc
image = image = clCreateImageWithProperties(context, NULL, CL_MEM_READ_WRITE,
clCreateImageWithProperties(context, NULL, CL_MEM_READ_WRITE, &img_format, &image_desc, NULL,
&img_format, &image_desc, NULL, &errNum); &errNum);
test_error(errNum, test_error(errNum,
"Unable to create image with NULL properties " "Unable to create image with NULL properties "
"with valid image format and image desc"); "with valid image format and image desc");
@@ -170,8 +179,8 @@ int test_consistency_external_for_1dimage(cl_device_id deviceID,
// 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,
NULL, &errNum); &image_desc, NULL, &errNum);
test_failure_error(errNum, CL_INVALID_IMAGE_FORMAT_DESCRIPTOR, test_failure_error(errNum, CL_INVALID_IMAGE_FORMAT_DESCRIPTOR,
"Image creation must fail with " "Image creation must fail with "
"CL_INVALID_IMAGE_FORMAT_DESCRIPTOR" "CL_INVALID_IMAGE_FORMAT_DESCRIPTOR"
@@ -181,16 +190,24 @@ int test_consistency_external_for_1dimage(cl_device_id deviceID,
// Passing image_desc as NULL // Passing image_desc as NULL
image = clCreateImageWithProperties(context, extMemProperties.data(), image = clCreateImageWithProperties(context, extMemProperties.data(),
CL_MEM_READ_WRITE, &img_format, NULL, CL_MEM_READ_WRITE, &img_format,
NULL, &errNum); NULL, NULL, &errNum);
test_failure_error(errNum, CL_INVALID_IMAGE_DESCRIPTOR, test_failure_error(errNum, CL_INVALID_IMAGE_DESCRIPTOR,
"Image creation must fail with " "Image creation must fail with "
"CL_INVALID_IMAGE_DESCRIPTOR " "CL_INVALID_IMAGE_DESCRIPTOR "
"when image desc passed as NULL"); "when image desc passed as NULL");
image.reset(); image.reset();
if (cmd_queue) clReleaseCommandQueue(cmd_queue);
if (context) clReleaseContext(context);
return TEST_PASS; return TEST_PASS;
}
};
}
int test_consistency_external_for_1dimage(cl_device_id deviceID,
cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
return MakeAndRunTest<ConsistencyExternalImage1DTest>(
deviceID, context, defaultQueue, num_elements);
} }

121
test_conformance/vulkan/test_vulkan_api_consistency_for_3dimages.cpp
View File

@@ -1,3 +1,19 @@
//
// Copyright (c) 2024 The Khronos Group Inc.
//
// Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License.
// You may obtain a copy of the License at
//
// http://www.apache.org/licenses/LICENSE-2.0
//
// Unless required by applicable law or agreed to in writing, software
// distributed under the License is distributed on an "AS IS" BASIS,
// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
// See the License for the specific language governing permissions and
// limitations under the License.
//
#include <vulkan_interop_common.hpp> #include <vulkan_interop_common.hpp>
#include <opencl_vulkan_wrapper.hpp> #include <opencl_vulkan_wrapper.hpp>
#include <vulkan_wrapper.hpp> #include <vulkan_wrapper.hpp>
@@ -18,43 +34,31 @@
#include "harness/deviceInfo.h" #include "harness/deviceInfo.h"
#include <string> #include <string>
#include "vulkan_test_base.h"
#include "opencl_vulkan_wrapper.hpp"
int test_consistency_external_for_3dimage(cl_device_id deviceID, namespace {
cl_context _context,
cl_command_queue _queue, struct ConsistencyExternalImage3DTest : public VulkanTestBase
int num_elements)
{ {
ConsistencyExternalImage3DTest(cl_device_id device, cl_context context,
cl_command_queue queue, cl_int nelems)
: VulkanTestBase(device, context, queue, nelems)
{}
cl_int Run() override
{
cl_int errNum; cl_int errNum;
VulkanDevice vkDevice;
// Context and command queue creation
cl_platform_id platform = NULL;
cl_context context = NULL;
cl_command_queue cmd_queue = NULL;
cl_context_properties contextProperties[] = { CL_CONTEXT_PLATFORM, 0, 0 };
errNum = clGetPlatformIDs(1, &platform, NULL);
test_error(errNum, "Failed to get platform id");
contextProperties[1] = (cl_context_properties)platform;
context = clCreateContextFromType(contextProperties, CL_DEVICE_TYPE_GPU,
NULL, NULL, &errNum);
test_error(errNum, "Unable to create context with properties");
cmd_queue = clCreateCommandQueue(context, deviceID, 0, &errNum);
test_error(errNum, "Unable to create command queue");
cl_device_id devList[] = { deviceID, NULL };
#ifdef _WIN32 #ifdef _WIN32
if (!is_extension_available(devList[0], "cl_khr_external_memory_win32")) if (!is_extension_available(device, "cl_khr_external_memory_win32"))
{ {
throw std::runtime_error("Device does not support" throw std::runtime_error(
"Device does not support"
"cl_khr_external_memory_win32 extension \n"); "cl_khr_external_memory_win32 extension \n");
} }
#else #else
if (!is_extension_available(devList[0], "cl_khr_external_memory_opaque_fd")) if (!is_extension_available(device, "cl_khr_external_memory_opaque_fd"))
{ {
throw std::runtime_error( throw std::runtime_error(
"Device does not support cl_khr_external_memory_opaque_fd " "Device does not support cl_khr_external_memory_opaque_fd "
@@ -73,14 +77,15 @@ int test_consistency_external_for_3dimage(cl_device_id deviceID,
VulkanImageTiling vulkanImageTiling = VulkanImageTiling vulkanImageTiling =
vkClExternalMemoryHandleTilingAssumption( vkClExternalMemoryHandleTilingAssumption(
deviceID, vkExternalMemoryHandleType, &errNum); device, vkExternalMemoryHandleType, &errNum);
ASSERT_SUCCESS(errNum, "Failed to query OpenCL tiling mode"); ASSERT_SUCCESS(errNum, "Failed to query OpenCL tiling mode");
VulkanImage3D vkImage3D = VulkanImage3D vkImage3D = VulkanImage3D(
VulkanImage3D(vkDevice, VULKAN_FORMAT_R8G8B8A8_UNORM, width, height, *vkDevice, VULKAN_FORMAT_R8G8B8A8_UNORM, width, height, depth,
depth, vulkanImageTiling, 1, vkExternalMemoryHandleType); vulkanImageTiling, 1, vkExternalMemoryHandleType);
const VulkanMemoryTypeList& memoryTypeList = vkImage3D.getMemoryTypeList(); const VulkanMemoryTypeList& memoryTypeList =
vkImage3D.getMemoryTypeList();
uint64_t totalImageMemSize = vkImage3D.getSize(); uint64_t totalImageMemSize = vkImage3D.getSize();
log_info("Memory type index: %u\n", (uint32_t)memoryTypeList[0]); log_info("Memory type index: %u\n", (uint32_t)memoryTypeList[0]);
@@ -88,15 +93,16 @@ int test_consistency_external_for_3dimage(cl_device_id deviceID,
memoryTypeList[0].getMemoryTypeProperty()); memoryTypeList[0].getMemoryTypeProperty());
log_info("Image size : %lu\n", totalImageMemSize); log_info("Image size : %lu\n", totalImageMemSize);
VulkanDeviceMemory* vkDeviceMem = new VulkanDeviceMemory( VulkanDeviceMemory* vkDeviceMem =
vkDevice, vkImage3D, memoryTypeList[0], vkExternalMemoryHandleType); new VulkanDeviceMemory(*vkDevice, vkImage3D, memoryTypeList[0],
vkExternalMemoryHandleType);
vkDeviceMem->bindImage(vkImage3D, 0); vkDeviceMem->bindImage(vkImage3D, 0);
void* handle = NULL; void* handle = NULL;
int fd; int fd;
std::vector<cl_mem_properties> extMemProperties{ std::vector<cl_mem_properties> extMemProperties{
(cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_KHR, (cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_KHR,
(cl_mem_properties)devList[0], (cl_mem_properties)device,
(cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_END_KHR, (cl_mem_properties)CL_MEM_DEVICE_HANDLE_LIST_END_KHR,
}; };
switch (vkExternalMemoryHandleType) switch (vkExternalMemoryHandleType)
@@ -105,15 +111,16 @@ int test_consistency_external_for_3dimage(cl_device_id deviceID,
case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32_NT: case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32_NT:
handle = vkDeviceMem->getHandle(vkExternalMemoryHandleType); handle = vkDeviceMem->getHandle(vkExternalMemoryHandleType);
errNum = check_external_memory_handle_type( errNum = check_external_memory_handle_type(
devList[0], CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KHR); device, CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KHR);
extMemProperties.push_back( extMemProperties.push_back(
(cl_mem_properties)CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KHR); (cl_mem_properties)
CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KHR);
extMemProperties.push_back((cl_mem_properties)handle); extMemProperties.push_back((cl_mem_properties)handle);
break; break;
case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32_KMT: case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_WIN32_KMT:
handle = vkDeviceMem->getHandle(vkExternalMemoryHandleType); handle = vkDeviceMem->getHandle(vkExternalMemoryHandleType);
errNum = check_external_memory_handle_type( errNum = check_external_memory_handle_type(
devList[0], CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KMT_KHR); device, CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KMT_KHR);
extMemProperties.push_back( extMemProperties.push_back(
(cl_mem_properties) (cl_mem_properties)
CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KMT_KHR); CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_WIN32_KMT_KHR);
@@ -123,7 +130,7 @@ int test_consistency_external_for_3dimage(cl_device_id deviceID,
case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_FD: case VULKAN_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_FD:
fd = (int)vkDeviceMem->getHandle(vkExternalMemoryHandleType); fd = (int)vkDeviceMem->getHandle(vkExternalMemoryHandleType);
errNum = check_external_memory_handle_type( errNum = check_external_memory_handle_type(
devList[0], CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR); device, CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR);
extMemProperties.push_back( extMemProperties.push_back(
(cl_mem_properties)CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR); (cl_mem_properties)CL_EXTERNAL_MEMORY_HANDLE_OPAQUE_FD_KHR);
extMemProperties.push_back((cl_mem_properties)fd); extMemProperties.push_back((cl_mem_properties)fd);
@@ -145,8 +152,9 @@ int test_consistency_external_for_3dimage(cl_device_id deviceID,
const VkImageCreateInfo VulkanImageCreateInfo = const VkImageCreateInfo VulkanImageCreateInfo =
vkImage3D.getVkImageCreateInfo(); vkImage3D.getVkImageCreateInfo();
errNum = getCLImageInfoFromVkImageInfo( errNum = getCLImageInfoFromVkImageInfo(&VulkanImageCreateInfo,
&VulkanImageCreateInfo, totalImageMemSize, &img_format, &image_desc); totalImageMemSize, &img_format,
&image_desc);
if (errNum != CL_SUCCESS) if (errNum != CL_SUCCESS)
{ {
log_error("getCLImageInfoFromVkImageInfo failed!!!"); log_error("getCLImageInfoFromVkImageInfo failed!!!");
@@ -163,9 +171,9 @@ int test_consistency_external_for_3dimage(cl_device_id deviceID,
image.reset(); image.reset();
// Passing NULL properties and a valid image_format and image_desc // Passing NULL properties and a valid image_format and image_desc
image = image = clCreateImageWithProperties(context, NULL, CL_MEM_READ_WRITE,
clCreateImageWithProperties(context, NULL, CL_MEM_READ_WRITE, &img_format, &image_desc, NULL,
&img_format, &image_desc, NULL, &errNum); &errNum);
test_error(errNum, test_error(errNum,
"Unable to create image with NULL properties " "Unable to create image with NULL properties "
"with valid image format and image desc"); "with valid image format and image desc");
@@ -174,8 +182,8 @@ int test_consistency_external_for_3dimage(cl_device_id deviceID,
// 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,
NULL, &errNum); &image_desc, NULL, &errNum);
test_failure_error(errNum, CL_INVALID_IMAGE_FORMAT_DESCRIPTOR, test_failure_error(errNum, CL_INVALID_IMAGE_FORMAT_DESCRIPTOR,
"Image creation must fail with " "Image creation must fail with "
"CL_INVALID_IMAGE_FORMAT_DESCRIPTOR" "CL_INVALID_IMAGE_FORMAT_DESCRIPTOR"
@@ -185,16 +193,25 @@ int test_consistency_external_for_3dimage(cl_device_id deviceID,
// Passing image_desc as NULL // Passing image_desc as NULL
image = clCreateImageWithProperties(context, extMemProperties.data(), image = clCreateImageWithProperties(context, extMemProperties.data(),
CL_MEM_READ_WRITE, &img_format, NULL, CL_MEM_READ_WRITE, &img_format,
NULL, &errNum); NULL, NULL, &errNum);
test_failure_error(errNum, CL_INVALID_IMAGE_DESCRIPTOR, test_failure_error(errNum, CL_INVALID_IMAGE_DESCRIPTOR,
"Image creation must fail with " "Image creation must fail with "
"CL_INVALID_IMAGE_DESCRIPTOR " "CL_INVALID_IMAGE_DESCRIPTOR "
"when image desc passed as NULL"); "when image desc passed as NULL");
image.reset(); image.reset();
if (cmd_queue) clReleaseCommandQueue(cmd_queue);
if (context) clReleaseContext(context);
return TEST_PASS; return TEST_PASS;
}
};
} // anonymous namespace
int test_consistency_external_for_3dimage(cl_device_id deviceID,
cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
return MakeAndRunTest<ConsistencyExternalImage3DTest>(
deviceID, context, defaultQueue, num_elements);
} }

506
test_conformance/vulkan/test_vulkan_interop_buffer.cpp
View File

@@ -1,5 +1,5 @@
// //
// Copyright (c) 2022 The Khronos Group Inc. // Copyright (c) 2024 The Khronos Group Inc.
// //
// Licensed under the Apache License, Version 2.0 (the "License"); // Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License. // you may not use this file except in compliance with the License.
@@ -24,42 +24,45 @@
#include <string.h> #include <string.h>
#include "harness/errorHelpers.h" #include "harness/errorHelpers.h"
#include "harness/os_helpers.h" #include "harness/os_helpers.h"
#include "deviceInfo.h"
#include "vulkan_test_base.h"
#include "opencl_vulkan_wrapper.hpp"
#define MAX_BUFFERS 5 #define MAX_BUFFERS 5
#define MAX_IMPORTS 5 #define MAX_IMPORTS 5
#define BUFFERSIZE 3000 #define BUFFERSIZE 3000
static cl_uchar uuid[CL_UUID_SIZE_KHR];
static cl_device_id deviceId = NULL;
namespace { namespace {
cl_uchar uuid[CL_UUID_SIZE_KHR];
cl_device_id deviceId = nullptr;
struct Params struct Params
{ {
uint32_t numBuffers; uint32_t numBuffers;
uint32_t bufferSize; uint32_t bufferSize;
uint32_t interBufferOffset; uint32_t interBufferOffset;
}; };
}
const char *kernel_text_numbuffer_1 = " \ const char *kernel_text_numbuffer_1 = " \
__kernel void clUpdateBuffer(int bufferSize, __global unsigned char *a) { \n\ __kernel void clUpdateBuffer(int bufferSize, __global unsigned char *a) { \n\
int gid = get_global_id(0); \n\ int gid = get_global_id(0); \n\
if (gid < bufferSize) { \n\ if (gid < bufferSize) { \n\
a[gid]++; \n\ a[gid]++; \n\
} \n\ } \n\
}"; }";
const char *kernel_text_numbuffer_2 = " \ const char *kernel_text_numbuffer_2 = " \
__kernel void clUpdateBuffer(int bufferSize, __global unsigned char *a, __global unsigned char *b) { \n\ __kernel void clUpdateBuffer(int bufferSize, __global unsigned char *a, __global unsigned char *b) { \n\
int gid = get_global_id(0); \n\ int gid = get_global_id(0); \n\
if (gid < bufferSize) { \n\ if (gid < bufferSize) { \n\
a[gid]++; \n\ a[gid]++; \n\
b[gid]++;\n\ b[gid]++;\n\
} \n\ } \n\
}"; }";
const char *kernel_text_numbuffer_4 = " \ const char *kernel_text_numbuffer_4 = " \
__kernel void clUpdateBuffer(int bufferSize, __global unsigned char *a, __global unsigned char *b, __global unsigned char *c, __global unsigned char *d) { \n\ __kernel void clUpdateBuffer(int bufferSize, __global unsigned char *a, __global unsigned char *b, __global unsigned char *c, __global unsigned char *d) { \n\
int gid = get_global_id(0); \n\ int gid = get_global_id(0); \n\
if (gid < bufferSize) { \n\ if (gid < bufferSize) { \n\
a[gid]++;\n\ a[gid]++;\n\
@@ -67,19 +70,20 @@ __kernel void clUpdateBuffer(int bufferSize, __global unsigned char *a, __global
c[gid]++; \n\ c[gid]++; \n\
d[gid]++; \n\ d[gid]++; \n\
} \n\ } \n\
}"; }";
const char *kernel_text_verify = " \ const char *kernel_text_verify = " \
__kernel void checkKernel(__global unsigned char *ptr, int size, int expVal, __global unsigned char *err) \n\ __kernel void checkKernel(__global unsigned char *ptr, int size, int expVal, __global unsigned char *err) \n\
{ \n\ { \n\
int idx = get_global_id(0); \n\ int idx = get_global_id(0); \n\
if ((idx < size) && (*err == 0)) { \n\ if ((idx < size) && (*err == 0)) { \n\
if (ptr[idx] != expVal){ \n\ if (ptr[idx] != expVal){ \n\
*err = 1; \n\ *err = 1; \n\
} \n\ } \n\
} \n\ } \n\
}"; }";
int run_test_with_two_queue( int run_test_with_two_queue(
cl_context &context, cl_command_queue &cmd_queue1, cl_context &context, cl_command_queue &cmd_queue1,
@@ -114,7 +118,8 @@ int run_test_with_two_queue(
VulkanSemaphore vkCl2VkSemaphore(vkDevice, vkExternalSemaphoreHandleType); VulkanSemaphore vkCl2VkSemaphore(vkDevice, vkExternalSemaphoreHandleType);
std::shared_ptr<VulkanFence> fence = nullptr; std::shared_ptr<VulkanFence> fence = nullptr;
VulkanQueue &vkQueue = vkDevice.getQueue(); VulkanQueue &vkQueue =
vkDevice.getQueue(getVulkanQueueFamily(vkDevice.getPhysicalDevice()));
std::vector<char> vkBufferShader = readFile("buffer.spv", exe_dir()); std::vector<char> vkBufferShader = readFile("buffer.spv", exe_dir());
@@ -150,6 +155,7 @@ int run_test_with_two_queue(
} }
const uint32_t maxIter = innerIterations; const uint32_t maxIter = innerIterations;
VulkanCommandPool vkCommandPool(vkDevice); VulkanCommandPool vkCommandPool(vkDevice);
VulkanCommandBuffer vkCommandBuffer(vkDevice, vkCommandPool); VulkanCommandBuffer vkCommandBuffer(vkDevice, vkCommandPool);
@@ -446,7 +452,8 @@ int run_test_with_one_queue(
VulkanSemaphore vkCl2VkSemaphore(vkDevice, vkExternalSemaphoreHandleType); VulkanSemaphore vkCl2VkSemaphore(vkDevice, vkExternalSemaphoreHandleType);
std::shared_ptr<VulkanFence> fence = nullptr; std::shared_ptr<VulkanFence> fence = nullptr;
VulkanQueue &vkQueue = vkDevice.getQueue(); VulkanQueue &vkQueue =
vkDevice.getQueue(getVulkanQueueFamily(vkDevice.getPhysicalDevice()));
std::vector<char> vkBufferShader = readFile("buffer.spv", exe_dir()); std::vector<char> vkBufferShader = readFile("buffer.spv", exe_dir());
@@ -482,6 +489,7 @@ int run_test_with_one_queue(
} }
const uint32_t maxIter = innerIterations; const uint32_t maxIter = innerIterations;
VulkanCommandPool vkCommandPool(vkDevice); VulkanCommandPool vkCommandPool(vkDevice);
VulkanCommandBuffer vkCommandBuffer(vkDevice, vkCommandPool); VulkanCommandBuffer vkCommandBuffer(vkDevice, vkCommandPool);
@@ -749,7 +757,7 @@ int run_test_with_multi_import_same_ctx(
VulkanSemaphore vkCl2VkSemaphore(vkDevice, vkExternalSemaphoreHandleType); VulkanSemaphore vkCl2VkSemaphore(vkDevice, vkExternalSemaphoreHandleType);
std::shared_ptr<VulkanFence> fence = nullptr; std::shared_ptr<VulkanFence> fence = nullptr;
VulkanQueue &vkQueue = vkDevice.getQueue(); VulkanQueue &vkQueue = vkDevice.getQueue(getVulkanQueueFamily());
std::vector<char> vkBufferShader = readFile("buffer.spv", exe_dir()); std::vector<char> vkBufferShader = readFile("buffer.spv", exe_dir());
@@ -833,10 +841,10 @@ int run_test_with_multi_import_same_ctx(
std::vector<clExternalMemory *> pExternalMemory; std::vector<clExternalMemory *> pExternalMemory;
for (size_t cl_bIdx = 0; cl_bIdx < numImports; cl_bIdx++) for (size_t cl_bIdx = 0; cl_bIdx < numImports; cl_bIdx++)
{ {
pExternalMemory.push_back(new clExternalMemory( pExternalMemory.push_back(
vkBufferListDeviceMemory[bIdx], new clExternalMemory(vkBufferListDeviceMemory[bIdx],
vkExternalMemoryHandleType, bufferSize, context, vkExternalMemoryHandleType,
deviceId)); bufferSize, context, deviceId));
} }
externalMemory.push_back(pExternalMemory); externalMemory.push_back(pExternalMemory);
} }
@@ -851,12 +859,11 @@ int run_test_with_multi_import_same_ctx(
for (size_t bIdx = 0; bIdx < vkBufferList.size(); bIdx++) for (size_t bIdx = 0; bIdx < vkBufferList.size(); bIdx++)
{ {
size_t buffer_size = vkBufferList[bIdx].getSize(); size_t buffer_size = vkBufferList[bIdx].getSize();
vkBufferListDeviceMemory[bIdx]->bindBuffer( vkBufferListDeviceMemory[bIdx]->bindBuffer(vkBufferList[bIdx],
vkBufferList[bIdx], 0); 0);
for (size_t cl_bIdx = 0; cl_bIdx < numImports; cl_bIdx++) for (size_t cl_bIdx = 0; cl_bIdx < numImports; cl_bIdx++)
{ {
buffers[bIdx][cl_bIdx] = buffers[bIdx][cl_bIdx] = externalMemory[bIdx][cl_bIdx]
externalMemory[bIdx][cl_bIdx]
->getExternalMemoryBuffer(); ->getExternalMemoryBuffer();
} }
} }
@@ -903,25 +910,24 @@ int run_test_with_multi_import_same_ctx(
else else
{ {
err = clVk2CLExternalSemaphore->wait(cmd_queue1); err = clVk2CLExternalSemaphore->wait(cmd_queue1);
test_error_and_cleanup( test_error_and_cleanup(err, CLEANUP,
err, CLEANUP, "Error: failed to wait on "
"Error: failed to wait on CL external semaphore\n"); "CL external semaphore\n");
} }
for (uint8_t launchIter = 0; launchIter < numImports; for (uint8_t launchIter = 0; launchIter < numImports;
launchIter++) launchIter++)
{ {
err = clSetKernelArg(update_buffer_kernel, 0, err = clSetKernelArg(update_buffer_kernel, 0,
sizeof(uint32_t), sizeof(uint32_t), (void *)&bufferSize);
(void *)&bufferSize);
for (int i = 0; i < numBuffers; i++) for (int i = 0; i < numBuffers; i++)
{ {
err |= clSetKernelArg( err |= clSetKernelArg(
update_buffer_kernel, i + 1, sizeof(cl_mem), update_buffer_kernel, i + 1, sizeof(cl_mem),
(void *)&(buffers[i][launchIter])); (void *)&(buffers[i][launchIter]));
err = clEnqueueAcquireExternalMemObjectsKHRptr( err = clEnqueueAcquireExternalMemObjectsKHRptr(
cmd_queue1, 1, &buffers[i][launchIter], 0, cmd_queue1, 1, &buffers[i][launchIter], 0, nullptr,
nullptr, nullptr); nullptr);
test_error_and_cleanup(err, CLEANUP, test_error_and_cleanup(err, CLEANUP,
"Failed to acquire buffers"); "Failed to acquire buffers");
} }
@@ -933,16 +939,15 @@ int run_test_with_multi_import_same_ctx(
err = clEnqueueNDRangeKernel( err = clEnqueueNDRangeKernel(
cmd_queue1, update_buffer_kernel, 1, NULL, cmd_queue1, update_buffer_kernel, 1, NULL,
global_work_size, NULL, 0, NULL, NULL); global_work_size, NULL, 0, NULL, NULL);
test_error_and_cleanup( test_error_and_cleanup(err, CLEANUP,
err, CLEANUP,
"Error: Failed to launch " "Error: Failed to launch "
"update_buffer_kernel, error\n "); "update_buffer_kernel, error\n ");
for (int i = 0; i < numBuffers; i++) for (int i = 0; i < numBuffers; i++)
{ {
err = clEnqueueReleaseExternalMemObjectsKHRptr( err = clEnqueueReleaseExternalMemObjectsKHRptr(
cmd_queue1, 1, &buffers[i][launchIter], 0, cmd_queue1, 1, &buffers[i][launchIter], 0, nullptr,
nullptr, nullptr); nullptr);
test_error_and_cleanup(err, CLEANUP, test_error_and_cleanup(err, CLEANUP,
"Failed to release buffers"); "Failed to release buffers");
} }
@@ -954,8 +959,8 @@ int run_test_with_multi_import_same_ctx(
else if (!use_fence && iter != (maxIter - 1)) else if (!use_fence && iter != (maxIter - 1))
{ {
err = clCl2VkExternalSemaphore->signal(cmd_queue1); err = clCl2VkExternalSemaphore->signal(cmd_queue1);
test_error_and_cleanup( test_error_and_cleanup(err, CLEANUP,
err, CLEANUP, "Failed to signal CL semaphore\n"); "Failed to signal CL semaphore\n");
} }
} }
@@ -968,12 +973,10 @@ int run_test_with_multi_import_same_ctx(
error_1 = clCreateBuffer(context, CL_MEM_WRITE_ONLY, error_1 = clCreateBuffer(context, CL_MEM_WRITE_ONLY,
sizeof(uint8_t), NULL, &err); sizeof(uint8_t), NULL, &err);
test_error_and_cleanup(err, CLEANUP, test_error_and_cleanup(err, CLEANUP, "Error: clCreateBuffer \n");
"Error: clCreateBuffer \n");
uint8_t val = 0; uint8_t val = 0;
err = err = clEnqueueWriteBuffer(cmd_queue1, error_1, CL_TRUE, 0,
clEnqueueWriteBuffer(cmd_queue1, error_1, CL_TRUE, 0,
sizeof(uint8_t), &val, 0, NULL, NULL); sizeof(uint8_t), &val, 0, NULL, NULL);
test_error_and_cleanup(err, CLEANUP, test_error_and_cleanup(err, CLEANUP,
"Error: clEnqueueWriteBuffer \n"); "Error: clEnqueueWriteBuffer \n");
@@ -984,20 +987,19 @@ int run_test_with_multi_import_same_ctx(
{ {
err = clSetKernelArg(verify_kernel, 0, sizeof(cl_mem), err = clSetKernelArg(verify_kernel, 0, sizeof(cl_mem),
(void *)&(buffers[i][0])); (void *)&(buffers[i][0]));
err |= clSetKernelArg(verify_kernel, 1, sizeof(int), err |=
&bufferSize); clSetKernelArg(verify_kernel, 1, sizeof(int), &bufferSize);
err |= clSetKernelArg(verify_kernel, 2, sizeof(int), err |= clSetKernelArg(verify_kernel, 2, sizeof(int),
&calc_max_iter); &calc_max_iter);
err |= clSetKernelArg(verify_kernel, 3, sizeof(cl_mem), err |= clSetKernelArg(verify_kernel, 3, sizeof(cl_mem),
(void *)&error_1); (void *)&error_1);
test_error_and_cleanup( test_error_and_cleanup(err, CLEANUP,
err, CLEANUP,
"Error: Failed to set arg values for " "Error: Failed to set arg values for "
"verify_kernel \n"); "verify_kernel \n");
err = clEnqueueNDRangeKernel(cmd_queue1, verify_kernel, 1, err = clEnqueueNDRangeKernel(cmd_queue1, verify_kernel, 1, NULL,
NULL, global_work_size, NULL, global_work_size, NULL, 0, NULL,
0, NULL, NULL); NULL);
test_error_and_cleanup( test_error_and_cleanup(
err, CLEANUP, err, CLEANUP,
"Error: Failed to launch verify_kernel, error\n"); "Error: Failed to launch verify_kernel, error\n");
@@ -1005,14 +1007,13 @@ int run_test_with_multi_import_same_ctx(
err = clEnqueueReadBuffer(cmd_queue1, error_1, CL_TRUE, 0, err = clEnqueueReadBuffer(cmd_queue1, error_1, CL_TRUE, 0,
sizeof(uint8_t), error_2, 0, NULL, sizeof(uint8_t), error_2, 0, NULL,
NULL); NULL);
test_error_and_cleanup( test_error_and_cleanup(err, CLEANUP,
err, CLEANUP, "Error: Failed read output, error \n"); "Error: Failed read output, error \n");
if (*error_2 == 1) if (*error_2 == 1)
{ {
test_fail_and_cleanup( test_fail_and_cleanup(
err, CLEANUP, err, CLEANUP, " vulkan_opencl_buffer test FAILED\n");
" vulkan_opencl_buffer test FAILED\n");
} }
} }
for (size_t i = 0; i < vkBufferList.size(); i++) for (size_t i = 0; i < vkBufferList.size(); i++)
@@ -1031,8 +1032,7 @@ int run_test_with_multi_import_same_ctx(
for (size_t i = 0; i < externalMemory.size(); i++) for (size_t i = 0; i < externalMemory.size(); i++)
{ {
externalMemory[i].erase(externalMemory[i].begin(), externalMemory[i].erase(externalMemory[i].begin(),
externalMemory[i].begin() externalMemory[i].begin() + numBuffers);
+ numBuffers);
} }
externalMemory.clear(); externalMemory.clear();
} }
@@ -1097,7 +1097,7 @@ int run_test_with_multi_import_diff_ctx(
VulkanSemaphore vkCl2VkSemaphore(vkDevice, vkExternalSemaphoreHandleType); VulkanSemaphore vkCl2VkSemaphore(vkDevice, vkExternalSemaphoreHandleType);
std::shared_ptr<VulkanFence> fence = nullptr; std::shared_ptr<VulkanFence> fence = nullptr;
VulkanQueue &vkQueue = vkDevice.getQueue(); VulkanQueue &vkQueue = vkDevice.getQueue(getVulkanQueueFamily());
std::vector<char> vkBufferShader = readFile("buffer.spv", exe_dir()); std::vector<char> vkBufferShader = readFile("buffer.spv", exe_dir());
@@ -1273,9 +1273,9 @@ int run_test_with_multi_import_diff_ctx(
else else
{ {
err = clVk2CLExternalSemaphore->wait(cmd_queue1); err = clVk2CLExternalSemaphore->wait(cmd_queue1);
test_error_and_cleanup( test_error_and_cleanup(err, CLEANUP,
err, CLEANUP, "Error: failed to wait on "
"Error: failed to wait on CL external semaphore\n"); "CL external semaphore\n");
} }
for (uint8_t launchIter = 0; launchIter < numImports; for (uint8_t launchIter = 0; launchIter < numImports;
@@ -1361,17 +1361,16 @@ int run_test_with_multi_import_diff_ctx(
else else
{ {
err = clVk2CLExternalSemaphore2->wait(cmd_queue2); err = clVk2CLExternalSemaphore2->wait(cmd_queue2);
test_error_and_cleanup( test_error_and_cleanup(err, CLEANUP,
err, CLEANUP, "Error: failed to wait on "
"Error: failed to wait on CL external semaphore\n"); "CL external semaphore\n");
} }
for (uint8_t launchIter = 0; launchIter < numImports; for (uint8_t launchIter = 0; launchIter < numImports;
launchIter++) launchIter++)
{ {
err = clSetKernelArg(update_buffer_kernel2[launchIter], err = clSetKernelArg(update_buffer_kernel2[launchIter], 0,
0, sizeof(uint32_t), sizeof(uint32_t), (void *)&bufferSize);
(void *)&bufferSize);
test_error_and_cleanup(err, CLEANUP, test_error_and_cleanup(err, CLEANUP,
"Failed to set kernel arg"); "Failed to set kernel arg");
@@ -1379,14 +1378,13 @@ int run_test_with_multi_import_diff_ctx(
{ {
err = clSetKernelArg( err = clSetKernelArg(
update_buffer_kernel2[launchIter], i + 1, update_buffer_kernel2[launchIter], i + 1,
sizeof(cl_mem), sizeof(cl_mem), (void *)&(buffers2[i][launchIter]));
(void *)&(buffers2[i][launchIter]));
test_error_and_cleanup(err, CLEANUP, test_error_and_cleanup(err, CLEANUP,
"Failed to set kernel arg"); "Failed to set kernel arg");
err = clEnqueueAcquireExternalMemObjectsKHRptr( err = clEnqueueAcquireExternalMemObjectsKHRptr(
cmd_queue2, 1, &buffers2[i][launchIter], 0, cmd_queue2, 1, &buffers2[i][launchIter], 0, nullptr,
nullptr, nullptr); nullptr);
test_error_and_cleanup(err, CLEANUP, test_error_and_cleanup(err, CLEANUP,
"Failed to acquire buffers"); "Failed to acquire buffers");
} }
@@ -1396,17 +1394,16 @@ int run_test_with_multi_import_diff_ctx(
"kernel\n "); "kernel\n ");
err = clEnqueueNDRangeKernel( err = clEnqueueNDRangeKernel(
cmd_queue2, update_buffer_kernel2[launchIter], 1, cmd_queue2, update_buffer_kernel2[launchIter], 1, NULL,
NULL, global_work_size, NULL, 0, NULL, NULL); global_work_size, NULL, 0, NULL, NULL);
test_error_and_cleanup( test_error_and_cleanup(err, CLEANUP,
err, CLEANUP,
"Error: Failed to launch " "Error: Failed to launch "
"update_buffer_kernel, error\n "); "update_buffer_kernel, error\n ");
for (int i = 0; i < numBuffers; i++) for (int i = 0; i < numBuffers; i++)
{ {
err = clEnqueueReleaseExternalMemObjectsKHRptr( err = clEnqueueReleaseExternalMemObjectsKHRptr(
cmd_queue2, 1, &buffers2[i][launchIter], 0, cmd_queue2, 1, &buffers2[i][launchIter], 0, nullptr,
nullptr, nullptr); nullptr);
test_error_and_cleanup(err, CLEANUP, test_error_and_cleanup(err, CLEANUP,
"Failed to release buffers"); "Failed to release buffers");
} }
@@ -1418,8 +1415,8 @@ int run_test_with_multi_import_diff_ctx(
else if (!use_fence && iter != (maxIter - 1)) else if (!use_fence && iter != (maxIter - 1))
{ {
err = clCl2VkExternalSemaphore2->signal(cmd_queue2); err = clCl2VkExternalSemaphore2->signal(cmd_queue2);
test_error_and_cleanup( test_error_and_cleanup(err, CLEANUP,
err, CLEANUP, "Failed to signal CL semaphore\n"); "Failed to signal CL semaphore\n");
} }
} }
clFinish(cmd_queue2); clFinish(cmd_queue2);
@@ -1432,23 +1429,19 @@ int run_test_with_multi_import_diff_ctx(
error_1 = clCreateBuffer(context, CL_MEM_WRITE_ONLY, error_1 = clCreateBuffer(context, CL_MEM_WRITE_ONLY,
sizeof(uint8_t), NULL, &err); sizeof(uint8_t), NULL, &err);
test_error_and_cleanup(err, CLEANUP, test_error_and_cleanup(err, CLEANUP, "Error: clCreateBuffer \n");
"Error: clCreateBuffer \n");
error_2 = clCreateBuffer(context2, CL_MEM_WRITE_ONLY, error_2 = clCreateBuffer(context2, CL_MEM_WRITE_ONLY,
sizeof(uint8_t), NULL, &err); sizeof(uint8_t), NULL, &err);
test_error_and_cleanup(err, CLEANUP, test_error_and_cleanup(err, CLEANUP, "Error: clCreateBuffer \n");
"Error: clCreateBuffer \n");
uint8_t val = 0; uint8_t val = 0;
err = err = clEnqueueWriteBuffer(cmd_queue1, error_1, CL_TRUE, 0,
clEnqueueWriteBuffer(cmd_queue1, error_1, CL_TRUE, 0,
sizeof(uint8_t), &val, 0, NULL, NULL); sizeof(uint8_t), &val, 0, NULL, NULL);
test_error_and_cleanup(err, CLEANUP, test_error_and_cleanup(err, CLEANUP,
"Error: Failed read output, error \n"); "Error: Failed read output, error \n");
err = err = clEnqueueWriteBuffer(cmd_queue2, error_2, CL_TRUE, 0,
clEnqueueWriteBuffer(cmd_queue2, error_2, CL_TRUE, 0,
sizeof(uint8_t), &val, 0, NULL, NULL); sizeof(uint8_t), &val, 0, NULL, NULL);
test_error_and_cleanup(err, CLEANUP, test_error_and_cleanup(err, CLEANUP,
"Error: Failed read output, error \n"); "Error: Failed read output, error \n");
@@ -1458,30 +1451,28 @@ int run_test_with_multi_import_diff_ctx(
{ {
err = clSetKernelArg(verify_kernel, 0, sizeof(cl_mem), err = clSetKernelArg(verify_kernel, 0, sizeof(cl_mem),
(void *)&(buffers1[i][0])); (void *)&(buffers1[i][0]));
err |= clSetKernelArg(verify_kernel, 1, sizeof(int), err |=
&pBufferSize); clSetKernelArg(verify_kernel, 1, sizeof(int), &pBufferSize);
err |= clSetKernelArg(verify_kernel, 2, sizeof(int), err |= clSetKernelArg(verify_kernel, 2, sizeof(int),
&calc_max_iter); &calc_max_iter);
err |= clSetKernelArg(verify_kernel, 3, sizeof(cl_mem), err |= clSetKernelArg(verify_kernel, 3, sizeof(cl_mem),
(void *)&error_1); (void *)&error_1);
test_error_and_cleanup( test_error_and_cleanup(err, CLEANUP,
err, CLEANUP,
"Error: Failed to set arg values for " "Error: Failed to set arg values for "
"verify_kernel \n"); "verify_kernel \n");
err = clEnqueueNDRangeKernel(cmd_queue1, verify_kernel, 1, err = clEnqueueNDRangeKernel(cmd_queue1, verify_kernel, 1, NULL,
NULL, global_work_size, NULL, global_work_size, NULL, 0, NULL,
0, NULL, NULL); NULL);
test_error_and_cleanup( test_error_and_cleanup(err, CLEANUP,
err, CLEANUP,
"Error: Failed to launch verify_kernel," "Error: Failed to launch verify_kernel,"
"error\n"); "error\n");
err = clEnqueueReadBuffer(cmd_queue1, error_1, CL_TRUE, 0, err = clEnqueueReadBuffer(cmd_queue1, error_1, CL_TRUE, 0,
sizeof(uint8_t), error_3, 0, NULL, sizeof(uint8_t), error_3, 0, NULL,
NULL); NULL);
test_error_and_cleanup( test_error_and_cleanup(err, CLEANUP,
err, CLEANUP, "Error: Failed read output, error\n"); "Error: Failed read output, error\n");
if (*error_3 == 1) if (*error_3 == 1)
{ {
@@ -1501,24 +1492,22 @@ int run_test_with_multi_import_diff_ctx(
&calc_max_iter); &calc_max_iter);
err |= clSetKernelArg(verify_kernel2, 3, sizeof(cl_mem), err |= clSetKernelArg(verify_kernel2, 3, sizeof(cl_mem),
(void *)&error_2); (void *)&error_2);
test_error_and_cleanup( test_error_and_cleanup(err, CLEANUP,
err, CLEANUP,
"Error: Failed to set arg values for " "Error: Failed to set arg values for "
"verify_kernel \n"); "verify_kernel \n");
err = clEnqueueNDRangeKernel(cmd_queue2, verify_kernel2, 1, err = clEnqueueNDRangeKernel(cmd_queue2, verify_kernel2, 1,
NULL, global_work_size, NULL, NULL, global_work_size, NULL, 0,
0, NULL, NULL); NULL, NULL);
test_error_and_cleanup( test_error_and_cleanup(err, CLEANUP,
err, CLEANUP,
"Error: Failed to launch verify_kernel," "Error: Failed to launch verify_kernel,"
"error\n"); "error\n");
err = clEnqueueReadBuffer(cmd_queue2, error_2, CL_TRUE, 0, err = clEnqueueReadBuffer(cmd_queue2, error_2, CL_TRUE, 0,
sizeof(uint8_t), error_3, 0, NULL, sizeof(uint8_t), error_3, 0, NULL,
NULL); NULL);
test_error_and_cleanup( test_error_and_cleanup(err, CLEANUP,
err, CLEANUP, "Error: Failed read output, error\n"); "Error: Failed read output, error\n");
if (*error_3 == 1) if (*error_3 == 1)
{ {
@@ -1597,80 +1586,52 @@ CLEANUP:
return err; return err;
} }
int test_buffer_common(cl_device_id device_, cl_context context_,
cl_command_queue queue_, int numElements_,
bool use_fence)
{
struct BufferTestBase : public VulkanTestBase
{
BufferTestBase(cl_device_id device, cl_context context,
cl_command_queue queue, cl_int nelems)
: VulkanTestBase(device, context, queue, nelems)
{}
int test_buffer_common(bool use_fence)
{
int current_device = 0; int current_device = 0;
int device_count = 0; int device_count = 0;
int devices_prohibited = 0; int devices_prohibited = 0;
cl_int errNum = CL_SUCCESS; cl_int errNum = CL_SUCCESS;
cl_platform_id platform = NULL;
size_t extensionSize = 0; size_t extensionSize = 0;
cl_uint num_devices = 0;
cl_uint device_no = 0;
const size_t bufsize = BUFFERSIZE; const size_t bufsize = BUFFERSIZE;
char buf[BUFFERSIZE]; char buf[BUFFERSIZE];
cl_device_id *devices;
char *extensions = NULL; char *extensions = NULL;
cl_kernel verify_kernel; clKernelWrapper verify_kernel;
cl_kernel verify_kernel2; clKernelWrapper verify_kernel2;
cl_kernel kernel[3] = { NULL, NULL, NULL }; clKernelWrapper kernel[3] = { NULL, NULL, NULL };
cl_kernel kernel2[3] = { NULL, NULL, NULL }; clKernelWrapper kernel2[3] = { NULL, NULL, NULL };
const char *program_source_const[3] = { kernel_text_numbuffer_1, const char *program_source_const[3] = { kernel_text_numbuffer_1,
kernel_text_numbuffer_2, kernel_text_numbuffer_2,
kernel_text_numbuffer_4 }; kernel_text_numbuffer_4 };
const char *program_source_const_verify; const char *program_source_const_verify;
size_t program_source_length; size_t program_source_length;
cl_command_queue cmd_queue1 = NULL; clCommandQueueWrapper cmd_queue1;
cl_command_queue cmd_queue2 = NULL; clCommandQueueWrapper cmd_queue2;
cl_command_queue cmd_queue3 = NULL; clCommandQueueWrapper cmd_queue3;
cl_context context = NULL;
cl_program program[3] = { NULL, NULL, NULL };
cl_program program_verify, program_verify2;
cl_context context2 = NULL;
clProgramWrapper program[3] = { NULL, NULL, NULL };
clProgramWrapper program_verify, program_verify2;
clContextWrapper context2;
VulkanDevice vkDevice;
uint32_t numBuffersList[] = { 1, 2, 4 }; uint32_t numBuffersList[] = { 1, 2, 4 };
uint32_t bufferSizeList[] = { 4 * 1024, 64 * 1024, 2 * 1024 * 1024 }; uint32_t bufferSizeList[] = { 4 * 1024, 64 * 1024, 2 * 1024 * 1024 };
uint32_t bufferSizeListforOffset[] = { 256, 512, 1024 }; uint32_t bufferSizeListforOffset[] = { 256, 512, 1024 };
cl_context_properties contextProperties[] = { CL_CONTEXT_PLATFORM, 0, 0 };
std::vector<VulkanExternalSemaphoreHandleType> supportedSemaphoreTypes; std::vector<VulkanExternalSemaphoreHandleType> supportedSemaphoreTypes;
errNum = clGetPlatformIDs(1, &platform, NULL);
test_error_and_cleanup(errNum, CLEANUP, "Error: Failed to get platform\n");
errNum =
clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, 0, NULL, &num_devices);
test_error_and_cleanup(errNum, CLEANUP,
"clGetDeviceIDs failed in returning of devices\n");
devices = (cl_device_id *)malloc(num_devices * sizeof(cl_device_id));
if (NULL == devices)
{
test_fail_and_cleanup(errNum, CLEANUP,
"Unable to allocate memory for devices\n");
}
errNum = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, num_devices, devices,
NULL);
test_error_and_cleanup(errNum, CLEANUP, "Failed to get deviceID.\n");
contextProperties[1] = (cl_context_properties)platform;
log_info("Assigned contextproperties for platform\n");
for (device_no = 0; device_no < num_devices; device_no++)
{
errNum = clGetDeviceInfo(devices[device_no], CL_DEVICE_UUID_KHR,
CL_UUID_SIZE_KHR, uuid, NULL);
test_error_and_cleanup(errNum, CLEANUP, "clGetDeviceInfo failed\n");
if (!use_fence) if (!use_fence)
{ {
supportedSemaphoreTypes = supportedSemaphoreTypes =
getSupportedInteropExternalSemaphoreHandleTypes( getSupportedInteropExternalSemaphoreHandleTypes(device,
devices[device_no], vkDevice); *vkDevice);
} }
else else
{ {
@@ -1678,48 +1639,24 @@ int test_buffer_common(cl_device_id device_, cl_context context_,
VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_NONE); VULKAN_EXTERNAL_SEMAPHORE_HANDLE_TYPE_NONE);
} }
// If device does not support any semaphores, try the next one // If device does not support any semaphores, try the next one
if (!use_fence && supportedSemaphoreTypes.empty()) if (!use_fence && supportedSemaphoreTypes.empty())
{ {
continue; return TEST_FAIL;
}
errNum =
memcmp(uuid, vkDevice.getPhysicalDevice().getUUID(), VK_UUID_SIZE);
if (errNum == 0)
{
break;
}
} }
if (!use_fence && supportedSemaphoreTypes.empty()) if (!use_fence && supportedSemaphoreTypes.empty())
{ {
test_fail_and_cleanup( test_error_fail(
errNum, CLEANUP, errNum, "No devices found that support OpenCL semaphores\n");
"No devices found that support OpenCL semaphores\n");
} }
if (device_no >= num_devices) deviceId = device;
{ cmd_queue1 = clCreateCommandQueue(context, device, 0, &errNum);
test_fail_and_cleanup(errNum, CLEANUP, test_error(errNum, "Error: Failed to create command queue!\n");
"OpenCL error: "
"No Vulkan-OpenCL Interop capable GPU found.\n");
}
deviceId = devices[device_no];
context = clCreateContextFromType(contextProperties, CL_DEVICE_TYPE_GPU,
NULL, NULL, &errNum);
test_error_and_cleanup(errNum, CLEANUP, "error creating context\n");
log_info("Successfully created context !!!\n"); cmd_queue2 = clCreateCommandQueue(context, device, 0, &errNum);
test_error(errNum, "Error: Failed to create command queue!\n");
cmd_queue1 = clCreateCommandQueue(context, devices[device_no], 0, &errNum);
test_error_and_cleanup(errNum, CLEANUP,
"Error: Failed to create command queue!\n");
cmd_queue2 = clCreateCommandQueue(context, devices[device_no], 0, &errNum);
test_error_and_cleanup(errNum, CLEANUP,
"Error: Failed to create command queue!\n");
log_info("clCreateCommandQueue successful\n"); log_info("clCreateCommandQueue successful\n");
for (int i = 0; i < 3; i++) for (int i = 0; i < 3; i++)
@@ -1729,12 +1666,11 @@ int test_buffer_common(cl_device_id device_, cl_context context_,
clCreateProgramWithSource(context, 1, &program_source_const[i], clCreateProgramWithSource(context, 1, &program_source_const[i],
&program_source_length, &errNum); &program_source_length, &errNum);
errNum = clBuildProgram(program[i], 0, NULL, NULL, NULL, NULL); errNum = clBuildProgram(program[i], 0, NULL, NULL, NULL, NULL);
test_error_and_cleanup(errNum, CLEANUP, test_error(errNum, "Error: Failed to build program \n");
"Error: Failed to build program \n");
// create the kernel // create the kernel
kernel[i] = clCreateKernel(program[i], "clUpdateBuffer", &errNum); kernel[i] = clCreateKernel(program[i], "clUpdateBuffer", &errNum);
test_error_and_cleanup(errNum, CLEANUP, "clCreateKernel failed \n"); test_error(errNum, "clCreateKernel failed \n");
} }
program_source_const_verify = kernel_text_verify; program_source_const_verify = kernel_text_verify;
@@ -1743,47 +1679,44 @@ int test_buffer_common(cl_device_id device_, cl_context context_,
clCreateProgramWithSource(context, 1, &program_source_const_verify, clCreateProgramWithSource(context, 1, &program_source_const_verify,
&program_source_length, &errNum); &program_source_length, &errNum);
errNum = clBuildProgram(program_verify, 0, NULL, NULL, NULL, NULL); errNum = clBuildProgram(program_verify, 0, NULL, NULL, NULL, NULL);
test_error_and_cleanup(errNum, CLEANUP, test_error(errNum, "Error: Failed to build program2\n");
"Error: Failed to build program2\n");
verify_kernel = clCreateKernel(program_verify, "checkKernel", &errNum); verify_kernel = clCreateKernel(program_verify, "checkKernel", &errNum);
test_error_and_cleanup(errNum, CLEANUP, "clCreateKernel failed \n"); test_error(errNum, "clCreateKernel failed \n");
if (multiCtx) // different context guard if (multiCtx) // different context guard
{ {
context2 = clCreateContextFromType( context2 =
contextProperties, CL_DEVICE_TYPE_GPU, NULL, NULL, &errNum); clCreateContext(0, 1, &device, nullptr, nullptr, &errNum);
test_error_and_cleanup(errNum, CLEANUP, "error creating context\n"); test_error(errNum, "error creating context\n");
cmd_queue3 = cmd_queue3 = clCreateCommandQueue(context2, device, 0, &errNum);
clCreateCommandQueue(context2, devices[device_no], 0, &errNum); test_error(errNum, "Error: Failed to create command queue!\n");
test_error_and_cleanup(errNum, CLEANUP,
"Error: Failed to create command queue!\n");
for (int i = 0; i < 3; i++) for (int i = 0; i < 3; i++)
{ {
program_source_length = strlen(program_source_const[i]); program_source_length = strlen(program_source_const[i]);
program[i] = program[i] = clCreateProgramWithSource(
clCreateProgramWithSource(context2, 1, &program_source_const[i], context2, 1, &program_source_const[i],
&program_source_length, &errNum); &program_source_length, &errNum);
errNum = clBuildProgram(program[i], 0, NULL, NULL, NULL, NULL); errNum = clBuildProgram(program[i], 0, NULL, NULL, NULL, NULL);
test_error_and_cleanup(errNum, CLEANUP, test_error(errNum, "Error: Failed to build program \n");
"Error: Failed to build program \n");
// create the kernel // create the kernel
kernel2[i] = clCreateKernel(program[i], "clUpdateBuffer", &errNum); kernel2[i] =
test_error_and_cleanup(errNum, CLEANUP, "clCreateKernel failed \n"); clCreateKernel(program[i], "clUpdateBuffer", &errNum);
test_error(errNum, "clCreateKernel failed \n");
} }
program_source_length = strlen(program_source_const_verify); program_source_length = strlen(program_source_const_verify);
program_verify = program_verify = clCreateProgramWithSource(
clCreateProgramWithSource(context2, 1, &program_source_const_verify, context2, 1, &program_source_const_verify,
&program_source_length, &errNum); &program_source_length, &errNum);
errNum = clBuildProgram(program_verify, 0, NULL, NULL, NULL, NULL); errNum = clBuildProgram(program_verify, 0, NULL, NULL, NULL, NULL);
test_error_and_cleanup(errNum, CLEANUP, test_error(errNum, "Error: Failed to build program2\n");
"Error: Failed to build program2\n");
verify_kernel2 = clCreateKernel(program_verify, "checkKernel", &errNum); verify_kernel2 =
test_error_and_cleanup(errNum, CLEANUP, "clCreateKernel failed \n"); clCreateKernel(program_verify, "checkKernel", &errNum);
test_error(errNum, "clCreateKernel failed \n");
} }
// TODO: Add support for empty list if use_fence enabled // TODO: Add support for empty list if use_fence enabled
@@ -1799,55 +1732,154 @@ int test_buffer_common(cl_device_id device_, cl_context context_,
sizeIdx++) sizeIdx++)
{ {
uint32_t bufferSize = bufferSizeList[sizeIdx]; uint32_t bufferSize = bufferSizeList[sizeIdx];
log_info( log_info("&&&& RUNNING vulkan_opencl_buffer test "
"&&&& RUNNING vulkan_opencl_buffer test for Buffer size: " "for Buffer size: "
"%d\n", "%d\n",
bufferSize); bufferSize);
if (multiImport && !multiCtx) if (multiImport && !multiCtx)
{ {
errNum = run_test_with_multi_import_same_ctx( errNum = run_test_with_multi_import_same_ctx(
context, cmd_queue1, kernel, verify_kernel, vkDevice, context, (cl_command_queue &)cmd_queue1,
numBuffers, bufferSize, use_fence, semaphoreType); (cl_kernel *)&kernel, (cl_kernel &)verify_kernel,
*vkDevice, numBuffers, bufferSize, use_fence,
semaphoreType);
} }
else if (multiImport && multiCtx) else if (multiImport && multiCtx)
{ {
errNum = run_test_with_multi_import_diff_ctx( errNum = run_test_with_multi_import_diff_ctx(
context, context2, cmd_queue1, cmd_queue3, kernel, context, (cl_context &)context2,
kernel2, verify_kernel, verify_kernel2, vkDevice, (cl_command_queue &)cmd_queue1,
numBuffers, bufferSize, use_fence, semaphoreType); (cl_command_queue &)cmd_queue3,
(cl_kernel *)&kernel, (cl_kernel *)&kernel2,
(cl_kernel &)verify_kernel, verify_kernel2,
*vkDevice, numBuffers, bufferSize, use_fence,
semaphoreType);
} }
else if (numCQ == 2) else if (numCQ == 2)
{ {
errNum = run_test_with_two_queue( errNum = run_test_with_two_queue(
context, cmd_queue1, cmd_queue2, kernel, verify_kernel, context, (cl_command_queue &)cmd_queue1,
vkDevice, numBuffers + 1, bufferSize, use_fence, (cl_command_queue &)cmd_queue2,
(cl_kernel *)&kernel, (cl_kernel &)verify_kernel,
*vkDevice, numBuffers + 1, bufferSize, use_fence,
semaphoreType); semaphoreType);
} }
else else
{ {
errNum = run_test_with_one_queue( errNum = run_test_with_one_queue(
context, cmd_queue1, kernel, verify_kernel, vkDevice, context, (cl_command_queue &)cmd_queue1,
numBuffers, bufferSize, semaphoreType, use_fence); (cl_kernel *)&kernel, (cl_kernel &)verify_kernel,
*vkDevice, numBuffers, bufferSize, semaphoreType,
use_fence);
} }
test_error_and_cleanup(errNum, CLEANUP, "func_name failed \n"); test_error(errNum, "func_name failed \n");
} }
} }
} }
CLEANUP:
for (int i = 0; i < 3; i++)
{
if (program[i]) clReleaseProgram(program[i]);
if (kernel[i]) clReleaseKernel(kernel[i]);
}
if (cmd_queue1) clReleaseCommandQueue(cmd_queue1);
if (cmd_queue2) clReleaseCommandQueue(cmd_queue2);
if (cmd_queue3) clReleaseCommandQueue(cmd_queue3);
if (context) clReleaseContext(context);
if (context2) clReleaseContext(context2);
if (devices) free(devices);
if (extensions) free(extensions);
return errNum; return errNum;
}
};
template <bool use_fence> struct BufferCommonBufferTest : public BufferTestBase
{
BufferCommonBufferTest(cl_device_id device, cl_context context,
cl_command_queue queue, cl_int nelems)
: BufferTestBase(device, context, queue, nelems)
{}
cl_int Run() override { return test_buffer_common(use_fence); }
};
} // anonymous namespace
int test_buffer_single_queue(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue, int num_elements)
{
params_reset();
log_info("RUNNING TEST WITH ONE QUEUE...... \n\n");
return MakeAndRunTest<BufferCommonBufferTest<false>>(
deviceID, context, defaultQueue, num_elements);
}
int test_buffer_multiple_queue(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue, int num_elements)
{
params_reset();
numCQ = 2;
log_info("RUNNING TEST WITH TWO QUEUE...... \n\n");
return MakeAndRunTest<BufferCommonBufferTest<false>>(
deviceID, context, defaultQueue, num_elements);
}
int test_buffer_multiImport_sameCtx(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
params_reset();
multiImport = true;
log_info("RUNNING TEST WITH MULTIPLE DEVICE MEMORY IMPORT "
"IN SAME CONTEXT...... \n\n");
return MakeAndRunTest<BufferCommonBufferTest<false>>(
deviceID, context, defaultQueue, num_elements);
}
int test_buffer_multiImport_diffCtx(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
params_reset();
multiImport = true;
multiCtx = true;
log_info("RUNNING TEST WITH MULTIPLE DEVICE MEMORY IMPORT "
"IN DIFFERENT CONTEXT...... \n\n");
return MakeAndRunTest<BufferCommonBufferTest<false>>(
deviceID, context, defaultQueue, num_elements);
}
int test_buffer_single_queue_fence(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
params_reset();
log_info("RUNNING TEST WITH ONE QUEUE...... \n\n");
return MakeAndRunTest<BufferCommonBufferTest<true>>(
deviceID, context, defaultQueue, num_elements);
}
int test_buffer_multiple_queue_fence(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
params_reset();
numCQ = 2;
log_info("RUNNING TEST WITH TWO QUEUE...... \n\n");
return MakeAndRunTest<BufferCommonBufferTest<true>>(
deviceID, context, defaultQueue, num_elements);
}
int test_buffer_multiImport_sameCtx_fence(cl_device_id deviceID,
cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
params_reset();
multiImport = true;
log_info("RUNNING TEST WITH MULTIPLE DEVICE MEMORY IMPORT "
"IN SAME CONTEXT...... \n\n");
return MakeAndRunTest<BufferCommonBufferTest<true>>(
deviceID, context, defaultQueue, num_elements);
}
int test_buffer_multiImport_diffCtx_fence(cl_device_id deviceID,
cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
params_reset();
multiImport = true;
multiCtx = true;
log_info("RUNNING TEST WITH MULTIPLE DEVICE MEMORY IMPORT "
"IN DIFFERENT CONTEXT...... \n\n");
return MakeAndRunTest<BufferCommonBufferTest<true>>(
deviceID, context, defaultQueue, num_elements);
} }

280
test_conformance/vulkan/test_vulkan_interop_image.cpp
View File

@@ -1,5 +1,5 @@
// //
// Copyright (c) 2022 The Khronos Group Inc. // Copyright (c) 2024 The Khronos Group Inc.
// //
// Licensed under the Apache License, Version 2.0 (the "License"); // Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License. // you may not use this file except in compliance with the License.
@@ -19,7 +19,11 @@
#include "harness/errorHelpers.h" #include "harness/errorHelpers.h"
#include "harness/os_helpers.h" #include "harness/os_helpers.h"
#include <algorithm> #include <algorithm>
#include "deviceInfo.h"
#include "vulkan_test_base.h"
#include "opencl_vulkan_wrapper.hpp"
namespace {
#define MAX_2D_IMAGES 5 #define MAX_2D_IMAGES 5
#define MAX_2D_IMAGE_WIDTH 1024 #define MAX_2D_IMAGE_WIDTH 1024
@@ -46,14 +50,13 @@
ASSERT(0); \ ASSERT(0); \
} }
namespace {
struct Params struct Params
{ {
uint32_t numImage2DDescriptors; uint32_t numImage2DDescriptors;
}; };
}
static cl_uchar uuid[CL_UUID_SIZE_KHR]; cl_uchar uuid[CL_UUID_SIZE_KHR];
static cl_device_id deviceId = NULL; cl_device_id deviceId = NULL;
size_t max_width = MAX_2D_IMAGE_WIDTH; size_t max_width = MAX_2D_IMAGE_WIDTH;
size_t max_height = MAX_2D_IMAGE_HEIGHT; size_t max_height = MAX_2D_IMAGE_HEIGHT;
@@ -245,7 +248,7 @@ int run_test_with_two_queue(
VulkanCommandPool vkCommandPool(vkDevice); VulkanCommandPool vkCommandPool(vkDevice);
VulkanCommandBuffer vkCopyCommandBuffer(vkDevice, vkCommandPool); VulkanCommandBuffer vkCopyCommandBuffer(vkDevice, vkCommandPool);
VulkanCommandBuffer vkShaderCommandBuffer(vkDevice, vkCommandPool); VulkanCommandBuffer vkShaderCommandBuffer(vkDevice, vkCommandPool);
VulkanQueue &vkQueue = vkDevice.getQueue(); VulkanQueue &vkQueue = vkDevice.getQueue(getVulkanQueueFamily());
VulkanSemaphore vkVk2CLSemaphore(vkDevice, vkExternalSemaphoreHandleType); VulkanSemaphore vkVk2CLSemaphore(vkDevice, vkExternalSemaphoreHandleType);
VulkanSemaphore vkCl2VkSemaphore(vkDevice, vkExternalSemaphoreHandleType); VulkanSemaphore vkCl2VkSemaphore(vkDevice, vkExternalSemaphoreHandleType);
@@ -857,7 +860,7 @@ int run_test_with_one_queue(
VulkanCommandPool vkCommandPool(vkDevice); VulkanCommandPool vkCommandPool(vkDevice);
VulkanCommandBuffer vkCopyCommandBuffer(vkDevice, vkCommandPool); VulkanCommandBuffer vkCopyCommandBuffer(vkDevice, vkCommandPool);
VulkanCommandBuffer vkShaderCommandBuffer(vkDevice, vkCommandPool); VulkanCommandBuffer vkShaderCommandBuffer(vkDevice, vkCommandPool);
VulkanQueue &vkQueue = vkDevice.getQueue(); VulkanQueue &vkQueue = vkDevice.getQueue(getVulkanQueueFamily());
VulkanSemaphore vkVk2CLSemaphore(vkDevice, vkExternalSemaphoreHandleType); VulkanSemaphore vkVk2CLSemaphore(vkDevice, vkExternalSemaphoreHandleType);
VulkanSemaphore vkCl2VkSemaphore(vkDevice, vkExternalSemaphoreHandleType); VulkanSemaphore vkCl2VkSemaphore(vkDevice, vkExternalSemaphoreHandleType);
@@ -1352,23 +1355,18 @@ CLEANUP:
return err; return err;
} }
int test_image_common(cl_device_id device_, cl_context context_, struct ImageCommonTest : public VulkanTestBase
cl_command_queue queue_, int numElements_)
{ {
int current_device = 0; ImageCommonTest(cl_device_id device, cl_context context,
int device_count = 0; cl_command_queue queue, cl_int nelems)
int devices_prohibited = 0; : VulkanTestBase(device, context, queue, nelems)
{}
int test_image_common()
{
cl_int err = CL_SUCCESS; cl_int err = CL_SUCCESS;
cl_platform_id platform = NULL; clCommandQueueWrapper cmd_queue1;
size_t extensionSize = 0; clCommandQueueWrapper cmd_queue2;
cl_uint num_devices = 0;
cl_uint device_no = 0;
cl_device_id *devices;
char *extensions = NULL;
const char *program_source_const;
cl_command_queue cmd_queue1 = NULL;
cl_command_queue cmd_queue2 = NULL;
cl_context context = NULL;
const uint32_t num_kernels = ARRAY_SIZE(num2DImagesList) + 1; const uint32_t num_kernels = ARRAY_SIZE(num2DImagesList) + 1;
// One kernel for Cross-CQ case // One kernel for Cross-CQ case
const uint32_t num_kernel_types = 3; const uint32_t num_kernel_types = 3;
@@ -1379,119 +1377,41 @@ 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] = { NULL }; clProgramWrapper program[num_kernel_types] = { NULL };
cl_kernel kernel_float[num_kernels] = { NULL }; clKernelWrapper kernel_float[num_kernels] = { NULL };
cl_kernel kernel_signed[num_kernels] = { NULL }; clKernelWrapper kernel_signed[num_kernels] = { NULL };
cl_kernel kernel_unsigned[num_kernels] = { NULL }; clKernelWrapper kernel_unsigned[num_kernels] = { NULL };
cl_mem external_mem_image1; clMemWrapper external_mem_image1;
cl_mem external_mem_image2; clMemWrapper external_mem_image2;
std::vector<VulkanExternalSemaphoreHandleType> supportedSemaphoreTypes; std::vector<VulkanExternalSemaphoreHandleType> supportedSemaphoreTypes;
VulkanDevice vkDevice;
cl_context_properties contextProperties[] = { CL_CONTEXT_PLATFORM, 0, 0 };
// get the platform ID
err = clGetPlatformIDs(1, &platform, NULL);
test_error_and_cleanup(err, CLEANUP, "Error: Failed to get platform\n");
err = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, 0, NULL, &num_devices);
test_error_and_cleanup(
err, CLEANUP, "clGetDeviceIDs failed in returning no. of devices\n");
devices = (cl_device_id *)malloc(num_devices * sizeof(cl_device_id));
if (NULL == devices)
{
test_fail_and_cleanup(err, CLEANUP,
"Unable to allocate memory for devices\n");
}
err = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, num_devices, devices,
NULL);
test_error_and_cleanup(err, CLEANUP, "Failed to get deviceID.\n");
contextProperties[1] = (cl_context_properties)platform;
log_info("Assigned contextproperties for platform\n");
for (device_no = 0; device_no < num_devices; device_no++)
{
err = clGetDeviceInfo(devices[device_no], CL_DEVICE_EXTENSIONS, 0, NULL,
&extensionSize);
if (CL_SUCCESS != err)
{
print_error(
err,
"Error in clGetDeviceInfo for getting device_extension size\n");
goto CLEANUP;
}
extensions = (char *)malloc(extensionSize);
if (NULL == extensions)
{
err = CL_OUT_OF_HOST_MEMORY;
print_error(err, "Unable to allocate memory for extensions\n");
goto CLEANUP;
}
err = clGetDeviceInfo(devices[device_no], CL_DEVICE_EXTENSIONS,
extensionSize, extensions, NULL);
if (CL_SUCCESS != err)
{
print_error(
err, "Error in clGetDeviceInfo for getting device_extension\n");
goto CLEANUP;
}
err = clGetDeviceInfo(devices[device_no], CL_DEVICE_UUID_KHR,
CL_UUID_SIZE_KHR, uuid, NULL);
test_error_and_cleanup(err, CLEANUP,
"clGetDeviceInfo failed with error");
supportedSemaphoreTypes = supportedSemaphoreTypes =
getSupportedInteropExternalSemaphoreHandleTypes(devices[device_no], getSupportedInteropExternalSemaphoreHandleTypes(device, *vkDevice);
vkDevice);
// If device does not support any semaphores, try the next one // If device does not support any semaphores, try the next one
if (supportedSemaphoreTypes.empty()) if (supportedSemaphoreTypes.empty())
{ {
continue; log_info("Device does not support any semaphores!\n");
return TEST_SKIPPED_ITSELF;
} }
err = deviceId = device;
memcmp(uuid, vkDevice.getPhysicalDevice().getUUID(), VK_UUID_SIZE);
if (err == 0)
{
break;
}
}
if (supportedSemaphoreTypes.empty())
{
test_fail_and_cleanup(
err, CLEANUP, "No devices found that support OpenCL semaphores\n");
}
if (device_no >= num_devices)
{
test_fail_and_cleanup(err, CLEANUP,
"OpenCL error:"
"No Vulkan-OpenCL Interop capable GPU found.\n");
}
deviceId = devices[device_no];
err = setMaxImageDimensions(deviceId, max_width, max_height); err = setMaxImageDimensions(deviceId, max_width, max_height);
test_error_and_cleanup(err, CLEANUP, "error setting max image dimensions"); test_error(err, "error setting max image dimensions");
log_info("Set max_width to %zu and max_height to %zu\n", max_width, log_info("Set max_width to %zu and max_height to %zu\n", max_width,
max_height); max_height);
context = clCreateContextFromType(contextProperties, CL_DEVICE_TYPE_GPU,
NULL, NULL, &err);
test_error_and_cleanup(err, CLEANUP, "error creating context");
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, deviceId, 0, &err);
test_error_and_cleanup(err, CLEANUP, test_error(err, "Error: Failed to create command queue!\n");
"Error: Failed to create command queue!\n");
log_info("clCreateCommandQueue successfull \n"); log_info("clCreateCommandQueue successfull \n");
cmd_queue2 = clCreateCommandQueue(context, devices[device_no], 0, &err); cmd_queue2 = clCreateCommandQueue(context, deviceId, 0, &err);
test_error_and_cleanup(err, CLEANUP, test_error(err, "Error: Failed to create command queue!\n");
"Error: Failed to create command queue!\n");
log_info("clCreateCommandQueue2 successful \n"); log_info("clCreateCommandQueue2 successful \n");
@@ -1502,44 +1422,46 @@ int test_image_common(cl_device_id device_, cl_context context_,
case 0: case 0:
sprintf(source_1, kernel_source[i], "float4", "f", "float4", sprintf(source_1, kernel_source[i], "float4", "f", "float4",
"f", "f", "f"); "f", "f", "f");
sprintf(source_2, kernel_source[i], "int4", "i", "int4", "i", sprintf(source_2, kernel_source[i], "int4", "i", "int4",
"i", "i"); "i", "i", "i");
sprintf(source_3, kernel_source[i], "uint4", "ui", "uint4", sprintf(source_3, kernel_source[i], "uint4", "ui", "uint4",
"ui", "ui", "ui"); "ui", "ui", "ui");
break; break;
case 1: case 1:
sprintf(source_1, kernel_source[i], "float4", "f", "float4", sprintf(source_1, kernel_source[i], "float4", "f", "float4",
"f", "float4", "f", "float4", "f", "f", "f", "f", "f"); "f", "float4", "f", "float4", "f", "f", "f", "f",
sprintf(source_2, kernel_source[i], "int4", "i", "int4", "i", "f");
"int4", "i", "int4", "i", "i", "i", "i", "i"); sprintf(source_2, kernel_source[i], "int4", "i", "int4",
"i", "int4", "i", "int4", "i", "i", "i", "i", "i");
sprintf(source_3, kernel_source[i], "uint4", "ui", "uint4", sprintf(source_3, kernel_source[i], "uint4", "ui", "uint4",
"ui", "uint4", "ui", "uint4", "ui", "ui", "ui", "ui", "ui", "uint4", "ui", "uint4", "ui", "ui", "ui",
"ui"); "ui", "ui");
break; break;
case 2: case 2:
sprintf(source_1, kernel_source[i], "float4", "f", "float4", sprintf(source_1, kernel_source[i], "float4", "f", "float4",
"f", "float4", "f", "float4", "f", "float4", "f", "f", "float4", "f", "float4", "f", "float4", "f",
"float4", "f", "float4", "f", "float4", "f", "f", "f", "float4", "f", "float4", "f", "float4", "f", "f",
"f", "f", "f", "f", "f", "f"); "f", "f", "f", "f", "f", "f", "f");
sprintf(source_2, kernel_source[i], "int4", "i", "int4", "i", sprintf(source_2, kernel_source[i], "int4", "i", "int4",
"int4", "i", "int4", "i", "int4", "i", "int4", "i", "i", "int4", "i", "int4", "i", "int4", "i", "int4",
"int4", "i", "int4", "i", "i", "i", "i", "i", "i", "i", "i", "int4", "i", "int4", "i", "i", "i", "i", "i",
"i", "i"); "i", "i", "i", "i");
sprintf(source_3, kernel_source[i], "uint4", "ui", "uint4", sprintf(source_3, kernel_source[i], "uint4", "ui", "uint4",
"ui", "uint4", "ui", "uint4", "ui", "uint4", "ui", "ui", "uint4", "ui", "uint4", "ui", "uint4", "ui",
"uint4", "ui", "uint4", "ui", "uint4", "ui", "ui", "ui", "uint4", "ui", "uint4", "ui", "uint4", "ui", "ui",
"ui", "ui", "ui", "ui", "ui", "ui"); "ui", "ui", "ui", "ui", "ui", "ui", "ui");
break; break;
case 3: case 3:
// Addtional case for creating updateKernelCQ2 which takes two // Addtional case for creating updateKernelCQ2 which takes
// images // two images
sprintf(source_1, kernel_source[1], "float4", "f", "float4", sprintf(source_1, kernel_source[1], "float4", "f", "float4",
"f", "float4", "f", "float4", "f", "f", "f", "f", "f"); "f", "float4", "f", "float4", "f", "f", "f", "f",
sprintf(source_2, kernel_source[1], "int4", "i", "int4", "i", "f");
"int4", "i", "int4", "i", "i", "i", "i", "i"); sprintf(source_2, kernel_source[1], "int4", "i", "int4",
"i", "int4", "i", "int4", "i", "i", "i", "i", "i");
sprintf(source_3, kernel_source[1], "uint4", "ui", "uint4", sprintf(source_3, kernel_source[1], "uint4", "ui", "uint4",
"ui", "uint4", "ui", "uint4", "ui", "ui", "ui", "ui", "ui", "uint4", "ui", "uint4", "ui", "ui", "ui",
"ui"); "ui", "ui");
break; break;
} }
const char *sourceTexts[num_kernel_types] = { source_1, source_2, const char *sourceTexts[num_kernel_types] = { source_1, source_2,
@@ -1551,17 +1473,19 @@ 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"); test_error(err, "Error: Failed to build program");
// create the kernel // create the kernel
kernel_float[i] = clCreateKernel(program[0], "image2DKernel", &err); kernel_float[i] = clCreateKernel(program[0], "image2DKernel", &err);
test_error_and_cleanup(err, CLEANUP, "clCreateKernel failed"); test_error(err, "clCreateKernel failed");
kernel_signed[i] = clCreateKernel(program[1], "image2DKernel", &err); kernel_signed[i] =
test_error_and_cleanup(err, CLEANUP, "clCreateKernel failed"); clCreateKernel(program[1], "image2DKernel", &err);
test_error(err, "clCreateKernel failed");
kernel_unsigned[i] = clCreateKernel(program[2], "image2DKernel", &err); kernel_unsigned[i] =
test_error_and_cleanup(err, CLEANUP, "clCreateKernel failed "); clCreateKernel(program[2], "image2DKernel", &err);
test_error(err, "clCreateKernel failed ");
} }
for (VulkanExternalSemaphoreHandleType externalSemaphoreType : for (VulkanExternalSemaphoreHandleType externalSemaphoreType :
supportedSemaphoreTypes) supportedSemaphoreTypes)
@@ -1569,45 +1493,47 @@ int test_image_common(cl_device_id device_, cl_context context_,
if (numCQ == 2) if (numCQ == 2)
{ {
err = run_test_with_two_queue( err = run_test_with_two_queue(
context, cmd_queue1, cmd_queue2, kernel_unsigned, kernel_signed, context, (cl_command_queue &)cmd_queue1,
kernel_float, vkDevice, externalSemaphoreType); (cl_command_queue &)cmd_queue2,
(cl_kernel *)kernel_unsigned, (cl_kernel *)kernel_signed,
(cl_kernel *)kernel_float, *vkDevice,
externalSemaphoreType);
} }
else else
{ {
err = run_test_with_one_queue(context, cmd_queue1, kernel_unsigned, err = run_test_with_one_queue(
kernel_signed, kernel_float, vkDevice, context, (cl_command_queue &)cmd_queue1,
(cl_kernel *)kernel_unsigned, (cl_kernel *)kernel_signed,
(cl_kernel *)kernel_float, *vkDevice,
externalSemaphoreType); externalSemaphoreType);
} }
test_error(err, "func_name failed \n");
} }
CLEANUP:
for (int i = 0; i < num_kernels; i++)
{
if (kernel_float[i])
{
clReleaseKernel(kernel_float[i]);
}
if (kernel_unsigned[i])
{
clReleaseKernel(kernel_unsigned[i]);
}
if (kernel_signed[i])
{
clReleaseKernel(kernel_signed[i]);
}
}
for (int i = 0; i < num_kernel_types; i++)
{
if (program[i])
{
clReleaseProgram(program[i]);
}
}
if (cmd_queue1) clReleaseCommandQueue(cmd_queue1);
if (cmd_queue2) clReleaseCommandQueue(cmd_queue2);
if (context) clReleaseContext(context);
if (extensions) free(extensions);
if (devices) free(devices);
return err; return err;
}
cl_int Run() override { return test_image_common(); }
};
} // anonymous namespace
int test_image_single_queue(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue, int num_elements)
{
params_reset();
log_info("RUNNING TEST WITH ONE QUEUE...... \n\n");
return MakeAndRunTest<ImageCommonTest>(deviceID, context, defaultQueue,
num_elements);
}
int test_image_multiple_queue(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue, int num_elements)
{
params_reset();
numCQ = 2;
log_info("RUNNING TEST WITH TWO QUEUE...... \n\n");
return MakeAndRunTest<ImageCommonTest>(deviceID, context, defaultQueue,
num_elements);
} }

101
test_conformance/vulkan/test_vulkan_platform_device_info.cpp
View File

@@ -1,5 +1,5 @@
// //
// Copyright (c) 2022 The Khronos Group Inc. // Copyright (c) 2024 The Khronos Group Inc.
// //
// Licensed under the Apache License, Version 2.0 (the "License"); // Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License. // you may not use this file except in compliance with the License.
@@ -22,6 +22,10 @@
#include <string> #include <string>
#include <vector> #include <vector>
#include "vulkan_test_base.h"
namespace {
typedef struct typedef struct
{ {
cl_uint info; cl_uint info;
@@ -29,38 +33,44 @@ typedef struct
} _info; } _info;
_info platform_info_table[] = { _info platform_info_table[] = {
#define STRING(x) \ #define PLATFORM_INFO_STRING(x) \
{ \ { \
x, #x \ x, #x \
} }
STRING(CL_PLATFORM_EXTERNAL_MEMORY_IMPORT_HANDLE_TYPES_KHR), PLATFORM_INFO_STRING(CL_PLATFORM_EXTERNAL_MEMORY_IMPORT_HANDLE_TYPES_KHR),
STRING(CL_PLATFORM_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR), PLATFORM_INFO_STRING(CL_PLATFORM_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR),
STRING(CL_PLATFORM_SEMAPHORE_IMPORT_HANDLE_TYPES_KHR) PLATFORM_INFO_STRING(CL_PLATFORM_SEMAPHORE_IMPORT_HANDLE_TYPES_KHR)
#undef STRING #undef PLATFORM_INFO_STRING
}; };
_info device_info_table[] = { _info device_info_table[] = {
#define STRING(x) \ #define DEVICE_INFO_STRING(x) \
{ \ { \
x, #x \ x, #x \
} }
STRING(CL_DEVICE_SEMAPHORE_IMPORT_HANDLE_TYPES_KHR), DEVICE_INFO_STRING(CL_DEVICE_SEMAPHORE_IMPORT_HANDLE_TYPES_KHR),
STRING(CL_DEVICE_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR), DEVICE_INFO_STRING(CL_DEVICE_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR),
STRING(CL_DEVICE_EXTERNAL_MEMORY_IMPORT_HANDLE_TYPES_KHR) DEVICE_INFO_STRING(CL_DEVICE_EXTERNAL_MEMORY_IMPORT_HANDLE_TYPES_KHR)
#undef STRING #undef DEVICE_INFO_STRING
}; };
int test_platform_info(cl_device_id deviceID, cl_context _context, struct PlatformInfoTest : public VulkanTestBase
cl_command_queue _queue, int num_elements)
{ {
PlatformInfoTest(cl_device_id device, cl_context context,
cl_command_queue queue, cl_int nelems)
: VulkanTestBase(device, context, queue, nelems)
{}
cl_int Run() override
{
cl_uint i; cl_uint i;
cl_platform_id platform = getPlatformFromDevice(deviceID); cl_platform_id platform = getPlatformFromDevice(device);
cl_int errNum; cl_int errNum;
cl_uint *handle_type; cl_uint *handle_type;
size_t handle_type_size = 0; size_t handle_type_size = 0;
cl_uint num_handles = 0; cl_uint num_handles = 0;
cl_bool external_mem_extn_available = cl_bool external_mem_extn_available = is_platform_extension_available(
is_platform_extension_available(platform, "cl_khr_external_semaphore"); platform, "cl_khr_external_semaphore");
cl_bool external_sema_extn_available = cl_bool external_sema_extn_available =
is_platform_extension_available(platform, "cl_khr_external_memory"); is_platform_extension_available(platform, "cl_khr_external_memory");
cl_bool supports_atleast_one_sema_query = false; cl_bool supports_atleast_one_sema_query = false;
@@ -75,7 +85,8 @@ int test_platform_info(cl_device_id deviceID, cl_context _context,
log_info("Platform (id %lu) info:\n", (unsigned long)platform); log_info("Platform (id %lu) info:\n", (unsigned long)platform);
for (i = 0; for (i = 0;
i < sizeof(platform_info_table) / sizeof(platform_info_table[0]); i++) i < sizeof(platform_info_table) / sizeof(platform_info_table[0]);
i++)
{ {
errNum = clGetPlatformInfo(platform, platform_info_table[i].info, 0, errNum = clGetPlatformInfo(platform, platform_info_table[i].info, 0,
NULL, &handle_type_size); NULL, &handle_type_size);
@@ -87,8 +98,8 @@ int test_platform_info(cl_device_id deviceID, cl_context _context,
== CL_PLATFORM_EXTERNAL_MEMORY_IMPORT_HANDLE_TYPES_KHR == CL_PLATFORM_EXTERNAL_MEMORY_IMPORT_HANDLE_TYPES_KHR
&& external_mem_extn_available) && external_mem_extn_available)
{ {
test_fail( test_fail("External memory import handle types should be "
"External memory import handle types should be reported if " "reported if "
"cl_khr_external_memory is available.\n"); "cl_khr_external_memory is available.\n");
} }
log_info("%s not supported. Skipping the query.\n", log_info("%s not supported. Skipping the query.\n",
@@ -123,26 +134,34 @@ int test_platform_info(cl_device_id deviceID, cl_context _context,
if (external_sema_extn_available && !supports_atleast_one_sema_query) if (external_sema_extn_available && !supports_atleast_one_sema_query)
{ {
log_info("External semaphore import/export or both should be supported " log_info(
"External semaphore import/export or both should be supported "
"if cl_khr_external_semaphore is available.\n"); "if cl_khr_external_semaphore is available.\n");
return TEST_FAIL; return TEST_FAIL;
} }
return TEST_PASS; return TEST_PASS;
} }
};
int test_device_info(cl_device_id deviceID, cl_context _context, struct DeviceInfoTest : public VulkanTestBase
cl_command_queue _queue, int num_elements)
{ {
DeviceInfoTest(cl_device_id device, cl_context context,
cl_command_queue queue, cl_int nelems)
: VulkanTestBase(device, context, queue, nelems)
{}
cl_int Run() override
{
cl_uint j; cl_uint j;
cl_uint *handle_type; cl_uint *handle_type;
size_t handle_type_size = 0; size_t handle_type_size = 0;
cl_uint num_handles = 0; cl_uint num_handles = 0;
cl_int errNum = CL_SUCCESS; cl_int errNum = CL_SUCCESS;
cl_bool external_mem_extn_available = cl_bool external_mem_extn_available =
is_extension_available(deviceID, "cl_khr_external_memory"); is_extension_available(device, "cl_khr_external_memory");
cl_bool external_sema_extn_available = cl_bool external_sema_extn_available =
is_extension_available(deviceID, "cl_khr_external_semaphore"); is_extension_available(device, "cl_khr_external_semaphore");
cl_bool supports_atleast_one_sema_query = false; cl_bool supports_atleast_one_sema_query = false;
if (!external_mem_extn_available && !external_sema_extn_available) if (!external_mem_extn_available && !external_sema_extn_available)
@@ -152,10 +171,10 @@ int test_device_info(cl_device_id deviceID, cl_context _context,
return TEST_SKIPPED_ITSELF; return TEST_SKIPPED_ITSELF;
} }
for (j = 0; j < sizeof(device_info_table) / sizeof(device_info_table[0]); for (j = 0;
j++) j < sizeof(device_info_table) / sizeof(device_info_table[0]); j++)
{ {
errNum = clGetDeviceInfo(deviceID, device_info_table[j].info, 0, NULL, errNum = clGetDeviceInfo(device, device_info_table[j].info, 0, NULL,
&handle_type_size); &handle_type_size);
test_error(errNum, "clGetDeviceInfo failed"); test_error(errNum, "clGetDeviceInfo failed");
@@ -165,8 +184,8 @@ int test_device_info(cl_device_id deviceID, cl_context _context,
== CL_DEVICE_EXTERNAL_MEMORY_IMPORT_HANDLE_TYPES_KHR == CL_DEVICE_EXTERNAL_MEMORY_IMPORT_HANDLE_TYPES_KHR
&& external_mem_extn_available) && external_mem_extn_available)
{ {
test_fail( test_fail("External memory import handle types should be "
"External memory import handle types should be reported if " "reported if "
"cl_khr_external_memory is available.\n"); "cl_khr_external_memory is available.\n");
} }
log_info("%s not supported. Skipping the query.\n", log_info("%s not supported. Skipping the query.\n",
@@ -185,7 +204,7 @@ int test_device_info(cl_device_id deviceID, cl_context _context,
num_handles = handle_type_size / sizeof(cl_uint); num_handles = handle_type_size / sizeof(cl_uint);
handle_type = (cl_uint *)malloc(handle_type_size); handle_type = (cl_uint *)malloc(handle_type_size);
errNum = clGetDeviceInfo(deviceID, device_info_table[j].info, errNum = clGetDeviceInfo(device, device_info_table[j].info,
handle_type_size, handle_type, NULL); handle_type_size, handle_type, NULL);
test_error(errNum, "clGetDeviceInfo failed"); test_error(errNum, "clGetDeviceInfo failed");
@@ -202,10 +221,28 @@ int test_device_info(cl_device_id deviceID, cl_context _context,
if (external_sema_extn_available && !supports_atleast_one_sema_query) if (external_sema_extn_available && !supports_atleast_one_sema_query)
{ {
log_info("External semaphore import/export or both should be supported " log_info(
"External semaphore import/export or both should be supported "
"if cl_khr_external_semaphore is available.\n"); "if cl_khr_external_semaphore is available.\n");
return TEST_FAIL; return TEST_FAIL;
} }
return TEST_PASS; return TEST_PASS;
}
};
} // anonymous namespace
int test_platform_info(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue, int num_elements)
{
return MakeAndRunTest<PlatformInfoTest>(deviceID, context, defaultQueue,
num_elements);
}
int test_device_info(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue, int num_elements)
{
return MakeAndRunTest<DeviceInfoTest>(deviceID, context, defaultQueue,
num_elements);
} }

129
test_conformance/vulkan/vulkan_test_base.h Normal file
View File

@@ -0,0 +1,129 @@
//
// Copyright (c) 2024 The Khronos Group Inc.
//
// Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License.
// You may obtain a copy of the License at
//
// http://www.apache.org/licenses/LICENSE-2.0
//
// Unless required by applicable law or agreed to in writing, software
// distributed under the License is distributed on an "AS IS" BASIS,
// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
// See the License for the specific language governing permissions and
// limitations under the License.
#ifndef CL_VULKAN_TEST_BASE_H
#define CL_VULKAN_TEST_BASE_H
#include <CL/cl_ext.h>
#include <memory>
#include <vector>
#include "vulkan_interop_common.hpp"
#include "harness/deviceInfo.h"
#include "harness/testHarness.h"
#include "harness/typeWrappers.h"
inline void params_reset()
{
numCQ = 1;
multiImport = false;
multiCtx = false;
}
struct VulkanTestBase
{
VulkanTestBase(cl_device_id device, cl_context context,
cl_command_queue queue, cl_int nelems)
: device(device), context(context), num_elems(nelems)
{
vkDevice.reset(
new VulkanDevice(getAssociatedVulkanPhysicalDevice(device)));
if (!(is_extension_available(device, "cl_khr_external_memory")
&& is_extension_available(device, "cl_khr_external_semaphore")))
{
log_info("Device does not support cl_khr_external_memory "
"or cl_khr_external_semaphore\n");
log_info(" TEST SKIPPED\n");
throw std::runtime_error("VulkanTestBase not supported");
}
cl_platform_id platform;
cl_int error = clGetDeviceInfo(device, CL_DEVICE_PLATFORM,
sizeof(cl_platform_id), &platform, NULL);
if (error != CL_SUCCESS)
throw std::runtime_error(
"clGetDeviceInfo for CL_DEVICE_PLATFORM failed");
// verify whether selected device is one of the type CL_DEVICE_TYPE_GPU
cl_uint num_devices = 0;
error =
clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, 0, NULL, &num_devices);
if (CL_SUCCESS != error)
throw std::runtime_error(
"clGetDeviceIDs failed in returning of devices");
std::vector<cl_device_id> devices(num_devices);
error = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, num_devices,
devices.data(), NULL);
bool found_gpu_match = false;
for (cl_uint i = 0; i < num_devices; i++)
if (devices[i] == device)
{
found_gpu_match = true;
break;
}
if (!found_gpu_match)
throw std::runtime_error(
"Vulkan tests can only run on a GPU device.");
init_cl_vk_ext(platform, 1, &device);
}
virtual cl_int Run() = 0;
protected:
cl_device_id device = nullptr;
cl_context context = nullptr;
clCommandQueueWrapper queue = nullptr;
cl_int num_elems = 0;
std::unique_ptr<VulkanDevice> vkDevice;
};
template <class T>
int MakeAndRunTest(cl_device_id device, cl_context context,
cl_command_queue queue, cl_int nelems)
{
if (!checkVkSupport())
{
log_info("Vulkan supported GPU not found \n");
log_info("TEST SKIPPED \n");
return TEST_SKIPPED_ITSELF;
}
cl_int status = TEST_PASS;
try
{
// moved from original test - do we want to stick to that ?
cl_int numElementsToUse = 1024;
auto test_fixture =
T(device, context, queue, /*nelems*/ numElementsToUse);
status = test_fixture.Run();
} catch (const std::runtime_error &e)
{
log_error("%s", e.what());
return TEST_FAIL;
}
return status;
}
#endif // CL_VULKAN_TEST_BASE_H