BRL-CAD/ispc
1
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

Mock driver and initial set of ISPCRT mock tests

Browse Source 
This infrastructure allows testing of scenarios
that are difficult to be tested on real hardware,
e.g. failuers, gpu hangs etc.

It consists of two parts:
- level zero mock that is pretending to be Level Zero NULL driver
- gtest library
- initial set of tests checking if ISPCRT correctly passes GPU hang info to the user
This commit is contained in:
Jan Zielinski
2020-12-08 21:53:45 -08:00
committed by Dmitry Babokin
parent c20d8ec50b
commit c5d179a601
16 changed files with 1063 additions and 38 deletions
Download Patch File Download Diff File Expand all files Collapse all files
.gitmodules
+3
View File
@@ -1,3 +1,6 @@
[submodule "benchmarks/vendor/google/benchmark"]
path = benchmarks/vendor/google/benchmark
url = https://github.com/google/benchmark
[submodule "ispcrt/tests/vendor/google/googletest"]
path = ispcrt/tests/vendor/google/googletest
url = https://github.com/google/googletest.git
examples/portable/genx/cmake/AddPerfExample.cmake
+1 -1
View File
@@ -81,7 +81,7 @@ function(add_perf_example)
add_executable(${GBENCH_TEST_NAME} ${parsed_GBENCH_SRC_NAME})
target_compile_definitions(${GBENCH_TEST_NAME} PRIVATE ISPCRT)
target_include_directories(${GBENCH_TEST_NAME} PRIVATE ${COMMON_PATH} ${GBENCH_INCLUDE_DIR})
link_directories(${CMAKE_SOURCE_DIR}/build/)
link_directories(${CMAKE_BINARY_DIR})
if (ISPC_BUILD)
target_link_libraries(${GBENCH_TEST_NAME} ispcrt benchmark)
else()
ispcrt/CMakeLists.txt
+7
View File
@@ -33,6 +33,7 @@ list(APPEND CMAKE_MODULE_PATH ${CMAKE_SOURCE_DIR}/cmake)
option(ISPCRT_BUILD_CPU "Enable CPU support in ispcrt" ON)
option(ISPCRT_BUILD_GPU "Enable Level0 GPU support in ispcrt" ON)
option(ISPCRT_BUILD_TASKING "Enable CPU tasking targets in ispcrt" ON)
option(ISPCRT_BUILD_TESTS "Enable ispcrt tests" ON)
if (NOT ISPCRT_BUILD_CPU AND NOT ISPCRT_BUILD_GPU)
message(FATAL_ERROR "You must enable either CPU or GPU support!")
@@ -114,6 +115,12 @@ endmacro()
build_ispcrt(SHARED ${PROJECT_NAME})
build_ispcrt(STATIC ${PROJECT_NAME}_static)
## Build tests ############################################################
if (ISPCRT_BUILD_TESTS)
add_subdirectory(tests)
endif()
## Install headers ############################################################
install(FILES ispcrt.h ispcrt.hpp ispcrt.isph
ispcrt/detail/Exception.h
+24
View File
@@ -0,0 +1,24 @@
// Copyright 2020 Intel Corporation
// SPDX-License-Identifier: BSD-3-Clause
#pragma once
// public
#include "../ispcrt.h"
// std
#include <stdexcept>
#include <string>
namespace ispcrt {
namespace base {
struct ispcrt_runtime_error : public std::runtime_error {
ISPCRTError e;
ispcrt_runtime_error(ISPCRTError _e, const std::string& msg) : std::runtime_error(msg), e {_e} {}
};
} // namespace base
} // namespace ispcrt
ispcrt/detail/gpu/GPUDevice.cpp
+77 -14
View File
@@ -20,19 +20,71 @@
#include <sstream>
#include <vector>
// ispcrt
#include "detail/Exception.h"
// level0
#include <level_zero/ze_api.h>
#define L0_SAFE_CALL(call) \
namespace ispcrt {
namespace gpu {
static const ISPCRTError getIspcrtError(ze_result_t err) {
auto res = ISPCRT_UNKNOWN_ERROR;
switch (err) {
case ZE_RESULT_SUCCESS:
res = ISPCRT_NO_ERROR;
break;
case ZE_RESULT_ERROR_DEVICE_LOST:
res = ISPCRT_DEVICE_LOST;
break;
case ZE_RESULT_ERROR_INVALID_ARGUMENT:
case ZE_RESULT_ERROR_INVALID_NULL_HANDLE:
case ZE_RESULT_ERROR_INVALID_NULL_POINTER:
case ZE_RESULT_ERROR_INVALID_SIZE:
case ZE_RESULT_ERROR_UNSUPPORTED_SIZE:
case ZE_RESULT_ERROR_UNSUPPORTED_ALIGNMENT:
case ZE_RESULT_ERROR_INVALID_ENUMERATION:
case ZE_RESULT_ERROR_UNSUPPORTED_ENUMERATION:
case ZE_RESULT_ERROR_UNSUPPORTED_IMAGE_FORMAT:
case ZE_RESULT_ERROR_INVALID_NATIVE_BINARY:
case ZE_RESULT_ERROR_INVALID_GLOBAL_NAME:
case ZE_RESULT_ERROR_INVALID_KERNEL_NAME:
case ZE_RESULT_ERROR_INVALID_FUNCTION_NAME:
case ZE_RESULT_ERROR_INVALID_GROUP_SIZE_DIMENSION:
case ZE_RESULT_ERROR_INVALID_GLOBAL_WIDTH_DIMENSION:
case ZE_RESULT_ERROR_INVALID_KERNEL_ARGUMENT_INDEX:
case ZE_RESULT_ERROR_INVALID_KERNEL_ARGUMENT_SIZE:
case ZE_RESULT_ERROR_INVALID_KERNEL_ATTRIBUTE_VALUE:
case ZE_RESULT_ERROR_INVALID_MODULE_UNLINKED:
case ZE_RESULT_ERROR_INVALID_COMMAND_LIST_TYPE:
case ZE_RESULT_ERROR_OVERLAPPING_REGIONS:
res = ISPCRT_INVALID_ARGUMENT;
break;
case ZE_RESULT_ERROR_HANDLE_OBJECT_IN_USE:
case ZE_RESULT_ERROR_INVALID_SYNCHRONIZATION_OBJECT:
res = ISPCRT_INVALID_OPERATION;
break;
default:
res = ISPCRT_UNKNOWN_ERROR;
}
return res;
}
#define L0_THROW_IF(status) \
{ \
auto status = (call); \
if (status != 0) { \
std::stringstream ss; \
ss << __FILE__ << ":" << __LINE__ << ": L0 error 0x" << std::hex << (int)status; \
throw std::runtime_error(ss.str()); \
throw ispcrt::base::ispcrt_runtime_error(ispcrt::gpu::getIspcrtError(status), ss.str()); \
} \
}
#define L0_SAFE_CALL(call) \
{ \
L0_THROW_IF((call)) \
}
#define L0_SAFE_CALL_NOEXCEPT(call) \
{ \
auto status = (call); \
@@ -43,9 +95,6 @@
} \
}
namespace ispcrt {
namespace gpu {
struct Future : public ispcrt::base::Future {
Future() {}
virtual ~Future() {}
@@ -107,7 +156,7 @@ struct EventPool {
// Create pool
ze_event_pool_desc_t eventPoolDesc = {};
eventPoolDesc.count = POOL_SIZE;
// JZTODO: shouldn't it be host-visible?
// TODO: shouldn't it be host-visible?
// ZE_EVENT_POOL_FLAG_HOST_VISIBLE
eventPoolDesc.flags = (ze_event_pool_flag_t)(ZE_EVENT_POOL_FLAG_KERNEL_TIMESTAMP);
L0_SAFE_CALL(zeEventPoolCreate(m_context, &eventPoolDesc, 1, &m_device, &m_pool));
@@ -181,7 +230,10 @@ struct MemoryView : public ispcrt::base::MemoryView {
private:
void allocate() {
ze_device_mem_alloc_desc_t allocDesc = {};
L0_SAFE_CALL(zeMemAllocDevice(m_context, &allocDesc, m_size, m_size, m_device, &m_devicePtr));
auto status = zeMemAllocDevice(m_context, &allocDesc, m_size, m_size, m_device, &m_devicePtr);
if (status != ZE_RESULT_SUCCESS)
m_devicePtr = nullptr;
L0_THROW_IF(status);
}
void *m_hostPtr{nullptr};
@@ -193,12 +245,16 @@ struct MemoryView : public ispcrt::base::MemoryView {
};
struct Module : public ispcrt::base::Module {
Module(ze_device_handle_t device, ze_context_handle_t context, const char *moduleFile) : m_file(moduleFile) {
Module(ze_device_handle_t device, ze_context_handle_t context, const char *moduleFile, bool is_mock_dev)
: m_file(moduleFile) {
std::ifstream is;
ze_module_format_t moduleFormat = ZE_MODULE_FORMAT_IL_SPIRV;
// Try to open spv file by default if ISPCRT_USE_ZEBIN is not set.
// TODO: change default to zebin when it gets more mature
const char *userZEBinFormatEnv = getenv("ISPCRT_USE_ZEBIN");
size_t codeSize = 0;
if (!is_mock_dev) {
if (userZEBinFormatEnv) {
is.open(m_file + ".bin", std::ios::binary);
moduleFormat = ZE_MODULE_FORMAT_NATIVE;
@@ -212,7 +268,7 @@ struct Module : public ispcrt::base::Module {
}
is.seekg(0, std::ios::end);
size_t codeSize = is.tellg();
codeSize = is.tellg();
is.seekg(0, std::ios::beg);
if (codeSize == 0)
@@ -222,6 +278,7 @@ struct Module : public ispcrt::base::Module {
is.read((char *)m_code.data(), codeSize);
is.close();
}
// Collect potential additional options for the compiler from the environment.
// We assume some default options for the compiler, but we also
@@ -312,16 +369,13 @@ struct TaskQueue : public ispcrt::base::TaskQueue {
TaskQueue(ze_device_handle_t device, ze_context_handle_t context) : m_ep(context, device) {
ze_command_list_desc_t commandListDesc = {};
L0_SAFE_CALL(zeCommandListCreate(context, device, &commandListDesc, &m_cl));
if (m_cl == nullptr)
throw std::runtime_error("Failed to create command list!");
// Create a command queue
ze_command_queue_desc_t commandQueueDesc = {};
commandQueueDesc.mode = ZE_COMMAND_QUEUE_MODE_ASYNCHRONOUS;
commandQueueDesc.priority = ZE_COMMAND_QUEUE_PRIORITY_NORMAL;
L0_SAFE_CALL(zeCommandQueueCreate(context, device, &commandQueueDesc, &m_q));
if (m_q == nullptr)
throw std::runtime_error("Failed to create command queue!");
}
@@ -431,6 +485,10 @@ GPUDevice::GPUDevice() {
std::istringstream(gpuDeviceEnv) >> gpuDeviceToGrab;
}
auto gpuDevice = 0;
// We can use a mock device driver for testing
m_is_mock = getenv("ISPCRT_MOCK_DEVICE") != nullptr;
for (auto &driver : allDrivers) {
uint32_t deviceCount = 0;
L0_SAFE_CALL(zeDeviceGet(driver, &deviceCount, nullptr));
@@ -440,6 +498,11 @@ GPUDevice::GPUDevice() {
for (auto &device : allDevices) {
ze_device_properties_t device_properties;
L0_SAFE_CALL(zeDeviceGetProperties(device, &device_properties));
if (m_is_mock && std::string(device_properties.name) == "ISPCRT Mock Device") {
m_device = device;
m_driver = driver;
break;
}
if (device_properties.type == ZE_DEVICE_TYPE_GPU && device_properties.vendorId == 0x8086) {
gpuDevice++;
if (gpuDevice == gpuDeviceToGrab + 1) {
@@ -474,7 +537,7 @@ base::TaskQueue *GPUDevice::newTaskQueue() const {
}
base::Module *GPUDevice::newModule(const char *moduleFile) const {
return new gpu::Module((ze_device_handle_t)m_device, (ze_context_handle_t)m_context, moduleFile);
return new gpu::Module((ze_device_handle_t)m_device, (ze_context_handle_t)m_context, moduleFile, m_is_mock);
}
base::Kernel *GPUDevice::newKernel(const base::Module &module, const char *name) const {
ispcrt/detail/gpu/GPUDevice.h
+1
View File
@@ -23,6 +23,7 @@ struct GPUDevice : public base::Device {
void *m_driver{nullptr};
void *m_device{nullptr};
void *m_context{nullptr};
bool m_is_mock{false};
};
} // namespace ispcrt
ispcrt/ispcrt.cpp
+9 -1
View File
@@ -6,6 +6,7 @@
#include <exception>
#include <iostream>
// ispcrt
#include "detail/Exception.h"
#include "detail/Module.h"
#include "detail/TaskQueue.h"
@@ -17,7 +18,10 @@
#include "detail/gpu/GPUDevice.h"
#endif
static void defaultErrorFcn(ISPCRTError e, const char *msg) { std::cerr << "ISPCRT Error: " << std::endl; }
static void defaultErrorFcn(ISPCRTError e, const char *msg) {
std::cerr << "ISPCRT Error (" << e << "): " << msg << std::endl;
exit(-1);
}
static ISPCRTErrorFunc g_errorFcn = &defaultErrorFcn;
@@ -36,6 +40,10 @@ static OBJECT_T &referenceFromHandle(HANDLE_T handle) {
#define ISPCRT_CATCH_BEGIN try {
#define ISPCRT_CATCH_END(a) \
} \
catch (const ispcrt::base::ispcrt_runtime_error &e) { \
handleError(e.e, e.what()); \
return a; \
} \
catch (const std::logic_error &e) { \
handleError(ISPCRT_INVALID_OPERATION, e.what()); \
return a; \
ispcrt/ispcrt.h
+2 -1
View File
@@ -45,7 +45,8 @@ typedef enum {
ISPCRT_NO_ERROR = 0,
ISPCRT_UNKNOWN_ERROR = 1,
ISPCRT_INVALID_ARGUMENT = 2,
ISPCRT_INVALID_OPERATION = 3
ISPCRT_INVALID_OPERATION = 3,
ISPCRT_DEVICE_LOST = 4
} ISPCRTError;
typedef void (*ISPCRTErrorFunc)(ISPCRTError, const char *errorMessage);
ispcrt/tests/CMakeLists.txt
+48
View File
@@ -0,0 +1,48 @@
## Copyright 2020 Intel Corporation
## SPDX-License-Identifier: BSD-3-Clause
cmake_minimum_required(VERSION 3.13)
set(ISPCRT_TESTS_PROJECT_NAME ispcrt_tests)
project(${ISPCRT_TESTS_PROJECT_NAME}
DESCRIPTION "ISPCRT tests"
LANGUAGES CXX)
#################### Add Google Test to the project #####################
# Tests depend on Google Test submodule
if(NOT EXISTS "${PROJECT_SOURCE_DIR}/ispcrt/tests/vendor/google/googletest/CMakeLists.txt")
message(STATUS "Google Test submodule update")
execute_process(COMMAND ${GIT_BINARY} submodule update --init --recursive
WORKING_DIRECTORY ${CMAKE_SOURCE_DIR}
RESULT_VARIABLE GIT_SUBMOD_RESULT)
if(NOT GIT_SUBMOD_RESULT EQUAL "0")
message(FATAL_ERROR "git submodule update --init failed with ${GIT_SUBMOD_RESULT}, please checkout submodules")
endif()
endif()
set_property(GLOBAL PROPERTY USE_FOLDERS ON)
# Build google test (target: gtest)
# Do not install testing lib, we don't need it, as we run them from build directory with correct rpath embedded.
set(INSTALL_GTEST OFF CACHE BOOL "Suppressing gtest install" FORCE)
# Link shared library
set(BUILD_SHARED_LIBS ON CACHE BOOL "Build shared gtest lib" FORCE)
add_subdirectory(vendor/google/googletest)
# Warn if build type is not release.
if(CMAKE_BUILD_TYPE STREQUAL "Release")
# NDEBUG may be excluded from build options globally, if LLVM build doesn't have that.
# But building Google Test, we need it to let it know that optimized build is used.
target_compile_definitions(gtest PRIVATE NDEBUG)
endif()
####################### Configure and add submodules ########################
# Meta target for tests
add_custom_target(${ISPCRT_TESTS_PROJECT_NAME})
# Level Zero API mock library
add_subdirectory(level_zero_mock)
# Tests using Level Zero mock library
add_subdirectory(mock_tests)
ispcrt/tests/level_zero_mock/CMakeLists.txt
+27
View File
@@ -0,0 +1,27 @@
## Copyright 2020 Intel Corporation
## SPDX-License-Identifier: BSD-3-Clause
set(TARGET_NAME ze_null)
add_library(${TARGET_NAME}
SHARED
${CMAKE_CURRENT_SOURCE_DIR}/ze_mock.h
${CMAKE_CURRENT_SOURCE_DIR}/ze_mock.cpp
${CMAKE_CURRENT_SOURCE_DIR}/ze_mock_driver.cpp
)
set_target_properties(${TARGET_NAME} PROPERTIES
# VERSION "${PROJECT_VERSION_MAJOR}.${PROJECT_VERSION_MINOR}.${PROJECT_VERSION_PATCH}"
VERSION "1.0.0"
SOVERSION "1"
)
target_include_directories(${TARGET_NAME}
PUBLIC
${CMAKE_CURRENT_SOURCE_DIR}
)
# if(UNIX)
# set(GCC_COVERAGE_COMPILE_FLAGS "-fvisibility=hidden -fvisibility-inlines-hidden")
# set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} ${GCC_COVERAGE_COMPILE_FLAGS}")
# endif()
ispcrt/tests/level_zero_mock/ze_mock.cpp
+42
View File
@@ -0,0 +1,42 @@
// Copyright 2020 Intel Corporation
// SPDX-License-Identifier: BSD-3-Clause
#include "ze_mock.h"
#include <cstring>
namespace ispcrt {
namespace testing {
namespace mock {
std::unordered_map<std::string, ze_result_t> Config::resultsMap;
std::vector<CmdListElem> Config::cmdList;
bool Config::cmdListOpened = true;
void Config::setRetValue(const std::string &fun, ze_result_t result) { resultsMap[fun] = result; }
ze_result_t Config::getRetValue(const std::string &fun) {
return resultsMap.count(fun) == 0 ? ZE_RESULT_SUCCESS : resultsMap[fun];
}
void Config::cleanup() {
resetCmdList();
resultsMap.clear();
}
void Config::addToCmdList(CmdListElem cle) { cmdList.push_back(cle); }
void Config::resetCmdList() {
cmdList.clear();
cmdListOpened = true;
}
void Config::closeCmdList() { cmdListOpened = false; }
bool Config::isCmdListClosed() { return !cmdListOpened; }
bool Config::checkCmdList(std::vector<CmdListElem> expected) { return expected == cmdList; }
} // namespace mock
} // namespace testing
} // namespace ispcrt
ispcrt/tests/level_zero_mock/ze_mock.h
+38
View File
@@ -0,0 +1,38 @@
// Copyright 2020 Intel Corporation
// SPDX-License-Identifier: BSD-3-Clause
#pragma once
#include <stdlib.h>
#include <unordered_map>
#include <vector>
#include <level_zero/ze_ddi.h>
#include <level_zero/zes_ddi.h>
#include <level_zero/zet_ddi.h>
namespace ispcrt {
namespace testing {
namespace mock {
enum class CmdListElem { MemoryCopy, KernelLaunch, Barrier };
class Config {
public:
static void setRetValue(const std::string &fun, ze_result_t result);
static ze_result_t getRetValue(const std::string &fun);
static void cleanup();
static void addToCmdList(CmdListElem cle);
static void resetCmdList();
static void closeCmdList();
static bool isCmdListClosed();
static bool checkCmdList(std::vector<CmdListElem> expected);
private:
static std::unordered_map<std::string, ze_result_t> resultsMap;
static std::vector<CmdListElem> cmdList;
static bool cmdListOpened;
};
} // namespace mock
} // namespace testing
} // namespace ispcrt
ispcrt/tests/level_zero_mock/ze_mock_driver.cpp
+366
View File
@@ -0,0 +1,366 @@
// Copyright 2020 Intel Corporation
// SPDX-License-Identifier: BSD-3-Clause
#include "ze_mock.h"
#include <cstring>
#include <iostream>
namespace ispcrt {
namespace testing {
namespace mock {
namespace driver {
#define MOCK_RET return Config::getRetValue(__FUNCTION__)
#define MOCK_SHOULD_SUCCEED (Config::getRetValue(__FUNCTION__) == ZE_RESULT_SUCCESS)
static unsigned MockHandleHandle = 1;
template <typename HT> struct MockHandle {
HT get() { return handle; }
MockHandle() { handle = reinterpret_cast<HT>(MockHandleHandle++); }
private:
HT handle;
};
MockHandle<ze_device_handle_t> DeviceHandle;
MockHandle<ze_context_handle_t> ContextHandle;
MockHandle<ze_module_handle_t> ModuleHandle;
MockHandle<ze_kernel_handle_t> KernelHandle;
MockHandle<ze_command_list_handle_t> CmdListHandle;
MockHandle<ze_command_queue_handle_t> CmdQueueHandle;
MockHandle<ze_event_pool_handle_t> EventPoolHandle;
MockHandle<ze_event_handle_t> LaunchEventHandle;
ze_result_t zeInit(ze_init_flags_t flags) { MOCK_RET; }
ze_result_t zeDriverGet(uint32_t *pCount, ze_driver_handle_t *phDrivers) {
if (*pCount == 0) {
*pCount = 1;
}
// Should always succeed - error handling is done by Level Zero loader
return ZE_RESULT_SUCCESS;
}
ze_result_t zeDeviceGet(ze_driver_handle_t hDriver, uint32_t *pCount, ze_device_handle_t *phDevices) {
*pCount = 1;
if (phDevices) {
*phDevices = DeviceHandle.get();
}
MOCK_RET;
}
ze_result_t zeDeviceGetProperties(ze_device_handle_t hDevice, ze_device_properties_t *pDeviceProperties) {
if (hDevice != DeviceHandle.get() || pDeviceProperties == nullptr)
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
pDeviceProperties->stype = ZE_STRUCTURE_TYPE_DEVICE_PROPERTIES;
pDeviceProperties->type = ZE_DEVICE_TYPE_GPU;
std::strcpy(pDeviceProperties->name, "ISPCRT Mock Device");
MOCK_RET;
}
ze_result_t zeContextCreate(ze_driver_handle_t hDriver, const ze_context_desc_t *desc, ze_context_handle_t *phContext) {
*phContext = ContextHandle.get();
MOCK_RET;
}
ze_result_t zeContextDestroy(ze_context_handle_t hContext) {
if (hContext != ContextHandle.get())
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
MOCK_RET;
}
ze_result_t zeCommandQueueCreate(ze_context_handle_t hContext, ze_device_handle_t hDevice,
const ze_command_queue_desc_t *desc, ze_command_queue_handle_t *phCommandQueue) {
if (hDevice != DeviceHandle.get() || hContext != ContextHandle.get() || desc == nullptr ||
phCommandQueue == nullptr)
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
*phCommandQueue = CmdQueueHandle.get();
MOCK_RET;
}
ze_result_t zeCommandQueueDestroy(ze_command_queue_handle_t hCommandQueue) {
if (hCommandQueue != CmdQueueHandle.get())
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
MOCK_RET;
}
ze_result_t zeCommandQueueExecuteCommandLists(ze_command_queue_handle_t hCommandQueue, uint32_t numCommandLists,
ze_command_list_handle_t *phCommandLists, ze_fence_handle_t hFence) {
if (hCommandQueue != CmdQueueHandle.get() || !Config::isCmdListClosed())
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
MOCK_RET;
}
ze_result_t zeCommandQueueSynchronize(ze_command_queue_handle_t hCommandQueue, uint64_t timeout) {
if (hCommandQueue != CmdQueueHandle.get() || !Config::isCmdListClosed())
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
MOCK_RET;
}
ze_result_t zeCommandListCreate(ze_context_handle_t hContext, ze_device_handle_t hDevice,
const ze_command_list_desc_t *desc, ze_command_list_handle_t *phCommandList) {
if (hDevice != DeviceHandle.get() || hContext != ContextHandle.get() || desc == nullptr || phCommandList == nullptr)
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
*phCommandList = CmdListHandle.get();
MOCK_RET;
}
ze_result_t zeCommandListDestroy(ze_command_list_handle_t hCommandList) {
if (hCommandList != CmdListHandle.get())
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
MOCK_RET;
}
ze_result_t zeCommandListClose(ze_command_list_handle_t hCommandList) {
if (hCommandList != CmdListHandle.get())
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
Config::closeCmdList();
MOCK_RET;
}
ze_result_t zeCommandListReset(ze_command_list_handle_t hCommandList) {
if (hCommandList != CmdListHandle.get())
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
Config::resetCmdList();
MOCK_RET;
}
ze_result_t zeCommandListAppendBarrier(ze_command_list_handle_t hCommandList, ze_event_handle_t hSignalEvent,
uint32_t numWaitEvents, ze_event_handle_t *phWaitEvents) {
if (hCommandList != CmdListHandle.get() || Config::isCmdListClosed())
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
Config::addToCmdList(CmdListElem::Barrier);
MOCK_RET;
}
ze_result_t zeCommandListAppendMemoryCopy(ze_command_list_handle_t hCommandList, void *dstptr, const void *srcptr,
size_t size, ze_event_handle_t hSignalEvent, uint32_t numWaitEvents,
ze_event_handle_t *phWaitEvents) {
if (hCommandList != CmdListHandle.get() || Config::isCmdListClosed())
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
if (MOCK_SHOULD_SUCCEED)
Config::addToCmdList(CmdListElem::MemoryCopy);
MOCK_RET;
}
ze_result_t zeEventPoolCreate(ze_context_handle_t hContext, const ze_event_pool_desc_t *desc, uint32_t numDevices,
ze_device_handle_t *phDevices, ze_event_pool_handle_t *phEventPool) {
if (hContext != ContextHandle.get() || numDevices == 0 || !phDevices || !phEventPool)
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
*phEventPool = EventPoolHandle.get();
MOCK_RET;
}
ze_result_t zeEventPoolDestroy(ze_event_pool_handle_t hEventPool) {
if (hEventPool != EventPoolHandle.get())
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
MOCK_RET;
}
ze_result_t zeEventCreate(ze_event_pool_handle_t hEventPool, const ze_event_desc_t *desc, ze_event_handle_t *phEvent) {
if (hEventPool != EventPoolHandle.get() || !desc || !phEvent)
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
*phEvent = LaunchEventHandle.get();
MOCK_RET;
}
ze_result_t zeEventDestroy(ze_event_handle_t hEvent) { MOCK_RET; }
ze_result_t zeEventQueryKernelTimestamp(ze_event_handle_t hEvent, ze_kernel_timestamp_result_t *dstptr) { MOCK_RET; }
ze_result_t zeMemAllocDevice(ze_context_handle_t hContext, const ze_device_mem_alloc_desc_t *device_desc, size_t size,
size_t alignment, ze_device_handle_t hDevice, void **pptr) {
if (hContext != ContextHandle.get())
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
if (MOCK_SHOULD_SUCCEED)
*pptr = new uint8_t[size];
MOCK_RET;
}
ze_result_t zeMemFree(ze_context_handle_t hContext, void *ptr) {
if (hContext != ContextHandle.get() || !ptr)
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
delete[](uint8_t *) ptr;
MOCK_RET;
}
ze_result_t zeModuleCreate(ze_context_handle_t hContext, ze_device_handle_t hDevice, const ze_module_desc_t *desc,
ze_module_handle_t *phModule, ze_module_build_log_handle_t *phBuildLog) {
if (hContext != ContextHandle.get() || hDevice != DeviceHandle.get())
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
*phModule = ModuleHandle.get();
MOCK_RET;
}
ze_result_t zeModuleDestroy(ze_module_handle_t hModule) {
if (hModule != ModuleHandle.get())
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
MOCK_RET;
}
ze_result_t zeKernelCreate(ze_module_handle_t hModule, const ze_kernel_desc_t *desc, ze_kernel_handle_t *phKernel) {
if (hModule != ModuleHandle.get())
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
*phKernel = KernelHandle.get();
MOCK_RET;
}
ze_result_t zeKernelDestroy(ze_kernel_handle_t hKernel) {
if (hKernel != KernelHandle.get())
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
MOCK_RET;
}
ze_result_t zeKernelSetArgumentValue(ze_kernel_handle_t hKernel, uint32_t argIndex, size_t argSize,
const void *pArgValue) {
MOCK_RET;
}
ze_result_t zeCommandListAppendLaunchKernel(ze_command_list_handle_t hCommandList, ze_kernel_handle_t hKernel,
const ze_group_count_t *pLaunchFuncArgs, ze_event_handle_t hSignalEvent,
uint32_t numWaitEvents, ze_event_handle_t *phWaitEvents) {
if (hCommandList != CmdListHandle.get() || hKernel != KernelHandle.get() || !pLaunchFuncArgs)
return ZE_RESULT_ERROR_INVALID_NULL_HANDLE;
Config::addToCmdList(CmdListElem::KernelLaunch);
MOCK_RET;
}
} // namespace driver
} // namespace mock
} // namespace testing
} // namespace ispcrt
#if defined(__cplusplus)
extern "C" {
#endif
ze_result_t zeGetGlobalProcAddrTable(ze_api_version_t version, ze_global_dditable_t *pDdiTable) {
pDdiTable->pfnInit = ispcrt::testing::mock::driver::zeInit;
return ZE_RESULT_SUCCESS;
}
ze_result_t zeGetDriverProcAddrTable(ze_api_version_t version, ze_driver_dditable_t *pDdiTable) {
pDdiTable->pfnGet = ispcrt::testing::mock::driver::zeDriverGet;
return ZE_RESULT_SUCCESS;
}
ze_result_t zeGetDeviceProcAddrTable(ze_api_version_t version, ze_device_dditable_t *pDdiTable) {
pDdiTable->pfnGet = ispcrt::testing::mock::driver::zeDeviceGet;
pDdiTable->pfnGetProperties = ispcrt::testing::mock::driver::zeDeviceGetProperties;
return ZE_RESULT_SUCCESS;
}
ze_result_t zeGetContextProcAddrTable(ze_api_version_t version, ze_context_dditable_t *pDdiTable) {
pDdiTable->pfnCreate = ispcrt::testing::mock::driver::zeContextCreate;
pDdiTable->pfnDestroy = ispcrt::testing::mock::driver::zeContextDestroy;
return ZE_RESULT_SUCCESS;
}
ze_result_t zeGetCommandQueueProcAddrTable(ze_api_version_t version, ze_command_queue_dditable_t *pDdiTable) {
pDdiTable->pfnCreate = ispcrt::testing::mock::driver::zeCommandQueueCreate;
pDdiTable->pfnDestroy = ispcrt::testing::mock::driver::zeCommandQueueDestroy;
pDdiTable->pfnExecuteCommandLists = ispcrt::testing::mock::driver::zeCommandQueueExecuteCommandLists;
pDdiTable->pfnSynchronize = ispcrt::testing::mock::driver::zeCommandQueueSynchronize;
return ZE_RESULT_SUCCESS;
}
ze_result_t zeGetCommandListProcAddrTable(ze_api_version_t version, ze_command_list_dditable_t *pDdiTable) {
pDdiTable->pfnCreate = ispcrt::testing::mock::driver::zeCommandListCreate;
pDdiTable->pfnDestroy = ispcrt::testing::mock::driver::zeCommandListDestroy;
pDdiTable->pfnClose = ispcrt::testing::mock::driver::zeCommandListClose;
pDdiTable->pfnReset = ispcrt::testing::mock::driver::zeCommandListReset;
pDdiTable->pfnAppendBarrier = ispcrt::testing::mock::driver::zeCommandListAppendBarrier;
pDdiTable->pfnAppendMemoryCopy = ispcrt::testing::mock::driver::zeCommandListAppendMemoryCopy;
pDdiTable->pfnAppendLaunchKernel = ispcrt::testing::mock::driver::zeCommandListAppendLaunchKernel;
return ZE_RESULT_SUCCESS;
}
ze_result_t zeGetEventProcAddrTable(ze_api_version_t version, ze_event_dditable_t *pDdiTable) {
pDdiTable->pfnCreate = ispcrt::testing::mock::driver::zeEventCreate;
pDdiTable->pfnDestroy = ispcrt::testing::mock::driver::zeEventDestroy;
pDdiTable->pfnQueryKernelTimestamp = ispcrt::testing::mock::driver::zeEventQueryKernelTimestamp;
return ZE_RESULT_SUCCESS;
}
ze_result_t zeGetEventPoolProcAddrTable(ze_api_version_t version, ze_event_pool_dditable_t *pDdiTable) {
pDdiTable->pfnCreate = ispcrt::testing::mock::driver::zeEventPoolCreate;
pDdiTable->pfnDestroy = ispcrt::testing::mock::driver::zeEventPoolDestroy;
return ZE_RESULT_SUCCESS;
}
ze_result_t zeGetKernelProcAddrTable(ze_api_version_t version, ze_kernel_dditable_t *pDdiTable) {
pDdiTable->pfnCreate = ispcrt::testing::mock::driver::zeKernelCreate;
pDdiTable->pfnDestroy = ispcrt::testing::mock::driver::zeKernelDestroy;
pDdiTable->pfnSetArgumentValue = ispcrt::testing::mock::driver::zeKernelSetArgumentValue;
return ZE_RESULT_SUCCESS;
}
ze_result_t zeGetMemProcAddrTable(ze_api_version_t version, ze_mem_dditable_t *pDdiTable) {
pDdiTable->pfnAllocDevice = ispcrt::testing::mock::driver::zeMemAllocDevice;
pDdiTable->pfnFree = ispcrt::testing::mock::driver::zeMemFree;
return ZE_RESULT_SUCCESS;
}
ze_result_t zeGetModuleProcAddrTable(ze_api_version_t version, ze_module_dditable_t *pDdiTable) {
pDdiTable->pfnCreate = ispcrt::testing::mock::driver::zeModuleCreate;
pDdiTable->pfnDestroy = ispcrt::testing::mock::driver::zeModuleDestroy;
return ZE_RESULT_SUCCESS;
}
#define MOCK_DDI_FUN(Fn, TT) \
ze_result_t Fn(ze_api_version_t version, TT *pDdiTable) { return ZE_RESULT_SUCCESS; }
MOCK_DDI_FUN(zeGetFenceProcAddrTable, ze_fence_dditable_t)
MOCK_DDI_FUN(zeGetImageProcAddrTable, ze_image_dditable_t)
MOCK_DDI_FUN(zeGetModuleBuildLogProcAddrTable, ze_module_build_log_dditable_t)
MOCK_DDI_FUN(zeGetPhysicalMemProcAddrTable, ze_physical_mem_dditable_t)
MOCK_DDI_FUN(zeGetSamplerProcAddrTable, ze_sampler_dditable_t)
MOCK_DDI_FUN(zeGetVirtualMemProcAddrTable, ze_virtual_mem_dditable_t)
// ZES
MOCK_DDI_FUN(zesGetDeviceProcAddrTable, zes_device_dditable_t)
MOCK_DDI_FUN(zesGetDriverProcAddrTable, zes_driver_dditable_t)
MOCK_DDI_FUN(zesGetDiagnosticsProcAddrTable, zes_diagnostics_dditable_t)
MOCK_DDI_FUN(zesGetEngineProcAddrTable, zes_engine_dditable_t)
MOCK_DDI_FUN(zesGetFabricPortProcAddrTable, zes_fabric_port_dditable_t)
MOCK_DDI_FUN(zesGetFanProcAddrTable, zes_fan_dditable_t)
MOCK_DDI_FUN(zesGetFirmwareProcAddrTable, zes_firmware_dditable_t)
MOCK_DDI_FUN(zesGetFrequencyProcAddrTable, zes_frequency_dditable_t)
MOCK_DDI_FUN(zesGetLedProcAddrTable, zes_led_dditable_t)
MOCK_DDI_FUN(zesGetMemoryProcAddrTable, zes_memory_dditable_t)
MOCK_DDI_FUN(zesGetPerformanceFactorProcAddrTable, zes_performance_factor_dditable_t)
MOCK_DDI_FUN(zesGetPowerProcAddrTable, zes_power_dditable_t)
MOCK_DDI_FUN(zesGetPsuProcAddrTable, zes_psu_dditable_t)
MOCK_DDI_FUN(zesGetRasProcAddrTable, zes_ras_dditable_t)
MOCK_DDI_FUN(zesGetSchedulerProcAddrTable, zes_scheduler_dditable_t)
MOCK_DDI_FUN(zesGetStandbyProcAddrTable, zes_standby_dditable_t)
MOCK_DDI_FUN(zesGetTemperatureProcAddrTable, zes_temperature_dditable_t)
// ZET
MOCK_DDI_FUN(zetGetDeviceProcAddrTable, zet_device_dditable_t)
MOCK_DDI_FUN(zetGetContextProcAddrTable, zet_context_dditable_t)
MOCK_DDI_FUN(zetGetCommandListProcAddrTable, zet_command_list_dditable_t)
MOCK_DDI_FUN(zetGetKernelProcAddrTable, zet_kernel_dditable_t)
MOCK_DDI_FUN(zetGetModuleProcAddrTable, zet_module_dditable_t)
MOCK_DDI_FUN(zetGetDebugProcAddrTable, zet_debug_dditable_t)
MOCK_DDI_FUN(zetGetMetricProcAddrTable, zet_metric_dditable_t)
MOCK_DDI_FUN(zetGetMetricGroupProcAddrTable, zet_metric_group_dditable_t)
MOCK_DDI_FUN(zetGetMetricQueryProcAddrTable, zet_metric_query_dditable_t)
MOCK_DDI_FUN(zetGetMetricQueryPoolProcAddrTable, zet_metric_query_pool_dditable_t)
MOCK_DDI_FUN(zetGetMetricStreamerProcAddrTable, zet_metric_streamer_dditable_t)
MOCK_DDI_FUN(zetGetTracerExpProcAddrTable, zet_tracer_exp_dditable_t)
#undef MOCK_DDI_FUN
#if defined(__cplusplus)
}
#endif
ispcrt/tests/mock_tests/CMakeLists.txt
+13
View File
@@ -0,0 +1,13 @@
## Copyright 2020 Intel Corporation
## SPDX-License-Identifier: BSD-3-Clause
# set the project name
project(ispcrt_mock_tests)
# add the executable
add_executable(ispcrt_mock_tests ispcrt_mock_main.cpp)
link_directories(${CMAKE_BINARY_DIR})
target_link_libraries(ispcrt_mock_tests PUBLIC gtest_main ispcrt ze_null)
target_include_directories(ispcrt_mock_tests PUBLIC ${CMAKE_SOURCE_DIR}/ispcrt/tests/level_zero_mock)
ispcrt/tests/mock_tests/ispcrt_mock_main.cpp
+383
View File
@@ -0,0 +1,383 @@
// Copyright 2020 Intel Corporation
// SPDX-License-Identifier: BSD-3-Clause
#include "ispcrt.hpp"
#include "ze_mock.h"
#include "gtest/gtest.h"
#include <stdlib.h>
namespace ispcrt {
namespace testing {
namespace mock {
// Base fixture for mock tests
class MockTest : public ::testing::Test {
protected:
void SetUp() override {
ResetError();
Config::cleanup();
setenv("ISPCRT_MOCK_DEVICE", "y", 1);
// hijak ispcrt errors - we need it to test error handling
ispcrtSetErrorFunc([](ISPCRTError e, const char *m) { sm_rt_error = e; });
EXPECT_EQ(m_device, 0);
m_device = Device(ISPCRT_DEVICE_TYPE_GPU);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
EXPECT_NE(m_device, 0);
}
void TearDown() override {
ResetError();
Config::cleanup();
}
void ResetError() { sm_rt_error = ISPCRT_NO_ERROR; }
ispcrt::Device m_device;
// TODO: not great it's a static, but for now the ISPCRT error reporting
// does not support any kind of context allowing to pass 'this' pointer
static ISPCRTError sm_rt_error;
};
ISPCRTError MockTest::sm_rt_error;
class MockTestWithModule : public MockTest {
protected:
void SetUp() override {
MockTest::SetUp();
EXPECT_EQ(m_module, 0);
m_module = Module(m_device, "");
EXPECT_NE(m_module, 0);
}
ispcrt::Module m_module;
};
class MockTestWithModuleQueueKernel : public MockTestWithModule {
protected:
void SetUp() override {
MockTestWithModule::SetUp();
EXPECT_EQ(m_task_queue, 0);
EXPECT_EQ(m_kernel, 0);
m_task_queue = TaskQueue(m_device);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
EXPECT_NE(m_task_queue, 0);
m_kernel = Kernel(m_device, m_module, "");
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
EXPECT_NE(m_kernel, 0);
}
ispcrt::TaskQueue m_task_queue;
ispcrt::Kernel m_kernel;
};
/////////////////////////////////////////////////////////////////////
// Device tests
TEST_F(MockTest, Device_Constructor_zeInit) {
// Make sure we can re-create a device even if first try failed
Config::setRetValue("zeInit", ZE_RESULT_ERROR_DEVICE_LOST);
ispcrt::Device d(ISPCRT_DEVICE_TYPE_GPU);
ASSERT_EQ(sm_rt_error, ISPCRT_DEVICE_LOST);
Config::setRetValue("zeInit", ZE_RESULT_SUCCESS);
ResetError();
ispcrt::Device d2(ISPCRT_DEVICE_TYPE_GPU);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
}
TEST_F(MockTest, Device_Constructor_zeDeviceGet) {
// Make sure we can re-create a device even if first try failed
Config::setRetValue("zeDeviceGet", ZE_RESULT_ERROR_DEVICE_LOST);
ispcrt::Device d(ISPCRT_DEVICE_TYPE_GPU);
ASSERT_EQ(sm_rt_error, ISPCRT_DEVICE_LOST);
Config::setRetValue("zeDeviceGet", ZE_RESULT_SUCCESS);
ResetError();
ispcrt::Device d2(ISPCRT_DEVICE_TYPE_GPU);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
}
TEST_F(MockTest, Device_Constructor_zeDeviceGetProperties) {
// Make sure we can re-create a device even if first try failed
Config::setRetValue("zeDeviceGetProperties", ZE_RESULT_ERROR_DEVICE_LOST);
ispcrt::Device d(ISPCRT_DEVICE_TYPE_GPU);
ASSERT_EQ(sm_rt_error, ISPCRT_DEVICE_LOST);
Config::setRetValue("zeDeviceGetProperties", ZE_RESULT_SUCCESS);
ResetError();
ispcrt::Device d2(ISPCRT_DEVICE_TYPE_GPU);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
}
TEST_F(MockTest, Device_Constructor_zeContextCreate) {
// Make sure we can re-create a device even if first try failed
Config::setRetValue("zeContextCreate", ZE_RESULT_ERROR_DEVICE_LOST);
ispcrt::Device d(ISPCRT_DEVICE_TYPE_GPU);
ASSERT_EQ(sm_rt_error, ISPCRT_DEVICE_LOST);
Config::setRetValue("zeContextCreate", ZE_RESULT_SUCCESS);
ResetError();
ispcrt::Device d2(ISPCRT_DEVICE_TYPE_GPU);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
}
/////////////////////////////////////////////////////////////////////
// Module tests
TEST_F(MockTest, Module_Constructor) {
// Simply create a module
ispcrt::Module m(m_device, "");
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
}
TEST_F(MockTest, Module_Constructor_zeModuleCreate) {
// Check if it's possible to create a module after first try failed
Config::setRetValue("zeModuleCreate", ZE_RESULT_ERROR_DEVICE_LOST);
ispcrt::Module m(m_device, "");
ASSERT_EQ(sm_rt_error, ISPCRT_DEVICE_LOST);
Config::setRetValue("zeModuleCreate", ZE_RESULT_SUCCESS);
ResetError();
ispcrt::Module m2(m_device, "");
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
}
/////////////////////////////////////////////////////////////////////
// Kernel tests
TEST_F(MockTestWithModule, Kernel_Constructor) {
// Simply create a kernel
ispcrt::Kernel k(m_device, m_module, "");
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
}
TEST_F(MockTestWithModule, Kernel_Constructor_zeKernelCreate) {
// Check if it's possible to create a kernel after the first try failed
Config::setRetValue("zeKernelCreate", ZE_RESULT_ERROR_DEVICE_LOST);
ispcrt::Kernel k(m_device, m_module, "");
ASSERT_EQ(sm_rt_error, ISPCRT_DEVICE_LOST);
Config::setRetValue("zeKernelCreate", ZE_RESULT_SUCCESS);
ResetError();
ispcrt::Kernel k2(m_device, m_module, "");
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
}
/////////////////////////////////////////////////////////////////////
// Memory allocation tests
TEST_F(MockTest, ArrayObj) {
// Simply create an allocation
std::vector<float> buf(64 * 1024);
ispcrt::Array<float> buf_dev(m_device, buf);
// devicePtr() does actual allocation
auto dev_buf_ptr = buf_dev.devicePtr();
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
}
TEST_F(MockTest, ArrayObj_zeMemAllocDevice) {
std::vector<float> buf(64 * 1024);
ispcrt::Array<float> buf_dev(m_device, buf);
Config::setRetValue("zeMemAllocDevice", ZE_RESULT_ERROR_DEVICE_LOST);
auto dev_buf_ptr = buf_dev.devicePtr();
ASSERT_EQ(sm_rt_error, ISPCRT_DEVICE_LOST);
// Check that nullptr is returned
ASSERT_EQ(dev_buf_ptr, nullptr);
Config::setRetValue("zeMemAllocDevice", ZE_RESULT_SUCCESS);
ResetError();
// Reset error and try again
dev_buf_ptr = buf_dev.devicePtr();
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_NE(dev_buf_ptr, nullptr);
}
TEST_F(MockTest, ArrayObj_zeMemAllocDevice_DiffObjs) {
std::vector<float> buf(64 * 1024);
ispcrt::Array<float> buf_dev(m_device, buf);
Config::setRetValue("zeMemAllocDevice", ZE_RESULT_ERROR_DEVICE_LOST);
auto dev_buf_ptr = buf_dev.devicePtr();
ASSERT_EQ(sm_rt_error, ISPCRT_DEVICE_LOST);
// Check that nullptr is returned
ASSERT_EQ(dev_buf_ptr, nullptr);
Config::setRetValue("zeMemAllocDevice", ZE_RESULT_SUCCESS);
ResetError();
// Reset error and try again, but create different object
ispcrt::Array<float> buf_dev2(m_device, buf);
auto dev_buf_ptr2 = buf_dev2.devicePtr();
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_EQ(dev_buf_ptr, nullptr);
ASSERT_NE(dev_buf_ptr2, nullptr);
}
/////////////////////////////////////////////////////////////////////
// TaskQueue tests
TEST_F(MockTest, TaskQueue_Constructor) {
// Simply create a task queue
ispcrt::TaskQueue tq(m_device);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
}
TEST_F(MockTest, TaskQueue_Constructor_zeEventPoolCreate) {
// Check if it's possible to create a task queue after the first try failed
Config::setRetValue("zeEventPoolCreate", ZE_RESULT_ERROR_DEVICE_LOST);
ispcrt::TaskQueue tq(m_device);
ASSERT_EQ(sm_rt_error, ISPCRT_DEVICE_LOST);
Config::setRetValue("zeEventPoolCreate", ZE_RESULT_SUCCESS);
ResetError();
ispcrt::TaskQueue tq2(m_device);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
}
TEST_F(MockTest, TaskQueue_Constructor_zeCommandListCreate) {
// Check if it's possible to create a task queue after the first try failed
Config::setRetValue("zeCommandListCreate", ZE_RESULT_ERROR_DEVICE_LOST);
ispcrt::TaskQueue tq(m_device);
ASSERT_EQ(sm_rt_error, ISPCRT_DEVICE_LOST);
Config::setRetValue("zeCommandListCreate", ZE_RESULT_SUCCESS);
ResetError();
ispcrt::TaskQueue tq2(m_device);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
}
TEST_F(MockTest, TaskQueue_Constructor_zeCommandQueueCreate) {
// Check if it's possible to create a task queue after the first try failed
Config::setRetValue("zeCommandQueueCreate", ZE_RESULT_ERROR_DEVICE_LOST);
ispcrt::TaskQueue tq(m_device);
ASSERT_EQ(sm_rt_error, ISPCRT_DEVICE_LOST);
Config::setRetValue("zeCommandQueueCreate", ZE_RESULT_SUCCESS);
ResetError();
ispcrt::TaskQueue tq2(m_device);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
}
TEST_F(MockTest, TaskQueue_CopyToDevice) {
ispcrt::TaskQueue tq(m_device);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
// Create an allocation
std::vector<float> buf(64 * 1024);
ispcrt::Array<float> buf_dev(m_device, buf);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
// "copy"
tq.copyToDevice(buf_dev);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList({CmdListElem::MemoryCopy}));
tq.barrier();
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList({CmdListElem::MemoryCopy, CmdListElem::Barrier}));
}
TEST_F(MockTest, TaskQueue_CopyToDevice_zeCommandListAppendMemoryCopy) {
ispcrt::TaskQueue tq(m_device);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
// Create an allocation
std::vector<float> buf(64 * 1024);
ispcrt::Array<float> buf_dev(m_device, buf);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
// "copy", but fail
Config::setRetValue("zeCommandListAppendMemoryCopy", ZE_RESULT_ERROR_DEVICE_LOST);
tq.copyToDevice(buf_dev);
ASSERT_EQ(sm_rt_error, ISPCRT_DEVICE_LOST);
ASSERT_TRUE(Config::checkCmdList({}));
Config::setRetValue("zeCommandListAppendMemoryCopy", ZE_RESULT_SUCCESS);
ResetError();
tq.copyToDevice(buf_dev);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList({CmdListElem::MemoryCopy}));
}
// TODO: Write test that will check if command list is cleaned up when zeCommandQueueCreate fails
// Normal kernel launch (plus a few memory transfers) - but no waiting on future
TEST_F(MockTestWithModuleQueueKernel, TaskQueue_FullKernelLaunchNoFuture) {
auto tq = m_task_queue;
// Create an allocation
std::vector<float> buf(64 * 1024);
ispcrt::Array<float> buf_dev(m_device, buf);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
// "copy"
tq.copyToDevice(buf_dev);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList({CmdListElem::MemoryCopy}));
tq.barrier();
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList({CmdListElem::MemoryCopy, CmdListElem::Barrier}));
tq.launch(m_kernel, 0);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList({CmdListElem::MemoryCopy, CmdListElem::Barrier, CmdListElem::KernelLaunch}));
tq.barrier();
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList(
{CmdListElem::MemoryCopy, CmdListElem::Barrier, CmdListElem::KernelLaunch, CmdListElem::Barrier}));
tq.sync();
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList({}));
}
// Normal kernel launch (plus a few memory transfers)
TEST_F(MockTestWithModuleQueueKernel, TaskQueue_FullKernelLaunch) {
auto tq = m_task_queue;
// Create an allocation
std::vector<float> buf(64 * 1024);
ispcrt::Array<float> buf_dev(m_device, buf);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
// "copy"
tq.copyToDevice(buf_dev);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList({CmdListElem::MemoryCopy}));
tq.barrier();
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList({CmdListElem::MemoryCopy, CmdListElem::Barrier}));
auto f = tq.launch(m_kernel, 0);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList({CmdListElem::MemoryCopy, CmdListElem::Barrier, CmdListElem::KernelLaunch}));
tq.barrier();
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList(
{CmdListElem::MemoryCopy, CmdListElem::Barrier, CmdListElem::KernelLaunch, CmdListElem::Barrier}));
tq.sync();
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList({}));
ASSERT_TRUE(f.valid());
}
// The test is disabled due to crash in EventPool destructor
// The fix for this crash is to properly deallocate all events
// that has been created in ::launch method and should be cleaned up
// in sync (but sync never happens due to - for example program exit)
TEST_F(MockTestWithModuleQueueKernel, DISABLED_TaskQueue_KernelLaunchNoSync) {
auto tq = m_task_queue;
auto f = tq.launch(m_kernel, 0);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList({CmdListElem::KernelLaunch}));
tq.barrier();
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList({CmdListElem::KernelLaunch, CmdListElem::Barrier}));
// Future should not be signaled
ASSERT_FALSE(f.valid());
}
TEST_F(MockTestWithModuleQueueKernel, TaskQueue_FullKernelLaunch1) {
auto tq = m_task_queue;
// Create an allocation
std::vector<float> buf(64 * 1024);
ispcrt::Array<float> buf_dev(m_device, buf);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
// "copy"
tq.copyToDevice(buf_dev);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList({CmdListElem::MemoryCopy}));
tq.barrier();
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList({CmdListElem::MemoryCopy, CmdListElem::Barrier}));
auto f = tq.launch(m_kernel, 0);
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList({CmdListElem::MemoryCopy, CmdListElem::Barrier, CmdListElem::KernelLaunch}));
tq.barrier();
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList(
{CmdListElem::MemoryCopy, CmdListElem::Barrier, CmdListElem::KernelLaunch, CmdListElem::Barrier}));
tq.sync();
ASSERT_EQ(sm_rt_error, ISPCRT_NO_ERROR);
ASSERT_TRUE(Config::checkCmdList({}));
ASSERT_TRUE(f.valid());
}
} // namespace mock
} // namespace testing
} // namespace ispcrt
ispcrt/tests/vendor/google/googletest
Vendored Submodule
+1
Submodule ispcrt/tests/vendor/google/googletest added at 6a7ed316a5