From 192528c7f26a693d96a2b8e172ae8e9cf64e8e75 Mon Sep 17 00:00:00 2001 From: Jaydeep Patel Date: Mon, 3 Jun 2024 11:00:55 +0000 Subject: [PATCH] SWDEV-465220 - Update Kernel Launch negative cases to match with current CLR code. Change-Id: I891544cdfdc1092467454189ea348f986fc12818 --- catch/hipTestMain/config/config_amd_linux | 3 +- catch/hipTestMain/config/config_amd_windows | 238 +++++++++++++++++- .../executionControl/hipExtLaunchKernel.cc | 10 +- .../unit/executionControl/hipLaunchKernel.cc | 10 +- catch/unit/module/hipExtModuleLaunchKernel.cc | 2 +- .../module/hip_module_launch_kernel_common.hh | 28 ++- 6 files changed, 266 insertions(+), 25 deletions(-) diff --git a/catch/hipTestMain/config/config_amd_linux b/catch/hipTestMain/config/config_amd_linux index 1b279d9f7e..762fdb3488 100644 --- a/catch/hipTestMain/config/config_amd_linux +++ b/catch/hipTestMain/config/config_amd_linux @@ -31,8 +31,6 @@ "NOTE: The following test is disabled due to defect - EXSWHTEC-241", "NOTE: The following test is disabled due to defect - EXSWHTEC-242", "Unit_hipFuncGetAttributes_Positive_Basic", - "NOTE: The following test is disabled due to defect - EXSWHTEC-243", - "Unit_hipExtLaunchKernel_Negative_Parameters", "NOTE: The following test is disabled due to defect - EXSWHTEC-244", "Unit_hipExtLaunchMultiKernelMultiDevice_Negative_Parameters", "Unit_hipOccupancyMaxActiveBlocksPerMultiprocessor_Negative_Parameters", @@ -213,6 +211,7 @@ "Unit_hipDrvGraphAddMemsetNode_hipMallocManaged", "Unit_hipExtModuleLaunchKernel_Negative_Parameters", "Unit_hipLaunchKernel_Negative_Parameters", + "Unit_hipModuleLaunchCooperativeKernel_Negative_Parameters", "Unit_Device_modf_modff_Negative_RTC", "SWDEV-446588 - Disable graph multi gpu testcases until graph has support for it", "Unit_hipGraphExecUpdate_Negative_MultiDevice_Context_Changed", diff --git a/catch/hipTestMain/config/config_amd_windows b/catch/hipTestMain/config/config_amd_windows index 91ba6d2a8a..db9c3ff75d 100644 --- a/catch/hipTestMain/config/config_amd_windows +++ b/catch/hipTestMain/config/config_amd_windows @@ -213,7 +213,6 @@ "NOTE: The following test is disabled due to defect - EXSWHTEC-242", "Unit_hipFuncGetAttributes_Positive_Basic", "NOTE: The following test is disabled due to defect - EXSWHTEC-243", - "Unit_hipExtLaunchKernel_Negative_Parameters", "NOTE: The following test is disabled due to defect - EXSWHTEC-244", "Unit_hipExtLaunchMultiKernelMultiDevice_Negative_Parameters", "Unit_hipMemAddressFree_negative", @@ -355,6 +354,243 @@ "Unit_hipModuleLaunchKernel_Negative_Parameters", "Unit_hipExtModuleLaunchKernel_Negative_Parameters", "Unit_hipLaunchKernel_Negative_Parameters", + "Unit_tex1Dfetch_Positive_ReadModeElementType - char", + "Unit_tex1Dfetch_Positive_ReadModeElementType - unsigned char", + "Unit_tex1Dfetch_Positive_ReadModeElementType - short", + "Unit_tex1Dfetch_Positive_ReadModeElementType - unsigned short", + "Unit_tex1Dfetch_Positive_ReadModeElementType - int", + "Unit_tex1Dfetch_Positive_ReadModeElementType - unsigned int", + "Unit_tex1Dfetch_Positive_ReadModeElementType - float", + "Unit_tex1Dfetch_Positive_ReadModeNormalizedFloat - char", + "Unit_tex1Dfetch_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_tex1Dfetch_Positive_ReadModeNormalizedFloat - short", + "Unit_tex1Dfetch_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_tex1D_Positive_ReadModeNormalizedFloat - char", + "Unit_tex1D_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_tex1D_Positive_ReadModeNormalizedFloat - short", + "Unit_tex1D_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_tex1DLayered_Positive_ReadModeNormalizedFloat - char", + "Unit_tex1DLayered_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_tex1DLayered_Positive_ReadModeNormalizedFloat - short", + "Unit_tex1DLayered_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_tex1DGrad_Positive_ReadModeElementType - char", + "Unit_tex1DGrad_Positive_ReadModeElementType - unsigned char", + "Unit_tex1DGrad_Positive_ReadModeElementType - short", + "Unit_tex1DGrad_Positive_ReadModeElementType - unsigned short", + "Unit_tex1DGrad_Positive_ReadModeElementType - int", + "Unit_tex1DGrad_Positive_ReadModeElementType - unsigned int", + "Unit_tex1DGrad_Positive_ReadModeElementType - float", + "Unit_tex1DGrad_Positive_ReadModeNormalizedFloat - char", + "Unit_tex1DGrad_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_tex1DGrad_Positive_ReadModeNormalizedFloat - short", + "Unit_tex1DGrad_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_tex1DLayeredGrad_Positive_ReadModeElementType - char", + "Unit_tex1DLayeredGrad_Positive_ReadModeElementType - unsigned char", + "Unit_tex1DLayeredGrad_Positive_ReadModeElementType - short", + "Unit_tex1DLayeredGrad_Positive_ReadModeElementType - unsigned short", + "Unit_tex1DLayeredGrad_Positive_ReadModeElementType - int", + "Unit_tex1DLayeredGrad_Positive_ReadModeElementType - unsigned int", + "Unit_tex1DLayeredGrad_Positive_ReadModeElementType - float", + "Unit_tex1DLayeredGrad_Positive_ReadModeNormalizedFloat - char", + "Unit_tex1DLayeredGrad_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_tex1DLayeredGrad_Positive_ReadModeNormalizedFloat - short", + "Unit_tex1DLayeredGrad_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_tex1DLayeredLod_Positive_ReadModeElementType - char", + "Unit_tex1DLayeredLod_Positive_ReadModeElementType - unsigned char", + "Unit_tex1DLayeredLod_Positive_ReadModeElementType - short", + "Unit_tex1DLayeredLod_Positive_ReadModeElementType - unsigned short", + "Unit_tex1DLayeredLod_Positive_ReadModeElementType - int", + "Unit_tex1DLayeredLod_Positive_ReadModeElementType - unsigned int", + "Unit_tex1DLayeredLod_Positive_ReadModeElementType - float", + "Unit_tex1DLayeredLod_Positive_ReadModeNormalizedFloat - char", + "Unit_tex1DLayeredLod_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_tex1DLayeredLod_Positive_ReadModeNormalizedFloat - short", + "Unit_tex1DLayeredLod_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_tex1DLod_Positive_ReadModeElementType - char", + "Unit_tex1DLod_Positive_ReadModeElementType - unsigned char", + "Unit_tex1DLod_Positive_ReadModeElementType - short", + "Unit_tex1DLod_Positive_ReadModeElementType - unsigned short", + "Unit_tex1DLod_Positive_ReadModeElementType - int", + "Unit_tex1DLod_Positive_ReadModeElementType - unsigned int", + "Unit_tex1DLod_Positive_ReadModeElementType - float", + "Unit_tex1DLod_Positive_ReadModeNormalizedFloat - char", + "Unit_tex1DLod_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_tex1DLod_Positive_ReadModeNormalizedFloat - short", + "Unit_tex1DLod_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_tex3D_Positive_ReadModeElementType - char", + "Unit_tex3D_Positive_ReadModeElementType - unsigned char", + "Unit_tex3D_Positive_ReadModeElementType - short", + "Unit_tex3D_Positive_ReadModeElementType - unsigned short", + "Unit_tex3D_Positive_ReadModeElementType - int", + "Unit_tex3D_Positive_ReadModeElementType - unsigned int", + "Unit_tex3D_Positive_ReadModeElementType - float", + "Unit_tex3D_Positive_ReadModeNormalizedFloat - char", + "Unit_tex3D_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_tex3D_Positive_ReadModeNormalizedFloat - short", + "Unit_tex3D_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_tex3DLod_Positive_ReadModeElementType - char", + "Unit_tex3DLod_Positive_ReadModeElementType - unsigned char", + "Unit_tex3DLod_Positive_ReadModeElementType - short", + "Unit_tex3DLod_Positive_ReadModeElementType - unsigned short", + "Unit_tex3DLod_Positive_ReadModeElementType - int", + "Unit_tex3DLod_Positive_ReadModeElementType - unsigned int", + "Unit_tex3DLod_Positive_ReadModeElementType - float", + "Unit_tex3DLod_Positive_ReadModeNormalizedFloat - char", + "Unit_tex3DLod_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_tex3DLod_Positive_ReadModeNormalizedFloat - short", + "Unit_tex3DLod_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_tex3DGrad_Positive_ReadModeElementType - char", + "Unit_tex3DGrad_Positive_ReadModeElementType - unsigned char", + "Unit_tex3DGrad_Positive_ReadModeElementType - short", + "Unit_tex3DGrad_Positive_ReadModeElementType - unsigned short", + "Unit_tex3DGrad_Positive_ReadModeElementType - int", + "Unit_tex3DGrad_Positive_ReadModeElementType - unsigned int", + "Unit_tex3DGrad_Positive_ReadModeElementType - float", + "Unit_tex3DGrad_Positive_ReadModeNormalizedFloat - char", + "Unit_tex3DGrad_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_tex3DGrad_Positive_ReadModeNormalizedFloat - short", + "Unit_tex3DGrad_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_texCubemap_Positive_ReadModeElementType - char", + "Unit_texCubemap_Positive_ReadModeElementType - unsigned char", + "Unit_texCubemap_Positive_ReadModeElementType - short", + "Unit_texCubemap_Positive_ReadModeElementType - unsigned short", + "Unit_texCubemap_Positive_ReadModeElementType - int", + "Unit_texCubemap_Positive_ReadModeElementType - unsigned int", + "Unit_texCubemap_Positive_ReadModeElementType - float", + "Unit_texCubemap_Positive_ReadModeNormalizedFloat - char", + "Unit_texCubemap_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_texCubemap_Positive_ReadModeNormalizedFloat - short", + "Unit_texCubemap_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_texCubemapLod_Positive_ReadModeElementType - char", + "Unit_texCubemapLod_Positive_ReadModeElementType - unsigned char", + "Unit_texCubemapLod_Positive_ReadModeElementType - short", + "Unit_texCubemapLod_Positive_ReadModeElementType - unsigned short", + "Unit_texCubemapLod_Positive_ReadModeElementType - int", + "Unit_texCubemapLod_Positive_ReadModeElementType - unsigned int", + "Unit_texCubemapLod_Positive_ReadModeElementType - float", + "Unit_texCubemapLod_Positive_ReadModeNormalizedFloat - char", + "Unit_texCubemapLod_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_texCubemapLod_Positive_ReadModeNormalizedFloat - short", + "Unit_texCubemapLod_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_texCubemapGrad_Positive_ReadModeElementType - char", + "Unit_texCubemapGrad_Positive_ReadModeElementType - unsigned char", + "Unit_texCubemapGrad_Positive_ReadModeElementType - short", + "Unit_texCubemapGrad_Positive_ReadModeElementType - unsigned short", + "Unit_texCubemapGrad_Positive_ReadModeElementType - int", + "Unit_texCubemapGrad_Positive_ReadModeElementType - unsigned int", + "Unit_texCubemapGrad_Positive_ReadModeElementType - float", + "Unit_texCubemapGrad_Positive_ReadModeNormalizedFloat - char", + "Unit_texCubemapGrad_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_texCubemapGrad_Positive_ReadModeNormalizedFloat - short", + "Unit_texCubemapGrad_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_texCubemapLayered_Positive_ReadModeElementType - char", + "Unit_texCubemapLayered_Positive_ReadModeElementType - unsigned char", + "Unit_texCubemapLayered_Positive_ReadModeElementType - short", + "Unit_texCubemapLayered_Positive_ReadModeElementType - unsigned short", + "Unit_texCubemapLayered_Positive_ReadModeElementType - int", + "Unit_texCubemapLayered_Positive_ReadModeElementType - unsigned int", + "Unit_texCubemapLayered_Positive_ReadModeElementType - float", + "Unit_texCubemapLayered_Positive_ReadModeNormalizedFloat - char", + "Unit_texCubemapLayered_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_texCubemapLayered_Positive_ReadModeNormalizedFloat - short", + "Unit_texCubemapLayered_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_texCubemapLayeredLod_Positive_ReadModeElementType - char", + "Unit_texCubemapLayeredLod_Positive_ReadModeElementType - unsigned char", + "Unit_texCubemapLayeredLod_Positive_ReadModeElementType - short", + "Unit_texCubemapLayeredLod_Positive_ReadModeElementType - unsigned short", + "Unit_texCubemapLayeredLod_Positive_ReadModeElementType - int", + "Unit_texCubemapLayeredLod_Positive_ReadModeElementType - unsigned int", + "Unit_texCubemapLayeredLod_Positive_ReadModeElementType - float", + "Unit_texCubemapLayeredLod_Positive_ReadModeNormalizedFloat - char", + "Unit_texCubemapLayeredLod_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_texCubemapLayeredLod_Positive_ReadModeNormalizedFloat - short", + "Unit_texCubemapLayeredLod_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_texCubemapLayeredGrad_Positive_ReadModeElementType - char", + "Unit_texCubemapLayeredGrad_Positive_ReadModeElementType - unsigned char", + "Unit_texCubemapLayeredGrad_Positive_ReadModeElementType - short", + "Unit_texCubemapLayeredGrad_Positive_ReadModeElementType - unsigned short", + "Unit_texCubemapLayeredGrad_Positive_ReadModeElementType - int", + "Unit_texCubemapLayeredGrad_Positive_ReadModeElementType - unsigned int", + "Unit_texCubemapLayeredGrad_Positive_ReadModeElementType - float", + "Unit_texCubemapLayeredGrad_Positive_ReadModeNormalizedFloat - char", + "Unit_texCubemapLayeredGrad_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_texCubemapLayeredGrad_Positive_ReadModeNormalizedFloat - short", + "Unit_texCubemapLayeredGrad_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_tex2Dgather_Positive_ReadModeElementType - char", + "Unit_tex2Dgather_Positive_ReadModeElementType - unsigned char", + "Unit_tex2Dgather_Positive_ReadModeElementType - short", + "Unit_tex2Dgather_Positive_ReadModeElementType - unsigned short", + "Unit_tex2Dgather_Positive_ReadModeElementType - int", + "Unit_tex2Dgather_Positive_ReadModeElementType - unsigned int", + "Unit_tex2Dgather_Positive_ReadModeElementType - float", + "Unit_tex2D_Positive_ReadModeElementType - char", + "Unit_tex2D_Positive_ReadModeElementType - unsigned char", + "Unit_tex2D_Positive_ReadModeElementType - short", + "Unit_tex2D_Positive_ReadModeElementType - unsigned short", + "Unit_tex2D_Positive_ReadModeElementType - int", + "Unit_tex2D_Positive_ReadModeElementType - unsigned int", + "Unit_tex2D_Positive_ReadModeElementType - float", + "Unit_tex2D_Positive_ReadModeNormalizedFloat - char", + "Unit_tex2D_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_tex2D_Positive_ReadModeNormalizedFloat - short", + "Unit_tex2D_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_tex2DLayered_Positive_ReadModeElementType - char", + "Unit_tex2DLayered_Positive_ReadModeElementType - unsigned char", + "Unit_tex2DLayered_Positive_ReadModeElementType - short", + "Unit_tex2DLayered_Positive_ReadModeElementType - unsigned short", + "Unit_tex2DLayered_Positive_ReadModeElementType - int", + "Unit_tex2DLayered_Positive_ReadModeElementType - unsigned int", + "Unit_tex2DLayered_Positive_ReadModeElementType - float", + "Unit_tex2DLayered_Positive_ReadModeNormalizedFloat - char", + "Unit_tex2DLayered_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_tex2DLayered_Positive_ReadModeNormalizedFloat - short", + "Unit_tex2DLayered_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_tex2DGrad_Positive_ReadModeElementType - char", + "Unit_tex2DGrad_Positive_ReadModeElementType - unsigned char", + "Unit_tex2DGrad_Positive_ReadModeElementType - short", + "Unit_tex2DGrad_Positive_ReadModeElementType - unsigned short", + "Unit_tex2DGrad_Positive_ReadModeElementType - int", + "Unit_tex2DGrad_Positive_ReadModeElementType - unsigned int", + "Unit_tex2DGrad_Positive_ReadModeElementType - float", + "Unit_tex2DGrad_Positive_ReadModeNormalizedFloat - char", + "Unit_tex2DGrad_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_tex2DGrad_Positive_ReadModeNormalizedFloat - short", + "Unit_tex2DGrad_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_tex2DLayeredGrad_Positive_ReadModeElementType - char", + "Unit_tex2DLayeredGrad_Positive_ReadModeElementType - unsigned char", + "Unit_tex2DLayeredGrad_Positive_ReadModeElementType - short", + "Unit_tex2DLayeredGrad_Positive_ReadModeElementType - unsigned short", + "Unit_tex2DLayeredGrad_Positive_ReadModeElementType - int", + "Unit_tex2DLayeredGrad_Positive_ReadModeElementType - unsigned int", + "Unit_tex2DLayeredGrad_Positive_ReadModeElementType - float", + "Unit_tex2DLayeredGrad_Positive_ReadModeNormalizedFloat - char", + "Unit_tex2DLayeredGrad_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_tex2DLayeredGrad_Positive_ReadModeNormalizedFloat - short", + "Unit_tex2DLayeredGrad_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_tex2DLod_Positive_ReadModeElementType - char", + "Unit_tex2DLod_Positive_ReadModeElementType - unsigned char", + "Unit_tex2DLod_Positive_ReadModeElementType - short", + "Unit_tex2DLod_Positive_ReadModeElementType - unsigned short", + "Unit_tex2DLod_Positive_ReadModeElementType - int", + "Unit_tex2DLod_Positive_ReadModeElementType - unsigned int", + "Unit_tex2DLod_Positive_ReadModeElementType - float", + "Unit_tex2DLod_Positive_ReadModeNormalizedFloat - char", + "Unit_tex2DLod_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_tex2DLod_Positive_ReadModeNormalizedFloat - short", + "Unit_tex2DLod_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_tex2DLayeredLod_Positive_ReadModeElementType - char", + "Unit_tex2DLayeredLod_Positive_ReadModeElementType - unsigned char", + "Unit_tex2DLayeredLod_Positive_ReadModeElementType - short", + "Unit_tex2DLayeredLod_Positive_ReadModeElementType - unsigned short", + "Unit_tex2DLayeredLod_Positive_ReadModeElementType - int", + "Unit_tex2DLayeredLod_Positive_ReadModeElementType - unsigned int", + "Unit_tex2DLayeredLod_Positive_ReadModeElementType - float", + "Unit_tex2DLayeredLod_Positive_ReadModeNormalizedFloat - char", + "Unit_tex2DLayeredLod_Positive_ReadModeNormalizedFloat - unsigned char", + "Unit_tex2DLayeredLod_Positive_ReadModeNormalizedFloat - short", + "Unit_tex2DLayeredLod_Positive_ReadModeNormalizedFloat - unsigned short", + "Unit_hipModuleLaunchKernel_Negative_Parameters", + "Unit_hipModuleGetTexRef_Positive_Basic", "Unit_Kernel_Launch_bounds_Negative_OutOfBounds", "Unit_Kernel_Launch_bounds_Negative_Parameters_RTC", "Unit_AtomicBuiltins_Negative_Parameters_RTC", diff --git a/catch/unit/executionControl/hipExtLaunchKernel.cc b/catch/unit/executionControl/hipExtLaunchKernel.cc index c0b4206733..db2118c71b 100644 --- a/catch/unit/executionControl/hipExtLaunchKernel.cc +++ b/catch/unit/executionControl/hipExtLaunchKernel.cc @@ -114,21 +114,21 @@ TEST_CASE("Unit_hipExtLaunchKernel_Negative_Parameters") { const unsigned int x = GetDeviceAttribute(hipDeviceAttributeMaxBlockDimX, 0) + 1u; HIP_CHECK_ERROR(hipExtLaunchKernel(reinterpret_cast(kernel), dim3{1, 1, 1}, dim3{x, 1, 1}, nullptr, 0, nullptr, nullptr, nullptr, 0u), - hipErrorInvalidConfiguration); + hipErrorInvalidValue); } SECTION("blockDim.y > maxBlockDimY") { const unsigned int y = GetDeviceAttribute(hipDeviceAttributeMaxBlockDimY, 0) + 1u; HIP_CHECK_ERROR(hipExtLaunchKernel(reinterpret_cast(kernel), dim3{1, 1, 1}, dim3{1, y, 1}, nullptr, 0, nullptr, nullptr, nullptr, 0u), - hipErrorInvalidConfiguration); + hipErrorInvalidValue); } SECTION("blockDim.z > maxBlockDimZ") { const unsigned int z = GetDeviceAttribute(hipDeviceAttributeMaxBlockDimZ, 0) + 1u; HIP_CHECK_ERROR(hipExtLaunchKernel(reinterpret_cast(kernel), dim3{1, 1, 1}, dim3{1, 1, z}, nullptr, 0, nullptr, nullptr, nullptr, 0u), - hipErrorInvalidConfiguration); + hipErrorInvalidValue); } SECTION("blockDim.x * blockDim.y * blockDim.z > maxThreadsPerBlock") { @@ -137,14 +137,14 @@ TEST_CASE("Unit_hipExtLaunchKernel_Negative_Parameters") { HIP_CHECK_ERROR( hipExtLaunchKernel(reinterpret_cast(kernel), dim3{1, 1, 1}, dim3{dim, dim, dim}, nullptr, 0, nullptr, nullptr, nullptr, 0u), - hipErrorInvalidConfiguration); + hipErrorInvalidValue); } SECTION("sharedMemBytes > maxSharedMemoryPerBlock") { const unsigned int max = GetDeviceAttribute(hipDeviceAttributeMaxSharedMemoryPerBlock, 0) + 1u; HIP_CHECK_ERROR(hipExtLaunchKernel(reinterpret_cast(kernel), dim3{1, 1, 1}, dim3{1, 1, 1}, nullptr, max, nullptr, nullptr, nullptr, 0u), - hipErrorOutOfMemory); + hipErrorInvalidValue); } SECTION("Invalid stream") { diff --git a/catch/unit/executionControl/hipLaunchKernel.cc b/catch/unit/executionControl/hipLaunchKernel.cc index d9272107eb..0dc072aed9 100644 --- a/catch/unit/executionControl/hipLaunchKernel.cc +++ b/catch/unit/executionControl/hipLaunchKernel.cc @@ -113,21 +113,21 @@ TEST_CASE("Unit_hipLaunchKernel_Negative_Parameters") { const unsigned int x = GetDeviceAttribute(hipDeviceAttributeMaxBlockDimX, 0) + 1u; HIP_CHECK_ERROR(hipLaunchKernel(reinterpret_cast(kernel), dim3{1, 1, 1}, dim3{x, 1, 1}, nullptr, 0, nullptr), - hipErrorInvalidConfiguration); + hipErrorInvalidValue); } SECTION("blockDim.y > maxBlockDimY") { const unsigned int y = GetDeviceAttribute(hipDeviceAttributeMaxBlockDimY, 0) + 1u; HIP_CHECK_ERROR(hipLaunchKernel(reinterpret_cast(kernel), dim3{1, 1, 1}, dim3{1, y, 1}, nullptr, 0, nullptr), - hipErrorInvalidConfiguration); + hipErrorInvalidValue); } SECTION("blockDim.z > maxBlockDimZ") { const unsigned int z = GetDeviceAttribute(hipDeviceAttributeMaxBlockDimZ, 0) + 1u; HIP_CHECK_ERROR(hipLaunchKernel(reinterpret_cast(kernel), dim3{1, 1, 1}, dim3{1, 1, z}, nullptr, 0, nullptr), - hipErrorInvalidConfiguration); + hipErrorInvalidValue); } SECTION("blockDim.x * blockDim.y * blockDim.z > maxThreadsPerBlock") { @@ -135,14 +135,14 @@ TEST_CASE("Unit_hipLaunchKernel_Negative_Parameters") { const unsigned int dim = std::ceil(std::cbrt(max)); HIP_CHECK_ERROR(hipLaunchKernel(reinterpret_cast(kernel), dim3{1, 1, 1}, dim3{dim, dim, dim}, nullptr, 0, nullptr), - hipErrorInvalidConfiguration); + hipErrorInvalidValue); } SECTION("sharedMemBytes > maxSharedMemoryPerBlock") { const unsigned int max = GetDeviceAttribute(hipDeviceAttributeMaxSharedMemoryPerBlock, 0) + 1u; HIP_CHECK_ERROR(hipLaunchKernel(reinterpret_cast(kernel), dim3{1, 1, 1}, dim3{1, 1, 1}, nullptr, max, nullptr), - hipErrorOutOfMemory); + hipErrorInvalidValue); } SECTION("Invalid stream") { diff --git a/catch/unit/module/hipExtModuleLaunchKernel.cc b/catch/unit/module/hipExtModuleLaunchKernel.cc index 2b3e9bc447..5121b17667 100644 --- a/catch/unit/module/hipExtModuleLaunchKernel.cc +++ b/catch/unit/module/hipExtModuleLaunchKernel.cc @@ -271,7 +271,7 @@ TEST_CASE("Unit_hipExtModuleLaunchKernel_Positive_Parameters") { } TEST_CASE("Unit_hipExtModuleLaunchKernel_Negative_Parameters") { - ModuleLaunchKernelNegativeParameters(); + ModuleLaunchKernelNegativeParameters(true); } /** * Test Description diff --git a/catch/unit/module/hip_module_launch_kernel_common.hh b/catch/unit/module/hip_module_launch_kernel_common.hh index 94f1fd57a3..56de72fea7 100644 --- a/catch/unit/module/hip_module_launch_kernel_common.hh +++ b/catch/unit/module/hip_module_launch_kernel_common.hh @@ -118,8 +118,13 @@ template void ModuleLaunchKernelPositiveParamet } } -template void ModuleLaunchKernelNegativeParameters() { +template void ModuleLaunchKernelNegativeParameters( + bool extLaunch = false) { hipFunction_t f = GetKernel(mg.module(), "NOPKernel"); + hipError_t expectedErrorZeroBlockDim = (extLaunch == true) ? hipErrorInvalidConfiguration + : hipErrorInvalidValue; + hipError_t expectedErrorOverCapacityGridDim = (extLaunch == true) ? hipSuccess + : hipErrorInvalidValue; SECTION("f == nullptr") { HIP_CHECK_ERROR( @@ -144,35 +149,35 @@ template void ModuleLaunchKernelNegativeParamet SECTION("blockDimX == 0") { HIP_CHECK_ERROR(func(f, 1, 1, 1, 0, 1, 1, 0, nullptr, nullptr, nullptr, nullptr, nullptr, 0u), - hipErrorInvalidValue); + expectedErrorZeroBlockDim); } SECTION("blockDimY == 0") { HIP_CHECK_ERROR(func(f, 1, 1, 1, 1, 0, 1, 0, nullptr, nullptr, nullptr, nullptr, nullptr, 0u), - hipErrorInvalidValue); + expectedErrorZeroBlockDim); } SECTION("blockDimZ == 0") { HIP_CHECK_ERROR(func(f, 1, 1, 1, 1, 1, 0, 0, nullptr, nullptr, nullptr, nullptr, nullptr, 0u), - hipErrorInvalidValue); + expectedErrorZeroBlockDim); } SECTION("gridDimX > maxGridDimX") { const unsigned int x = GetDeviceAttribute(hipDeviceAttributeMaxGridDimX, 0) + 1u; HIP_CHECK_ERROR(func(f, x, 1, 1, 1, 1, 1, 0, nullptr, nullptr, nullptr, nullptr, nullptr, 0u), - hipErrorInvalidValue); + expectedErrorOverCapacityGridDim); } SECTION("gridDimY > maxGridDimY") { const unsigned int y = GetDeviceAttribute(hipDeviceAttributeMaxGridDimY, 0) + 1u; HIP_CHECK_ERROR(func(f, 1, y, 1, 1, 1, 1, 0, nullptr, nullptr, nullptr, nullptr, nullptr, 0u), - hipErrorInvalidValue); + expectedErrorOverCapacityGridDim); } SECTION("gridDimZ > maxGridDimZ") { const unsigned int z = GetDeviceAttribute(hipDeviceAttributeMaxGridDimZ, 0) + 1u; HIP_CHECK_ERROR(func(f, 1, 1, z, 1, 1, 1, 0, nullptr, nullptr, nullptr, nullptr, nullptr, 0u), - hipErrorInvalidValue); + expectedErrorOverCapacityGridDim); } SECTION("blockDimX > maxBlockDimX") { @@ -207,16 +212,17 @@ template void ModuleLaunchKernelNegativeParamet hipErrorInvalidValue); } -// Disabled on AMD due to defect - EXSWHTEC-160 -#if HT_NVIDIA SECTION("Invalid stream") { hipStream_t stream = nullptr; HIP_CHECK(hipStreamCreate(&stream)); HIP_CHECK(hipStreamDestroy(stream)); + hipError_t err = hipErrorInvalidValue; + #if HT_NVIDIA + err = hipErrorContextIsDestroyed; + #endif HIP_CHECK_ERROR(func(f, 1, 1, 1, 1, 1, 0, 0, stream, nullptr, nullptr, nullptr, nullptr, 0u), - hipErrorContextIsDestroyed); + err); } -#endif SECTION("Passing kernel_args and extra simultaneously") { hipFunction_t f = GetKernel(mg.module(), "Kernel42");