From f55328116ac4cb64341dd3c7c8dc79b8a1ee11af Mon Sep 17 00:00:00 2001
From: Jun Liu <Liu.Jun@amd.com>
Date: Sun, 17 Dec 2023 22:11:46 -0800
Subject: [PATCH 1/3] Revert "[gtest] conversion for code coverage tests
 (#2580)"

This reverts commit c5a2384dc0f29682ed51aeccf9b981dbdf7e058f.
---
 test/CMakeLists.txt                   |  22 +++
 test/driver.hpp                       |   7 +-
 test/gtest/CMakeLists.txt             |   2 +-
 test/gtest/conv3d_codecov.cpp         | 172 ---------------------
 test/gtest/conv_2d.hpp                |  66 +++++++++
 test/gtest/conv_igemm_mlir_xdlops.cpp |   2 +-
 test/gtest/immed_conv2d_codecov.cpp   | 206 --------------------------
 test/gtest/immed_conv3d_codecov.cpp   | 206 --------------------------
 test/gtest/pooling2d_asymmetric.cpp   | 136 -----------------
 test/gtest/pooling2d_codecov.cpp      | 136 -----------------
 test/gtest/pooling2d_wide.cpp         | 136 -----------------
 test/immed_conv2d.cpp                 |  62 ++++++++
 test/{conv3d.hpp => immed_conv3d.cpp} |  12 +-
 test/{pooling2d.hpp => pooling2d.cpp} |   2 +
 14 files changed, 162 insertions(+), 1005 deletions(-)
 delete mode 100644 test/gtest/conv3d_codecov.cpp
 create mode 100644 test/gtest/conv_2d.hpp
 delete mode 100644 test/gtest/immed_conv2d_codecov.cpp
 delete mode 100644 test/gtest/immed_conv3d_codecov.cpp
 delete mode 100644 test/gtest/pooling2d_asymmetric.cpp
 delete mode 100644 test/gtest/pooling2d_codecov.cpp
 delete mode 100644 test/gtest/pooling2d_wide.cpp
 create mode 100644 test/immed_conv2d.cpp
 rename test/{conv3d.hpp => immed_conv3d.cpp} (91%)
 rename test/{pooling2d.hpp => pooling2d.cpp} (98%)

diff --git a/test/CMakeLists.txt b/test/CMakeLists.txt
index ffaeb93228..d35ed0c43f 100644
--- a/test/CMakeLists.txt
+++ b/test/CMakeLists.txt
@@ -673,6 +673,20 @@ function(add_custom_test NAME)
         set_tests_properties(${NAME} PROPERTIES DISABLED On)
     endif()
 endfunction()
+if(${CODECOV_TEST})
+    add_custom_test(test_conv3d_codecov
+        COMMAND $<TARGET_FILE:test_conv3d> ${MIOPEN_TEST_FLOAT_ARG} --input 2 4 4 4 4 --weights 2 4 1 1 1 --pads_strides_dilations 0 0 0 1 1 1 1 1 1 ${MIOPEN_TEST_FLAGS_ARGS}
+    )
+    add_custom_test(test_immed_conv2d_codecov
+        COMMAND $<TARGET_FILE:test_immed_conv2d> ${MIOPEN_TEST_FLOAT_ARG} --input  2 2 14 14 --weights 8 2 3 3 --pads_strides_dilations 0 0 1 1 1 1 ${MIOPEN_TEST_FLAGS_ARGS}
+    )
+    add_custom_test(test_immed_conv3d_codecov
+        COMMAND $<TARGET_FILE:test_immed_conv3d> ${MIOPEN_TEST_FLOAT_ARG} --input 1 4 4 4 4 --weights 2 4 3 3 3 --pads_strides_dilations 0 0 0 1 1 1 1 1 1 ${MIOPEN_TEST_FLAGS_ARGS}
+    )
+    add_custom_test(test_pooling2d_codecov
+        COMMAND $<TARGET_FILE:test_pooling2d> ${MIOPEN_TEST_FLOAT_ARG} --input 1, 192, 28, 28 --lens 2 2 --strides 2 2 --pads 0 0 ${MIOPEN_TEST_FLAGS_ARGS}
+    )
+endif()
 
 if(${MIOPEN_TEST_WITH_MIOPENDRIVER})
     add_custom_test(test_miopendriver_regression_issue_1576 SKIP_UNLESS_ALL GFX94X_ENABLED GFX103X_ENABLED GFX110X_ENABLED FLOAT_DISABLED HALF_ENABLED
@@ -762,6 +776,14 @@ set(TEST_CONV_VERBOSE_F ${MIOPEN_TEST_FLOAT_ARG} --verbose --disable-backward-da
 set(TEST_CONV_VERBOSE_B ${MIOPEN_TEST_FLOAT_ARG} --verbose --disable-forward --disable-backward-weights)
 set(TEST_CONV_VERBOSE_W ${MIOPEN_TEST_FLOAT_ARG} --verbose --disable-forward --disable-backward-data)
 
+add_custom_test(test_pooling2d_asymmetric SKIP_UNLESS_ALL HALF_ENABLED GFX94X_ENABLED GFX103X_ENABLED GFX110X_ENABLED
+    COMMAND $<TARGET_FILE:test_pooling2d> ${MIOPEN_TEST_FLOAT_ARG} --all --dataset 1 --limit 0 ${MIOPEN_TEST_FLAGS_ARGS}
+)
+
+add_custom_test(test_pooling2d_wide SKIP_UNLESS_ALL HALF_ENABLED GFX94X_ENABLED GFX103X_ENABLED GFX110X_ENABLED
+    COMMAND $<TARGET_FILE:test_pooling2d> ${MIOPEN_TEST_FLOAT_ARG} --all --dataset 2 --limit 0 ${MIOPEN_TEST_FLAGS_ARGS}
+)
+
 set(IMPLICITGEMM_MLIR_ENV_F_XDLOPS ${IMPLICITGEMM_MLIR_ENV_BASE} MIOPEN_DEBUG_FIND_ONLY_SOLVER=ConvMlirIgemmFwdXdlops)
 set(IMPLICITGEMM_MLIR_ENV_B_XDLOPS ${IMPLICITGEMM_MLIR_ENV_BASE} MIOPEN_DEBUG_FIND_ONLY_SOLVER=ConvMlirIgemmBwdXdlops)
 set(IMPLICITGEMM_MLIR_ENV_W_XDLOPS ${IMPLICITGEMM_MLIR_ENV_BASE} MIOPEN_DEBUG_FIND_ONLY_SOLVER=ConvMlirIgemmWrWXdlops)
diff --git a/test/driver.hpp b/test/driver.hpp
index 97896d6c5d..825a0d4511 100644
--- a/test/driver.hpp
+++ b/test/driver.hpp
@@ -1242,11 +1242,6 @@ void test_drive_impl_1(std::string program_name, std::vector<std::string> as)
     Driver d{};
     d.program_name = program_name;
 
-    std::cout << program_name << " ";
-    for(const auto& str : as)
-        std::cout << str << " ";
-    std::cout << std::endl;
-
     std::set<std::string> keywords{
         "--help", "-h", "--half", "--float", "--double", "--int8", "--bfloat16"};
     d.parse(keyword_set{keywords});
@@ -1386,7 +1381,7 @@ template <template <class...> class Driver>
 void test_drive(int argc, const char* argv[])
 {
     std::vector<std::string> as(argv + 1, argv + argc);
-    // as.emplace_back("--float");
+    as.emplace_back("--float");
     for(auto&& arg : as)
     {
         if(arg == "--half")
diff --git a/test/gtest/CMakeLists.txt b/test/gtest/CMakeLists.txt
index 1e3fb206f3..2e80fb5eac 100644
--- a/test/gtest/CMakeLists.txt
+++ b/test/gtest/CMakeLists.txt
@@ -30,7 +30,7 @@ function(add_gtest TEST_NAME)
     endif()
     if(NOT WIN32) # TODO: cannot run on Windows due to missing DLL dependencies
       # Enable CMake to discover the test binary
-      gtest_discover_tests(test_${TEST_NAME} PROPERTIES ENVIRONMENT "MIOPEN_USER_DB_PATH=${CMAKE_CURRENT_BINARY_DIR};MIOPEN_TEST_FLOAT_ARG=${MIOPEN_TEST_FLOAT_ARG};MIOPEN_TEST_ALL=${MIOPEN_TEST_ALL};MIOPEN_TEST_MLIR=${MIOPEN_TEST_MLIR};MIOPEN_TEST_COMPOSABLEKERNEL=${MIOPEN_TEST_COMPOSABLEKERNEL};CODECOV_TEST=${CODECOV_TEST}")
+      gtest_discover_tests(test_${TEST_NAME} PROPERTIES ENVIRONMENT "MIOPEN_USER_DB_PATH=${CMAKE_CURRENT_BINARY_DIR};MIOPEN_TEST_FLOAT_ARG=${MIOPEN_TEST_FLOAT_ARG};MIOPEN_TEST_ALL=${MIOPEN_TEST_ALL};MIOPEN_TEST_MLIR=${MIOPEN_TEST_MLIR};MIOPEN_TEST_COMPOSABLEKERNEL=${MIOPEN_TEST_COMPOSABLEKERNEL}")
     endif()
 
     if(WIN32)
diff --git a/test/gtest/conv3d_codecov.cpp b/test/gtest/conv3d_codecov.cpp
deleted file mode 100644
index f85b072722..0000000000
--- a/test/gtest/conv3d_codecov.cpp
+++ /dev/null
@@ -1,172 +0,0 @@
-/*******************************************************************************
- *
- * MIT License
- *
- * Copyright (c) 2023 Advanced Micro Devices, Inc.
- *
- * Permission is hereby granted, free of charge, to any person obtaining a copy
- * of this software and associated documentation files (the "Software"), to deal
- * in the Software without restriction, including without limitation the rights
- * to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
- * copies of the Software, and to permit persons to whom the Software is
- * furnished to do so, subject to the following conditions:
- *
- * The above copyright notice and this permission notice shall be included in all
- * copies or substantial portions of the Software.
- *
- * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
- * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
- * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
- * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
- * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
- * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
- * SOFTWARE.
- *
- *******************************************************************************/
-#include "conv3d.hpp"
-#include "get_handle.hpp"
-#include <miopen/env.hpp>
-#include <gtest/gtest.h>
-
-MIOPEN_DECLARE_ENV_VAR_BOOL(CODECOV_TEST)
-MIOPEN_DECLARE_ENV_VAR_STR(MIOPEN_TEST_FLAGS_ARGS)
-
-static bool SkipTest(void) { return !miopen::IsEnabled(ENV(CODECOV_TEST)); }
-
-void GetArgs(const std::string& param, std::vector<std::string>& tokens)
-{
-    std::stringstream ss(param);
-    std::istream_iterator<std::string> begin(ss);
-    std::istream_iterator<std::string> end;
-    while(begin != end)
-        tokens.push_back(*begin++);
-}
-
-class Conv3dFloat : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-class Conv3dHalf : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-class Conv3dBFloat16 : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-class Conv3dInt8 : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-void Run3dDriver(miopenDataType_t prec)
-{
-
-    std::vector<std::string> params;
-    switch(prec)
-    {
-    case miopenHalf: params = Conv3dHalf::GetParam(); break;
-    case miopenBFloat16: params = Conv3dBFloat16::GetParam(); break;
-    case miopenFloat: params = Conv3dFloat::GetParam(); break;
-    case miopenInt8: params = Conv3dInt8::GetParam(); break;
-    case miopenFloat8:
-    case miopenBFloat8:
-    case miopenInt32:
-    case miopenDouble:
-        FAIL() << "miopenInt32, miopenDouble, miopenFloat8, miopenBFloat8 "
-                  "data type not supported by "
-                  "conv3d_codecov test";
-
-    default: params = Conv3dFloat::GetParam();
-    }
-
-    for(const auto& test_value : params)
-    {
-        std::vector<std::string> tokens;
-        GetArgs(test_value, tokens);
-        std::vector<const char*> ptrs;
-
-        std::transform(tokens.begin(), tokens.end(), std::back_inserter(ptrs), [](const auto& str) {
-            return str.data();
-        });
-
-        testing::internal::CaptureStderr();
-        test_drive<conv3d_driver>(ptrs.size(), ptrs.data());
-        auto capture = testing::internal::GetCapturedStderr();
-        std::cout << capture;
-    }
-};
-
-bool IsTestSupportedForDevice(const miopen::Handle& handle) { return true; }
-
-TEST_P(Conv3dFloat, FloatTest)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run3dDriver(miopenFloat);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-TEST_P(Conv3dHalf, HalfTest)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run3dDriver(miopenHalf);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-TEST_P(Conv3dBFloat16, BFloat16Test)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run3dDriver(miopenBFloat16);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-TEST_P(Conv3dInt8, Int8Test)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run3dDriver(miopenInt8);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-std::vector<std::string> GetTestCases(const std::string& precision)
-{
-    const auto& flag_arg = miopen::GetStringEnv(ENV(MIOPEN_TEST_FLAGS_ARGS));
-
-    const std::vector<std::string> test_cases = {
-        // clang-format off
-    {"test_conv3d " + precision + " --input 2 4 4 4 4 --weights 2 4 1 1 1 --pads_strides_dilations 0 0 0 1 1 1 1 1 1 "+flag_arg}
-        // clang-format on
-    };
-
-    return test_cases;
-}
-
-INSTANTIATE_TEST_SUITE_P(Conv3D, Conv3dFloat, testing::Values(GetTestCases("--float")));
-
-INSTANTIATE_TEST_SUITE_P(Conv3D, Conv3dHalf, testing::Values(GetTestCases("--half")));
-
-INSTANTIATE_TEST_SUITE_P(Conv3D, Conv3dBFloat16, testing::Values(GetTestCases("--bfloat16")));
-
-INSTANTIATE_TEST_SUITE_P(Conv3D, Conv3dInt8, testing::Values(GetTestCases("--int8")));
diff --git a/test/gtest/conv_2d.hpp b/test/gtest/conv_2d.hpp
new file mode 100644
index 0000000000..df73453ef2
--- /dev/null
+++ b/test/gtest/conv_2d.hpp
@@ -0,0 +1,66 @@
+/*******************************************************************************
+ *
+ * MIT License
+ *
+ * Copyright (c) 2023 Advanced Micro Devices, Inc.
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a copy
+ * of this software and associated documentation files (the "Software"), to deal
+ * in the Software without restriction, including without limitation the rights
+ * to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
+ * copies of the Software, and to permit persons to whom the Software is
+ * furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included in all
+ * copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+ * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
+ * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+ * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+ * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
+ * SOFTWARE.
+ *
+ *******************************************************************************/
+#include "conv_common.hpp"
+
+template <class T>
+struct conv2d_driver : conv_driver<T>
+{
+    conv2d_driver() : conv_driver<T>()
+    {
+        this->add(this->input_dims, "input");
+        this->add(this->weight_tensor_dims, "weights");
+        this->add(this->batch_size,
+                  "batch_size",
+                  this->generate_data_limited(this->get_batch_sizes(), 1));
+        this->add(this->input_channels,
+                  "input_channels",
+                  this->generate_data_limited(this->get_input_channels(), 1, {32}));
+        this->add(this->output_channels,
+                  "output_channels",
+                  this->generate_data_limited(this->get_output_channels(), 1, {64}));
+        this->add(this->spatial_dim_elements,
+                  "spatial_dim_elements",
+                  this->generate_data_limited(this->get_2d_spatial_dims(), 1, {28, 28}));
+        this->add(this->filter_dims,
+                  "filter_dims",
+                  this->generate_data_limited(this->get_2d_filter_dims(), 2, {3, 3}));
+        this->add(this->pads_strides_dilations,
+                  "pads_strides_dilations",
+                  this->generate_data_limited(this->get_2d_pads_strides_dilations(), 2));
+        this->add(this->trans_output_pads,
+                  "trans_output_pads",
+                  this->generate_data(this->get_2d_trans_output_pads()));
+        this->add(this->in_layout, "in_layout", this->generate_data({"NCHW"}));
+        this->add(this->fil_layout, "fil_layout", this->generate_data({"NCHW"}));
+        this->add(this->out_layout, "out_layout", this->generate_data({"NCHW"}));
+        this->add(this->deterministic, "deterministic", this->generate_data({false}));
+        this->add(this->tensor_vect, "tensor_vect", this->generate_data({0}));
+        this->add(this->vector_length, "vector_length", this->generate_data({1}));
+        // Only valid for int8 input and weights
+        this->add(this->output_type, "output_type", this->generate_data({"int32"}));
+        this->add(this->int8_vectorize, "int8_vectorize", this->generate_data({false}));
+    }
+};
diff --git a/test/gtest/conv_igemm_mlir_xdlops.cpp b/test/gtest/conv_igemm_mlir_xdlops.cpp
index 3d339d9c07..bcc9495204 100644
--- a/test/gtest/conv_igemm_mlir_xdlops.cpp
+++ b/test/gtest/conv_igemm_mlir_xdlops.cpp
@@ -4,7 +4,7 @@
 #include <gtest/gtest.h>
 #include <miopen/miopen.h>
 #include <miopen/env.hpp>
-#include "conv2d.hpp"
+#include "conv_2d.hpp"
 #include "get_handle.hpp"
 
 MIOPEN_DECLARE_ENV_VAR_STR(MIOPEN_TEST_FLOAT_ARG)
diff --git a/test/gtest/immed_conv2d_codecov.cpp b/test/gtest/immed_conv2d_codecov.cpp
deleted file mode 100644
index ba2a9de09c..0000000000
--- a/test/gtest/immed_conv2d_codecov.cpp
+++ /dev/null
@@ -1,206 +0,0 @@
-/*******************************************************************************
- *
- * MIT License
- *
- * Copyright (c) 2023 Advanced Micro Devices, Inc.
- *
- * Permission is hereby granted, free of charge, to any person obtaining a copy
- * of this software and associated documentation files (the "Software"), to deal
- * in the Software without restriction, including without limitation the rights
- * to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
- * copies of the Software, and to permit persons to whom the Software is
- * furnished to do so, subject to the following conditions:
- *
- * The above copyright notice and this permission notice shall be included in all
- * copies or substantial portions of the Software.
- *
- * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
- * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
- * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
- * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
- * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
- * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
- * SOFTWARE.
- *
- *******************************************************************************/
-#include "conv_common.hpp"
-#include "get_handle.hpp"
-#include <miopen/env.hpp>
-#include <gtest/gtest.h>
-
-MIOPEN_DECLARE_ENV_VAR_BOOL(CODECOV_TEST)
-MIOPEN_DECLARE_ENV_VAR_STR(MIOPEN_TEST_FLAGS_ARGS)
-
-template <class T>
-struct conv2d_driver : conv_driver<T, ConvApi::Immediate>
-{
-    conv2d_driver() : conv_driver<T, ConvApi::Immediate>()
-    {
-        this->add(this->input_dims, "input");
-        this->add(this->weight_tensor_dims, "weights");
-        this->add(this->batch_size,
-                  "batch_size",
-                  this->generate_data_limited(this->get_batch_sizes(), 1, {16}));
-        this->add(this->input_channels,
-                  "input_channels",
-                  this->generate_data_limited(this->get_input_channels(), 1, {32}));
-        this->add(this->output_channels,
-                  "output_channels",
-                  this->generate_data_limited(this->get_output_channels(), 1, {32}));
-        this->add(this->spatial_dim_elements,
-                  "spatial_dim_elements",
-                  this->generate_data_limited(this->get_2d_spatial_dims(), 1, {56, 56}));
-        this->add(this->filter_dims,
-                  "filter_dims",
-                  this->generate_data_limited(this->get_2d_filter_dims(), 2, {3, 3}));
-        this->add(this->pads_strides_dilations,
-                  "pads_strides_dilations",
-                  this->generate_data_limited(this->get_2d_pads_strides_dilations(), 2));
-        this->add(this->trans_output_pads,
-                  "trans_output_pads",
-                  this->generate_data_limited(this->get_2d_trans_output_pads(), 1));
-        this->add(this->in_layout, "in_layout", this->generate_data({"NCHW"}));
-        this->add(this->fil_layout, "fil_layout", this->generate_data({"NCHW"}));
-        this->add(this->out_layout, "out_layout", this->generate_data({"NCHW"}));
-    }
-};
-
-static bool SkipTest(void) { return !miopen::IsEnabled(ENV(CODECOV_TEST)); }
-
-void GetArgs(const std::string& param, std::vector<std::string>& tokens)
-{
-    std::stringstream ss(param);
-    std::istream_iterator<std::string> begin(ss);
-    std::istream_iterator<std::string> end;
-    while(begin != end)
-        tokens.push_back(*begin++);
-}
-
-class Conv2dFloat : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-class Conv2dHalf : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-class Conv2dBFloat16 : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-class Conv2dInt8 : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-void Run2dDriver(miopenDataType_t prec)
-{
-
-    std::vector<std::string> params;
-    switch(prec)
-    {
-    case miopenHalf: params = Conv2dHalf::GetParam(); break;
-    case miopenBFloat16: params = Conv2dBFloat16::GetParam(); break;
-    case miopenFloat: params = Conv2dFloat::GetParam(); break;
-    case miopenInt8: params = Conv2dInt8::GetParam(); break;
-    case miopenFloat8:
-    case miopenBFloat8:
-    case miopenInt32:
-    case miopenDouble:
-        FAIL() << "miopenInt32, miopenDouble, miopenFloat8, miopenBFloat8 "
-                  "data type not supported by "
-                  "immed_conv2d_codecov test";
-
-    default: params = Conv2dFloat::GetParam();
-    }
-
-    for(const auto& test_value : params)
-    {
-        std::vector<std::string> tokens;
-        GetArgs(test_value, tokens);
-        std::vector<const char*> ptrs;
-
-        std::transform(tokens.begin(), tokens.end(), std::back_inserter(ptrs), [](const auto& str) {
-            return str.data();
-        });
-
-        testing::internal::CaptureStderr();
-        test_drive<conv2d_driver>(ptrs.size(), ptrs.data());
-        auto capture = testing::internal::GetCapturedStderr();
-        std::cout << capture;
-    }
-};
-
-bool IsTestSupportedForDevice(const miopen::Handle& handle) { return true; }
-
-TEST_P(Conv2dFloat, FloatTest)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run2dDriver(miopenFloat);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-TEST_P(Conv2dHalf, HalfTest)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run2dDriver(miopenHalf);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-TEST_P(Conv2dBFloat16, BFloat16Test)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run2dDriver(miopenBFloat16);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-TEST_P(Conv2dInt8, Int8Test)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run2dDriver(miopenInt8);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-std::vector<std::string> GetTestCases(const std::string& precision)
-{
-    const auto& flag_arg = miopen::GetStringEnv(ENV(MIOPEN_TEST_FLAGS_ARGS));
-
-    const std::vector<std::string> test_cases = {
-        // clang-format off
-    {"test_immed_conv2d " + precision + " --input  2 2 14 14 --weights 8 2 3 3 --pads_strides_dilations 0 0 1 1 1 1 "+flag_arg}
-        // clang-format on
-    };
-
-    return test_cases;
-}
-
-INSTANTIATE_TEST_SUITE_P(ImmedConv2D, Conv2dFloat, testing::Values(GetTestCases("--float")));
-
-INSTANTIATE_TEST_SUITE_P(ImmedConv2D, Conv2dHalf, testing::Values(GetTestCases("--half")));
-
-INSTANTIATE_TEST_SUITE_P(ImmedConv2D, Conv2dBFloat16, testing::Values(GetTestCases("--bfloat16")));
-
-INSTANTIATE_TEST_SUITE_P(ImmedConv2D, Conv2dInt8, testing::Values(GetTestCases("--int8")));
diff --git a/test/gtest/immed_conv3d_codecov.cpp b/test/gtest/immed_conv3d_codecov.cpp
deleted file mode 100644
index 3f22479d91..0000000000
--- a/test/gtest/immed_conv3d_codecov.cpp
+++ /dev/null
@@ -1,206 +0,0 @@
-/*******************************************************************************
- *
- * MIT License
- *
- * Copyright (c) 2023 Advanced Micro Devices, Inc.
- *
- * Permission is hereby granted, free of charge, to any person obtaining a copy
- * of this software and associated documentation files (the "Software"), to deal
- * in the Software without restriction, including without limitation the rights
- * to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
- * copies of the Software, and to permit persons to whom the Software is
- * furnished to do so, subject to the following conditions:
- *
- * The above copyright notice and this permission notice shall be included in all
- * copies or substantial portions of the Software.
- *
- * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
- * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
- * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
- * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
- * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
- * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
- * SOFTWARE.
- *
- *******************************************************************************/
-#include "conv_common.hpp"
-#include "get_handle.hpp"
-#include <miopen/env.hpp>
-#include <gtest/gtest.h>
-
-MIOPEN_DECLARE_ENV_VAR_BOOL(CODECOV_TEST)
-MIOPEN_DECLARE_ENV_VAR_STR(MIOPEN_TEST_FLAGS_ARGS)
-
-template <class T>
-struct conv3d_driver : conv_driver<T, ConvApi::Immediate>
-{
-    conv3d_driver() : conv_driver<T, ConvApi::Immediate>()
-    {
-        this->add(this->input_dims, "input");
-        this->add(this->weight_tensor_dims, "weights");
-        this->add(this->batch_size,
-                  "batch_size",
-                  this->generate_data_limited(this->get_batch_sizes(), 1, {8}));
-        this->add(this->input_channels,
-                  "input_channels",
-                  this->generate_data_limited(this->get_input_channels(), 1, {2}));
-        this->add(this->output_channels,
-                  "output_channels",
-                  this->generate_data_limited(this->get_output_channels(), 1, {16}));
-        this->add(this->spatial_dim_elements,
-                  "spatial_dim_elements",
-                  this->generate_data_limited(this->get_3d_spatial_dims(), 1, {16, 16, 16}));
-        this->add(this->filter_dims,
-                  "filter_dims",
-                  this->generate_data_limited(this->get_3d_filter_dims(), 2, {5, 5, 5}));
-        this->add(this->pads_strides_dilations,
-                  "pads_strides_dilations",
-                  this->generate_data_limited(this->get_3d_pads_strides_dilations(), 2));
-        this->add(this->trans_output_pads,
-                  "trans_output_pads",
-                  this->generate_data_limited(this->get_3d_trans_output_pads(), 1));
-        this->add(this->in_layout, "in_layout", this->generate_data({"NCDHW"}));
-        this->add(this->fil_layout, "fil_layout", this->generate_data({"NCDHW"}));
-        this->add(this->out_layout, "out_layout", this->generate_data({"NCDHW"}));
-    }
-};
-
-static bool SkipTest(void) { return !miopen::IsEnabled(ENV(CODECOV_TEST)); }
-
-void GetArgs(const std::string& param, std::vector<std::string>& tokens)
-{
-    std::stringstream ss(param);
-    std::istream_iterator<std::string> begin(ss);
-    std::istream_iterator<std::string> end;
-    while(begin != end)
-        tokens.push_back(*begin++);
-}
-
-class Conv3dFloat : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-class Conv3dHalf : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-class Conv3dBFloat16 : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-class Conv3dInt8 : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-void Run3dDriver(miopenDataType_t prec)
-{
-
-    std::vector<std::string> params;
-    switch(prec)
-    {
-    case miopenHalf: params = Conv3dHalf::GetParam(); break;
-    case miopenBFloat16: params = Conv3dBFloat16::GetParam(); break;
-    case miopenFloat: params = Conv3dFloat::GetParam(); break;
-    case miopenInt8: params = Conv3dInt8::GetParam(); break;
-    case miopenFloat8:
-    case miopenBFloat8:
-    case miopenInt32:
-    case miopenDouble:
-        FAIL() << "miopenInt32, miopenDouble, miopenFloat8, miopenBFloat8 "
-                  "data type not supported by "
-                  "immed_conv3d_codecov test";
-
-    default: params = Conv3dFloat::GetParam();
-    }
-
-    for(const auto& test_value : params)
-    {
-        std::vector<std::string> tokens;
-        GetArgs(test_value, tokens);
-        std::vector<const char*> ptrs;
-
-        std::transform(tokens.begin(), tokens.end(), std::back_inserter(ptrs), [](const auto& str) {
-            return str.data();
-        });
-
-        testing::internal::CaptureStderr();
-        test_drive<conv3d_driver>(ptrs.size(), ptrs.data());
-        auto capture = testing::internal::GetCapturedStderr();
-        std::cout << capture;
-    }
-};
-
-bool IsTestSupportedForDevice(const miopen::Handle& handle) { return true; }
-
-TEST_P(Conv3dFloat, FloatTest)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run3dDriver(miopenFloat);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-TEST_P(Conv3dHalf, HalfTest)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run3dDriver(miopenHalf);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-TEST_P(Conv3dBFloat16, BFloat16Test)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run3dDriver(miopenBFloat16);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-TEST_P(Conv3dInt8, Int8Test)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run3dDriver(miopenInt8);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-std::vector<std::string> GetTestCases(const std::string& precision)
-{
-    const auto& flag_arg = miopen::GetStringEnv(ENV(MIOPEN_TEST_FLAGS_ARGS));
-
-    const std::vector<std::string> test_cases = {
-        // clang-format off
-    {"test_immed_conv3d " + precision + " --input 1 4 4 4 4 --weights 2 4 3 3 3 --pads_strides_dilations 0 0 0 1 1 1 1 1 1 "+flag_arg}
-        // clang-format on
-    };
-
-    return test_cases;
-}
-
-INSTANTIATE_TEST_SUITE_P(ImmedConv3D, Conv3dFloat, testing::Values(GetTestCases("--float")));
-
-INSTANTIATE_TEST_SUITE_P(ImmedConv3D, Conv3dHalf, testing::Values(GetTestCases("--half")));
-
-INSTANTIATE_TEST_SUITE_P(ImmedConv3D, Conv3dBFloat16, testing::Values(GetTestCases("--bfloat16")));
-
-INSTANTIATE_TEST_SUITE_P(ImmedConv3D, Conv3dInt8, testing::Values(GetTestCases("--int8")));
diff --git a/test/gtest/pooling2d_asymmetric.cpp b/test/gtest/pooling2d_asymmetric.cpp
deleted file mode 100644
index 894210b813..0000000000
--- a/test/gtest/pooling2d_asymmetric.cpp
+++ /dev/null
@@ -1,136 +0,0 @@
-/*******************************************************************************
- *
- * MIT License
- *
- * Copyright (c) 2019 Advanced Micro Devices, Inc.
- *
- * Permission is hereby granted, free of charge, to any person obtaining a copy
- * of this software and associated documentation files (the "Software"), to deal
- * in the Software without restriction, including without limitation the rights
- * to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
- * copies of the Software, and to permit persons to whom the Software is
- * furnished to do so, subject to the following conditions:
- *
- * The above copyright notice and this permission notice shall be included in all
- * copies or substantial portions of the Software.
- *
- * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
- * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
- * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
- * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
- * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
- * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
- * SOFTWARE.
- *
- *******************************************************************************/
-
-#include "pooling2d.hpp"
-#include "get_handle.hpp"
-#include <miopen/env.hpp>
-#include <gtest/gtest.h>
-
-MIOPEN_DECLARE_ENV_VAR_BOOL(MIOPEN_TEST_ALL)
-MIOPEN_DECLARE_ENV_VAR_STR(MIOPEN_TEST_FLAGS_ARGS)
-
-static bool SkipTest(void) { return miopen::IsDisabled(ENV(MIOPEN_TEST_ALL)); }
-
-void GetArgs(const std::string& param, std::vector<std::string>& tokens)
-{
-    std::stringstream ss(param);
-    std::istream_iterator<std::string> begin(ss);
-    std::istream_iterator<std::string> end;
-    while(begin != end)
-        tokens.push_back(*begin++);
-}
-
-class Pooling2dFloat : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-class Pooling2dHalf : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-void Run2dDriver(miopenDataType_t prec)
-{
-
-    std::vector<std::string> params;
-    switch(prec)
-    {
-    case miopenFloat: params = Pooling2dFloat::GetParam(); break;
-    case miopenHalf: params = Pooling2dHalf::GetParam(); break;
-    case miopenBFloat16:
-    case miopenInt8:
-    case miopenFloat8:
-    case miopenBFloat8:
-    case miopenInt32:
-    case miopenDouble:
-        FAIL()
-            << "miopenBFloat16, miopenInt8, miopenInt32, miopenDouble, miopenFloat8, miopenBFloat8 "
-               "data type not supported by "
-               "immed_conv2d_codecov test";
-
-    default: params = Pooling2dFloat::GetParam();
-    }
-
-    for(const auto& test_value : params)
-    {
-        std::vector<std::string> tokens;
-        GetArgs(test_value, tokens);
-        std::vector<const char*> ptrs;
-
-        std::transform(tokens.begin(), tokens.end(), std::back_inserter(ptrs), [](const auto& str) {
-            return str.data();
-        });
-
-        testing::internal::CaptureStderr();
-        test_drive<pooling2d_driver>(ptrs.size(), ptrs.data());
-        auto capture = testing::internal::GetCapturedStderr();
-        std::cout << capture;
-    }
-};
-
-bool IsTestSupportedForDevice(const miopen::Handle& handle) { return true; }
-
-TEST_P(Pooling2dFloat, FloatTest)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run2dDriver(miopenFloat);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-TEST_P(Pooling2dHalf, HalfTest)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run2dDriver(miopenHalf);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-std::vector<std::string> GetTestCases(const std::string& precision)
-{
-    const auto& flag_arg = miopen::GetStringEnv(ENV(MIOPEN_TEST_FLAGS_ARGS));
-
-    const std::vector<std::string> test_cases = {
-        // clang-format off
-    {"test_pooling2d " + precision + " --all --dataset 1 --limit 0 "+flag_arg}
-        // clang-format on
-    };
-
-    return test_cases;
-}
-
-INSTANTIATE_TEST_SUITE_P(Pooling2D, Pooling2dFloat, testing::Values(GetTestCases("--float")));
-
-INSTANTIATE_TEST_SUITE_P(Pooling2D, Pooling2dHalf, testing::Values(GetTestCases("--half")));
diff --git a/test/gtest/pooling2d_codecov.cpp b/test/gtest/pooling2d_codecov.cpp
deleted file mode 100644
index 52daa50f3f..0000000000
--- a/test/gtest/pooling2d_codecov.cpp
+++ /dev/null
@@ -1,136 +0,0 @@
-/*******************************************************************************
- *
- * MIT License
- *
- * Copyright (c) 2019 Advanced Micro Devices, Inc.
- *
- * Permission is hereby granted, free of charge, to any person obtaining a copy
- * of this software and associated documentation files (the "Software"), to deal
- * in the Software without restriction, including without limitation the rights
- * to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
- * copies of the Software, and to permit persons to whom the Software is
- * furnished to do so, subject to the following conditions:
- *
- * The above copyright notice and this permission notice shall be included in all
- * copies or substantial portions of the Software.
- *
- * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
- * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
- * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
- * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
- * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
- * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
- * SOFTWARE.
- *
- *******************************************************************************/
-
-#include "pooling2d.hpp"
-#include "get_handle.hpp"
-#include <miopen/env.hpp>
-#include <gtest/gtest.h>
-
-MIOPEN_DECLARE_ENV_VAR_BOOL(CODECOV_TEST)
-MIOPEN_DECLARE_ENV_VAR_STR(MIOPEN_TEST_FLAGS_ARGS)
-
-static bool SkipTest(void) { return !miopen::IsEnabled(ENV(CODECOV_TEST)); }
-
-void GetArgs(const std::string& param, std::vector<std::string>& tokens)
-{
-    std::stringstream ss(param);
-    std::istream_iterator<std::string> begin(ss);
-    std::istream_iterator<std::string> end;
-    while(begin != end)
-        tokens.push_back(*begin++);
-}
-
-class Pooling2dFloat : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-class Pooling2dHalf : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-void Run2dDriver(miopenDataType_t prec)
-{
-
-    std::vector<std::string> params;
-    switch(prec)
-    {
-    case miopenHalf: params = Pooling2dHalf::GetParam(); break;
-    case miopenFloat: params = Pooling2dFloat::GetParam(); break;
-    case miopenBFloat16:
-    case miopenInt8:
-    case miopenFloat8:
-    case miopenBFloat8:
-    case miopenInt32:
-    case miopenDouble:
-        FAIL()
-            << "miopenBFloat16, miopenInt8, miopenInt32, miopenDouble, miopenFloat8, miopenBFloat8 "
-               "data type not supported by "
-               "immed_conv2d_codecov test";
-
-    default: params = Pooling2dFloat::GetParam();
-    }
-
-    for(const auto& test_value : params)
-    {
-        std::vector<std::string> tokens;
-        GetArgs(test_value, tokens);
-        std::vector<const char*> ptrs;
-
-        std::transform(tokens.begin(), tokens.end(), std::back_inserter(ptrs), [](const auto& str) {
-            return str.data();
-        });
-
-        testing::internal::CaptureStderr();
-        test_drive<pooling2d_driver>(ptrs.size(), ptrs.data());
-        auto capture = testing::internal::GetCapturedStderr();
-        std::cout << capture;
-    }
-};
-
-bool IsTestSupportedForDevice(const miopen::Handle& handle) { return true; }
-
-TEST_P(Pooling2dFloat, FloatTest)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run2dDriver(miopenFloat);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-TEST_P(Pooling2dHalf, HalfTest)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run2dDriver(miopenHalf);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-std::vector<std::string> GetTestCases(const std::string& precision)
-{
-    const auto& flag_arg = miopen::GetStringEnv(ENV(MIOPEN_TEST_FLAGS_ARGS));
-
-    const std::vector<std::string> test_cases = {
-        // clang-format off
-    {"test_pooling2d " + precision + "--input 1, 192, 28, 28 --lens 2 2 --strides 2 2 --pads 0 0 "+flag_arg}
-        // clang-format on
-    };
-
-    return test_cases;
-}
-
-INSTANTIATE_TEST_SUITE_P(Pooling2D, Pooling2dFloat, testing::Values(GetTestCases("--float")));
-
-INSTANTIATE_TEST_SUITE_P(Pooling2D, Pooling2dHalf, testing::Values(GetTestCases("--half")));
diff --git a/test/gtest/pooling2d_wide.cpp b/test/gtest/pooling2d_wide.cpp
deleted file mode 100644
index 159da6f608..0000000000
--- a/test/gtest/pooling2d_wide.cpp
+++ /dev/null
@@ -1,136 +0,0 @@
-/*******************************************************************************
- *
- * MIT License
- *
- * Copyright (c) 2019 Advanced Micro Devices, Inc.
- *
- * Permission is hereby granted, free of charge, to any person obtaining a copy
- * of this software and associated documentation files (the "Software"), to deal
- * in the Software without restriction, including without limitation the rights
- * to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
- * copies of the Software, and to permit persons to whom the Software is
- * furnished to do so, subject to the following conditions:
- *
- * The above copyright notice and this permission notice shall be included in all
- * copies or substantial portions of the Software.
- *
- * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
- * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
- * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
- * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
- * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
- * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
- * SOFTWARE.
- *
- *******************************************************************************/
-
-#include "pooling2d.hpp"
-#include "get_handle.hpp"
-#include <miopen/env.hpp>
-#include <gtest/gtest.h>
-
-MIOPEN_DECLARE_ENV_VAR_BOOL(MIOPEN_TEST_ALL)
-MIOPEN_DECLARE_ENV_VAR_STR(MIOPEN_TEST_FLAGS_ARGS)
-
-static bool SkipTest(void) { return miopen::IsDisabled(ENV(MIOPEN_TEST_ALL)); }
-
-void GetArgs(const std::string& param, std::vector<std::string>& tokens)
-{
-    std::stringstream ss(param);
-    std::istream_iterator<std::string> begin(ss);
-    std::istream_iterator<std::string> end;
-    while(begin != end)
-        tokens.push_back(*begin++);
-}
-
-class Pooling2dFloat : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-class Pooling2dHalf : public testing::TestWithParam<std::vector<std::string>>
-{
-};
-
-void Run2dDriver(miopenDataType_t prec)
-{
-
-    std::vector<std::string> params;
-    switch(prec)
-    {
-    case miopenFloat: params = Pooling2dFloat::GetParam(); break;
-    case miopenHalf: params = Pooling2dHalf::GetParam(); break;
-    case miopenBFloat16:
-    case miopenInt8:
-    case miopenFloat8:
-    case miopenBFloat8:
-    case miopenInt32:
-    case miopenDouble:
-        FAIL()
-            << "miopenBFloat16, miopenInt8, miopenInt32, miopenDouble, miopenFloat8, miopenBFloat8 "
-               "data type not supported by "
-               "immed_conv2d_codecov test";
-
-    default: params = Pooling2dFloat::GetParam();
-    }
-
-    for(const auto& test_value : params)
-    {
-        std::vector<std::string> tokens;
-        GetArgs(test_value, tokens);
-        std::vector<const char*> ptrs;
-
-        std::transform(tokens.begin(), tokens.end(), std::back_inserter(ptrs), [](const auto& str) {
-            return str.data();
-        });
-
-        testing::internal::CaptureStderr();
-        test_drive<pooling2d_driver>(ptrs.size(), ptrs.data());
-        auto capture = testing::internal::GetCapturedStderr();
-        std::cout << capture;
-    }
-};
-
-bool IsTestSupportedForDevice(const miopen::Handle& handle) { return true; }
-
-TEST_P(Pooling2dFloat, FloatTest)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run2dDriver(miopenFloat);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-TEST_P(Pooling2dHalf, HalfTest)
-{
-    const auto& handle = get_handle();
-    if(IsTestSupportedForDevice(handle) && !SkipTest())
-    {
-        Run2dDriver(miopenHalf);
-    }
-    else
-    {
-        GTEST_SKIP();
-    }
-};
-
-std::vector<std::string> GetTestCases(const std::string precision)
-{
-    const auto& flag_arg = miopen::GetStringEnv(ENV(MIOPEN_TEST_FLAGS_ARGS));
-
-    const std::vector<std::string> test_cases = {
-        // clang-format off
-    {"test_pooling2d " + precision + " --all --dataset 2 --limit 0 "+flag_arg}
-        // clang-format on
-    };
-
-    return test_cases;
-}
-
-INSTANTIATE_TEST_SUITE_P(Pooling2D, Pooling2dFloat, testing::Values(GetTestCases("--float")));
-
-INSTANTIATE_TEST_SUITE_P(Pooling2D, Pooling2dHalf, testing::Values(GetTestCases("--half")));
diff --git a/test/immed_conv2d.cpp b/test/immed_conv2d.cpp
new file mode 100644
index 0000000000..6649c69d65
--- /dev/null
+++ b/test/immed_conv2d.cpp
@@ -0,0 +1,62 @@
+/*******************************************************************************
+ *
+ * MIT License
+ *
+ * Copyright (c) 2019 Advanced Micro Devices, Inc.
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a copy
+ * of this software and associated documentation files (the "Software"), to deal
+ * in the Software without restriction, including without limitation the rights
+ * to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
+ * copies of the Software, and to permit persons to whom the Software is
+ * furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included in all
+ * copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+ * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
+ * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+ * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+ * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
+ * SOFTWARE.
+ *
+ *******************************************************************************/
+#include "conv_common.hpp"
+
+template <class T>
+struct conv2d_driver : conv_driver<T, ConvApi::Immediate>
+{
+    conv2d_driver() : conv_driver<T, ConvApi::Immediate>()
+    {
+        this->add(this->input_dims, "input");
+        this->add(this->weight_tensor_dims, "weights");
+        this->add(this->batch_size,
+                  "batch_size",
+                  this->generate_data_limited(this->get_batch_sizes(), 1, {16}));
+        this->add(this->input_channels,
+                  "input_channels",
+                  this->generate_data_limited(this->get_input_channels(), 1, {32}));
+        this->add(this->output_channels,
+                  "output_channels",
+                  this->generate_data_limited(this->get_output_channels(), 1, {32}));
+        this->add(this->spatial_dim_elements,
+                  "spatial_dim_elements",
+                  this->generate_data_limited(this->get_2d_spatial_dims(), 1, {56, 56}));
+        this->add(this->filter_dims,
+                  "filter_dims",
+                  this->generate_data_limited(this->get_2d_filter_dims(), 2, {3, 3}));
+        this->add(this->pads_strides_dilations,
+                  "pads_strides_dilations",
+                  this->generate_data_limited(this->get_2d_pads_strides_dilations(), 2));
+        this->add(this->trans_output_pads,
+                  "trans_output_pads",
+                  this->generate_data_limited(this->get_2d_trans_output_pads(), 1));
+        this->add(this->in_layout, "in_layout", this->generate_data({"NCHW"}));
+        this->add(this->fil_layout, "fil_layout", this->generate_data({"NCHW"}));
+        this->add(this->out_layout, "out_layout", this->generate_data({"NCHW"}));
+    }
+};
+
+int main(int argc, const char* argv[]) { test_drive<conv2d_driver>(argc, argv); }
diff --git a/test/conv3d.hpp b/test/immed_conv3d.cpp
similarity index 91%
rename from test/conv3d.hpp
rename to test/immed_conv3d.cpp
index 2b49c0f10e..a968cd6100 100644
--- a/test/conv3d.hpp
+++ b/test/immed_conv3d.cpp
@@ -26,9 +26,9 @@
 #include "conv_common.hpp"
 
 template <class T>
-struct conv3d_driver : conv_driver<T>
+struct conv3d_driver : conv_driver<T, ConvApi::Immediate>
 {
-    conv3d_driver() : conv_driver<T>()
+    conv3d_driver() : conv_driver<T, ConvApi::Immediate>()
     {
         this->add(this->input_dims, "input");
         this->add(this->weight_tensor_dims, "weights");
@@ -37,16 +37,16 @@ struct conv3d_driver : conv_driver<T>
                   this->generate_data_limited(this->get_batch_sizes(), 1, {8}));
         this->add(this->input_channels,
                   "input_channels",
-                  this->generate_data_limited(this->get_input_channels(), 1, {32}));
+                  this->generate_data_limited(this->get_input_channels(), 1, {2}));
         this->add(this->output_channels,
                   "output_channels",
-                  this->generate_data_limited(this->get_output_channels(), 1, {32}));
+                  this->generate_data_limited(this->get_output_channels(), 1, {16}));
         this->add(this->spatial_dim_elements,
                   "spatial_dim_elements",
                   this->generate_data_limited(this->get_3d_spatial_dims(), 1, {16, 16, 16}));
         this->add(this->filter_dims,
                   "filter_dims",
-                  this->generate_data_limited(this->get_3d_filter_dims(), 2, {3, 3, 3}));
+                  this->generate_data_limited(this->get_3d_filter_dims(), 2, {5, 5, 5}));
         this->add(this->pads_strides_dilations,
                   "pads_strides_dilations",
                   this->generate_data_limited(this->get_3d_pads_strides_dilations(), 2));
@@ -58,3 +58,5 @@ struct conv3d_driver : conv_driver<T>
         this->add(this->out_layout, "out_layout", this->generate_data({"NCDHW"}));
     }
 };
+
+int main(int argc, const char* argv[]) { test_drive<conv3d_driver>(argc, argv); }
diff --git a/test/pooling2d.hpp b/test/pooling2d.cpp
similarity index 98%
rename from test/pooling2d.hpp
rename to test/pooling2d.cpp
index 128e81cce2..c0df88c481 100644
--- a/test/pooling2d.hpp
+++ b/test/pooling2d.cpp
@@ -105,3 +105,5 @@ struct pooling2d_driver : pooling_driver<T>
         this->add(this->wsidx, "wsidx", this->generate_data({0, 1}));
     }
 };
+
+int main(int argc, const char* argv[]) { test_drive<pooling2d_driver>(argc, argv); }

From bfecff655e9556a9a89a36a759516c964d489116 Mon Sep 17 00:00:00 2001
From: Jun Liu <Liu.Jun@amd.com>
Date: Sun, 17 Dec 2023 22:12:54 -0800
Subject: [PATCH 2/3] re-enable smoke tests in CI

---
 Jenkinsfile | 6 +++---
 1 file changed, 3 insertions(+), 3 deletions(-)

diff --git a/Jenkinsfile b/Jenkinsfile
index 7e07d3deba..a103509f23 100644
--- a/Jenkinsfile
+++ b/Jenkinsfile
@@ -436,15 +436,15 @@ pipeline {
             description: "")
         booleanParam(
             name: "BUILD_SMOKE_FP32",
-            defaultValue: env.BRANCH_NAME == env.NIGHTLY_BRANCH ? true : false,
+            defaultValue: true,
             description: "")
         booleanParam(
             name: "BUILD_SMOKE_AUX1",
-            defaultValue: env.BRANCH_NAME == env.NIGHTLY_BRANCH ? true : false,
+            defaultValue: true,
             description: "")
         booleanParam(
             name: "BUILD_SMOKE_FP16_BF16_INT8",
-            defaultValue: env.BRANCH_NAME == env.NIGHTLY_BRANCH ? true : false,
+            defaultValue: true,
             description: "")
         booleanParam(
             name: "BUILD_FULL_TESTS",

From 3d2a0a31d43d185b024203894805179a9928d7c5 Mon Sep 17 00:00:00 2001
From: Jun Liu <Liu.Jun@amd.com>
Date: Sun, 17 Dec 2023 22:47:16 -0800
Subject: [PATCH 3/3] remove problematic github action

---
 .github/workflows/update_develop_nightly.yml | 20 --------------------
 1 file changed, 20 deletions(-)
 delete mode 100644 .github/workflows/update_develop_nightly.yml

diff --git a/.github/workflows/update_develop_nightly.yml b/.github/workflows/update_develop_nightly.yml
deleted file mode 100644
index a8b5aabdb9..0000000000
--- a/.github/workflows/update_develop_nightly.yml
+++ /dev/null
@@ -1,20 +0,0 @@
-name: Sync branch
-
-on:
-  pull_request:
-  workflow_dispatch:
-  push:
-    branches:
-      - develop
-jobs:
-  sync-branch:
-    name: Update nightly branch
-    runs-on: ubuntu-latest
-    steps:
-      - name: Checkout repository
-        uses: actions/checkout@main
-      - uses: connor-baer/action-sync-branch@main
-        with:
-          branch: develop_nightly
-          token: ${{ secrets.GITHUB_TOKEN }}
-          force: false
\ No newline at end of file