(tvm) branch test_all_cases_on_unity updated: try disable stream

[syfeng]

This is an automated email from the ASF dual-hosted git repository.

syfeng pushed a commit to branch test_all_cases_on_unity
in repository https://gitbox.apache.org/repos/asf/tvm.git


The following commit(s) were added to refs/heads/test_all_cases_on_unity by 
this push:
     new fc5e0346f9 try disable stream
fc5e0346f9 is described below

commit fc5e0346f9a4ecd57e60d0856e7049c7915dcd65
Author: Siyuan Feng <hzfen...@sjtu.edu.cn>
AuthorDate: Fri Dec 15 16:48:10 2023 +0800

    try disable stream
---
 gallery/how_to/work_with_relay/using_pipeline_executor.py | 3 ++-
 python/tvm/contrib/cutlass/gemm_operation.py              | 7 +------
 tests/scripts/task_config_build_gpu.sh                    | 1 -
 3 files changed, 3 insertions(+), 8 deletions(-)

diff --git a/gallery/how_to/work_with_relay/using_pipeline_executor.py 
b/gallery/how_to/work_with_relay/using_pipeline_executor.py
index 8bb53cc743..97496b86dc 100644
--- a/gallery/how_to/work_with_relay/using_pipeline_executor.py
+++ b/gallery/how_to/work_with_relay/using_pipeline_executor.py
@@ -109,7 +109,8 @@ subgraphs = graph_split(net["main"], split_config, params)
 cutlass = tvm.target.Target(
     {
         "kind": "cutlass",
-        "sm": int(tvm.target.Target("cuda").arch.split("_")[1]),
+        "sm": 75,
+        # "sm": int(tvm.target.Target("cuda").arch.split("_")[1]),
         "use_3xtf32": True,
         "split_k_slices": [1],
         "profile_all_alignments": False,
diff --git a/python/tvm/contrib/cutlass/gemm_operation.py 
b/python/tvm/contrib/cutlass/gemm_operation.py
index 2639a0359a..4f5d7e5322 100644
--- a/python/tvm/contrib/cutlass/gemm_operation.py
+++ b/python/tvm/contrib/cutlass/gemm_operation.py
@@ -344,12 +344,7 @@ def instantiate_gemm_template(attrs):
   CHECK(status == cutlass::Status::kSuccess);
   status = gemm_op.initialize(arguments, workspace.get());
   CHECK(status == cutlass::Status::kSuccess);
-
-  auto func = tvm::runtime::Registry::Get("runtime.get_cuda_stream");
-  ICHECK(func != nullptr);
-  cudaStream_t stream = static_cast<cudaStream_t>((*func)().operator void*());
-
-  status = gemm_op(stream);
+  status = gemm_op();
   CHECK(status == cutlass::Status::kSuccess);
 """
     op_type = attrs["op_type"]
diff --git a/tests/scripts/task_config_build_gpu.sh 
b/tests/scripts/task_config_build_gpu.sh
index e2ed950628..37ab0a87f1 100755
--- a/tests/scripts/task_config_build_gpu.sh
+++ b/tests/scripts/task_config_build_gpu.sh
@@ -54,4 +54,3 @@ echo set\(USE_PIPELINE_EXECUTOR ON\) >> config.cmake
 echo set\(USE_CUTLASS ON\) >> config.cmake
 echo set\(USE_CMSISNN ON\) >> config.cmake
 echo set\(USE_MSC ON\) >> config.cmake
-echo set\(USE_GRAPH_EXECUTOR_CUDA_GRAPH ON\) >> config.cmake