diff --git a/onnxruntime/core/providers/cuda/cudnn_common.cc b/onnxruntime/core/providers/cuda/cudnn_common.cc
index dbc45f13e1..a76bd46be1 100644
--- a/onnxruntime/core/providers/cuda/cudnn_common.cc
+++ b/onnxruntime/core/providers/cuda/cudnn_common.cc
@@ -137,15 +137,9 @@ cudnnDataType_t CudnnTensor::GetDataType<half>() {
 }
 
 template <>
-
 cudnnDataType_t CudnnTensor::GetDataType<BFloat16>() {
-#if CUDNN_VERSION >= 8100
-  return CUDNN_DATA_BFLOAT16;
-#else
-  ORT_THROW("cuDNN version is too low to support BFloat16.");
-  // Not reachable but GCC complains
+  ORT_THROW("cuDNN doesn't support BFloat16.");
   return CUDNN_DATA_FLOAT;
-#endif
 }
 
 template <>
diff --git a/onnxruntime/core/providers/cuda/reduction/reduction_ops.cc b/onnxruntime/core/providers/cuda/reduction/reduction_ops.cc
index da8be441ea..7ffd4bb185 100644
--- a/onnxruntime/core/providers/cuda/reduction/reduction_ops.cc
+++ b/onnxruntime/core/providers/cuda/reduction/reduction_ops.cc
@@ -511,15 +511,16 @@ Status ReduceComputeCore(CUDAExecutionProvider& cuda_ep, const Tensor& input, Pr
   CUDA_RETURN_IF_ERROR(cudaMemsetAsync(output.MutableDataRaw(), 0, output.SizeInBytes(), stream));
 
   IAllocatorUniquePtr<float> temp_X;
-  cudnnDataType_t cudnn_type_X = CudnnTensor::GetDataType<CudaT>();
+  cudnnDataType_t cudnn_type_X = CUDNN_DATA_FLOAT;
 
   // Reducesum with BFP16 is not supported by cudnn, so convert input to fp32 then call cudnn
   if ((ReduceTensorIndices == CUDNN_REDUCE_TENSOR_FLATTENED_INDICES && std::is_same<T, MLFloat16>::value) || 
       (ReduceTensorIndices == CUDNN_REDUCE_TENSOR_NO_INDICES && std::is_same<T, BFloat16>::value)) {
     // ArgMax/ArgMin with FP16 are not supported by cudnn, so convert input to fp32 then call cudnn
     temp_X = cuda_ep.GetScratchBuffer<float>(input_count);
-    cudnn_type_X = CUDNN_DATA_FLOAT;
     Impl_Cast<CudaT, float>(stream, reinterpret_cast<const CudaT*>(input.template Data<T>()), temp_X.get(), input_shape.Size());
+  } else {
+    cudnn_type_X = CudnnTensor::GetDataType<CudaT>();
   }
 
   CudnnReduceDescriptor reduce_desc;
diff --git a/onnxruntime/core/providers/rocm/miopen_common.cc b/onnxruntime/core/providers/rocm/miopen_common.cc
index 24f23853f1..73656984d4 100644
--- a/onnxruntime/core/providers/rocm/miopen_common.cc
+++ b/onnxruntime/core/providers/rocm/miopen_common.cc
@@ -93,7 +93,8 @@ miopenDataType_t MiopenTensor::GetDataType<half>() {
 
 template <>
 miopenDataType_t MiopenTensor::GetDataType<BFloat16>() {
-  return miopenBFloat16;
+  ORT_THROW("miopen doesn't support BFloat16.");
+  return miopenFloat;
 }
 
 template <>
diff --git a/onnxruntime/core/providers/rocm/reduction/reduction_ops.cc b/onnxruntime/core/providers/rocm/reduction/reduction_ops.cc
index dbc4bfd6aa..a064eb70b8 100644
--- a/onnxruntime/core/providers/rocm/reduction/reduction_ops.cc
+++ b/onnxruntime/core/providers/rocm/reduction/reduction_ops.cc
@@ -509,7 +509,7 @@ Status ReduceComputeCore(ROCMExecutionProvider& rocm_ep, const Tensor& input, Pr
   HIP_RETURN_IF_ERROR(hipMemsetAsync(output.MutableDataRaw(), 0, output.SizeInBytes(), stream));
 
   IAllocatorUniquePtr<float> temp_X;
-  miopenDataType_t miopen_type_X = MiopenTensor::GetDataType<HipT>();
+  miopenDataType_t miopen_type_X = miopenFloat;
 
   // unlike bfp16 not supported in cudnn, miopen call for bfp16 succeeded below, however, UT shows data error
   // so for now, follow the same logic in cudnn and convert input to fp32 then call miopen
@@ -517,8 +517,9 @@ Status ReduceComputeCore(ROCMExecutionProvider& rocm_ep, const Tensor& input, Pr
       (ReduceTensorIndices == MIOPEN_REDUCE_TENSOR_NO_INDICES && std::is_same<T, BFloat16>::value)) {
     // ArgMax/ArgMin with FP16 are not supported by miopen, so convert input to fp32 then call miopen
     temp_X = rocm_ep.GetScratchBuffer<float>(input_count);
-    miopen_type_X = miopenFloat;
     Impl_Cast<HipT, float>(stream, reinterpret_cast<const HipT*>(input.template Data<T>()), temp_X.get(), input_shape.Size());
+  } else {
+    miopen_type_X = MiopenTensor::GetDataType<HipT>();
   }
 
   MiopenReduceDescriptor reduce_desc;
diff --git a/onnxruntime/test/providers/cpu/reduction/reduction_ops_test.cc b/onnxruntime/test/providers/cpu/reduction/reduction_ops_test.cc
index 5dbfc40fcb..e7de0ecc7e 100644
--- a/onnxruntime/test/providers/cpu/reduction/reduction_ops_test.cc
+++ b/onnxruntime/test/providers/cpu/reduction/reduction_ops_test.cc
@@ -1536,9 +1536,10 @@ TEST(ReductionOpTest, ReduceSumBFloat16) {
 
 // on CUDA - this UT, with axes {0,2}, will go thru cudnn lib only if ATenOp is not initialized
 // on ROCM - miopen call succeeded, but results in data error, thus follow the same logic done in cudnn for now
+//           4.2 doesn't run properly (data error), thus enable the UT only above 4.3
 // TODO - try ROCm 4.5.2 and/or double check the source code on BFloat16 support 
-#if defined(USE_CUDA) || defined(USE_ROCM)
-TEST(ReductionOpTest, DISABLED_ReduceSumBFloat16_2) {
+#if defined(USE_CUDA) || (defined(USE_ROCM)  && ROCM_VERSION >= 40300)
+TEST(ReductionOpTest, ReduceSumBFloat16_2) {
   OpTester test("ReduceSum", 14);
   test.AddAttribute("keepdims", (int64_t)0);
   test.AddInput<BFloat16>("data", {3, 2, 2},