diff --git a/layers/99_svmplusplus/emulate.cpp b/layers/99_svmplusplus/emulate.cpp index b13b555..6f0b65b 100644 --- a/layers/99_svmplusplus/emulate.cpp +++ b/layers/99_svmplusplus/emulate.cpp @@ -1,5 +1,5 @@ /* -// Copyright (c) 2022 Ben Ashbaugh +// Copyright (c) 2023 Ben Ashbaugh // // SPDX-License-Identifier: MIT */ @@ -82,7 +82,7 @@ static bool isUSMPtr( type == CL_SVM_MEM_TYPE_SHARED_EXP; } -static cl_device_id getAssociatedDeviceFromPropertie( +static cl_device_id getAssociatedDeviceFromProperties( const cl_svm_mem_properties_exp* props) { if (props) { @@ -112,7 +112,7 @@ void* CL_API_CALL clSVMAllocWithPropertiesEXP_EMU( cl_uint alignment, cl_int* errcode_ret) { - cl_device_id device = getAssociatedDeviceFromPropertie(properties); + cl_device_id device = getAssociatedDeviceFromProperties(properties); if (flags & CL_MEM_SVM_DEVICE_EXP) { return clDeviceMemAllocINTEL( @@ -202,6 +202,53 @@ cl_int CL_API_CALL clGetDeviceInfo_override( size_t* param_value_size_ret) { switch(param_name) { + case CL_DEVICE_SVM_CAPABILITIES: + { + cl_device_unified_shared_memory_capabilities_intel deviceCaps = 0; + g_pNextDispatch->clGetDeviceInfo( + device, + CL_DEVICE_DEVICE_MEM_CAPABILITIES_INTEL, + sizeof(deviceCaps), + &deviceCaps, + nullptr ); + + cl_device_unified_shared_memory_capabilities_intel hostCaps = 0; + g_pNextDispatch->clGetDeviceInfo( + device, + CL_DEVICE_HOST_MEM_CAPABILITIES_INTEL, + sizeof(hostCaps), + &hostCaps, + nullptr ); + + // We can just check the single device shared capabilities: + cl_device_unified_shared_memory_capabilities_intel sharedCaps = 0; + g_pNextDispatch->clGetDeviceInfo( + device, + CL_DEVICE_SINGLE_DEVICE_SHARED_MEM_CAPABILITIES_INTEL, + sizeof(sharedCaps), + &sharedCaps, + nullptr ); + + cl_device_svm_capabilities svmCaps = 0; + g_pNextDispatch->clGetDeviceInfo( + device, + CL_DEVICE_SVM_CAPABILITIES, + sizeof(svmCaps), + &svmCaps, + nullptr ); + + svmCaps |= (deviceCaps != 0) ? CL_DEVICE_SVM_DEVICE_ALLOC_EXP : 0; + svmCaps |= (hostCaps != 0) ? CL_DEVICE_SVM_HOST_ALLOC_EXP : 0; + svmCaps |= (sharedCaps != 0) ? CL_DEVICE_SVM_SHARED_ALLOC_EXP : 0; + + auto ptr = (cl_device_svm_capabilities*)param_value; + return writeParamToMemory( + param_value_size, + svmCaps, + param_value_size_ret, + ptr ); + } + break; case CL_DEVICE_EXTENSIONS: { size_t size = 0; @@ -222,10 +269,9 @@ cl_int CL_API_CALL clGetDeviceInfo_override( if( checkStringForExtension( deviceExtensions.data(), - CL_EXP_NEW_SVM_EXTENSION_NAME ) == false ) + CL_EXP_UNIFIED_SVM_EXTENSION_NAME ) == false ) { std::string newExtensions; - newExtensions += CL_EXP_NEW_SVM_EXTENSION_NAME; std::string oldExtensions(deviceExtensions.data()); @@ -274,7 +320,7 @@ cl_int CL_API_CALL clGetDeviceInfo_override( bool found = false; for( const auto& extension : extensions ) { - if( strcmp(extension.name, CL_EXP_NEW_SVM_EXTENSION_NAME) == 0 ) + if( strcmp(extension.name, CL_EXP_UNIFIED_SVM_EXTENSION_NAME) == 0 ) { found = true; break; @@ -287,7 +333,7 @@ cl_int CL_API_CALL clGetDeviceInfo_override( cl_name_version& extension = extensions.back(); memset(extension.name, 0, CL_NAME_VERSION_MAX_NAME_SIZE); - strcpy(extension.name, CL_EXP_NEW_SVM_EXTENSION_NAME); + strcpy(extension.name, CL_EXP_UNIFIED_SVM_EXTENSION_NAME); extension.version = version_cl_exp_new_svm_extension; @@ -301,8 +347,7 @@ cl_int CL_API_CALL clGetDeviceInfo_override( } } break; - // TODO - case CL_DEVICE_SVM_CAPABILITIES: + // USM aliases - pass through. case CL_DEVICE_HOST_MEM_CAPABILITIES_EXP: case CL_DEVICE_DEVICE_MEM_CAPABILITIES_EXP: case CL_DEVICE_SINGLE_DEVICE_SHARED_MEM_CAPABILITIES_EXP: @@ -347,10 +392,10 @@ cl_int CL_API_CALL clGetPlatformInfo_override( if( checkStringForExtension( platformExtensions.data(), - CL_EXP_NEW_SVM_EXTENSION_NAME ) == false ) + CL_EXP_UNIFIED_SVM_EXTENSION_NAME ) == false ) { std::string newExtensions; - newExtensions += CL_EXP_NEW_SVM_EXTENSION_NAME; + newExtensions += CL_EXP_UNIFIED_SVM_EXTENSION_NAME; std::string oldExtensions(platformExtensions.data()); @@ -399,7 +444,7 @@ cl_int CL_API_CALL clGetPlatformInfo_override( bool found = false; for( const auto& extension : extensions ) { - if( strcmp(extension.name, CL_EXP_NEW_SVM_EXTENSION_NAME) == 0 ) + if( strcmp(extension.name, CL_EXP_UNIFIED_SVM_EXTENSION_NAME) == 0 ) { found = true; break; @@ -412,7 +457,7 @@ cl_int CL_API_CALL clGetPlatformInfo_override( cl_name_version& extension = extensions.back(); memset(extension.name, 0, CL_NAME_VERSION_MAX_NAME_SIZE); - strcpy(extension.name, CL_EXP_NEW_SVM_EXTENSION_NAME); + strcpy(extension.name, CL_EXP_UNIFIED_SVM_EXTENSION_NAME); extension.version = version_cl_exp_new_svm_extension; @@ -457,6 +502,17 @@ cl_int CL_API_CALL clSetKernelArgSVMPointer_override( arg_value); } +void CL_API_CALL clSVMFree_override( + cl_context context, + void* ptr) +{ + if (isUSMPtr(context, ptr)) { + clMemFreeINTEL(context, ptr); + } + + g_pNextDispatch->clSVMFree(context, ptr); +} + cl_int CL_API_CALL clEnqueueSVMMemAdviseEXP_EMU( cl_command_queue command_queue, const void* ptr, @@ -486,7 +542,7 @@ cl_int CL_API_CALL clEnqueueSVMMemcpy_override( { cl_context context = getContext(command_queue); - if (isUSMPtr(context, dst_ptr)) { + if (isUSMPtr(context, dst_ptr) || isUSMPtr(context, src_ptr)) { return clEnqueueMemcpyINTEL( command_queue, blocking_copy, diff --git a/layers/99_svmplusplus/emulate.h b/layers/99_svmplusplus/emulate.h index 7c4c631..f7de9e8 100644 --- a/layers/99_svmplusplus/emulate.h +++ b/layers/99_svmplusplus/emulate.h @@ -7,170 +7,6 @@ #include #include -#include - -// TODO: Move this to a shared header. - -#define CL_EXP_NEW_SVM_EXTENSION_NAME \ - "cl_exp_unified_svm" - -// New-ish types and enums: - -typedef cl_bitfield cl_device_unified_shared_memory_capabilities_exp; // analogous to cl_device_unified_shared_memory_capabilities_intel -typedef cl_properties cl_svm_mem_properties_exp; // analogous to cl_mem_properties_intel -typedef cl_uint cl_svm_mem_info_exp; // analogous to cl_mem_info_intel -typedef cl_uint cl_svm_mem_type_exp; // analogous to cl_unified_shared_memory_type_intel -typedef cl_uint cl_svm_mem_advice_exp; // analogous to cl_mem_advice_intel -typedef cl_bitfield cl_svm_free_flags_exp; // new -typedef cl_properties cl_svm_free_properties_exp; // new - -/* cl_svm_mem_flags */ -#define CL_MEM_SVM_DEVICE_EXP (1 << 16) -#define CL_MEM_SVM_HOST_EXP (1 << 17) -#define CL_MEM_SVM_SHARED_EXP (1 << 18) - -/* cl_device_svm_capabilities */ -// These may not be needed - can be derived from specific device queries! -#define CL_DEVICE_SVM_DEVICE_ALLOC_EXP (1 << 4) -#define CL_DEVICE_SVM_HOST_ALLOC_EXP (1 << 5) -#define CL_DEVICE_SVM_SHARED_ALLOC_EXP (1 << 6) - -/* cl_svm_free_flags_exp */ -#define CL_SVM_FREE_NON_BLOCKING_EXP (1 << 0) -#define CL_SVM_FREE_BLOCKING_EXP (1 << 1) - -/* cl_svm_mem_properties_exp */ -#define CL_SVM_MEM_ASSOCIATED_DEVICE_HANDLE_EXP 0x10100 // note: placeholder! -// consider: CL_SVM_MEM_DEVICE_HANDLE_LIST for cross-device allocations? - -// Aliased types and enums: - -/* cl_device_info - aliases for USM */ -#define CL_DEVICE_HOST_MEM_CAPABILITIES_EXP 0x4190 -#define CL_DEVICE_DEVICE_MEM_CAPABILITIES_EXP 0x4191 -#define CL_DEVICE_SINGLE_DEVICE_SHARED_MEM_CAPABILITIES_EXP 0x4192 -#define CL_DEVICE_CROSS_DEVICE_SHARED_MEM_CAPABILITIES_EXP 0x4193 -#define CL_DEVICE_SHARED_SYSTEM_MEM_CAPABILITIES_EXP 0x4194 - -/* cl_unified_shared_memory_capabilities_intel - bitfield - aliases for USM */ -#define CL_UNIFIED_SHARED_MEMORY_ACCESS_EXP (1 << 0) -#define CL_UNIFIED_SHARED_MEMORY_ATOMIC_ACCESS_EXP (1 << 1) -#define CL_UNIFIED_SHARED_MEMORY_CONCURRENT_ACCESS_EXP (1 << 2) -#define CL_UNIFIED_SHARED_MEMORY_CONCURRENT_ATOMIC_ACCESS_EXP (1 << 3) - -// TODO: should these be cl_mem_svm_flags? -// CL_MEM_ALLOC_WRITE_COMBINED_INTEL (1 << 0) -// CL_MEM_ALLOC_INITIAL_PLACEMENT_DEVICE_INTEL (1 << 1) -// CL_MEM_ALLOC_INITIAL_PLACEMENT_HOST_INTEL (1 << 2) - -/* cl_svm_mem_info_exp */ -#define CL_SVM_MEM_TYPE_EXP 0x419A -#define CL_SVM_MEM_BASE_PTR_EXP 0x419B -#define CL_SVM_MEM_SIZE_EXP 0x419C -#define CL_SVM_MEM_DEVICE_EXP 0x419D - -/* cl_svm_mem_type_exp */ -#define CL_SVM_MEM_TYPE_UNKNOWN_EXP 0x4196 -#define CL_SVM_MEM_TYPE_HOST_EXP 0x4197 -#define CL_SVM_MEM_TYPE_DEVICE_EXP 0x4198 -#define CL_SVM_MEM_TYPE_SHARED_EXP 0x4199 -// TODO: do we need types for SVM buffer, SVM coarse grain buffer, SVM fine grain buffer, ... ? - -/* cl_kernel_exec_info */ -#define CL_KERNEL_EXEC_INFO_INDIRECT_HOST_ACCESS_EXP 0x4200 -#define CL_KERNEL_EXEC_INFO_INDIRECT_DEVICE_ACCESS_EXP 0x4201 -#define CL_KERNEL_EXEC_INFO_INDIRECT_SHARED_ACCESS_EXP 0x4202 -// TODO: do we need indirect access flags for SVM buffer, ... ? - -/* cl_command_type */ -#define CL_COMMAND_MEMADVISE_EXP 0x4207 - -// New functions: - -typedef void* CL_API_CALL -clSVMAllocWithPropertiesEXP_t( - cl_context context, - const cl_svm_mem_properties_exp* properties, - cl_svm_mem_flags flags, - size_t size, - cl_uint alignment, - cl_int* errcode_ret); - -typedef clSVMAllocWithPropertiesEXP_t * -clSVMAllocWithPropertiesEXP_fn ; - -typedef cl_int CL_API_CALL -clSVMFreeWithPropertiesEXP_t( - cl_context context, - const cl_svm_free_properties_exp* properties, - cl_svm_free_flags_exp flags, - void* ptr); - -typedef clSVMFreeWithPropertiesEXP_t * -clSVMFreeWithPropertiesEXP_fn ; - -typedef cl_int CL_API_CALL -clGetSVMMemInfoEXP_t( - const void* ptr, - cl_svm_mem_info_exp param_name, - size_t param_value_size, - void* param_value, - size_t* param_value_size_ret); - -typedef clGetSVMMemInfoEXP_t * -clGetSVMMemInfoEXP_fn ; - -typedef cl_int CL_API_CALL -clEnqueueSVMMemAdviseEXP_t( - cl_command_queue command_queue, - const void* ptr, - size_t size, - cl_svm_mem_advice_exp advice, - cl_uint num_events_in_wait_list, - const cl_event* event_wait_list, - cl_event* event); - -typedef clEnqueueMemAdviseINTEL_t * -clEnqueueMemAdviseINTEL_fn ; - -#if !defined(CL_NO_NON_ICD_DISPATCH_EXTENSION_PROTOTYPES) - -extern CL_API_ENTRY void* CL_API_CALL -clSVMAllocWithPropertiesEXP( - cl_context context, - const cl_svm_mem_properties_exp* properties, - cl_svm_mem_flags flags, - size_t size, - cl_uint alignment, - cl_int* errcode_ret); - -extern CL_API_ENTRY cl_int CL_API_CALL -clSVMFreeWithPropertiesEXP( - cl_context context, - const cl_svm_free_properties_exp* properties, - cl_svm_free_flags_exp flags, - void* ptr); - -extern CL_API_ENTRY cl_int CL_API_CALL -clGetSVMMemInfoEXP( - const void* ptr, - cl_svm_mem_info_exp param_name, - size_t param_value_size, - void* param_value, - size_t* param_value_size_ret); - -extern CL_API_ENTRY cl_int CL_API_CALL -clEnqueueSVMMemAdviseEXP( - cl_command_queue command_queue, - const void* ptr, - size_t size, - cl_svm_mem_advice_exp advice, - cl_uint num_events_in_wait_list, - const cl_event* event_wait_list, - cl_event* event); - -#endif /* !defined(CL_NO_NON_ICD_DISPATCH_EXTENSION_PROTOTYPES) */ - extern const struct _cl_icd_dispatch* g_pNextDispatch; void* CL_API_CALL clSVMAllocWithPropertiesEXP_EMU( @@ -214,6 +50,10 @@ cl_int CL_API_CALL clSetKernelArgSVMPointer_override( cl_uint arg_index, const void* arg_value); +void CL_API_CALL clSVMFree_override( + cl_context context, + void* ptr); + cl_int CL_API_CALL clEnqueueSVMMemAdviseEXP_EMU( cl_command_queue command_queue, const void* ptr, diff --git a/layers/99_svmplusplus/main.cpp b/layers/99_svmplusplus/main.cpp index 71c737e..20e2673 100644 --- a/layers/99_svmplusplus/main.cpp +++ b/layers/99_svmplusplus/main.cpp @@ -56,6 +56,7 @@ static void _init_dispatch() dispatch.clGetExtensionFunctionAddressForPlatform = clGetExtensionFunctionAddressForPlatform_override; dispatch.clGetPlatformInfo = clGetPlatformInfo_override; dispatch.clSetKernelArgSVMPointer = clSetKernelArgSVMPointer_override; + dispatch.clSVMFree = clSVMFree_override; dispatch.clEnqueueSVMMemcpy = clEnqueueSVMMemcpy_override; dispatch.clEnqueueSVMMemFill = clEnqueueSVMMemFill_override; dispatch.clEnqueueSVMMigrateMem = clEnqueueSVMMigrateMem_override; diff --git a/samples/CMakeLists.txt b/samples/CMakeLists.txt index b8419ee..b50d546 100644 --- a/samples/CMakeLists.txt +++ b/samples/CMakeLists.txt @@ -72,6 +72,7 @@ add_subdirectory( opengl ) add_subdirectory( python ) add_subdirectory( vulkan ) add_subdirectory( usm ) +add_subdirectory( usvm ) add_subdirectory( 00_enumopencl ) add_subdirectory( 00_enumopenclpp ) diff --git a/samples/usm/01_usmmeminfo/CMakeLists.txt b/samples/usm/01_usmmeminfo/CMakeLists.txt index 565281b..47c4e8d 100644 --- a/samples/usm/01_usmmeminfo/CMakeLists.txt +++ b/samples/usm/01_usmmeminfo/CMakeLists.txt @@ -20,7 +20,7 @@ add_opencl_sample( TEST - NUMBER 00 + NUMBER 01 TARGET usmmeminfo VERSION 120 CATEGORY usm diff --git a/samples/usvm/00_usvmqueries/CMakeLists.txt b/samples/usvm/00_usvmqueries/CMakeLists.txt new file mode 100644 index 0000000..6d856d1 --- /dev/null +++ b/samples/usvm/00_usvmqueries/CMakeLists.txt @@ -0,0 +1,12 @@ +# Copyright (c) 2023 Ben Ashbaugh +# +# SPDX-License-Identifier: MIT + +add_opencl_sample( + TEST + NUMBER 00 + TARGET usvmqueries + VERSION 300 + CATEGORY usvm + SOURCES main.cpp + LIBS OpenCLExt) diff --git a/samples/usvm/00_usvmqueries/main.cpp b/samples/usvm/00_usvmqueries/main.cpp new file mode 100644 index 0000000..4990858 --- /dev/null +++ b/samples/usvm/00_usvmqueries/main.cpp @@ -0,0 +1,137 @@ +/* +// Copyright (c) 2023 Ben Ashbaugh +// +// SPDX-License-Identifier: MIT +*/ + +#include + +#include + +void PrintSVMCaps( + const char* label, + cl_device_svm_capabilities svmcaps ) +{ + printf("\t%s: %s%s%s%s%s%s%s\n", + label, + ( svmcaps & CL_DEVICE_SVM_COARSE_GRAIN_BUFFER) ? "\n\t\tCL_DEVICE_SVM_COARSE_GRAIN_BUFFER" : "", + ( svmcaps & CL_DEVICE_SVM_FINE_GRAIN_BUFFER ) ? "\n\t\tCL_DEVICE_SVM_FINE_GRAIN_BUFFER" : "", + ( svmcaps & CL_DEVICE_SVM_FINE_GRAIN_SYSTEM ) ? "\n\t\tCL_DEVICE_SVM_FINE_GRAIN_SYSTEM" : "", + ( svmcaps & CL_DEVICE_SVM_ATOMICS ) ? "\n\t\tCL_DEVICE_SVM_ATOMICS" : "", + ( svmcaps & CL_DEVICE_SVM_DEVICE_ALLOC_EXP ) ? "\n\t\tCL_DEVICE_SVM_DEVICE_ALLOC_EXP" : "", + ( svmcaps & CL_DEVICE_SVM_HOST_ALLOC_EXP ) ? "\n\t\tCL_DEVICE_SVM_HOST_ALLOC_EXP" : "", + ( svmcaps & CL_DEVICE_SVM_SHARED_ALLOC_EXP ) ? "\n\t\tCL_DEVICE_SVM_SHARED_ALLOC_EXP" : "" ); +} + +void PrintUSMCaps( + const char* label, + cl_device_unified_shared_memory_capabilities_intel usmcaps ) +{ + printf("\t%s: %s%s%s%s\n", + label, + ( usmcaps & CL_UNIFIED_SHARED_MEMORY_ACCESS_INTEL ) ? "\n\t\tCL_UNIFIED_SHARED_MEMORY_ACCESS_INTEL" : "", + ( usmcaps & CL_UNIFIED_SHARED_MEMORY_ATOMIC_ACCESS_INTEL ) ? "\n\t\tCL_UNIFIED_SHARED_MEMORY_ATOMIC_ACCESS_INTEL" : "", + ( usmcaps & CL_UNIFIED_SHARED_MEMORY_CONCURRENT_ACCESS_INTEL ) ? "\n\t\tCL_UNIFIED_SHARED_MEMORY_CONCURRENT_ACCESS_INTEL" : "", + ( usmcaps & CL_UNIFIED_SHARED_MEMORY_CONCURRENT_ATOMIC_ACCESS_INTEL ) ? "\n\t\tCL_UNIFIED_SHARED_MEMORY_CONCURRENT_ATOMIC_ACCESS_INTEL" : "" ); +} + +int main( + int argc, + char** argv ) +{ + { + popl::OptionParser op("Supported Options"); + + bool printUsage = false; + try { + op.parse(argc, argv); + } catch (std::exception& e) { + fprintf(stderr, "Error: %s\n\n", e.what()); + printUsage = true; + } + + if (printUsage || !op.unknown_options().empty() || !op.non_option_args().empty()) { + fprintf(stderr, + "Usage: usvmqueries [options]\n" + "%s", op.help().c_str()); + return -1; + } + } + + std::vector platforms; + cl::Platform::get(&platforms); + + for( size_t i = 0; i < platforms.size(); i++ ) + { + printf( "Platform[%zu]: %s\n", + i, + platforms[i].getInfo().c_str()); + + std::vector devices; + platforms[i].getDevices(CL_DEVICE_TYPE_ALL, &devices); + + for( size_t d = 0; d < devices.size(); d++ ) + { + printf("\tDevice[%zu]: %s\n", + d, + devices[d].getInfo().c_str()); + + cl_device_svm_capabilities svmcaps = 0; + clGetDeviceInfo( + devices[d](), + CL_DEVICE_SVM_CAPABILITIES, + sizeof(svmcaps), + &svmcaps, + nullptr ); + PrintSVMCaps( "CL_DEVICE_SVM_CAPABILITIES", svmcaps ); + + cl_device_unified_shared_memory_capabilities_intel usmcaps = 0; + + clGetDeviceInfo( + devices[d](), + CL_DEVICE_HOST_MEM_CAPABILITIES_INTEL, + sizeof(usmcaps), + &usmcaps, + nullptr ); + PrintUSMCaps( "CL_DEVICE_HOST_MEM_CAPABILITIES_INTEL", usmcaps ); + + clGetDeviceInfo( + devices[d](), + CL_DEVICE_DEVICE_MEM_CAPABILITIES_INTEL, + sizeof(usmcaps), + &usmcaps, + nullptr ); + PrintUSMCaps( "CL_DEVICE_DEVICE_MEM_CAPABILITIES_INTEL", usmcaps ); + + clGetDeviceInfo( + devices[d](), + CL_DEVICE_SINGLE_DEVICE_SHARED_MEM_CAPABILITIES_INTEL, + sizeof(usmcaps), + &usmcaps, + nullptr ); + PrintUSMCaps( "CL_DEVICE_SINGLE_DEVICE_SHARED_MEM_CAPABILITIES_INTEL", usmcaps ); + + clGetDeviceInfo( + devices[d](), + CL_DEVICE_CROSS_DEVICE_SHARED_MEM_CAPABILITIES_INTEL, + sizeof(usmcaps), + &usmcaps, + nullptr ); + PrintUSMCaps( "CL_DEVICE_CROSS_DEVICE_SHARED_MEM_CAPABILITIES_INTEL", usmcaps ); + + clGetDeviceInfo( + devices[d](), + CL_DEVICE_SHARED_SYSTEM_MEM_CAPABILITIES_INTEL, + sizeof(usmcaps), + &usmcaps, + nullptr ); + PrintUSMCaps( "CL_DEVICE_SHARED_SYSTEM_MEM_CAPABILITIES_INTEL", usmcaps ); + + printf( "\n" ); + } + } + + printf("Cleaning up...\n"); + + return 0; +} \ No newline at end of file diff --git a/samples/usvm/01_usvmmeminfo/CMakeLists.txt b/samples/usvm/01_usvmmeminfo/CMakeLists.txt new file mode 100644 index 0000000..2c377b7 --- /dev/null +++ b/samples/usvm/01_usvmmeminfo/CMakeLists.txt @@ -0,0 +1,12 @@ +# Copyright (c) 2023 Ben Ashbaugh +# +# SPDX-License-Identifier: MIT + +add_opencl_sample( + TEST + NUMBER 01 + TARGET usvmmeminfo + VERSION 300 + CATEGORY usvm + SOURCES main.cpp + LIBS OpenCLExt) diff --git a/samples/usvm/01_usvmmeminfo/main.cpp b/samples/usvm/01_usvmmeminfo/main.cpp new file mode 100644 index 0000000..0bd3eee --- /dev/null +++ b/samples/usvm/01_usvmmeminfo/main.cpp @@ -0,0 +1,336 @@ +/* +// Copyright (c) 2023 Ben Ashbaugh +// +// SPDX-License-Identifier: MIT +*/ + +#include + +#include + +// Each of these functions should eventually move into opencl.hpp: + +static cl_unified_shared_memory_type_intel +getSVM_MEM_INFO_TYPE_EXP( cl::Context& context, const void* ptr ) +{ + cl_unified_shared_memory_type_intel type = 0; + clGetMemAllocInfoINTEL( + context(), + ptr, + CL_SVM_MEM_INFO_TYPE_EXP, + sizeof(type), + &type, + nullptr ); + return type; +} + +static const void* +getSVM_MEM_INFO_BASE_PTR_EXP( cl::Context& context, const void* ptr ) +{ + const void* base = nullptr; + clGetMemAllocInfoINTEL( + context(), + ptr, + CL_SVM_MEM_INFO_BASE_PTR_EXP, + sizeof(base), + &base, + nullptr ); + return base; +} + +static size_t +getSVM_MEM_INFO_SIZE_EXP( cl::Context& context, const void* ptr ) +{ + size_t size = 0; + clGetMemAllocInfoINTEL( + context(), + ptr, + CL_SVM_MEM_INFO_SIZE_EXP, + sizeof(size), + &size, + nullptr ); + return size; +} + +static cl_device_id +getSVM_MEM_INFO_DEVICE_EXP( cl::Context& context, const void* ptr ) +{ + cl_device_id device = 0; + clGetMemAllocInfoINTEL( + context(), + ptr, + CL_SVM_MEM_INFO_DEVICE_EXP, + sizeof(device), + &device, + nullptr ); + return device; +} + +static const char* +usm_type_to_string( cl_unified_shared_memory_type_intel type ) +{ + switch( type ) + { + case CL_SVM_MEM_TYPE_UNKNOWN_EXP: return "CL_SVM_MEM_TYPE_UNKNOWN_EXP"; + case CL_SVM_MEM_TYPE_HOST_EXP: return "CL_SVM_MEM_TYPE_HOST_EXP"; + case CL_SVM_MEM_TYPE_DEVICE_EXP: return "CL_SVM_MEM_TYPE_DEVICE_EXP"; + case CL_SVM_MEM_TYPE_SHARED_EXP: return "CL_SVM_MEM_TYPE_SHARED_EXP"; + default: break; + } + return "***Unknown SVM Type***"; +} + +int main( + int argc, + char** argv ) +{ + int platformIndex = 0; + int deviceIndex = 0; + + { + popl::OptionParser op("Supported Options"); + op.add>("p", "platform", "Platform Index", platformIndex, &platformIndex); + op.add>("d", "device", "Device Index", deviceIndex, &deviceIndex); + bool printUsage = false; + try { + op.parse(argc, argv); + } catch (std::exception& e) { + fprintf(stderr, "Error: %s\n\n", e.what()); + printUsage = true; + } + + if (printUsage || !op.unknown_options().empty() || !op.non_option_args().empty()) { + fprintf(stderr, + "Usage: usmmeminfo [options]\n" + "%s", op.help().c_str()); + return -1; + } + } + + std::vector platforms; + cl::Platform::get(&platforms); + + printf("Running on platform: %s\n", + platforms[platformIndex].getInfo().c_str() ); + + std::vector devices; + platforms[platformIndex].getDevices(CL_DEVICE_TYPE_ALL, &devices); + + printf("Running on device: %s\n", + devices[deviceIndex].getInfo().c_str() ); + + cl::Context context{devices[deviceIndex]}; + + cl_device_unified_shared_memory_capabilities_intel usmcaps = 0; + cl_int errCode; + + errCode = clGetDeviceInfo( + devices[deviceIndex](), + CL_DEVICE_HOST_MEM_CAPABILITIES_INTEL, + sizeof(usmcaps), + &usmcaps, + nullptr ); + if( errCode == CL_SUCCESS && usmcaps != 0 ) + { + printf("\nTesting Host Allocations:\n"); + char* ptr0 = (char*)clHostMemAllocINTEL( + context(), + nullptr, + 16, + 0, + nullptr ); + printf("Allocated Host pointer 0: ptr = %p\n", ptr0); + char* ptr1 = (char*)clHostMemAllocINTEL( + context(), + nullptr, + 16, + 0, + nullptr ); + printf("Allocated Host pointer 1: ptr = %p\n", ptr1); + + cl_unified_shared_memory_type_intel type = 0; + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr1); + printf("Queried base pointer 1: type = %s (%X)\n", usm_type_to_string(type), type); + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr1 + 4); + printf("Queried offset pointer 1: type = %s (%X)\n", usm_type_to_string(type), type); + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr1 + 64); + printf("Queried out of range pointer 1: type = %s (%X)\n", usm_type_to_string(type), type); + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr0); + printf("Queried base pointer 0: type = %s (%X)\n", usm_type_to_string(type), type); + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr0 + 4); + printf("Queried offset pointer 0: type = %s (%X)\n", usm_type_to_string(type), type); + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr0 + 64); + printf("Queried out of range pointer 0: type = %s (%X)\n", usm_type_to_string(type), type); + + const void* base = getSVM_MEM_INFO_BASE_PTR_EXP(context, ptr0 + 4); + printf("Queried offset pointer 0: base = %p\n", base); + + size_t size = getSVM_MEM_INFO_SIZE_EXP(context, ptr0 + 4); + printf("Queried offset pointer 0: size = %u\n", (unsigned)size); + + cl_device_id device = getSVM_MEM_INFO_DEVICE_EXP(context, ptr0 + 4); + printf("Queried offset pointer 0: device = %p\n", device); + + clMemFreeINTEL( + context(), + ptr0 ); + clMemFreeINTEL( + context(), + ptr1 ); + printf("Freed pointers and done!\n"); + } + else + { + printf("\nThis device does not support HOST allocations.\n"); + } + + errCode = clGetDeviceInfo( + devices[deviceIndex](), + CL_DEVICE_DEVICE_MEM_CAPABILITIES_INTEL, + sizeof(usmcaps), + &usmcaps, + nullptr ); + if( errCode == CL_SUCCESS && usmcaps != 0 ) + { + printf("\nTesting Device Allocations:\n"); + printf("Associated Device is: %p (%s)\n", + devices[deviceIndex](), + devices[deviceIndex].getInfo().c_str()); + char* ptr0 = (char*)clDeviceMemAllocINTEL( + context(), + devices[deviceIndex](), + nullptr, + 16, + 0, + nullptr ); + printf("Allocated Device pointer 0: ptr = %p\n", ptr0); + char* ptr1 = (char*)clDeviceMemAllocINTEL( + context(), + devices[deviceIndex](), + nullptr, + 16, + 0, + nullptr ); + printf("Allocated Device pointer 1: ptr = %p\n", ptr1); + + cl_unified_shared_memory_type_intel type = 0; + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr1); + printf("Queried base pointer 1: type = %s (%X)\n", usm_type_to_string(type), type); + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr1 + 4); + printf("Queried offset pointer 1: type = %s (%X)\n", usm_type_to_string(type), type); + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr1 + 64); + printf("Queried out of range pointer 1: type = %s (%X)\n", usm_type_to_string(type), type); + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr0); + printf("Queried base pointer 0: type = %s (%X)\n", usm_type_to_string(type), type); + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr0 + 4); + printf("Queried offset pointer 0: type = %s (%X)\n", usm_type_to_string(type), type); + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr0 + 64); + printf("Queried out of range pointer 0: type = %s (%X)\n", usm_type_to_string(type), type); + + const void* base = getSVM_MEM_INFO_BASE_PTR_EXP(context, ptr0 + 4); + printf("Queried offset pointer 0: base = %p\n", base); + + size_t size = getSVM_MEM_INFO_SIZE_EXP(context, ptr0 + 4); + printf("Queried offset pointer 0: size = %u\n", (unsigned)size); + + cl_device_id device = getSVM_MEM_INFO_DEVICE_EXP(context, ptr0 + 4); + printf("Queried offset pointer 0: device = %p\n", device); + + clMemFreeINTEL( + context(), + ptr0 ); + clMemFreeINTEL( + context(), + ptr1 ); + printf("Freed pointers and done!\n"); + } + else + { + printf("\nThis device does not support DEVICE allocations.\n"); + } + + errCode = clGetDeviceInfo( + devices[deviceIndex](), + CL_DEVICE_SINGLE_DEVICE_SHARED_MEM_CAPABILITIES_INTEL, + sizeof(usmcaps), + &usmcaps, + nullptr ); + if( errCode == CL_SUCCESS && usmcaps != 0 ) + { + printf("\nTesting Shared Allocations:\n"); + printf("Associated Device is: %p (%s)\n", + devices[deviceIndex](), + devices[deviceIndex].getInfo().c_str()); + char* ptr0 = (char*)clSharedMemAllocINTEL( + context(), + devices[deviceIndex](), + nullptr, + 16, + 0, + nullptr ); + printf("Allocated Shared pointer 0: ptr = %p\n", ptr0); + char* ptr1 = (char*)clSharedMemAllocINTEL( + context(), + devices[deviceIndex](), + nullptr, + 16, + 0, + nullptr ); + printf("Allocated Shared pointer 1: ptr = %p\n", ptr1); + + cl_unified_shared_memory_type_intel type = 0; + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr1); + printf("Queried base pointer 1: type = %s (%X)\n", usm_type_to_string(type), type); + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr1 + 4); + printf("Queried offset pointer 1: type = %s (%X)\n", usm_type_to_string(type), type); + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr1 + 64); + printf("Queried out of range pointer 1: type = %s (%X)\n", usm_type_to_string(type), type); + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr0); + printf("Queried base pointer 0: type = %s (%X)\n", usm_type_to_string(type), type); + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr0 + 4); + printf("Queried offset pointer 0: type = %s (%X)\n", usm_type_to_string(type), type); + + type = getSVM_MEM_INFO_TYPE_EXP(context, ptr0 + 64); + printf("Queried out of range pointer 0: type = %s (%X)\n", usm_type_to_string(type), type); + + const void* base = getSVM_MEM_INFO_BASE_PTR_EXP(context, ptr0 + 4); + printf("Queried offset pointer 0: base = %p\n", base); + + size_t size = getSVM_MEM_INFO_SIZE_EXP(context, ptr0 + 4); + printf("Queried offset pointer 0: size = %u\n", (unsigned)size); + + cl_device_id device = getSVM_MEM_INFO_DEVICE_EXP(context, ptr0 + 4); + printf("Queried offset pointer 0: device = %p\n", device); + + clMemFreeINTEL( + context(), + ptr0 ); + clMemFreeINTEL( + context(), + ptr1 ); + printf("Freed pointers and done!\n"); + } + else + { + printf("\nThis device does not support SHARED allocations.\n"); + } + + return 0; +} \ No newline at end of file diff --git a/samples/usvm/100_udmemhelloworld/CMakeLists.txt b/samples/usvm/100_udmemhelloworld/CMakeLists.txt new file mode 100644 index 0000000..5d7b151 --- /dev/null +++ b/samples/usvm/100_udmemhelloworld/CMakeLists.txt @@ -0,0 +1,12 @@ +# Copyright (c) 2023 Ben Ashbaugh +# +# SPDX-License-Identifier: MIT + +add_opencl_sample( + TEST + NUMBER 100 + TARGET udmemhelloworld + VERSION 300 + CATEGORY usvm + SOURCES main.cpp + LIBS OpenCLExt) diff --git a/samples/usvm/100_udmemhelloworld/main.cpp b/samples/usvm/100_udmemhelloworld/main.cpp new file mode 100644 index 0000000..1495809 --- /dev/null +++ b/samples/usvm/100_udmemhelloworld/main.cpp @@ -0,0 +1,178 @@ +/* +// Copyright (c) 2023 Ben Ashbaugh +// +// SPDX-License-Identifier: MIT +*/ + +#include + +#include + +const size_t gwx = 1024*1024; + +static const char kernelString[] = R"CLC( +kernel void CopyBuffer( global uint* dst, global uint* src ) +{ + uint id = get_global_id(0); + dst[id] = src[id]; +} +)CLC"; + +int main( + int argc, + char** argv ) +{ + int platformIndex = 0; + int deviceIndex = 0; + + { + popl::OptionParser op("Supported Options"); + op.add>("p", "platform", "Platform Index", platformIndex, &platformIndex); + op.add>("d", "device", "Device Index", deviceIndex, &deviceIndex); + bool printUsage = false; + try { + op.parse(argc, argv); + } catch (std::exception& e) { + fprintf(stderr, "Error: %s\n\n", e.what()); + printUsage = true; + } + + if (printUsage || !op.unknown_options().empty() || !op.non_option_args().empty()) { + fprintf(stderr, + "Usage: udmemhelloworld [options]\n" + "%s", op.help().c_str()); + return -1; + } + } + + std::vector platforms; + cl::Platform::get(&platforms); + + printf("Running on platform: %s\n", + platforms[platformIndex].getInfo().c_str() ); + + std::vector devices; + platforms[platformIndex].getDevices(CL_DEVICE_TYPE_ALL, &devices); + + printf("Running on device: %s\n", + devices[deviceIndex].getInfo().c_str() ); + + cl::Context context{devices[deviceIndex]}; + cl::CommandQueue commandQueue = cl::CommandQueue{context, devices[deviceIndex]}; + + cl::Program program{ context, kernelString }; + program.build(); + cl::Kernel kernel = cl::Kernel{ program, "CopyBuffer" }; + + cl_uint* h_buf = new cl_uint[gwx]; + + const cl_svm_mem_properties_exp props[] = { + CL_SVM_MEM_ASSOCIATED_DEVICE_HANDLE_EXP, (cl_svm_mem_properties_exp)devices[deviceIndex](), + 0, + }; + cl_uint* d_src = (cl_uint*)clSVMAllocWithPropertiesEXP( + context(), + props, + CL_MEM_SVM_DEVICE_EXP, + gwx * sizeof(cl_uint), + 0, + nullptr ); + cl_uint* d_dst = (cl_uint*)clSVMAllocWithPropertiesEXP( + context(), + props, + CL_MEM_SVM_DEVICE_EXP, + gwx * sizeof(cl_uint), + 0, + nullptr ); + + if( h_buf && d_src && d_dst ) + { + // initialization + { + for( size_t i = 0; i < gwx; i++ ) + { + h_buf[i] = (cl_uint)(i); + } + + clEnqueueSVMMemcpy( + commandQueue(), + CL_TRUE, + d_src, + h_buf, + gwx * sizeof(cl_uint), + 0, + nullptr, + nullptr ); + + memset( h_buf, 0, gwx * sizeof(cl_uint) ); + } + + // execution + kernel.setArg(0, d_dst); + kernel.setArg(1, d_src); + commandQueue.enqueueNDRangeKernel( + kernel, + cl::NullRange, + cl::NDRange{gwx} ); + + // verification + { + clEnqueueSVMMemcpy( + commandQueue(), + CL_TRUE, + h_buf, + d_dst, + gwx * sizeof(cl_uint), + 0, + nullptr, + nullptr ); + + unsigned int mismatches = 0; + + for( size_t i = 0; i < gwx; i++ ) + { + if( h_buf[i] != i ) + { + if( mismatches < 16 ) + { + fprintf(stderr, "MisMatch! dst[%d] == %08X, want %08X\n", + (unsigned int)i, + h_buf[i], + (unsigned int)i ); + } + mismatches++; + } + } + + if( mismatches ) + { + fprintf(stderr, "Error: Found %d mismatches / %d values!!!\n", + mismatches, + (unsigned int)gwx ); + } + else + { + printf("Success.\n"); + } + } + } + else + { + printf("Allocation failed - does this device support Unified Shared Memory?\n"); + } + + printf("Cleaning up...\n"); + + delete [] h_buf; + + clSVMFree( + context(), + d_src ); + clSVMFreeWithPropertiesEXP( + context(), + nullptr, + CL_SVM_FREE_BLOCKING_EXP, + d_dst ); + + return 0; +} diff --git a/samples/usvm/101_udmemlinkedlist/CMakeLists.txt b/samples/usvm/101_udmemlinkedlist/CMakeLists.txt new file mode 100644 index 0000000..c99cc7f --- /dev/null +++ b/samples/usvm/101_udmemlinkedlist/CMakeLists.txt @@ -0,0 +1,12 @@ +# Copyright (c) 2023 Ben Ashbaugh +# +# SPDX-License-Identifier: MIT + +add_opencl_sample( + TEST + NUMBER 101 + TARGET udmemlinkedlist + VERSION 300 + CATEGORY usvm + SOURCES main.cpp + LIBS OpenCLExt) diff --git a/samples/usvm/101_udmemlinkedlist/main.cpp b/samples/usvm/101_udmemlinkedlist/main.cpp new file mode 100644 index 0000000..0cb8f16 --- /dev/null +++ b/samples/usvm/101_udmemlinkedlist/main.cpp @@ -0,0 +1,255 @@ +/* +// Copyright (c) 2023 Ben Ashbaugh +// +// SPDX-License-Identifier: MIT +*/ + +#include + +#include + +cl::CommandQueue commandQueue; +cl::Kernel kernel; + +cl_uint numNodes = 4; + +struct Node { + Node() : + pNext( nullptr ), + Num( 0xDEADBEEF ) {} + + Node* pNext; + cl_uint Num; +}; + +Node* d_head; + +static const char kernelString[] = R"CLC( +struct Node { + global struct Node* pNext; + uint Num; +}; +kernel void WalkLinkedList( global struct Node* pHead ) +{ + uint count = 0; + while( pHead ) + { + ++count; + pHead->Num = pHead->Num * 2 + 1; + pHead = pHead->pNext; + } +} +)CLC"; + +static void init( cl::Context& context, cl::Device& device ) +{ + const cl_svm_mem_properties_exp props[] = { + CL_SVM_MEM_ASSOCIATED_DEVICE_HANDLE_EXP, (cl_svm_mem_properties_exp)device(), + 0, + }; + + Node* d_cur = nullptr; + Node h_cur; + + for( cl_uint i = 0; i < numNodes; i++ ) + { + if( i == 0 ) + { + d_head = (Node*)clSVMAllocWithPropertiesEXP( + context(), + props, + CL_MEM_SVM_DEVICE_EXP, + sizeof(Node), + 0, + nullptr ); + d_cur = d_head; + } + + if( d_cur != nullptr ) + { + h_cur.Num = i * 2; + + if( i != numNodes - 1 ) + { + h_cur.pNext = (Node*)clSVMAllocWithPropertiesEXP( + context(), + props, + CL_MEM_SVM_DEVICE_EXP, + sizeof(Node), + 0, + nullptr ); + } + else + { + h_cur.pNext = nullptr; + } + + clEnqueueSVMMemcpy( + commandQueue(), + CL_TRUE, + d_cur, + &h_cur, + sizeof(Node), + 0, + nullptr, + nullptr ); + + d_cur = h_cur.pNext; + } + } +} + +static void go() +{ + kernel.setArg(0, d_head); + commandQueue.enqueueNDRangeKernel( + kernel, + cl::NullRange, + cl::NDRange{1} ); +} + +static void checkResults() +{ + const Node* d_cur = d_head; + Node h_cur; + + unsigned int mismatches = 0; + for( cl_uint i = 0; i < numNodes; i++ ) + { + clEnqueueSVMMemcpy( + commandQueue(), + CL_TRUE, + &h_cur, + d_cur, + sizeof(Node), + 0, + nullptr, + nullptr ); + + const cl_uint want = i * 4 + 1; + if( h_cur.Num != want ) + { + if( mismatches < 16 ) + { + fprintf(stderr, "MisMatch at node %u! got %08X, want %08X\n", + i, + h_cur.Num, + want ); + } + mismatches++; + } + + d_cur = h_cur.pNext; + } + + if( mismatches ) + { + fprintf(stderr, "Error: Found %d mismatches / %d values!!!\n", + mismatches, + numNodes ); + } + else + { + printf("Success.\n"); + } +} + +void cleanup( cl::Context& context ) +{ + Node* d_cur = d_head; + Node h_cur; + + while( d_cur != nullptr ) + { + clEnqueueSVMMemcpy( + commandQueue(), + CL_TRUE, + &h_cur, + d_cur, + sizeof(Node), + 0, + nullptr, + nullptr ); + + clSVMFree( + context(), + d_cur ); + + d_cur = h_cur.pNext; + } + + d_head = nullptr; +} + +int main( + int argc, + char** argv ) +{ + int platformIndex = 0; + int deviceIndex = 0; + + { + popl::OptionParser op("Supported Options"); + op.add>("p", "platform", "Platform Index", platformIndex, &platformIndex); + op.add>("d", "device", "Device Index", deviceIndex, &deviceIndex); + op.add>("n", "nodes", "Number of Linked List Nodes", numNodes, &numNodes); + + bool printUsage = false; + try { + op.parse(argc, argv); + } catch (std::exception& e) { + fprintf(stderr, "Error: %s\n\n", e.what()); + printUsage = true; + } + + if (printUsage || !op.unknown_options().empty() || !op.non_option_args().empty()) { + fprintf(stderr, + "Usage: udmemlinkedlist [options]\n" + "%s", op.help().c_str()); + return -1; + } + } + + std::vector platforms; + cl::Platform::get(&platforms); + + printf("Running on platform: %s\n", + platforms[platformIndex].getInfo().c_str() ); + + std::vector devices; + platforms[platformIndex].getDevices(CL_DEVICE_TYPE_ALL, &devices); + + printf("Running on device: %s\n", + devices[deviceIndex].getInfo().c_str() ); + + cl::Context context{devices[deviceIndex]}; + commandQueue = cl::CommandQueue{context, devices[deviceIndex]}; + + cl::Program program{ context, kernelString }; + program.build(); +#if 0 + for( auto& device : program.getInfo() ) + { + printf("Program build log for device %s:\n", + device.getInfo().c_str() ); + printf("%s\n", + program.getBuildInfo(device).c_str() ); + } +#endif + kernel = cl::Kernel{ program, "WalkLinkedList" }; + cl_bool enableIndirectAccess = CL_TRUE; + clSetKernelExecInfo( + kernel(), + CL_KERNEL_EXEC_INFO_INDIRECT_DEVICE_ACCESS_EXP, + sizeof(enableIndirectAccess), + &enableIndirectAccess ); + + init( context, devices[deviceIndex] ); + go(); + checkResults(); + + printf("Cleaning up...\n"); + cleanup( context ); + + return 0; +} \ No newline at end of file diff --git a/samples/usvm/200_uhmemhelloworld/CMakeLists.txt b/samples/usvm/200_uhmemhelloworld/CMakeLists.txt new file mode 100644 index 0000000..5f5e989 --- /dev/null +++ b/samples/usvm/200_uhmemhelloworld/CMakeLists.txt @@ -0,0 +1,12 @@ +# Copyright (c) 2023 Ben Ashbaugh +# +# SPDX-License-Identifier: MIT + +add_opencl_sample( + TEST + NUMBER 200 + TARGET uhmemhelloworld + VERSION 300 + CATEGORY usvm + SOURCES main.cpp + LIBS OpenCLExt) diff --git a/samples/usvm/200_uhmemhelloworld/main.cpp b/samples/usvm/200_uhmemhelloworld/main.cpp new file mode 100644 index 0000000..f15ba12 --- /dev/null +++ b/samples/usvm/200_uhmemhelloworld/main.cpp @@ -0,0 +1,152 @@ +/* +// Copyright (c) 2023 Ben Ashbaugh +// +// SPDX-License-Identifier: MIT +*/ + +#include + +#include + +const size_t gwx = 1024*1024; + +static const char kernelString[] = R"CLC( +kernel void CopyBuffer( global uint* dst, global uint* src ) +{ + uint id = get_global_id(0); + dst[id] = src[id]; +} +)CLC"; + +int main( + int argc, + char** argv ) +{ + int platformIndex = 0; + int deviceIndex = 0; + + { + popl::OptionParser op("Supported Options"); + op.add>("p", "platform", "Platform Index", platformIndex, &platformIndex); + op.add>("d", "device", "Device Index", deviceIndex, &deviceIndex); + bool printUsage = false; + try { + op.parse(argc, argv); + } catch (std::exception& e) { + fprintf(stderr, "Error: %s\n\n", e.what()); + printUsage = true; + } + + if (printUsage || !op.unknown_options().empty() || !op.non_option_args().empty()) { + fprintf(stderr, + "Usage: uhmemhelloworld [options]\n" + "%s", op.help().c_str()); + return -1; + } + } + + std::vector platforms; + cl::Platform::get(&platforms); + + printf("Running on platform: %s\n", + platforms[platformIndex].getInfo().c_str() ); + + std::vector devices; + platforms[platformIndex].getDevices(CL_DEVICE_TYPE_ALL, &devices); + + printf("Running on device: %s\n", + devices[deviceIndex].getInfo().c_str() ); + + cl::Context context{devices[deviceIndex]}; + cl::CommandQueue commandQueue = cl::CommandQueue{context, devices[deviceIndex]}; + + cl::Program program{ context, kernelString }; + program.build(); + cl::Kernel kernel = cl::Kernel{ program, "CopyBuffer" }; + + cl_uint* h_src = (cl_uint*)clSVMAllocWithPropertiesEXP( + context(), + nullptr, + CL_MEM_SVM_HOST_EXP, + gwx * sizeof(cl_uint), + 0, + nullptr ); + cl_uint* h_dst = (cl_uint*)clSVMAllocWithPropertiesEXP( + context(), + nullptr, + CL_MEM_SVM_HOST_EXP, + gwx * sizeof(cl_uint), + 0, + nullptr ); + + if( h_src && h_dst ) + { + // initialization + { + for( size_t i = 0; i < gwx; i++ ) + { + h_src[i] = (cl_uint)(i); + } + + memset( h_dst, 0, gwx * sizeof(cl_uint) ); + } + + // execution + kernel.setArg(0, h_dst); + kernel.setArg(1, h_src); + commandQueue.enqueueNDRangeKernel( + kernel, + cl::NullRange, + cl::NDRange{gwx} ); + + // verification + { + commandQueue.finish(); + + unsigned int mismatches = 0; + + for( size_t i = 0; i < gwx; i++ ) + { + if( h_dst[i] != i ) + { + if( mismatches < 16 ) + { + fprintf(stderr, "MisMatch! dst[%d] == %08X, want %08X\n", + (unsigned int)i, + h_dst[i], + (unsigned int)i ); + } + mismatches++; + } + } + + if( mismatches ) + { + fprintf(stderr, "Error: Found %d mismatches / %d values!!!\n", + mismatches, + (unsigned int)gwx ); + } + else + { + printf("Success.\n"); + } + } + } + else + { + printf("Allocation failed - does this device support Unified Shared Memory?\n"); + } + + printf("Cleaning up...\n"); + + clSVMFree( + context(), + h_src ); + clSVMFreeWithPropertiesEXP( + context(), + nullptr, + CL_SVM_FREE_BLOCKING_EXP, + h_dst ); + + return 0; +} diff --git a/samples/usvm/201_uhmemlinkedlist/CMakeLists.txt b/samples/usvm/201_uhmemlinkedlist/CMakeLists.txt new file mode 100644 index 0000000..5f3a797 --- /dev/null +++ b/samples/usvm/201_uhmemlinkedlist/CMakeLists.txt @@ -0,0 +1,12 @@ +# Copyright (c) 2023 Ben Ashbaugh +# +# SPDX-License-Identifier: MIT + +add_opencl_sample( + TEST + NUMBER 201 + TARGET uhmemlinkedlist + VERSION 300 + CATEGORY usvm + SOURCES main.cpp + LIBS OpenCLExt) diff --git a/samples/usvm/201_uhmemlinkedlist/main.cpp b/samples/usvm/201_uhmemlinkedlist/main.cpp new file mode 100644 index 0000000..ea58267 --- /dev/null +++ b/samples/usvm/201_uhmemlinkedlist/main.cpp @@ -0,0 +1,232 @@ +/* +// Copyright (c) 2023 Ben Ashbaugh +// +// SPDX-License-Identifier: MIT +*/ + +#include + +#include + +cl::CommandQueue commandQueue; +cl::Kernel kernel; + +cl_uint numNodes = 4; + +struct Node { + Node() : + pNext( nullptr ), + Num( 0xDEADBEEF ) {} + + Node* pNext; + cl_uint Num; +}; + +Node* h_head; + +static const char kernelString[] = R"CLC( +struct Node { + global struct Node* pNext; + uint Num; +}; +kernel void WalkLinkedList( global struct Node* pHead ) +{ + uint count = 0; + while( pHead ) + { + ++count; + pHead->Num = pHead->Num * 2 + 1; + pHead = pHead->pNext; + } +} +)CLC"; + +static void init( cl::Context& context, cl::Device& device ) +{ + Node* h_cur = nullptr; + + for( cl_uint i = 0; i < numNodes; i++ ) + { + if( i == 0 ) + { + h_head = (Node*)clSVMAllocWithPropertiesEXP( + context(), + nullptr, + CL_MEM_SVM_HOST_EXP, + sizeof(Node), + 0, + nullptr ); + h_cur = h_head; + } + + if( h_cur != nullptr ) + { + h_cur->Num = i * 2; + + if( i != numNodes - 1 ) + { + h_cur->pNext = (Node*)clSVMAllocWithPropertiesEXP( + context(), + nullptr, + CL_MEM_SVM_HOST_EXP, + sizeof(Node), + 0, + nullptr ); + } + else + { + h_cur->pNext = nullptr; + } + + h_cur = h_cur->pNext; + } + } +} + +static void go() +{ + kernel.setArg(0, h_head); + commandQueue.enqueueNDRangeKernel( + kernel, + cl::NullRange, + cl::NDRange{1} ); +} + +static void checkResults() +{ + commandQueue.finish(); + + const Node* h_cur = h_head; + + unsigned int mismatches = 0; + for( cl_uint i = 0; i < numNodes; i++ ) + { + const cl_uint want = i * 4 + 1; + if( h_cur == nullptr ) + { + if( mismatches < 16 ) + { + fprintf(stderr, "Node %u is NULL!\n", i); + } + mismatches++; + } + else + { + if( h_cur->Num != want ) + { + if( mismatches < 16 ) + { + fprintf(stderr, "MisMatch at node %u! got %08X, want %08X\n", + i, + h_cur->Num, + want ); + } + mismatches++; + } + + h_cur = h_cur->pNext; + } + } + + if( mismatches ) + { + fprintf(stderr, "Error: Found %d mismatches / %d values!!!\n", + mismatches, + numNodes ); + } + else + { + printf("Success.\n"); + } +} + +void cleanup( cl::Context& context ) +{ + Node* h_cur = h_head; + + while( h_cur != nullptr ) + { + Node* h_next = h_cur->pNext; + + clSVMFree( + context(), + h_cur ); + + h_cur = h_next; + } + + h_head = nullptr; +} + +int main( + int argc, + char** argv ) +{ + int platformIndex = 0; + int deviceIndex = 0; + + { + popl::OptionParser op("Supported Options"); + op.add>("p", "platform", "Platform Index", platformIndex, &platformIndex); + op.add>("d", "device", "Device Index", deviceIndex, &deviceIndex); + op.add>("n", "nodes", "Number of Linked List Nodes", numNodes, &numNodes); + + bool printUsage = false; + try { + op.parse(argc, argv); + } catch (std::exception& e) { + fprintf(stderr, "Error: %s\n\n", e.what()); + printUsage = true; + } + + if (printUsage || !op.unknown_options().empty() || !op.non_option_args().empty()) { + fprintf(stderr, + "Usage: hmemlinkedlist [options]\n" + "%s", op.help().c_str()); + return -1; + } + } + + std::vector platforms; + cl::Platform::get(&platforms); + + printf("Running on platform: %s\n", + platforms[platformIndex].getInfo().c_str() ); + + std::vector devices; + platforms[platformIndex].getDevices(CL_DEVICE_TYPE_ALL, &devices); + + printf("Running on device: %s\n", + devices[deviceIndex].getInfo().c_str() ); + + cl::Context context{devices[deviceIndex]}; + commandQueue = cl::CommandQueue{context, devices[deviceIndex]}; + + cl::Program program{ context, kernelString }; + program.build(); +#if 0 + for( auto& device : program.getInfo() ) + { + printf("Program build log for device %s:\n", + device.getInfo().c_str() ); + printf("%s\n", + program.getBuildInfo(device).c_str() ); + } +#endif + kernel = cl::Kernel{ program, "WalkLinkedList" }; + cl_bool enableIndirectAccess = CL_TRUE; + clSetKernelExecInfo( + kernel(), + CL_KERNEL_EXEC_INFO_INDIRECT_HOST_ACCESS_EXP, + sizeof(enableIndirectAccess), + &enableIndirectAccess ); + + init( context, devices[deviceIndex] ); + go(); + checkResults(); + + printf("Cleaning up...\n"); + cleanup( context ); + + return 0; +} \ No newline at end of file diff --git a/samples/usvm/300_usmemhelloworld/CMakeLists.txt b/samples/usvm/300_usmemhelloworld/CMakeLists.txt new file mode 100644 index 0000000..75c7d77 --- /dev/null +++ b/samples/usvm/300_usmemhelloworld/CMakeLists.txt @@ -0,0 +1,12 @@ +# Copyright (c) 2023 Ben Ashbaugh +# +# SPDX-License-Identifier: MIT + +add_opencl_sample( + TEST + NUMBER 300 + TARGET usmemhelloworld + VERSION 300 + CATEGORY usvm + SOURCES main.cpp + LIBS OpenCLExt) diff --git a/samples/usvm/300_usmemhelloworld/main.cpp b/samples/usvm/300_usmemhelloworld/main.cpp new file mode 100644 index 0000000..8c76770 --- /dev/null +++ b/samples/usvm/300_usmemhelloworld/main.cpp @@ -0,0 +1,156 @@ +/* +// Copyright (c) 2023 Ben Ashbaugh +// +// SPDX-License-Identifier: MIT +*/ + +#include + +#include + +const size_t gwx = 1024*1024; + +static const char kernelString[] = R"CLC( +kernel void CopyBuffer( global uint* dst, global uint* src ) +{ + uint id = get_global_id(0); + dst[id] = src[id]; +} +)CLC"; + +int main( + int argc, + char** argv ) +{ + int platformIndex = 0; + int deviceIndex = 0; + + { + popl::OptionParser op("Supported Options"); + op.add>("p", "platform", "Platform Index", platformIndex, &platformIndex); + op.add>("d", "device", "Device Index", deviceIndex, &deviceIndex); + bool printUsage = false; + try { + op.parse(argc, argv); + } catch (std::exception& e) { + fprintf(stderr, "Error: %s\n\n", e.what()); + printUsage = true; + } + + if (printUsage || !op.unknown_options().empty() || !op.non_option_args().empty()) { + fprintf(stderr, + "Usage: usmemhelloworld [options]\n" + "%s", op.help().c_str()); + return -1; + } + } + + std::vector platforms; + cl::Platform::get(&platforms); + + printf("Running on platform: %s\n", + platforms[platformIndex].getInfo().c_str() ); + + std::vector devices; + platforms[platformIndex].getDevices(CL_DEVICE_TYPE_ALL, &devices); + + printf("Running on device: %s\n", + devices[deviceIndex].getInfo().c_str() ); + + cl::Context context{devices[deviceIndex]}; + cl::CommandQueue commandQueue = cl::CommandQueue{context, devices[deviceIndex]}; + + cl::Program program{ context, kernelString }; + program.build(); + cl::Kernel kernel = cl::Kernel{ program, "CopyBuffer" }; + + const cl_svm_mem_properties_exp props[] = { + CL_SVM_MEM_ASSOCIATED_DEVICE_HANDLE_EXP, (cl_svm_mem_properties_exp)devices[deviceIndex](), + 0, + }; + cl_uint* s_src = (cl_uint*)clSVMAllocWithPropertiesEXP( + context(), + props, + CL_MEM_SVM_SHARED_EXP, + gwx * sizeof(cl_uint), + 0, + nullptr ); + cl_uint* s_dst = (cl_uint*)clSVMAllocWithPropertiesEXP( + context(), + props, + CL_MEM_SVM_SHARED_EXP, + gwx * sizeof(cl_uint), + 0, + nullptr ); + + if( s_src && s_dst ) + { + // initialization + { + for( size_t i = 0; i < gwx; i++ ) + { + s_src[i] = (cl_uint)(i); + } + + memset( s_dst, 0, gwx * sizeof(cl_uint) ); + } + + // execution + kernel.setArg(0, s_dst); + kernel.setArg(1, s_src); + commandQueue.enqueueNDRangeKernel( + kernel, + cl::NullRange, + cl::NDRange{gwx} ); + + // verification + { + commandQueue.finish(); + + unsigned int mismatches = 0; + + for( size_t i = 0; i < gwx; i++ ) + { + if( s_dst[i] != i ) + { + if( mismatches < 16 ) + { + fprintf(stderr, "MisMatch! dst[%d] == %08X, want %08X\n", + (unsigned int)i, + s_dst[i], + (unsigned int)i ); + } + mismatches++; + } + } + + if( mismatches ) + { + fprintf(stderr, "Error: Found %d mismatches / %d values!!!\n", + mismatches, + (unsigned int)gwx ); + } + else + { + printf("Success.\n"); + } + } + } + else + { + printf("Allocation failed - does this device support Unified Shared Memory?\n"); + } + + printf("Cleaning up...\n"); + + clSVMFree( + context(), + s_src ); + clSVMFreeWithPropertiesEXP( + context(), + nullptr, + CL_SVM_FREE_BLOCKING_EXP, + s_dst ); + + return 0; +} diff --git a/samples/usvm/301_usmemlinkedlist/CMakeLists.txt b/samples/usvm/301_usmemlinkedlist/CMakeLists.txt new file mode 100644 index 0000000..0215cdd --- /dev/null +++ b/samples/usvm/301_usmemlinkedlist/CMakeLists.txt @@ -0,0 +1,12 @@ +# Copyright (c) 2023 Ben Ashbaugh +# +# SPDX-License-Identifier: MIT + +add_opencl_sample( + TEST + NUMBER 301 + TARGET usmemlinkedlist + VERSION 200 + CATEGORY usvm + SOURCES main.cpp + LIBS OpenCLExt) diff --git a/samples/usvm/301_usmemlinkedlist/main.cpp b/samples/usvm/301_usmemlinkedlist/main.cpp new file mode 100644 index 0000000..f4fc0fe --- /dev/null +++ b/samples/usvm/301_usmemlinkedlist/main.cpp @@ -0,0 +1,237 @@ +/* +// Copyright (c) 2023 Ben Ashbaugh +// +// SPDX-License-Identifier: MIT +*/ + +#include + +#include + +cl::CommandQueue commandQueue; +cl::Kernel kernel; + +cl_uint numNodes = 4; + +struct Node { + Node() : + pNext( nullptr ), + Num( 0xDEADBEEF ) {} + + Node* pNext; + cl_uint Num; +}; + +Node* s_head; + +static const char kernelString[] = R"CLC( +struct Node { + global struct Node* pNext; + uint Num; +}; +kernel void WalkLinkedList( global struct Node* pHead ) +{ + uint count = 0; + while( pHead ) + { + ++count; + pHead->Num = pHead->Num * 2 + 1; + pHead = pHead->pNext; + } +} +)CLC"; + +static void init( cl::Context& context, cl::Device& device ) +{ + const cl_svm_mem_properties_exp props[] = { + CL_SVM_MEM_ASSOCIATED_DEVICE_HANDLE_EXP, (cl_svm_mem_properties_exp)device(), + 0, + }; + + Node* s_cur = nullptr; + + for( cl_uint i = 0; i < numNodes; i++ ) + { + if( i == 0 ) + { + s_head = (Node*)clSVMAllocWithPropertiesEXP( + context(), + props, + CL_MEM_SVM_SHARED_EXP, + sizeof(Node), + 0, + nullptr ); + s_cur = s_head; + } + + if( s_cur != nullptr ) + { + s_cur->Num = i * 2; + + if( i != numNodes - 1 ) + { + s_cur->pNext = (Node*)clSVMAllocWithPropertiesEXP( + context(), + props, + CL_MEM_SVM_SHARED_EXP, + sizeof(Node), + 0, + nullptr ); + } + else + { + s_cur->pNext = nullptr; + } + + s_cur = s_cur->pNext; + } + } +} + +static void go() +{ + kernel.setArg(0, s_head); + commandQueue.enqueueNDRangeKernel( + kernel, + cl::NullRange, + cl::NDRange{1} ); +} + +static void checkResults() +{ + commandQueue.finish(); + + const Node* s_cur = s_head; + + unsigned int mismatches = 0; + for( cl_uint i = 0; i < numNodes; i++ ) + { + const cl_uint want = i * 4 + 1; + if( s_cur == nullptr ) + { + if( mismatches < 16 ) + { + fprintf(stderr, "Node %u is NULL!\n", i); + } + mismatches++; + } + else + { + if( s_cur->Num != want ) + { + if( mismatches < 16 ) + { + fprintf(stderr, "MisMatch at node %u! got %08X, want %08X\n", + i, + s_cur->Num, + want ); + } + mismatches++; + } + + s_cur = s_cur->pNext; + } + } + + if( mismatches ) + { + fprintf(stderr, "Error: Found %d mismatches / %d values!!!\n", + mismatches, + numNodes ); + } + else + { + printf("Success.\n"); + } +} + +void cleanup( cl::Context& context ) +{ + Node* s_cur = s_head; + + while( s_cur != nullptr ) + { + Node* s_next = s_cur->pNext; + + clSVMFree( + context(), + s_cur ); + + s_cur = s_next; + } + + s_head = nullptr; +} + +int main( + int argc, + char** argv ) +{ + int platformIndex = 0; + int deviceIndex = 0; + + { + popl::OptionParser op("Supported Options"); + op.add>("p", "platform", "Platform Index", platformIndex, &platformIndex); + op.add>("d", "device", "Device Index", deviceIndex, &deviceIndex); + op.add>("n", "nodes", "Number of Linked List Nodes", numNodes, &numNodes); + + bool printUsage = false; + try { + op.parse(argc, argv); + } catch (std::exception& e) { + fprintf(stderr, "Error: %s\n\n", e.what()); + printUsage = true; + } + + if (printUsage || !op.unknown_options().empty() || !op.non_option_args().empty()) { + fprintf(stderr, + "Usage: usmemlinkedlist [options]\n" + "%s", op.help().c_str()); + return -1; + } + } + + std::vector platforms; + cl::Platform::get(&platforms); + + printf("Running on platform: %s\n", + platforms[platformIndex].getInfo().c_str() ); + + std::vector devices; + platforms[platformIndex].getDevices(CL_DEVICE_TYPE_ALL, &devices); + + printf("Running on device: %s\n", + devices[deviceIndex].getInfo().c_str() ); + + cl::Context context{devices[deviceIndex]}; + commandQueue = cl::CommandQueue{context, devices[deviceIndex]}; + + cl::Program program{ context, kernelString }; + program.build(); +#if 0 + for( auto& device : program.getInfo() ) + { + printf("Program build log for device %s:\n", + device.getInfo().c_str() ); + printf("%s\n", + program.getBuildInfo(device).c_str() ); + } +#endif + kernel = cl::Kernel{ program, "WalkLinkedList" }; + cl_bool enableIndirectAccess = CL_TRUE; + clSetKernelExecInfo( + kernel(), + CL_KERNEL_EXEC_INFO_INDIRECT_SHARED_ACCESS_EXP, + sizeof(enableIndirectAccess), + &enableIndirectAccess ); + + init( context, devices[deviceIndex] ); + go(); + checkResults(); + + printf("Cleaning up...\n"); + cleanup( context ); + + return 0; +} \ No newline at end of file diff --git a/samples/usvm/CMakeLists.txt b/samples/usvm/CMakeLists.txt new file mode 100644 index 0000000..c4e0bfa --- /dev/null +++ b/samples/usvm/CMakeLists.txt @@ -0,0 +1,23 @@ +# Copyright (c) 2023 Ben Ashbaugh +# +# SPDX-License-Identifier: MIT + +set(BUILD_USVM_SAMPLES TRUE) +if(NOT TARGET OpenCLExt) + message(STATUS "Skipping USM Samples - OpenCL Extension Loader is not found.") + set(BUILD_USVM_SAMPLES FALSE) +endif() + +if(BUILD_USVM_SAMPLES) + add_subdirectory( 00_usvmqueries ) + add_subdirectory( 01_usvmmeminfo ) + + add_subdirectory( 100_udmemhelloworld ) + add_subdirectory( 101_udmemlinkedlist ) + + add_subdirectory( 200_uhmemhelloworld ) + add_subdirectory( 201_uhmemlinkedlist ) + + add_subdirectory( 300_usmemhelloworld ) + add_subdirectory( 301_usmemlinkedlist ) +endif() diff --git a/xml/cl.xml b/xml/cl.xml new file mode 100644 index 0000000..e7c67a0 --- /dev/null +++ b/xml/cl.xml @@ -0,0 +1,7514 @@ + + + +Copyright 2013-2023 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. + +---- Exceptions to the Apache 2.0 License: ---- + +As an exception, if you use this Software to generate code and portions of +this Software are embedded into the generated code as a result, you may +redistribute such product without providing attribution as would otherwise +be required by Sections 4(a), 4(b) and 4(d) of the License. + +In addition, if you combine or link code generated by this Software with +software that is licensed under the GPLv2 or the LGPL v2.0 or 2.1 +("`Combined Software`") and if a court of competent jurisdiction determines +that the patent provision (Section 3), the indemnity provision (Section 9) +or other Section of the License conflicts with the conditions of the +applicable GPL or LGPL license, you may retroactively and prospectively +choose to deem waived or otherwise exclude such Section(s) of the License, +but only in their entirety and only with respect to the Combined Software. + + + +This file, cl.xml, is the OpenCL API Registry. It is a critically important +and normative part of the OpenCL API Specification, including a canonical +machine-readable definition of the API, parameter and member validation +language incorporated into the Specification and reference pages, and other +material which is registered by Khronos. + +The authoritative public version of cl.xml is maintained in the `main` +branch of the Khronos OpenCL-Docs GitHub repository. The authoritative +private version is maintained in the master branch of the member gitlab +server's OpenCL/api-docs repository. + + + + These are dependencies CL types require to be declared legally + #include <stdint.h> + + Windows headers and types + #include <d3d10.h> + #include <d3d11.h> + #include <d3d9.h> + #include <d3d9types.h> + #include <dxgiformat.h> + #include <dxvahd.h> + #include <wtypes.h> + + + + + + + + + + + + + Non-CL headers and types + #include <va/va.h> + + + + CL headers and types + #include <CL/cl.h> + #include <CL/cl_platform.h> + #include <CL/cl_gl.h> + #include <CL/cl_version.h> + + Basic C types, pulled in via platform header in some cases + + + + + + + + + + + + + + + + + + + Scalar types + typedef double cl_double __attribute__((aligned(8))); + typedef float cl_float __attribute__((aligned(4))); + typedef int16_t cl_short __attribute__((aligned(2))); + typedef int32_t cl_int __attribute__((aligned(4))); + typedef int64_t cl_long __attribute__((aligned(8))); + typedef int8_t cl_char; + typedef uint8_t cl_uchar; + typedef uint16_t cl_half __attribute__((aligned(2))); + typedef uint16_t cl_ushort __attribute__((aligned(2))); + typedef uint32_t cl_uint __attribute__((aligned(4))); + typedef uint64_t cl_ulong __attribute__((aligned(8))); + typedef int cl_GLint; + typedef unsigned int cl_GLenum; + typedef unsigned int cl_GLuint; + typedef cl_uint cl_d3d11_device_source_khr; + typedef cl_uint cl_d3d11_device_set_khr; + typedef cl_uint cl_dx9_media_adapter_type_khr; + typedef cl_uint cl_dx9_media_adapter_set_khr; + typedef cl_uint cl_d3d10_device_source_khr; + typedef cl_uint cl_d3d10_device_set_khr; + typedef cl_uint cl_dx9_device_source_intel; + typedef cl_uint cl_dx9_device_set_intel; + typedef struct _cl_accelerator_intel* cl_accelerator_intel; + typedef cl_uint cl_accelerator_type_intel; + typedef cl_uint cl_accelerator_info_intel; + typedef cl_uint cl_diagnostics_verbose_level; + typedef cl_uint cl_va_api_device_source_intel; + typedef cl_uint cl_va_api_device_set_intel; + typedef struct __GLsync * cl_GLsync; + typedef void* CLeglImageKHR; + typedef void* CLeglDisplayKHR; + typedef void* CLeglSyncKHR; + typedef intptr_t cl_egl_image_properties_khr; + typedef cl_ulong cl_device_partition_property_ext; + typedef cl_bitfield cl_mem_migration_flags_ext; + typedef cl_uint cl_image_pitch_info_qcom; + typedef cl_uint cl_queue_priority_khr; + typedef cl_uint cl_queue_throttle_khr; + typedef intptr_t cl_import_properties_arm; + typedef cl_bitfield cl_svm_mem_flags_arm; + typedef cl_uint cl_kernel_exec_info_arm; + typedef cl_bitfield cl_device_svm_capabilities_arm; + typedef cl_uint cl_gl_context_info; + typedef cl_uint cl_gl_object_type; + typedef cl_uint cl_gl_texture_info; + typedef cl_uint cl_gl_platform_info; + typedef struct _cl_platform_id * cl_platform_id; + typedef struct _cl_device_id * cl_device_id; + typedef struct _cl_context * cl_context; + typedef struct _cl_command_queue * cl_command_queue; + typedef struct _cl_mem * cl_mem; + typedef struct _cl_program * cl_program; + typedef struct _cl_kernel * cl_kernel; + typedef struct _cl_event * cl_event; + typedef struct _cl_sampler * cl_sampler; + typedef struct _cl_semaphore_khr * cl_semaphore_khr; + typedef cl_uint cl_bool; + typedef cl_ulong cl_bitfield; + typedef cl_ulong cl_properties; + typedef cl_bitfield cl_device_type; + typedef cl_uint cl_platform_info; + typedef cl_uint cl_device_info; + typedef cl_bitfield cl_device_fp_config; + typedef cl_uint cl_device_mem_cache_type; + typedef cl_uint cl_device_local_mem_type; + typedef cl_bitfield cl_device_exec_capabilities; + typedef cl_bitfield cl_device_svm_capabilities; + typedef cl_bitfield cl_command_queue_properties; + typedef intptr_t cl_device_partition_property; + typedef cl_bitfield cl_device_affinity_domain; + typedef intptr_t cl_context_properties; + typedef cl_uint cl_context_info; + typedef cl_properties cl_queue_properties; + typedef cl_properties cl_queue_properties_khr; + typedef cl_uint cl_command_queue_info; + typedef cl_uint cl_channel_order; + typedef cl_uint cl_channel_type; + typedef cl_bitfield cl_mem_flags; + typedef cl_bitfield cl_svm_mem_flags; + typedef cl_uint cl_mem_object_type; + typedef cl_uint cl_mem_info; + typedef cl_bitfield cl_mem_migration_flags; + typedef cl_properties cl_mem_properties; + typedef cl_uint cl_image_info; + typedef cl_uint cl_buffer_create_type; + typedef cl_uint cl_addressing_mode; + typedef cl_uint cl_filter_mode; + typedef cl_uint cl_sampler_info; + typedef cl_bitfield cl_map_flags; + typedef intptr_t cl_pipe_properties; + typedef cl_uint cl_pipe_info; + typedef cl_uint cl_program_info; + typedef cl_uint cl_program_build_info; + typedef cl_uint cl_program_binary_type; + typedef cl_int cl_build_status; + typedef cl_uint cl_kernel_info; + typedef cl_uint cl_kernel_arg_info; + typedef cl_uint cl_kernel_arg_address_qualifier; + typedef cl_uint cl_kernel_arg_access_qualifier; + typedef cl_bitfield cl_kernel_arg_type_qualifier; + typedef cl_uint cl_kernel_work_group_info; + typedef cl_uint cl_kernel_sub_group_info; + typedef cl_uint cl_event_info; + typedef cl_uint cl_command_type; + typedef cl_uint cl_profiling_info; + typedef cl_properties cl_sampler_properties; + typedef cl_uint cl_kernel_exec_info; + typedef cl_bitfield cl_context_memory_initialize_khr; + typedef cl_bitfield cl_device_terminate_capability_khr; + typedef cl_bitfield cl_device_unified_shared_memory_capabilities_intel; + typedef cl_properties cl_mem_properties_intel; + typedef cl_bitfield cl_mem_alloc_flags_intel; + typedef cl_uint cl_mem_info_intel; + typedef cl_uint cl_unified_shared_memory_type_intel; + typedef cl_uint cl_mem_advice_intel; + typedef cl_bitfield cl_device_atomic_capabilities; + typedef cl_uint cl_khronos_vendor_id; + typedef cl_uint cl_version; + typedef cl_uint cl_version_khr; + typedef cl_bitfield cl_device_device_enqueue_capabilities; + typedef cl_uint cl_mipmap_filter_mode_img; + typedef cl_bitfield cl_mem_alloc_flags_img; + typedef cl_uint cl_layer_info; + typedef cl_uint cl_layer_api_version; + typedef cl_uint cl_icdl_info; + typedef struct _cl_icd_dispatch cl_icd_dispatch; + typedef cl_bitfield cl_device_scheduling_controls_capabilities_arm; + typedef cl_bitfield cl_device_controlled_termination_capabilities_arm; + typedef cl_bitfield cl_command_queue_capabilities_intel; + typedef cl_bitfield cl_device_feature_capabilities_intel; + typedef cl_bitfield cl_device_integer_dot_product_capabilities_khr; + typedef cl_properties cl_semaphore_properties_khr; + typedef cl_uint cl_semaphore_info_khr; + typedef cl_uint cl_semaphore_type_khr; + typedef cl_ulong cl_semaphore_payload_khr; + typedef cl_uint cl_external_semaphore_handle_type_khr; + typedef cl_uint cl_external_memory_handle_type_khr; + typedef cl_bitfield cl_device_command_buffer_capabilities_khr; + typedef struct _cl_command_buffer_khr* cl_command_buffer_khr; + typedef cl_uint cl_sync_point_khr; + typedef cl_uint cl_command_buffer_info_khr; + typedef cl_uint cl_command_buffer_state_khr; + typedef cl_properties cl_command_buffer_properties_khr; + typedef cl_bitfield cl_command_buffer_flags_khr; + typedef cl_properties cl_ndrange_kernel_command_properties_khr; + typedef struct _cl_mutable_command_khr* cl_mutable_command_khr; + typedef cl_bitfield cl_mutable_dispatch_fields_khr; + typedef cl_uint cl_mutable_command_info_khr; + typedef cl_uint cl_command_buffer_structure_type_khr; + typedef cl_bitfield cl_device_fp_atomic_capabilities_ext; + typedef cl_uint cl_image_requirements_info_ext; + typedef cl_bitfield cl_platform_command_buffer_capabilities_khr; + typedef cl_bitfield cl_device_unified_shared_memory_capabilities_exp; + typedef cl_properties cl_svm_mem_properties_exp; + typedef cl_uint cl_svm_mem_info_exp; + typedef cl_uint cl_svm_mem_type_exp; + typedef cl_uint cl_svm_mem_advice_exp; + typedef cl_bitfield cl_svm_free_flags_exp; + typedef cl_properties cl_svm_free_properties_exp; + + Structure types + + IDirect3DSurface9* resource + HANDLE shared_handle + + + cl_uint mb_block_type + cl_uint subpixel_mode + cl_uint sad_adjust_mode + cl_uint search_path_type + + + cl_uint allocation_type + cl_uint host_cache_policy + + + cl_mem_ext_host_ptr ext_host_ptr + int ion_filedesc + void* ion_hostptr + + + cl_mem_ext_host_ptr ext_host_ptr + void* anb_ptr + + + cl_channel_order image_channel_order + cl_channel_type image_channel_data_type + + + cl_mem_object_type image_type + size_t image_width + size_t image_height + size_t image_depth + size_t image_array_size + size_t image_row_pitch + size_t image_slice_pitch + cl_uint num_mip_levels + cl_uint num_samples + + union { + cl_mem buffer; + cl_mem mem_object; + } + + + size_t origin + size_t size + + + cl_version version + charname[CL_NAME_VERSION_MAX_NAME_SIZE] + + + cl_version_khr version + charname[CL_NAME_VERSION_MAX_NAME_SIZE_KHR] + + + cl_uint pci_domain + cl_uint pci_bus + cl_uint pci_device + cl_uint pci_function + + + cl_command_queue_properties properties + cl_command_queue_capabilities_intel capabilities + cl_uint count + charname[CL_QUEUE_FAMILY_MAX_NAME_SIZE_INTEL] + + #define CL_VERSION_MAJOR_MASK_KHR ((1 << CL_VERSION_MAJOR_BITS_KHR) - 1) + #define CL_VERSION_MINOR_MASK_KHR ((1 << CL_VERSION_MINOR_BITS_KHR) - 1) + #define CL_VERSION_PATCH_MASK_KHR ((1 << CL_VERSION_PATCH_BITS_KHR) - 1) + + #define CL_VERSION_MAJOR_KHR(version) ((version) >> (CL_VERSION_MINOR_BITS_KHR + CL_VERSION_PATCH_BITS_KHR)) + #define CL_VERSION_MINOR_KHR(version) (((version) >> CL_VERSION_PATCH_BITS_KHR) & CL_VERSION_MINOR_MASK_KHR) + #define CL_VERSION_PATCH_KHR(version) ((version) & CL_VERSION_PATCH_MASK_KHR) + + #define CL_MAKE_VERSION_KHR(major, minor, patch) \ + ((((major) & CL_VERSION_MAJOR_MASK_KHR) << (CL_VERSION_MINOR_BITS_KHR + CL_VERSION_PATCH_BITS_KHR)) | \ + (((minor) & CL_VERSION_MINOR_MASK_KHR) << CL_VERSION_PATCH_BITS_KHR) | \ + ((patch) & CL_VERSION_PATCH_MASK_KHR)) + + cl_boolsigned_accelerated + cl_boolunsigned_accelerated + cl_boolmixed_signedness_accelerated + cl_boolaccumulating_saturating_signed_accelerated + cl_boolaccumulating_saturating_unsigned_accelerated + cl_boolaccumulating_saturating_mixed_signedness_accelerated + + + cl_uint arg_index + size_t arg_size + const void* arg_value + + + cl_uint param_name + size_t param_value_size + const void* param_value + + + cl_command_buffer_structure_type_khr type + const void* next + cl_mutable_command_khr command + cl_uint num_args + cl_uint num_svm_args + cl_uint num_exec_infos + cl_uint work_dim + const cl_mutable_dispatch_arg_khr* arg_list + const cl_mutable_dispatch_arg_khr* arg_svm_list + const cl_mutable_dispatch_exec_info_khr* exec_info_list + const size_t* global_work_offset + const size_t* global_work_size + const size_t* local_work_size + + + + cl_command_buffer_structure_type_khr type + const void* next + cl_uint num_mutable_dispatch + const cl_mutable_dispatch_config_khr* mutable_dispatch_list + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + These values are defined differently along the + #if (defined (_WIN32) && defined(_MSC_VER)) + compilation path, as follows: + <enum value="179769313486231570814527423731704356798070567525844996598917476803157260780028538760589558632766878171540458953514382464234321326889464182768467546703537516986049910576551282076245490090389328944075868508455133942304583236903222948165808559332123348274797826204144723168738177180919299881250404026184124858368.0" name="CL_DBL_MAX"/> + #if defined( __GNUC__ ) + #define CL_HUGE_VALF __builtin_huge_valf() + #define CL_HUGE_VAL __builtin_huge_val() + #define CL_NAN __builtin_nanf( "" ) + #else + #define CL_HUGE_VALF ((cl_float) 1e50) + #define CL_HUGE_VAL ((cl_double) 1e500) + float nanf( const char * ); + #define CL_NAN nanf( "" ) + #endif + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + In order to synchronize vendor IDs across Khronos APIs, Vulkan's vk.xml + is used as the central Khronos vendor ID registry. To obtain a vendor + ID for use in OpenCL, first follow the process defined Vulkan's "Procedures and Conventions" + document under the section "Registering a Vendor ID with Khronos". + Only once the ID has been reserved should a new enum entry be added here. + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + cl_int clGetDeviceIDsFromD3D10KHR + cl_platform_id platform + cl_d3d10_device_source_khr d3d_device_source + void* d3d_object + cl_d3d10_device_set_khr d3d_device_set + cl_uint num_entries + cl_device_id* devices + cl_uint* num_devices + + + cl_mem clCreateFromD3D10BufferKHR + cl_context context + cl_mem_flags flags + ID3D10Buffer* resource + cl_int* errcode_ret + + + cl_mem clCreateFromD3D10Texture2DKHR + cl_context context + cl_mem_flags flags + ID3D10Texture2D* resource + UINT subresource + cl_int* errcode_ret + + + cl_mem clCreateFromD3D10Texture3DKHR + cl_context context + cl_mem_flags flags + ID3D10Texture3D* resource + UINT subresource + cl_int* errcode_ret + + + cl_int clEnqueueAcquireD3D10ObjectsKHR + cl_command_queue command_queue + cl_uint num_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueReleaseD3D10ObjectsKHR + cl_command_queue command_queue + cl_uint num_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clGetDeviceIDsFromD3D11KHR + cl_platform_id platform + cl_d3d11_device_source_khr d3d_device_source + void* d3d_object + cl_d3d11_device_set_khr d3d_device_set + cl_uint num_entries + cl_device_id* devices + cl_uint* num_devices + + + cl_mem clCreateFromD3D11BufferKHR + cl_context context + cl_mem_flags flags + ID3D11Buffer* resource + cl_int* errcode_ret + + + cl_mem clCreateFromD3D11Texture2DKHR + cl_context context + cl_mem_flags flags + ID3D11Texture2D* resource + UINT subresource + cl_int* errcode_ret + + + cl_mem clCreateFromD3D11Texture3DKHR + cl_context context + cl_mem_flags flags + ID3D11Texture3D* resource + UINT subresource + cl_int* errcode_ret + + + cl_int clEnqueueAcquireD3D11ObjectsKHR + cl_command_queue command_queue + cl_uint num_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueReleaseD3D11ObjectsKHR + cl_command_queue command_queue + cl_uint num_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clGetDeviceIDsFromDX9MediaAdapterKHR + cl_platform_id platform + cl_uint num_media_adapters + cl_dx9_media_adapter_type_khr* media_adapter_type + void* media_adapters + cl_dx9_media_adapter_set_khr media_adapter_set + cl_uint num_entries + cl_device_id* devices + cl_uint* num_devices + + + cl_mem clCreateFromDX9MediaSurfaceKHR + cl_context context + cl_mem_flags flags + cl_dx9_media_adapter_type_khr adapter_type + void* surface_info + cl_uint plane + cl_int* errcode_ret + + + cl_int clEnqueueAcquireDX9MediaSurfacesKHR + cl_command_queue command_queue + cl_uint num_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueReleaseDX9MediaSurfacesKHR + cl_command_queue command_queue + cl_uint num_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clGetDeviceIDsFromDX9INTEL + cl_platform_id platform + cl_dx9_device_source_intel dx9_device_source + void* dx9_object + cl_dx9_device_set_intel dx9_device_set + cl_uint num_entries + cl_device_id* devices + cl_uint* num_devices + + + cl_mem clCreateFromDX9MediaSurfaceINTEL + cl_context context + cl_mem_flags flags + IDirect3DSurface9* resource + HANDLE sharedHandle + UINT plane + cl_int* errcode_ret + + + cl_int clEnqueueAcquireDX9ObjectsINTEL + cl_command_queue command_queue + cl_uint num_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueReleaseDX9ObjectsINTEL + cl_command_queue command_queue + cl_uint num_objects + cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_event clCreateEventFromEGLSyncKHR + cl_context context + CLeglSyncKHR sync + CLeglDisplayKHR display + cl_int* errcode_ret + + + cl_mem clCreateFromEGLImageKHR + cl_context context + CLeglDisplayKHR egldisplay + CLeglImageKHR eglimage + cl_mem_flags flags + const cl_egl_image_properties_khr* properties + cl_int* errcode_ret + + + cl_int clEnqueueAcquireEGLObjectsKHR + cl_command_queue command_queue + cl_uint num_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueReleaseEGLObjectsKHR + cl_command_queue command_queue + cl_uint num_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + void clLogMessagesToSystemLogAPPLE + const char* errstr + const void* private_info + size_t cb + void* user_data + + + void clLogMessagesToStdoutAPPLE + const char* errstr + const void* private_info + size_t cb + void* user_data + + + void clLogMessagesToStderrAPPLE + const char* errstr + const void* private_info + size_t cb + void* user_data + + + cl_int clIcdGetPlatformIDsKHR + cl_uint num_entries + cl_platform_id* platforms + cl_uint* num_platforms + + + cl_program clCreateProgramWithILKHR + cl_context context + const void* il + size_t length + cl_int* errcode_ret + + + cl_int clTerminateContextKHR + cl_context context + + + cl_command_queue clCreateCommandQueueWithPropertiesKHR + cl_context context + cl_device_id device + const cl_queue_properties_khr* properties + cl_int* errcode_ret + + + cl_int clReleaseDeviceEXT + cl_device_id device + + + cl_int clRetainDeviceEXT + cl_device_id device + + + cl_int clCreateSubDevicesEXT + cl_device_id in_device + const cl_device_partition_property_ext* properties + cl_uint num_entries + cl_device_id* out_devices + cl_uint* num_devices + + + cl_int clEnqueueMigrateMemObjectEXT + cl_command_queue command_queue + cl_uint num_mem_objects + const cl_mem* mem_objects + cl_mem_migration_flags_ext flags + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clGetDeviceImageInfoQCOM + cl_device_id device + size_t image_width + size_t image_height + const cl_image_format* image_format + cl_image_pitch_info_qcom param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clEnqueueAcquireGrallocObjectsIMG + cl_command_queue command_queue + cl_uint num_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueGenerateMipmapIMG + cl_command_queue command_queue + cl_mem src_image + cl_mem dst_image + cl_mipmap_filter_mode_img mipmap_filter_mode + const size_t* array_region + const size_t* mip_region + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueReleaseGrallocObjectsIMG + cl_command_queue command_queue + cl_uint num_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clGetKernelSubGroupInfoKHR + cl_kernel in_kernel + cl_device_id in_device + cl_kernel_sub_group_info param_name + size_t input_value_size + const void* input_value + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clGetKernelSuggestedLocalWorkSizeKHR + cl_command_queue command_queue + cl_kernel kernel + cl_uint work_dim + const size_t* global_work_offset + const size_t* global_work_size + size_t* suggested_local_work_size + + + cl_semaphore_khr clCreateSemaphoreWithPropertiesKHR + cl_context context + const cl_semaphore_properties_khr* sema_props + cl_int* errcode_ret + + + cl_int clEnqueueWaitSemaphoresKHR + cl_command_queue command_queue + cl_uint num_sema_objects + const cl_semaphore_khr* sema_objects + const cl_semaphore_payload_khr* sema_payload_list + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueSignalSemaphoresKHR + cl_command_queue command_queue + cl_uint num_sema_objects + const cl_semaphore_khr* sema_objects + const cl_semaphore_payload_khr* sema_payload_list + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clGetSemaphoreInfoKHR + cl_semaphore_khr sema_object + cl_semaphore_info_khr param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clReleaseSemaphoreKHR + cl_semaphore_khr sema_object + + + cl_int clRetainSemaphoreKHR + cl_semaphore_khr sema_object + + + cl_int clGetSemaphoreHandleForTypeKHR + cl_semaphore_khr sema_object + cl_device_id device + cl_external_semaphore_handle_type_khr handle_type + size_t handle_size + void* handle_ptr + size_t* handle_size_ret + + + cl_int clEnqueueAcquireExternalMemObjectsKHR + cl_command_queue command_queue + cl_uint num_mem_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueReleaseExternalMemObjectsKHR + cl_command_queue command_queue + cl_uint num_mem_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_mem clImportMemoryARM + cl_context context + cl_mem_flags flags + const cl_import_properties_arm* properties + void* memory + size_t size + cl_int* errcode_ret + + + void* clSVMAllocARM + cl_context context + cl_svm_mem_flags_arm flags + size_t size + cl_uint alignment + + + void clSVMFreeARM + cl_context context + void* svm_pointer + + + cl_int clEnqueueSVMFreeARM + cl_command_queue command_queue + cl_uint num_svm_pointers + void* svm_pointers[] + void (CL_CALLBACK* pfn_free_func)(cl_command_queue queue, cl_uint num_svm_pointers, void * svm_pointers[], void *user_data) + void* user_data + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueSVMMemcpyARM + cl_command_queue command_queue + cl_bool blocking_copy + void* dst_ptr + const void* src_ptr + size_t size + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueSVMMemFillARM + cl_command_queue command_queue + void* svm_ptr + const void* pattern + size_t pattern_size + size_t size + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueSVMMapARM + cl_command_queue command_queue + cl_bool blocking_map + cl_map_flags flags + void* svm_ptr + size_t size + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueSVMUnmapARM + cl_command_queue command_queue + void* svm_ptr + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clSetKernelArgSVMPointerARM + cl_kernel kernel + cl_uint arg_index + const void* arg_value + + + cl_int clSetKernelExecInfoARM + cl_kernel kernel + cl_kernel_exec_info_arm param_name + size_t param_value_size + const void* param_value + + + cl_accelerator_intel clCreateAcceleratorINTEL + cl_context context + cl_accelerator_type_intel accelerator_type + size_t descriptor_size + const void* descriptor + cl_int* errcode_ret + + + cl_int clGetAcceleratorInfoINTEL + cl_accelerator_intel accelerator + cl_accelerator_info_intel param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clRetainAcceleratorINTEL + cl_accelerator_intel accelerator + + + cl_int clReleaseAcceleratorINTEL + cl_accelerator_intel accelerator + + + cl_event clCreateEventFromGLsyncKHR + cl_context context + cl_GLsync sync + cl_int* errcode_ret + + + cl_int clGetGLContextInfoKHR + const cl_context_properties* properties + cl_gl_context_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_mem clCreateFromGLBuffer + cl_context context + cl_mem_flags flags + cl_GLuint bufobj + cl_int* errcode_ret + + + cl_mem clCreateFromGLTexture + cl_context context + cl_mem_flags flags + cl_GLenum target + cl_GLint miplevel + cl_GLuint texture + cl_int* errcode_ret + + + cl_mem clCreateFromGLRenderbuffer + cl_context context + cl_mem_flags flags + cl_GLuint renderbuffer + cl_int* errcode_ret + + + cl_int clGetGLObjectInfo + cl_mem memobj + cl_gl_object_type* gl_object_type + cl_GLuint* gl_object_name + + + cl_int clGetGLTextureInfo + cl_mem memobj + cl_gl_texture_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clEnqueueAcquireGLObjects + cl_command_queue command_queue + cl_uint num_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueReleaseGLObjects + cl_command_queue command_queue + cl_uint num_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_mem clCreateFromGLTexture2D + cl_context context + cl_mem_flags flags + cl_GLenum target + cl_GLint miplevel + cl_GLuint texture + cl_int* errcode_ret + + + cl_mem clCreateFromGLTexture3D + cl_context context + cl_mem_flags flags + cl_GLenum target + cl_GLint miplevel + cl_GLuint texture + cl_int* errcode_ret + + + cl_int clGetDeviceIDsFromVA_APIMediaAdapterINTEL + cl_platform_id platform + cl_va_api_device_source_intel media_adapter_type + void* media_adapter + cl_va_api_device_set_intel media_adapter_set + cl_uint num_entries + cl_device_id* devices + cl_uint* num_devices + + + cl_mem clCreateFromVA_APIMediaSurfaceINTEL + cl_context context + cl_mem_flags flags + VASurfaceID* surface + cl_uint plane + cl_int* errcode_ret + + + cl_int clEnqueueAcquireVA_APIMediaSurfacesINTEL + cl_command_queue command_queue + cl_uint num_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueReleaseVA_APIMediaSurfacesINTEL + cl_command_queue command_queue + cl_uint num_objects + const cl_mem* mem_objects + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + void* clHostMemAllocINTEL + cl_context context + const cl_mem_properties_intel* properties + size_t size + cl_uint alignment + cl_int* errcode_ret + + + void* clDeviceMemAllocINTEL + cl_context context + cl_device_id device + const cl_mem_properties_intel* properties + size_t size + cl_uint alignment + cl_int* errcode_ret + + + void* clSharedMemAllocINTEL + cl_context context + cl_device_id device + const cl_mem_properties_intel* properties + size_t size + cl_uint alignment + cl_int* errcode_ret + + + cl_int clMemFreeINTEL + cl_context context + void* ptr + + + cl_int clMemBlockingFreeINTEL + cl_context context + void* ptr + + + cl_int clGetMemAllocInfoINTEL + cl_context context + const void* ptr + cl_mem_info_intel param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clSetKernelArgMemPointerINTEL + cl_kernel kernel + cl_uint arg_index + const void* arg_value + + + cl_int clEnqueueMemsetINTEL + cl_command_queue command_queue + void* dst_ptr + cl_int value + size_t size + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueMemFillINTEL + cl_command_queue command_queue + void* dst_ptr + const void* pattern + size_t pattern_size + size_t size + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueMemcpyINTEL + cl_command_queue command_queue + cl_bool blocking + void* dst_ptr + const void* src_ptr + size_t size + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueMigrateMemINTEL + cl_command_queue command_queue + const void* ptr + size_t size + cl_mem_migration_flags flags + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueMemAdviseINTEL + cl_command_queue command_queue + const void* ptr + size_t size + cl_mem_advice_intel advice + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_mem clCreateBufferWithPropertiesINTEL + cl_context context + const cl_mem_properties_intel* properties + cl_mem_flags flags + size_t size + void* host_ptr + cl_int* errcode_ret + + + cl_command_buffer_khr clCreateCommandBufferKHR + cl_uint num_queues + const cl_command_queue* queues + const cl_command_buffer_properties_khr* properties + cl_int* errcode_ret + + + cl_int clFinalizeCommandBufferKHR + cl_command_buffer_khr command_buffer + + + cl_int clRetainCommandBufferKHR + cl_command_buffer_khr command_buffer + + + cl_int clReleaseCommandBufferKHR + cl_command_buffer_khr command_buffer + + + cl_int clEnqueueCommandBufferKHR + cl_uint num_queues + cl_command_queue* queues + cl_command_buffer_khr command_buffer + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clCommandBarrierWithWaitListKHR + cl_command_buffer_khr command_buffer + cl_command_queue command_queue + cl_uint num_sync_points_in_wait_list + const cl_sync_point_khr* sync_point_wait_list + cl_sync_point_khr* sync_point + cl_mutable_command_khr* mutable_handle + + + cl_int clCommandCopyBufferKHR + cl_command_buffer_khr command_buffer + cl_command_queue command_queue + cl_mem src_buffer + cl_mem dst_buffer + size_t src_offset + size_t dst_offset + size_t size + cl_uint num_sync_points_in_wait_list + const cl_sync_point_khr* sync_point_wait_list + cl_sync_point_khr* sync_point + cl_mutable_command_khr* mutable_handle + + + cl_int clCommandCopyBufferRectKHR + cl_command_buffer_khr command_buffer + cl_command_queue command_queue + cl_mem src_buffer + cl_mem dst_buffer + const size_t* src_origin + const size_t* dst_origin + const size_t* region + size_t src_row_pitch + size_t src_slice_pitch + size_t dst_row_pitch + size_t dst_slice_pitch + cl_uint num_sync_points_in_wait_list + const cl_sync_point_khr* sync_point_wait_list + cl_sync_point_khr* sync_point + cl_mutable_command_khr* mutable_handle + + + cl_int clCommandCopyBufferToImageKHR + cl_command_buffer_khr command_buffer + cl_command_queue command_queue + cl_mem src_buffer + cl_mem dst_image + size_t src_offset + const size_t* dst_origin + const size_t* region + cl_uint num_sync_points_in_wait_list + const cl_sync_point_khr* sync_point_wait_list + cl_sync_point_khr* sync_point + cl_mutable_command_khr* mutable_handle + + + cl_int clCommandCopyImageKHR + cl_command_buffer_khr command_buffer + cl_command_queue command_queue + cl_mem src_image + cl_mem dst_image + const size_t* src_origin + const size_t* dst_origin + const size_t* region + cl_uint num_sync_points_in_wait_list + const cl_sync_point_khr* sync_point_wait_list + cl_sync_point_khr* sync_point + cl_mutable_command_khr* mutable_handle + + + cl_int clCommandCopyImageToBufferKHR + cl_command_buffer_khr command_buffer + cl_command_queue command_queue + cl_mem src_image + cl_mem dst_buffer + const size_t* src_origin + const size_t* region + size_t dst_offset + cl_uint num_sync_points_in_wait_list + const cl_sync_point_khr* sync_point_wait_list + cl_sync_point_khr* sync_point + cl_mutable_command_khr* mutable_handle + + + cl_int clCommandFillBufferKHR + cl_command_buffer_khr command_buffer + cl_command_queue command_queue + cl_mem buffer + const void* pattern + size_t pattern_size + size_t offset + size_t size + cl_uint num_sync_points_in_wait_list + const cl_sync_point_khr* sync_point_wait_list + cl_sync_point_khr* sync_point + cl_mutable_command_khr* mutable_handle + + + cl_int clCommandFillImageKHR + cl_command_buffer_khr command_buffer + cl_command_queue command_queue + cl_mem image + const void* fill_color + const size_t* origin + const size_t* region + cl_uint num_sync_points_in_wait_list + const cl_sync_point_khr* sync_point_wait_list + cl_sync_point_khr* sync_point + cl_mutable_command_khr* mutable_handle + + + cl_int clCommandNDRangeKernelKHR + cl_command_buffer_khr command_buffer + cl_command_queue command_queue + const cl_ndrange_kernel_command_properties_khr* properties + cl_kernel kernel + cl_uint work_dim + const size_t* global_work_offset + const size_t* global_work_size + const size_t* local_work_size + cl_uint num_sync_points_in_wait_list + const cl_sync_point_khr* sync_point_wait_list + cl_sync_point_khr* sync_point + cl_mutable_command_khr* mutable_handle + + + cl_int clCommandSVMMemcpyKHR + cl_command_buffer_khr command_buffer + cl_command_queue command_queue + void* dst_ptr + const void* src_ptr + size_t size + cl_uint num_sync_points_in_wait_list + const cl_sync_point_khr* sync_point_wait_list + cl_sync_point_khr* sync_point + cl_mutable_command_khr* mutable_handle + + + cl_int clCommandSVMMemFillKHR + cl_command_buffer_khr command_buffer + cl_command_queue command_queue + void* svm_ptr + const void* pattern + size_t pattern_size + size_t size + cl_uint num_sync_points_in_wait_list + const cl_sync_point_khr* sync_point_wait_list + cl_sync_point_khr* sync_point + cl_mutable_command_khr* mutable_handle + + + cl_int clGetCommandBufferInfoKHR + cl_command_buffer_khr command_buffer + cl_command_buffer_info_khr param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clUpdateMutableCommandsKHR + cl_command_buffer_khr command_buffer + const cl_mutable_base_config_khr* mutable_config + + + cl_int clGetMutableCommandInfoKHR + cl_mutable_command_khr command + cl_mutable_command_info_khr param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_command_buffer_khr clRemapCommandBufferKHR + cl_command_buffer_khr command_buffer + cl_bool automatic + cl_uint num_queues + const cl_command_queue* queues + cl_uint num_handles + const cl_mutable_command_khr* handles + cl_mutable_command_khr* handles_ret + cl_int* errcode_ret + + + cl_int clSetContentSizeBufferPoCL + cl_mem buffer + cl_mem content_size_buffer + + + void* clSVMAllocWithPropertiesEXP + cl_context context + const cl_svm_mem_properties_exp* properties + cl_svm_mem_flags flags + size_t size + cl_uint alignment + cl_int* errcode_ret + + + cl_int clSVMFreeWithPropertiesEXP + cl_context context + const cl_svm_free_properties_exp* properties + cl_svm_free_flags_exp flags + void* ptr + + + cl_int clGetSVMMemInfoEXP + cl_context context + const void* ptr + cl_svm_mem_info_exp param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clEnqueueSVMMemAdviseEXP + cl_command_queue command_queue + const void* ptr + size_t size + cl_svm_mem_advice_exp advice + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clGetPlatformIDs + cl_uint num_entries + cl_platform_id* platforms + cl_uint* num_platforms + + + cl_int clGetPlatformInfo + cl_platform_id platform + cl_platform_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clGetDeviceIDs + cl_platform_id platform + cl_device_type device_type + cl_uint num_entries + cl_device_id* devices + cl_uint* num_devices + + + cl_int clGetDeviceInfo + cl_device_id device + cl_device_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clCreateSubDevices + cl_device_id in_device + const cl_device_partition_property* properties + cl_uint num_devices + cl_device_id* out_devices + cl_uint* num_devices_ret + + + cl_int clRetainDevice + cl_device_id device + + + cl_int clReleaseDevice + cl_device_id device + + + cl_int clSetDefaultDeviceCommandQueue + cl_context context + cl_device_id device + cl_command_queue command_queue + + + cl_int clGetDeviceAndHostTimer + cl_device_id device + cl_ulong* device_timestamp + cl_ulong* host_timestamp + + + cl_int clGetHostTimer + cl_device_id device + cl_ulong* host_timestamp + + + cl_context clCreateContext + const cl_context_properties* properties + cl_uint num_devices + const cl_device_id* devices + void (CL_CALLBACK* pfn_notify)(const char* errinfo, const void* private_info, size_t cb, void* user_data) + void* user_data + cl_int* errcode_ret + + + cl_context clCreateContextFromType + const cl_context_properties* properties + cl_device_type device_type + void (CL_CALLBACK* pfn_notify)(const char* errinfo, const void* private_info, size_t cb, void* user_data) + void* user_data + cl_int* errcode_ret + + + cl_int clRetainContext + cl_context context + + + cl_int clReleaseContext + cl_context context + + + cl_int clGetContextInfo + cl_context context + cl_context_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clSetContextDestructorCallback + cl_context context + void (CL_CALLBACK* pfn_notify)(cl_context context, void* user_data) + void* user_data + + + cl_command_queue clCreateCommandQueueWithProperties + cl_context context + cl_device_id device + const cl_queue_properties* properties + cl_int* errcode_ret + + + cl_int clRetainCommandQueue + cl_command_queue command_queue + + + cl_int clReleaseCommandQueue + cl_command_queue command_queue + + + cl_int clGetCommandQueueInfo + cl_command_queue command_queue + cl_command_queue_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_mem clCreateBuffer + cl_context context + cl_mem_flags flags + size_t size + void* host_ptr + cl_int* errcode_ret + + + cl_mem clCreateBufferWithProperties + cl_context context + const cl_mem_properties* properties + cl_mem_flags flags + size_t size + void* host_ptr + cl_int* errcode_ret + + + cl_mem clCreateSubBuffer + cl_mem buffer + cl_mem_flags flags + cl_buffer_create_type buffer_create_type + const void* buffer_create_info + cl_int* errcode_ret + + + cl_mem clCreateImage + cl_context context + cl_mem_flags flags + const cl_image_format* image_format + const cl_image_desc* image_desc + void* host_ptr + cl_int* errcode_ret + + + cl_mem clCreateImageWithProperties + cl_context context + const cl_mem_properties* properties + cl_mem_flags flags + const cl_image_format* image_format + const cl_image_desc* image_desc + void* host_ptr + cl_int* errcode_ret + + + cl_mem clCreatePipe + cl_context context + cl_mem_flags flags + cl_uint pipe_packet_size + cl_uint pipe_max_packets + const cl_pipe_properties* properties + cl_int* errcode_ret + + + cl_int clRetainMemObject + cl_mem memobj + + + cl_int clReleaseMemObject + cl_mem memobj + + + cl_int clGetSupportedImageFormats + cl_context context + cl_mem_flags flags + cl_mem_object_type image_type + cl_uint num_entries + cl_image_format* image_formats + cl_uint* num_image_formats + + + cl_int clGetMemObjectInfo + cl_mem memobj + cl_mem_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clGetImageInfo + cl_mem image + cl_image_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clGetPipeInfo + cl_mem pipe + cl_pipe_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clSetMemObjectDestructorCallback + cl_mem memobj + void (CL_CALLBACK* pfn_notify)(cl_mem memobj, void* user_data) + void* user_data + + + cl_int clSetMemObjectDestructorAPPLE + cl_mem memobj + void (CL_CALLBACK* pfn_notify)(cl_mem memobj, void* user_data) + void* user_data + + + void* clSVMAlloc + cl_context context + cl_svm_mem_flags flags + size_t size + cl_uint alignment + + + void clSVMFree + cl_context context + void* svm_pointer + + + cl_sampler clCreateSamplerWithProperties + cl_context context + const cl_sampler_properties* sampler_properties + cl_int* errcode_ret + + + cl_int clRetainSampler + cl_sampler sampler + + + cl_int clReleaseSampler + cl_sampler sampler + + + cl_int clGetSamplerInfo + cl_sampler sampler + cl_sampler_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_program clCreateProgramWithSource + cl_context context + cl_uint count + const char** strings + const size_t* lengths + cl_int* errcode_ret + + + cl_program clCreateProgramWithBinary + cl_context context + cl_uint num_devices + const cl_device_id* device_list + const size_t* lengths + const unsigned char** binaries + cl_int* binary_status + cl_int* errcode_ret + + + cl_program clCreateProgramWithBuiltInKernels + cl_context context + cl_uint num_devices + const cl_device_id* device_list + const char* kernel_names + cl_int* errcode_ret + + + cl_program clCreateProgramWithIL + cl_context context + const void* il + size_t length + cl_int* errcode_ret + + + cl_int clRetainProgram + cl_program program + + + cl_int clReleaseProgram + cl_program program + + + cl_int clBuildProgram + cl_program program + cl_uint num_devices + const cl_device_id* device_list + const char* options + void (CL_CALLBACK* pfn_notify)(cl_program program, void* user_data) + void* user_data + + + cl_int clCompileProgram + cl_program program + cl_uint num_devices + const cl_device_id* device_list + const char* options + cl_uint num_input_headers + const cl_program* input_headers + const char** header_include_names + void (CL_CALLBACK* pfn_notify)(cl_program program, void* user_data) + void* user_data + + + cl_program clLinkProgram + cl_context context + cl_uint num_devices + const cl_device_id* device_list + const char* options + cl_uint num_input_programs + const cl_program* input_programs + void (CL_CALLBACK* pfn_notify)(cl_program program, void* user_data) + void* user_data + cl_int* errcode_ret + + + cl_int clSetProgramReleaseCallback + cl_program program + void (CL_CALLBACK* pfn_notify)(cl_program program, void* user_data) + void* user_data + + + cl_int clSetProgramSpecializationConstant + cl_program program + cl_uint spec_id + size_t spec_size + const void* spec_value + + + cl_int clUnloadPlatformCompiler + cl_platform_id platform + + + cl_int clGetProgramInfo + cl_program program + cl_program_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clGetProgramBuildInfo + cl_program program + cl_device_id device + cl_program_build_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_kernel clCreateKernel + cl_program program + const char* kernel_name + cl_int* errcode_ret + + + cl_int clCreateKernelsInProgram + cl_program program + cl_uint num_kernels + cl_kernel* kernels + cl_uint* num_kernels_ret + + + cl_kernel clCloneKernel + cl_kernel source_kernel + cl_int* errcode_ret + + + cl_int clRetainKernel + cl_kernel kernel + + + cl_int clReleaseKernel + cl_kernel kernel + + + cl_int clSetKernelArg + cl_kernel kernel + cl_uint arg_index + size_t arg_size + const void* arg_value + + + cl_int clSetKernelArgSVMPointer + cl_kernel kernel + cl_uint arg_index + const void* arg_value + + + cl_int clSetKernelExecInfo + cl_kernel kernel + cl_kernel_exec_info param_name + size_t param_value_size + const void* param_value + + + cl_int clGetKernelInfo + cl_kernel kernel + cl_kernel_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clGetKernelArgInfo + cl_kernel kernel + cl_uint arg_index + cl_kernel_arg_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clGetKernelWorkGroupInfo + cl_kernel kernel + cl_device_id device + cl_kernel_work_group_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clGetKernelSubGroupInfo + cl_kernel kernel + cl_device_id device + cl_kernel_sub_group_info param_name + size_t input_value_size + const void* input_value + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clWaitForEvents + cl_uint num_events + const cl_event* event_list + + + cl_int clGetEventInfo + cl_event event + cl_event_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_event clCreateUserEvent + cl_context context + cl_int* errcode_ret + + + cl_int clRetainEvent + cl_event event + + + cl_int clReleaseEvent + cl_event event + + + cl_int clSetUserEventStatus + cl_event event + cl_int execution_status + + + cl_int clSetEventCallback + cl_event event + cl_int command_exec_callback_type + void (CL_CALLBACK* pfn_notify)(cl_event event, cl_int event_command_status, void *user_data) + void* user_data + + + cl_int clGetEventProfilingInfo + cl_event event + cl_profiling_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clFlush + cl_command_queue command_queue + + + cl_int clFinish + cl_command_queue command_queue + + + cl_int clEnqueueReadBuffer + cl_command_queue command_queue + cl_mem buffer + cl_bool blocking_read + size_t offset + size_t size + void* ptr + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueReadBufferRect + cl_command_queue command_queue + cl_mem buffer + cl_bool blocking_read + const size_t* buffer_origin + const size_t* host_origin + const size_t* region + size_t buffer_row_pitch + size_t buffer_slice_pitch + size_t host_row_pitch + size_t host_slice_pitch + void* ptr + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueWriteBuffer + cl_command_queue command_queue + cl_mem buffer + cl_bool blocking_write + size_t offset + size_t size + const void* ptr + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueWriteBufferRect + cl_command_queue command_queue + cl_mem buffer + cl_bool blocking_write + const size_t* buffer_origin + const size_t* host_origin + const size_t* region + size_t buffer_row_pitch + size_t buffer_slice_pitch + size_t host_row_pitch + size_t host_slice_pitch + const void* ptr + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueFillBuffer + cl_command_queue command_queue + cl_mem buffer + const void* pattern + size_t pattern_size + size_t offset + size_t size + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueCopyBuffer + cl_command_queue command_queue + cl_mem src_buffer + cl_mem dst_buffer + size_t src_offset + size_t dst_offset + size_t size + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueCopyBufferRect + cl_command_queue command_queue + cl_mem src_buffer + cl_mem dst_buffer + const size_t* src_origin + const size_t* dst_origin + const size_t* region + size_t src_row_pitch + size_t src_slice_pitch + size_t dst_row_pitch + size_t dst_slice_pitch + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueReadImage + cl_command_queue command_queue + cl_mem image + cl_bool blocking_read + const size_t* origin + const size_t* region + size_t row_pitch + size_t slice_pitch + void* ptr + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueWriteImage + cl_command_queue command_queue + cl_mem image + cl_bool blocking_write + const size_t* origin + const size_t* region + size_t input_row_pitch + size_t input_slice_pitch + const void* ptr + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueFillImage + cl_command_queue command_queue + cl_mem image + const void* fill_color + const size_t* origin + const size_t* region + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueCopyImage + cl_command_queue command_queue + cl_mem src_image + cl_mem dst_image + const size_t* src_origin + const size_t* dst_origin + const size_t* region + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueCopyImageToBuffer + cl_command_queue command_queue + cl_mem src_image + cl_mem dst_buffer + const size_t* src_origin + const size_t* region + size_t dst_offset + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueCopyBufferToImage + cl_command_queue command_queue + cl_mem src_buffer + cl_mem dst_image + size_t src_offset + const size_t* dst_origin + const size_t* region + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + void* clEnqueueMapBuffer + cl_command_queue command_queue + cl_mem buffer + cl_bool blocking_map + cl_map_flags map_flags + size_t offset + size_t size + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + cl_int* errcode_ret + + + void* clEnqueueMapImage + cl_command_queue command_queue + cl_mem image + cl_bool blocking_map + cl_map_flags map_flags + const size_t* origin + const size_t* region + size_t* image_row_pitch + size_t* image_slice_pitch + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + cl_int* errcode_ret + + + cl_int clEnqueueUnmapMemObject + cl_command_queue command_queue + cl_mem memobj + void* mapped_ptr + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueMigrateMemObjects + cl_command_queue command_queue + cl_uint num_mem_objects + const cl_mem* mem_objects + cl_mem_migration_flags flags + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueNDRangeKernel + cl_command_queue command_queue + cl_kernel kernel + cl_uint work_dim + const size_t* global_work_offset + const size_t* global_work_size + const size_t* local_work_size + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueNativeKernel + cl_command_queue command_queue + void (CL_CALLBACK* user_func)(void*) + void* args + size_t cb_args + cl_uint num_mem_objects + const cl_mem* mem_list + const void** args_mem_loc + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueMarkerWithWaitList + cl_command_queue command_queue + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueBarrierWithWaitList + cl_command_queue command_queue + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueSVMFree + cl_command_queue command_queue + cl_uint num_svm_pointers + void* svm_pointers[] + void (CL_CALLBACK* pfn_free_func)(cl_command_queue queue, cl_uint num_svm_pointers, void* svm_pointers[], void* user_data) + void* user_data + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueSVMMemcpy + cl_command_queue command_queue + cl_bool blocking_copy + void* dst_ptr + const void* src_ptr + size_t size + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueSVMMemFill + cl_command_queue command_queue + void* svm_ptr + const void* pattern + size_t pattern_size + size_t size + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueSVMMap + cl_command_queue command_queue + cl_bool blocking_map + cl_map_flags flags + void* svm_ptr + size_t size + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueSVMUnmap + cl_command_queue command_queue + void* svm_ptr + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueSVMMigrateMem + cl_command_queue command_queue + cl_uint num_svm_pointers + const void** svm_pointers + const size_t* sizes + cl_mem_migration_flags flags + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + void* clGetExtensionFunctionAddressForPlatform + cl_platform_id platform + const char* func_name + + + cl_int clSetCommandQueueProperty + cl_command_queue command_queue + cl_command_queue_properties properties + cl_bool enable + cl_command_queue_properties* old_properties + + + cl_mem clCreateImage2D + cl_context context + cl_mem_flags flags + const cl_image_format* image_format + size_t image_width + size_t image_height + size_t image_row_pitch + void* host_ptr + cl_int* errcode_ret + + + cl_mem clCreateImage3D + cl_context context + cl_mem_flags flags + const cl_image_format* image_format + size_t image_width + size_t image_height + size_t image_depth + size_t image_row_pitch + size_t image_slice_pitch + void* host_ptr + cl_int* errcode_ret + + + cl_int clEnqueueMarker + cl_command_queue command_queue + cl_event* event + + + cl_int clEnqueueWaitForEvents + cl_command_queue command_queue + cl_uint num_events + const cl_event* event_list + + + cl_int clEnqueueBarrier + cl_command_queue command_queue + + + cl_int clUnloadCompiler + + + void* clGetExtensionFunctionAddress + const char* func_name + + + cl_command_queue clCreateCommandQueue + cl_context context + cl_device_id device + cl_command_queue_properties properties + cl_int* errcode_ret + + + cl_sampler clCreateSampler + cl_context context + cl_bool normalized_coords + cl_addressing_mode addressing_mode + cl_filter_mode filter_mode + cl_int* errcode_ret + + + cl_int clEnqueueTask + cl_command_queue command_queue + cl_kernel kernel + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clGetLayerInfo + cl_layer_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clInitLayer + cl_uint num_entries + const cl_icd_dispatch* target_dispatch + cl_uint* num_entries_ret + const cl_icd_dispatch** layer_dispatch_ret + + + cl_int clGetICDLoaderInfoOCLICD + cl_icdl_info param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + cl_int clGetSupportedGLTextureFormatsINTEL + cl_context context + cl_mem_flags flags + cl_mem_object_type image_type + cl_uint num_entries + cl_GLenum* gl_formats + cl_uint* num_texture_formats + + + cl_int clGetSupportedDX9MediaSurfaceFormatsINTEL + cl_context context + cl_mem_flags flags + cl_mem_object_type image_type + cl_uint plane + cl_uint num_entries + D3DFORMAT* dx9_formats + cl_uint* num_surface_formats + + + cl_int clGetSupportedD3D10TextureFormatsINTEL + cl_context context + cl_mem_flags flags + cl_mem_object_type image_type + cl_uint num_entries + DXGI_FORMAT* d3d10_formats + cl_uint* num_texture_formats + + + cl_int clGetSupportedD3D11TextureFormatsINTEL + cl_context context + cl_mem_flags flags + cl_mem_object_type image_type + cl_uint plane + cl_uint num_entries + DXGI_FORMAT* d3d11_formats + cl_uint* num_texture_formats + + + cl_int clGetSupportedVA_APIMediaSurfaceFormatsINTEL + cl_context context + cl_mem_flags flags + cl_mem_object_type image_type + cl_uint plane + cl_uint num_entries + VAImageFormat* va_api_formats + cl_uint* num_surface_formats + + + cl_int clEnqueueReadHostPipeINTEL + cl_command_queue command_queue + cl_program program + const char* pipe_symbol + cl_bool blocking_read + void* ptr + size_t size + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clEnqueueWriteHostPipeINTEL + cl_command_queue command_queue + cl_program program + const char* pipe_symbol + cl_bool blocking_write + const void* ptr + size_t size + cl_uint num_events_in_wait_list + const cl_event* event_wait_list + cl_event* event + + + cl_int clGetImageRequirementsInfoEXT + cl_context context + const cl_mem_properties* properties + cl_mem_flags flags + const cl_image_format* image_format + const cl_image_desc* image_desc + cl_image_requirements_info_ext param_name + size_t param_value_size + void* param_value + size_t* param_value_size_ret + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + #ifdef CL_USE_DEPRECATED_OPENCL_1_0_APIS + /* + * WARNING: + * This API introduces mutable state into the OpenCL implementation. It has been REMOVED + * to better facilitate thread safety. The 1.0 API is not thread-safe. It is not tested by the + * OpenCL 1.1 conformance test, and consequently may not work or may not work dependably. + * It is likely to be non-performant. Use of this API is not advised. Use at your own risk. + * + * Software developers previously relying on this API are instructed to set the command-queue + * properties when creating the queue, instead. + */ + #endif /* CL_USE_DEPRECATED_OPENCL_1_0_APIS */ + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + Allows for direct memory import into OpenCL via the clImportMemoryARM function. + + Memory imported through this interface will be mapped into the device's page + tables directly, providing zero copy access. It will never fall back to copy + operations and aliased buffers. + + Types of memory supported for import are specified as additional extension + strings. + + This extension produces cl_mem allocations which are compatible with all other + users of cl_mem in the standard API. + + This extension maps pages with the same properties as the normal buffer creation + function clCreateBuffer. + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + + +