diff --git a/.github/dependabot.yml b/.github/dependabot.yml new file mode 100644 index 00000000..2390d8c8 --- /dev/null +++ b/.github/dependabot.yml @@ -0,0 +1,10 @@ +version: 2 +updates: + - package-ecosystem: "github-actions" + directory: "/" + schedule: + interval: "monthly" + groups: + github-actions: + patterns: + - "*" diff --git a/.github/workflows/presubmit.yml b/.github/workflows/presubmit.yml index 0cf3d423..94fe9e76 100644 --- a/.github/workflows/presubmit.yml +++ b/.github/workflows/presubmit.yml @@ -79,17 +79,17 @@ jobs: steps: - name: Install system CMake - if: ${{matrix.CMAKE}} == 'system' + if: matrix.CMAKE == 'system' run: apt-get update -qq && apt-get install -y cmake && echo "CMAKE_EXE=cmake" >> "$GITHUB_ENV" && echo "CPACK_EXE=cpack" >> "$GITHUB_ENV" && echo "CTEST_EXE=ctest" >> "$GITHUB_ENV" - name: Checkout OpenCL-ICD-Loader - uses: actions/checkout@v4 + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 - name: Checkout OpenCL-Headers - uses: actions/checkout@v4 + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 with: path: external/OpenCL-Headers repository: KhronosGroup/OpenCL-Headers @@ -267,7 +267,7 @@ jobs: - name: Cache Ninja install if: matrix.GEN == 'Ninja Multi-Config' id: ninja-install - uses: actions/cache@v4 + uses: actions/cache@27d5ce7f107fe9357f9df03efb73ab90386fccae # v5.0.5 with: path: | C:\Tools\Ninja @@ -281,10 +281,10 @@ jobs: Remove-Item ~\Downloads\* - name: Checkout OpenCL-ICD-Loader - uses: actions/checkout@v4 + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 - name: Checkout OpenCL-Headers - uses: actions/checkout@v4 + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 with: repository: KhronosGroup/OpenCL-Headers path: external/OpenCL-Headers @@ -623,10 +623,10 @@ jobs: steps: - name: Checkout OpenCL-ICD-Loader - uses: actions/checkout@v4 + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 - name: Checkout OpenCL-Headers - uses: actions/checkout@v4 + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 with: repository: KhronosGroup/OpenCL-Headers path: external/OpenCL-Headers @@ -728,10 +728,10 @@ jobs: CFLAGS: -Wall -Wextra -pedantic -Werror steps: - name: Checkout OpenCL-ICD-Loader - uses: actions/checkout@v4 + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 - name: Checkout OpenCL-Headers - uses: actions/checkout@v4 + uses: actions/checkout@de0fac2e4500dabe0009e67214ff5f5447ce83dd # v6.0.2 with: repository: KhronosGroup/OpenCL-Headers path: external/OpenCL-Headers diff --git a/CMakeLists.txt b/CMakeLists.txt index eded6f3f..f4669f5c 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -165,7 +165,7 @@ else () endif () set (OPENCL_COMPILE_DEFINITIONS - CL_TARGET_OPENCL_VERSION=300 + CL_TARGET_OPENCL_VERSION=310 CL_NO_NON_ICD_DISPATCH_EXTENSION_PROTOTYPES OPENCL_ICD_LOADER_VERSION_MAJOR=3 OPENCL_ICD_LOADER_VERSION_MINOR=0 diff --git a/loader/icd_dispatch_generated.c b/loader/icd_dispatch_generated.c index 2e75b260..43060bc1 100644 --- a/loader/icd_dispatch_generated.c +++ b/loader/icd_dispatch_generated.c @@ -5191,6 +5191,57 @@ static cl_mem CL_API_CALL clCreateImageWithProperties_disp( /////////////////////////////////////////////////////////////////////////////// +CL_API_ENTRY cl_int CL_API_CALL clGetKernelSuggestedLocalWorkSize( + 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) +{ +#if defined(CL_ENABLE_LAYERS) + if (khrFirstLayer) + return khrFirstLayer->dispatch.clGetKernelSuggestedLocalWorkSize( + command_queue, + kernel, + work_dim, + global_work_offset, + global_work_size, + suggested_local_work_size); +#endif // defined(CL_ENABLE_LAYERS) + KHR_ICD_VALIDATE_HANDLE_RETURN_ERROR(command_queue, CL_INVALID_COMMAND_QUEUE); + return KHR_ICD2_DISPATCH(command_queue)->clGetKernelSuggestedLocalWorkSize( + command_queue, + kernel, + work_dim, + global_work_offset, + global_work_size, + suggested_local_work_size); +} + +/////////////////////////////////////////////////////////////////////////////// +#if defined(CL_ENABLE_LAYERS) +static cl_int CL_API_CALL clGetKernelSuggestedLocalWorkSize_disp( + 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) +{ + KHR_ICD_VALIDATE_HANDLE_RETURN_ERROR(command_queue, CL_INVALID_COMMAND_QUEUE); + return KHR_ICD2_DISPATCH(command_queue)->clGetKernelSuggestedLocalWorkSize( + command_queue, + kernel, + work_dim, + global_work_offset, + global_work_size, + suggested_local_work_size); +} +#endif // defined(CL_ENABLE_LAYERS) + +/////////////////////////////////////////////////////////////////////////////// + // cl_ext_device_fission CL_API_ENTRY cl_int CL_API_CALL clReleaseDeviceEXT( @@ -7026,7 +7077,10 @@ const struct _cl_icd_dispatch khrMainDispatch = { /* OpenCL 3.0 */ &clCreateBufferWithProperties_disp, &clCreateImageWithProperties_disp, - &clSetContextDestructorCallback_disp + &clSetContextDestructorCallback_disp, + + /* OpenCL 3.1 */ + &clGetKernelSuggestedLocalWorkSize_disp, } ; #endif // defined(CL_ENABLE_LAYERS) @@ -8621,6 +8675,22 @@ static cl_mem CL_API_CALL clCreateImageWithProperties_unsupp( (void)errcode_ret; KHR_ICD_ERROR_RETURN_HANDLE(CL_INVALID_OPERATION); } +static cl_int CL_API_CALL clGetKernelSuggestedLocalWorkSize_unsupp( + 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) +{ + (void)command_queue; + (void)kernel; + (void)work_dim; + (void)global_work_offset; + (void)global_work_size; + (void)suggested_local_work_size; + KHR_ICD_ERROR_RETURN_ERROR(CL_INVALID_OPERATION); +} /////////////////////////////////////////////////////////////////////////////// // cl_ext_device_fission @@ -9373,7 +9443,10 @@ const struct _cl_icd_dispatch khrDeinitDispatch = { /* OpenCL 3.0 */ &clCreateBufferWithProperties_unsupp, &clCreateImageWithProperties_unsupp, - &clSetContextDestructorCallback_unsupp + &clSetContextDestructorCallback_unsupp, + + /* OpenCL 3.1 */ + &clGetKernelSuggestedLocalWorkSize_unsupp, } ; #endif // defined(CL_ENABLE_LAYERS) @@ -9728,6 +9801,9 @@ void khrIcd2PopulateDispatchTable( dispatch->clCreateImageWithProperties = (clCreateImageWithProperties_t *)(size_t)p_clIcdGetFunctionAddressForPlatform(platform, "clCreateImageWithProperties"); if (!dispatch->clCreateImageWithProperties) dispatch->clCreateImageWithProperties = &clCreateImageWithProperties_unsupp; + dispatch->clGetKernelSuggestedLocalWorkSize = (clGetKernelSuggestedLocalWorkSize_t *)(size_t)p_clIcdGetFunctionAddressForPlatform(platform, "clGetKernelSuggestedLocalWorkSize"); + if (!dispatch->clGetKernelSuggestedLocalWorkSize) + dispatch->clGetKernelSuggestedLocalWorkSize = &clGetKernelSuggestedLocalWorkSize_unsupp; /////////////////////////////////////////////////////////////////////////////// // cl_ext_device_fission diff --git a/loader/icd_version.h b/loader/icd_version.h index 24f4d7ad..9b0436e7 100644 --- a/loader/icd_version.h +++ b/loader/icd_version.h @@ -50,6 +50,9 @@ #if CL_TARGET_OPENCL_VERSION == 300 #define OPENCL_ICD_LOADER_OCL_VERSION_NUMBER "3.0" #endif +#if CL_TARGET_OPENCL_VERSION == 310 +#define OPENCL_ICD_LOADER_OCL_VERSION_NUMBER "3.1" +#endif #define OPENCL_ICD_LOADER_OCL_VERSION_STRING \ "OpenCL " OPENCL_ICD_LOADER_OCL_VERSION_NUMBER diff --git a/loader/linux/icd_exports.map b/loader/linux/icd_exports.map index 978a22d6..e90400c2 100644 --- a/loader/linux/icd_exports.map +++ b/loader/linux/icd_exports.map @@ -170,3 +170,8 @@ OPENCL_3.0 { clCreateImageWithProperties; clSetContextDestructorCallback; } OPENCL_2.2; + +OPENCL_3.1 { + global: + clGetKernelSuggestedLocalWorkSize; +} OPENCL_3.0; diff --git a/loader/windows/OpenCL-mingw-i686.def b/loader/windows/OpenCL-mingw-i686.def index 8ccf942f..8b0d83ca 100644 --- a/loader/windows/OpenCL-mingw-i686.def +++ b/loader/windows/OpenCL-mingw-i686.def @@ -160,3 +160,6 @@ clSetProgramSpecializationConstant@16 == clSetProgramSpecializationConstant clCreateBufferWithProperties@28 == clCreateBufferWithProperties clCreateImageWithProperties@32 == clCreateImageWithProperties clSetContextDestructorCallback@12 == clSetContextDestructorCallback + +; OpenCL 3.1 API +clGetKernelSuggestedLocalWorkSize@24 == clGetKernelSuggestedLocalWorkSize diff --git a/loader/windows/OpenCL.def b/loader/windows/OpenCL.def index 2ae2f357..6490e540 100644 --- a/loader/windows/OpenCL.def +++ b/loader/windows/OpenCL.def @@ -164,3 +164,6 @@ clSetProgramSpecializationConstant clCreateBufferWithProperties clCreateImageWithProperties clSetContextDestructorCallback + +; OpenCL 3.1 API +clGetKernelSuggestedLocalWorkSize diff --git a/scripts/dispatch_table.mako b/scripts/dispatch_table.mako index 4949d919..96c7d740 100644 --- a/scripts/dispatch_table.mako +++ b/scripts/dispatch_table.mako @@ -200,5 +200,8 @@ /* OpenCL 3.0 */ &clCreateBufferWithProperties_${suffix}, &clCreateImageWithProperties_${suffix}, - &clSetContextDestructorCallback_${suffix} + &clSetContextDestructorCallback_${suffix}, + + /* OpenCL 3.1 */ + &clGetKernelSuggestedLocalWorkSize_${suffix}, } diff --git a/test/driver_stub/CMakeLists.txt b/test/driver_stub/CMakeLists.txt index 765ca1e4..e1d44c51 100644 --- a/test/driver_stub/CMakeLists.txt +++ b/test/driver_stub/CMakeLists.txt @@ -9,7 +9,7 @@ add_library (OpenCLDriverStub SHARED ${OPENCL_DRIVER_STUB_SOURCES}) target_link_libraries (OpenCLDriverStub IcdLog OpenCL::Headers) -target_compile_definitions (OpenCLDriverStub PRIVATE CL_TARGET_OPENCL_VERSION=300) +target_compile_definitions (OpenCLDriverStub PRIVATE CL_TARGET_OPENCL_VERSION=310) set (OPENCL_DRIVER_STUB_ICD2_SOURCES cl.c cl_ext.c cl_gl.c icd.c) @@ -21,4 +21,4 @@ add_library (OpenCLDriverStubICD2 SHARED ${OPENCL_DRIVER_STUB_ICD2_SOURCES}) target_link_libraries (OpenCLDriverStubICD2 IcdLog OpenCL::Headers) -target_compile_definitions (OpenCLDriverStubICD2 PRIVATE CL_TARGET_OPENCL_VERSION=300 CL_ENABLE_ICD2=1) +target_compile_definitions (OpenCLDriverStubICD2 PRIVATE CL_TARGET_OPENCL_VERSION=310 CL_ENABLE_ICD2=1) diff --git a/test/driver_stub/cl.c b/test/driver_stub/cl.c index 3e5d134d..9b615a59 100644 --- a/test/driver_stub/cl.c +++ b/test/driver_stub/cl.c @@ -1090,6 +1090,27 @@ clGetKernelInfo(cl_kernel kernel , return return_value; } +CL_API_ENTRY cl_int CL_API_CALL +clGetKernelSuggestedLocalWorkSize(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_API_SUFFIX__VERSION_3_1 +{ + cl_int return_value = CL_OUT_OF_RESOURCES; + test_icd_stub_log("clGetKernelSuggestedLocalWorkSize(%p, %p, %u, %p, %p, %p)\n", + command_queue, + kernel, + work_dim, + global_work_offset, + global_work_size, + suggested_local_work_size); + + test_icd_stub_log("Value returned: %d\n", return_value); + return return_value; +} + CL_API_ENTRY cl_int CL_API_CALL clGetKernelArgInfo(cl_kernel kernel , cl_uint arg_indx , diff --git a/test/driver_stub/icd.c b/test/driver_stub/icd.c index 1c815bbf..116efde6 100644 --- a/test/driver_stub/icd.c +++ b/test/driver_stub/icd.c @@ -248,6 +248,9 @@ cl_int cliIcdDispatchTableCreate(CLIicdDispatchTable **outDispatchTable) ICD_DISPATCH_TABLE_ENTRY ( clCreateImageWithProperties ); ICD_DISPATCH_TABLE_ENTRY ( clSetContextDestructorCallback ); + /* OpenCL 3.1 */ + ICD_DISPATCH_TABLE_ENTRY ( clGetKernelSuggestedLocalWorkSize ); + // return success *outDispatchTable = dispatchTable; return CL_SUCCESS; @@ -471,6 +474,9 @@ void * CL_API_CALL clIcdGetFunctionAddressForPlatformKHR( ICD_GET_FUNCTION_ADDRESS ( clCreateImageWithProperties ); ICD_GET_FUNCTION_ADDRESS ( clSetContextDestructorCallback ); + /* OpenCL 3.1 */ + ICD_GET_FUNCTION_ADDRESS ( clGetKernelSuggestedLocalWorkSize ); + return NULL; } diff --git a/test/driver_stub/rename_api.h b/test/driver_stub/rename_api.h index 555695d4..cd4169ed 100644 --- a/test/driver_stub/rename_api.h +++ b/test/driver_stub/rename_api.h @@ -105,5 +105,6 @@ #define clCreateBufferWithProperties ___clCreateBufferWithProperties #define clCreateImageWithProperties ___clCreateImageWithProperties #define clSetContextDestructorCallback ___clSetContextDestructorCallback +#define clGetKernelSuggestedLocalWorkSize ___clGetKernelSuggestedLocalWorkSize #endif /* __RENAME_API_H__ */ diff --git a/test/layer/CMakeLists.txt b/test/layer/CMakeLists.txt index ea4e0e4b..2006ce0e 100644 --- a/test/layer/CMakeLists.txt +++ b/test/layer/CMakeLists.txt @@ -15,7 +15,7 @@ add_library (PrintLayer SHARED ${OPENCL_PRINT_LAYER_SOURCES}) target_include_directories(PrintLayer PRIVATE ${PARENT_DIR}/include) target_link_libraries(PrintLayer PUBLIC OpenCL::Headers) -target_compile_definitions (PrintLayer PRIVATE CL_TARGET_OPENCL_VERSION=300) +target_compile_definitions (PrintLayer PRIVATE CL_TARGET_OPENCL_VERSION=310) if (NOT WIN32 AND NOT APPLE) set_target_properties (PrintLayer PROPERTIES LINK_FLAGS "-Wl,--version-script -Wl,${CMAKE_CURRENT_SOURCE_DIR}/icd_print_layer.map") diff --git a/test/layer/icd_print_layer_generated.c b/test/layer/icd_print_layer_generated.c index 16652164..a5549cc2 100644 --- a/test/layer/icd_print_layer_generated.c +++ b/test/layer/icd_print_layer_generated.c @@ -1949,6 +1949,25 @@ return tdispatch->clCreateImageWithProperties( errcode_ret); } +/////////////////////////////////////////////////////////////////////////////// +static cl_int CL_API_CALL clGetKernelSuggestedLocalWorkSize_wrap( + 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_API_SUFFIX__VERSION_3_1 +{ +printf("clGetKernelSuggestedLocalWorkSize\n"); +return tdispatch->clGetKernelSuggestedLocalWorkSize( + command_queue, + kernel, + work_dim, + global_work_offset, + global_work_size, + suggested_local_work_size); +} + /////////////////////////////////////////////////////////////////////////////// // cl_ext_device_fission @@ -2742,6 +2761,9 @@ struct _cl_icd_dispatch dispatch = { /* OpenCL 3.0 */ &clCreateBufferWithProperties_wrap, &clCreateImageWithProperties_wrap, - &clSetContextDestructorCallback_wrap + &clSetContextDestructorCallback_wrap, + + /* OpenCL 3.1 */ + &clGetKernelSuggestedLocalWorkSize_wrap, } ; diff --git a/test/loader_test/CMakeLists.txt b/test/loader_test/CMakeLists.txt index 3a401210..f5d645ec 100644 --- a/test/loader_test/CMakeLists.txt +++ b/test/loader_test/CMakeLists.txt @@ -14,7 +14,7 @@ add_executable (icd_loader_test target_compile_definitions(icd_loader_test PRIVATE - CL_TARGET_OPENCL_VERSION=300 + CL_TARGET_OPENCL_VERSION=310 ) target_link_libraries (icd_loader_test OpenCL IcdLog OpenCL::Headers) diff --git a/test/loader_test/param_struct.h b/test/loader_test/param_struct.h index 00d96a97..2b80f505 100644 --- a/test/loader_test/param_struct.h +++ b/test/loader_test/param_struct.h @@ -745,6 +745,7 @@ struct clReleaseKernel_st #define NUM_ITEMS_clGetKernelInfo 1 #define NUM_ITEMS_clGetKernelArgInfo 1 #define NUM_ITEMS_clGetKernelWorkGroupInfo 1 +#define NUM_ITEMS_clGetKernelSuggestedLocalWorkSize 1 struct clSetKernelArg_st { @@ -783,6 +784,16 @@ struct clGetKernelWorkGroupInfo_st size_t *param_value_size_ret; }; +struct clGetKernelSuggestedLocalWorkSize_st +{ + 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; +}; + #define NUM_ITEMS_clEnqueueMigrateMemObjects 1 #define NUM_ITEMS_clEnqueueNDRangeKernel 1 #define NUM_ITEMS_clEnqueueTask 1 diff --git a/test/loader_test/test_kernel.c b/test/loader_test/test_kernel.c index e76c77ff..22e8fc19 100644 --- a/test/loader_test/test_kernel.c +++ b/test/loader_test/test_kernel.c @@ -138,6 +138,33 @@ int test_clGetKernelWorkGroupInfo(const struct clGetKernelWorkGroupInfo_st* data return 0; } +struct clGetKernelSuggestedLocalWorkSize_st clGetKernelSuggestedLocalWorkSizeData[NUM_ITEMS_clGetKernelSuggestedLocalWorkSize] = +{ + {NULL, NULL, 0, NULL, NULL, NULL} +}; + +int test_clGetKernelSuggestedLocalWorkSize(const struct clGetKernelSuggestedLocalWorkSize_st* data) +{ + test_icd_app_log("clGetKernelSuggestedLocalWorkSize(%p, %p, %u, %p, %p, %p)\n", + command_queue, + kernel, + data->work_dim, + data->global_work_offset, + data->global_work_size, + data->suggested_local_work_size); + + ret_val=clGetKernelSuggestedLocalWorkSize(command_queue, + kernel, + data->work_dim, + data->global_work_offset, + data->global_work_size, + data->suggested_local_work_size); + + test_icd_app_log("Value returned: %d\n", ret_val); + + return 0; +} + struct clEnqueueMigrateMemObjects_st clEnqueueMigrateMemObjectsData[NUM_ITEMS_clEnqueueMigrateMemObjects] = { {NULL, 0, NULL, 0x0, 0, NULL, NULL} diff --git a/test/log/CMakeLists.txt b/test/log/CMakeLists.txt index 341d19ed..e3d8f9a6 100644 --- a/test/log/CMakeLists.txt +++ b/test/log/CMakeLists.txt @@ -2,7 +2,7 @@ add_library (IcdLog SHARED icd_test_log.c) target_link_libraries (IcdLog OpenCL::Headers) -target_compile_definitions (IcdLog PRIVATE CL_TARGET_OPENCL_VERSION=300) +target_compile_definitions (IcdLog PRIVATE CL_TARGET_OPENCL_VERSION=310) if (WIN32) target_compile_definitions (IcdLog PRIVATE _CRT_SECURE_NO_WARNINGS) endif() diff --git a/test/plugin_test/CMakeLists.txt b/test/plugin_test/CMakeLists.txt index 96fa5644..c1b70e27 100644 --- a/test/plugin_test/CMakeLists.txt +++ b/test/plugin_test/CMakeLists.txt @@ -12,7 +12,7 @@ if (WIN32) endif () add_library (CLPlugin MODULE ${CL_PLUGIN_SOURCES}) -target_compile_definitions(CLPlugin PRIVATE CL_TARGET_OPENCL_VERSION=300) +target_compile_definitions(CLPlugin PRIVATE CL_TARGET_OPENCL_VERSION=310) target_link_libraries (CLPlugin OpenCL OpenCL::Headers) set (CL_PLUGIN_LOADER_TEST_SOURCES