[Bug] MIOpen Error: No invoker was registered for convolution forward. Was find executed?

[WsqRichards]

Expected behavior

The yolov5s network runs through relay.build, calls export_library() to save the model, and then runs load_module again for a second time, expecting to run successfully, but actually reporting an error.

Actual behavior

The yolov5s network runs through relay.build, calls export_library() to save the model, and then runs load_module again for a second time, expecting to run successfully, but actually reporting an error.

Error log

MIOpen Error: /long_pathname_so_that_rpms_can_package_the_debug_info/src/extlibs/MLOpen/src/ocl/convolutionocl.cpp:456: No invoker was registered for convolution forward. Was find executed?
^[[ATraceback (most recent call last):
  File "/home1/richard/tvm-demo/network/yolo/tvm_yolov5.py", line 107, in <module>
    tvm_output = run_tvm_module(module, image, 0, dtype, args.perf)
                 ^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^^
  File "/home1/richard/tvm-demo/network/yolo/../tvm_func.py", line 91, in run_tvm_module
    module.run()
  File "/home1/richard/tvm/python/tvm/contrib/graph_executor.py", line 264, in run
    self._run()
  File "/home1/richard/tvm/python/tvm/_ffi/_ctypes/packed_func.py", line 239, in __call__
    raise_last_ffi_error()
  File "/home1/richard/tvm/python/tvm/_ffi/base.py", line 476, in raise_last_ffi_error
    raise py_err
  File "/home/richard/h1/tvm/src/runtime/graph_executor/graph_executor.cc", line 680, in tvm::runtime::GraphExecutor::GetFunction(tvm::runtime::String const&, tvm::runtime::ObjectPtr<tvm::runtime::Object> const&)::{lambda(tvm::runtime::TVMArgs, tvm::runtime::TVMRetValue*)#8}::operator()(tvm::runtime::TVMArgs, tvm::runtime::TVMRetValue*) const
    return PackedFunc([sptr_to_self, this](TVMArgs args, TVMRetValue* rv) { this->Run(); });
                  ^^^^^^^^^^^^^^^^^^^^^^^^^^^
  File "/home/richard/h1/tvm/src/runtime/graph_executor/graph_executor.cc", line 64, in tvm::runtime::GraphExecutor::Run()
    if (op_execs_[i]) op_execs_[i]();
                  ^^^^^^^^^^^^^^^^^^^^
  File "/home/richard/h1/tvm/src/runtime/graph_executor/graph_executor.cc", line 607, in tvm::runtime::GraphExecutor::CreateTVMOp(tvm::runtime::TVMOpParam const&, std::vector<DLTensor*, std::allocator<DLTensor*> > const&)::{lambda()#3}::operator()() const
    pf.CallPacked(targs, &rv);
                  ^^^^^^^^^^^^^
  File "/home/richard/h1/tvm/src/runtime/library_module.cc", line 76, in tvm::runtime::WrapPackedFunc(int (*)(TVMValue*, int*, int, TVMValue*, int*, void*), tvm::runtime::ObjectPtr<tvm::runtime::Object> const&)::{lambda(tvm::runtime::TVMArgs, tvm::runtime::TVMRetValue*)#1}::operator()(tvm::runtime::TVMArgs, tvm::runtime::TVMRetValue*) const
    ICHECK_EQ(ret, 0) << TVMGetLastError();
                  ^^^^^^^^^^^^^^^^^^^^^^^^^^
tvm.error.InternalError: Traceback (most recent call last):
  3: tvm::runtime::GraphExecutor::GetFunction(tvm::runtime::String const&, tvm::runtime::ObjectPtr<tvm::runtime::Object> const&)::{lambda(tvm::runtime::TVMArgs, tvm::runtime::TVMRetValue*)#8}::operator()(tvm::runtime::TVMArgs, tvm::runtime::TVMRetValue*) const
        at /home/richard/h1/tvm/src/runtime/graph_executor/graph_executor.cc:680
  2: tvm::runtime::GraphExecutor::Run()
        at /home/richard/h1/tvm/src/runtime/graph_executor/graph_executor.cc:64
  1: tvm::runtime::GraphExecutor::CreateTVMOp(tvm::runtime::TVMOpParam const&, std::vector<DLTensor*, std::allocator<DLTensor*> > const&)::{lambda()#3}::operator()() const
        at /home/richard/h1/tvm/src/runtime/graph_executor/graph_executor.cc:607
  0: tvm::runtime::WrapPackedFunc(int (*)(TVMValue*, int*, int, TVMValue*, int*, void*), tvm::runtime::ObjectPtr<tvm::runtime::Object> const&)::{lambda(tvm::runtime::TVMArgs, tvm::runtime::TVMRetValue*)#1}::operator()(tvm::runtime::TVMArgs, tvm::runtime::TVMRetValue*) const
        at /home/richard/h1/tvm/src/runtime/library_module.cc:76
  0: tvm::contrib::miopen::{lambda(tvm::runtime::TVMArgs, tvm::runtime::TVMRetValue*)#2}::operator()(tvm::runtime::TVMArgs, tvm::runtime::TVMRetValue*) const
        at /home/richard/h1/tvm/src/runtime/contrib/miopen/conv_forward.cc:194
  File "/home/richard/h1/tvm/src/runtime/contrib/miopen/conv_forward.cc", line 198
  File "/home/richard/h1/tvm/src/runtime/library_module.cc", line 76
InternalError: Check failed: ret == 0 (-1 vs. 0) : InternalError: Check failed: e == miopenStatusSuccess (7 vs. 0) : miopen error: miopenStatusUnknownError

Environment

• AMD Radeon RX 7600 : target = tvm.target.Target("rocm -libs=miopen,rocblas")
• tvm 0.14.0
• ubuntu22.04
• network: yolov5s
• rocm 5.7.1
• miopen 2.18.0

[WsqRichards]

https://rocm.docs.amd.com/projects/MIOpen/en/latest/convolution.html#miopenconvolutionforward

miopenConvolutionForward(miopenHandle_t handle, 
const void *alpha, 
const miopenTensorDescriptor_t xDesc, 
const void *x, 
const miopenTensorDescriptor_t wDesc, 
const void *w, 
const miopenConvolutionDescriptor_t convDesc, 
miopenConvFwdAlgorithm_t algo, 
const void *beta, 
const miopenTensorDescriptor_t yDesc, 
void *y,
void *workSpace, 
size_t workSpaceSize)

Execute a forward convolution layer.

Runs the forward convolution layer based on the selected algorithm. The function [miopenFindConvolutionForwardAlgorithm()]must have been executed previously to determine the required memory needed for the workspace and the best convolutional algorithm.

If using Group/Depthwise convolution mode, call [miopenSetConvolutionGroupCount()] before running this.