diff --git a/projects/hip-tests/catch/hipTestMain/config/config_amd_linux b/projects/hip-tests/catch/hipTestMain/config/config_amd_linux index 35c2c27603..2d4079a23f 100644 --- a/projects/hip-tests/catch/hipTestMain/config/config_amd_linux +++ b/projects/hip-tests/catch/hipTestMain/config/config_amd_linux @@ -192,24 +192,12 @@ "Unit_hipModuleOccupancyMaxActiveBlocksPerMultiprocessor_Positive_RangeValidation", "Unit_hipModuleOccupancyMaxActiveBlocksPerMultiprocessorWithFlags_Positive_RangeValidation", "=== SWDEV-435667: Below tests failing randomly in stress test on 01/12/23 ===", - "Unit_atomicExch_Positive_Multi_Kernel - int", - "Unit_atomicExch_Positive_Multi_Kernel - unsigned int", - "Unit_atomicExch_Positive_Multi_Kernel - unsigned long", - "Unit_atomicExch_Positive_Multi_Kernel - unsigned long long", - "Unit_atomicExch_Positive_Multi_Kernel - float", - "Unit_atomicExch_Positive_Multi_Kernel - double", "Unit_atomicExch_system_Positive_Peer_GPUs - int", "Unit_atomicExch_system_Positive_Peer_GPUs - unsigned int", "Unit_atomicExch_system_Positive_Peer_GPUs - unsigned long", "Unit_atomicExch_system_Positive_Peer_GPUs - unsigned long long", "Unit_atomicExch_system_Positive_Peer_GPUs - float", "Unit_atomicExch_system_Positive_Peer_GPUs - double", - "Unit_atomicExch_system_Positive_Host_And_GPU - int", - "Unit_atomicExch_system_Positive_Host_And_GPU - unsigned int", - "Unit_atomicExch_system_Positive_Host_And_GPU - unsigned long", - "Unit_atomicExch_system_Positive_Host_And_GPU - unsigned long long", - "Unit_atomicExch_system_Positive_Host_And_GPU - float", - "Unit_atomicExch_system_Positive_Host_And_GPU - double", "Unit_atomicExch_system_Positive_Host_And_Peer_GPUs - int", "Unit_atomicExch_system_Positive_Host_And_Peer_GPUs - unsigned int", "Unit_atomicExch_system_Positive_Host_And_Peer_GPUs - unsigned long", @@ -648,12 +636,6 @@ "Unit___hip_atomic_fetch_max_Positive_Workgroup_Scattered_Addresses - unsigned long long", "Unit___hip_atomic_fetch_max_Positive_Workgroup_Scattered_Addresses - float", "Unit___hip_atomic_fetch_max_Positive_Workgroup_Scattered_Addresses - double", - "Unit_atomicExch_Positive - int", - "Unit_atomicExch_Positive - unsigned int", - "Unit_atomicExch_Positive - unsigned long", - "Unit_atomicExch_Positive - unsigned long long", - "Unit_atomicExch_Positive - float", - "Unit_atomicExch_Positive - double", "Unit___hip_atomic_fetch_and_Positive_Wavefront_SameAddress - int", "Unit___hip_atomic_fetch_and_Positive_Wavefront_SameAddress - unsigned int", "Unit___hip_atomic_fetch_and_Positive_Wavefront_SameAddress - unsigned long", @@ -726,18 +708,6 @@ "Unit___hip_atomic_fetch_xor_Positive_Workgroup_Scattered_Addresses - unsigned int", "Unit___hip_atomic_fetch_xor_Positive_Workgroup_Scattered_Addresses - unsigned long", "Unit___hip_atomic_fetch_xor_Positive_Workgroup_Scattered_Addresses - unsigned long long", - "Unit___hip_atomic_exchange_Positive_Wavefront - int", - "Unit___hip_atomic_exchange_Positive_Wavefront - unsigned int", - "Unit___hip_atomic_exchange_Positive_Wavefront - unsigned long", - "Unit___hip_atomic_exchange_Positive_Wavefront - unsigned long long", - "Unit___hip_atomic_exchange_Positive_Wavefront - float", - "Unit___hip_atomic_exchange_Positive_Wavefront - double", - "Unit___hip_atomic_exchange_Positive_Workgroup - int", - "Unit___hip_atomic_exchange_Positive_Workgroup - unsigned int", - "Unit___hip_atomic_exchange_Positive_Workgroup - unsigned long", - "Unit___hip_atomic_exchange_Positive_Workgroup - unsigned long long", - "Unit___hip_atomic_exchange_Positive_Workgroup - float", - "Unit___hip_atomic_exchange_Positive_Workgroup - double", "Unit_Kernel_Launch_bounds_Negative_OutOfBounds", "Unit_Kernel_Launch_bounds_Negative_Parameters_RTC", "Unit___threadfence_block_Positive_Basic_Peer", diff --git a/projects/hip-tests/catch/hipTestMain/config/config_amd_windows b/projects/hip-tests/catch/hipTestMain/config/config_amd_windows index 7497f4a35a..7697feba50 100644 --- a/projects/hip-tests/catch/hipTestMain/config/config_amd_windows +++ b/projects/hip-tests/catch/hipTestMain/config/config_amd_windows @@ -280,24 +280,12 @@ "Unit_hipModuleOccupancyMaxActiveBlocksPerMultiprocessor_Positive_RangeValidation", "Unit_hipModuleOccupancyMaxActiveBlocksPerMultiprocessorWithFlags_Positive_RangeValidation", "=== SWDEV-435667: Below tests failing randomly in stress test on 01/12/23 ===", - "Unit_atomicExch_Positive_Multi_Kernel - int", - "Unit_atomicExch_Positive_Multi_Kernel - unsigned int", - "Unit_atomicExch_Positive_Multi_Kernel - unsigned long", - "Unit_atomicExch_Positive_Multi_Kernel - unsigned long long", - "Unit_atomicExch_Positive_Multi_Kernel - float", - "Unit_atomicExch_Positive_Multi_Kernel - double", "Unit_atomicExch_system_Positive_Peer_GPUs - int", "Unit_atomicExch_system_Positive_Peer_GPUs - unsigned int", "Unit_atomicExch_system_Positive_Peer_GPUs - unsigned long", "Unit_atomicExch_system_Positive_Peer_GPUs - unsigned long long", "Unit_atomicExch_system_Positive_Peer_GPUs - float", "Unit_atomicExch_system_Positive_Peer_GPUs - double", - "Unit_atomicExch_system_Positive_Host_And_GPU - int", - "Unit_atomicExch_system_Positive_Host_And_GPU - unsigned int", - "Unit_atomicExch_system_Positive_Host_And_GPU - unsigned long", - "Unit_atomicExch_system_Positive_Host_And_GPU - unsigned long long", - "Unit_atomicExch_system_Positive_Host_And_GPU - float", - "Unit_atomicExch_system_Positive_Host_And_GPU - double", "Unit_atomicExch_system_Positive_Host_And_Peer_GPUs - int", "Unit_atomicExch_system_Positive_Host_And_Peer_GPUs - unsigned int", "Unit_atomicExch_system_Positive_Host_And_Peer_GPUs - unsigned long", @@ -1007,12 +995,6 @@ "Unit___hip_atomic_fetch_max_Positive_Workgroup_Scattered_Addresses - unsigned long long", "Unit___hip_atomic_fetch_max_Positive_Workgroup_Scattered_Addresses - float", "Unit___hip_atomic_fetch_max_Positive_Workgroup_Scattered_Addresses - double", - "Unit_atomicExch_Positive - int", - "Unit_atomicExch_Positive - unsigned int", - "Unit_atomicExch_Positive - unsigned long", - "Unit_atomicExch_Positive - unsigned long long", - "Unit_atomicExch_Positive - float", - "Unit_atomicExch_Positive - double", "Unit___hip_atomic_fetch_and_Positive_Wavefront_SameAddress - int", "Unit___hip_atomic_fetch_and_Positive_Wavefront_SameAddress - unsigned int", "Unit___hip_atomic_fetch_and_Positive_Wavefront_SameAddress - unsigned long", @@ -1085,18 +1067,6 @@ "Unit___hip_atomic_fetch_xor_Positive_Workgroup_Scattered_Addresses - unsigned int", "Unit___hip_atomic_fetch_xor_Positive_Workgroup_Scattered_Addresses - unsigned long", "Unit___hip_atomic_fetch_xor_Positive_Workgroup_Scattered_Addresses - unsigned long long", - "Unit___hip_atomic_exchange_Positive_Wavefront - int", - "Unit___hip_atomic_exchange_Positive_Wavefront - unsigned int", - "Unit___hip_atomic_exchange_Positive_Wavefront - unsigned long", - "Unit___hip_atomic_exchange_Positive_Wavefront - unsigned long long", - "Unit___hip_atomic_exchange_Positive_Wavefront - float", - "Unit___hip_atomic_exchange_Positive_Wavefront - double", - "Unit___hip_atomic_exchange_Positive_Workgroup - int", - "Unit___hip_atomic_exchange_Positive_Workgroup - unsigned int", - "Unit___hip_atomic_exchange_Positive_Workgroup - unsigned long", - "Unit___hip_atomic_exchange_Positive_Workgroup - unsigned long long", - "Unit___hip_atomic_exchange_Positive_Workgroup - float", - "Unit___hip_atomic_exchange_Positive_Workgroup - double", "=== Below tests cause timeout in stress test of 09/02/24 ===", "Unit_Device___half2half2_Accuracy_Positive", "Unit_Device_make_half2_Accuracy_Positive", diff --git a/projects/hip-tests/catch/unit/atomics/atomicExch_common.hh b/projects/hip-tests/catch/unit/atomics/atomicExch_common.hh index 96238a089b..fcb1fc770e 100644 --- a/projects/hip-tests/catch/unit/atomics/atomicExch_common.hh +++ b/projects/hip-tests/catch/unit/atomics/atomicExch_common.hh @@ -97,7 +97,7 @@ __global__ void atomic_exch_kernel(T* const global_mem, T* const old_vals, const __syncthreads(); } - const auto n = cooperative_groups::this_grid().size() - width; + const auto n = cooperative_groups::this_grid().size(); T* atomic_addr = pitched_offset(mem, pitch, tid % width); @@ -335,7 +335,7 @@ void AtomicExchSingleDeviceSingleKernelTest(const unsigned int width, const unsi } using LA = LinearAllocs; for (const auto alloc_type : - {LA::hipMalloc, LA::hipHostMalloc, LA::hipMallocManaged, LA::mallocAndRegister}) { + {LA::hipMalloc, LA::hipHostMalloc, LA::hipMallocManaged}) { params.alloc_type = alloc_type; DYNAMIC_SECTION("Allocation type: " << to_string(alloc_type)) { AtomicExch().run(params); @@ -370,7 +370,7 @@ void AtomicExchSingleDeviceMultipleKernelTest(const unsigned int kernel_count, using LA = LinearAllocs; for (const auto alloc_type : - {LA::hipMalloc, LA::hipHostMalloc, LA::hipMallocManaged, LA::mallocAndRegister}) { + {LA::hipMalloc, LA::hipHostMalloc, LA::hipMallocManaged}) { params.alloc_type = alloc_type; DYNAMIC_SECTION("Allocation type: " << to_string(alloc_type)) { AtomicExch().run(params); @@ -413,7 +413,7 @@ void AtomicExchMultipleDeviceMultipleKernelAndHostTest(const unsigned int num_de params.host_thread_count = host_thread_count; using LA = LinearAllocs; - for (const auto alloc_type : {LA::hipHostMalloc, LA::hipMallocManaged, LA::mallocAndRegister}) { + for (const auto alloc_type : {LA::hipHostMalloc , LA::hipMallocManaged}) { params.alloc_type = alloc_type; DYNAMIC_SECTION("Allocation type: " << to_string(alloc_type)) { AtomicExch().run(params);