code: def conv2d_spatial_pack_nchw(
      cfg, data, kernel, strides, padding, dilation, out_dtype, num_tile
  ):
      """compute define for Conv2d Spatial Pack with NCHW layout"""
      out_dtype = out_dtype or data.dtype
      N, CI, IH, IW = get_const_tuple(data.shape)
      if isinstance(N, tvm.tir.Any):
          N = tvm.te.size_var("n")
      if not isinstance(IH, int) or not isinstance(IW, int):
          raise RuntimeError(
              "ARM winograd conv2d doesn't support dynamic input height or width."
          )
  
      if isinstance(dilation, int):
          dilation_h = dilation_w = dilation
      else:
          dilation_h, dilation_w = dilation
  
      if len(kernel.shape) == 4:
          pre_packed = False
          CO, _, KH, KW = get_const_tuple(kernel.shape)
      else:  # kernel tensor is pre packed
          pre_packed = True
          CO, _, KH, KW, VC = get_const_tuple(kernel.shape)
          CO = CO * VC
  
      dilated_kernel_h = (KH - 1) * dilation_h + 1
      dilated_kernel_w = (KW - 1) * dilation_w + 1
      pad_top, pad_left, pad_bottom, pad_right = get_pad_tuple(
          padding, (dilated_kernel_h, dilated_kernel_w)
      )
      HSTR, WSTR = (
          strides
          if isinstance(strides, (tuple, list))
          else (strides, strides)
      )
      OH = (IH + pad_top + pad_bottom - dilated_kernel_h) // HSTR + 1
      OW = (IW + pad_left + pad_right - dilated_kernel_w) // WSTR + 1
      data_pad = nn.pad(
          data, [0, 0, pad_top, pad_left], [0, 0, pad_bottom, pad_right]
      )
  
      # ==================== define configuration space ====================
      # TODO(@kevinthesun): Support tuning/optimization for dynamic shape.
      n_tuning_axis = N if isinstance(N, int) else 1
      n, co, oh, ow = (
          cfg.axis(n_tuning_axis),
          cfg.axis(CO),
          cfg.axis(OH),
          cfg.axis(OW),
      )
      ci, kh, kw = (
          cfg.reduce_axis(CI),
          cfg.reduce_axis(KH),
          cfg.reduce_axis(KW),
      )
  
      if num_tile == 2:  # for arm cpu
          co, vc = cfg.define_split("tile_co", co, num_outputs=2)
          oh, vh = cfg.define_split("tile_oh", oh, num_outputs=2)
          ow, vw = cfg.define_split("tile_ow", ow, num_outputs=2)
      elif num_tile == 3:  # for mali gpu
          co, _, vc = cfg.define_split("tile_co", co, num_outputs=3)
          oh, _, vh = cfg.define_split("tile_oh", oh, num_outputs=3)
          ow, _, vw = cfg.define_split("tile_ow", ow, num_outputs=3)
      else:
          raise RuntimeError("Invalid num_tile")
  
      cfg.define_reorder(
          "reorder_0",
          [n, co, oh, ow, ci, kh, kw, vh, vw, vc],
          policy="candidate",
          candidate=[
              [n, co, oh, ow, ci, kh, kw, vh, vw, vc],
              [n, co, oh, ow, ci, kh, kw, vc, vh, vw],
          ],
      )
  
      cfg.define_annotate("ann_reduce", [kh, kw], policy="try_unroll")
      cfg.define_annotate(
          "ann_spatial", [vh, vw, vc], policy="try_unroll_vec"
      )
  
      # fallback support
      if cfg.is_fallback:
          if num_tile == 2:  # arm cpu
              ref_log = autotvm.tophub.load_reference_log(
                  "arm_cpu",
                  "rk3399",
                  "conv2d_nchw_spatial_pack.arm_cpu",
              )
              cfg.fallback_with_reference_log(ref_log)
          elif num_tile == 3:  # mali gpu
              ref_log = autotvm.tophub.load_reference_log(
                  "mali", "rk3399", "conv2d_nchw_spatial_pack.mali"
              )
              cfg.fallback_with_reference_log(ref_log)
      # ====================================================================
  
      VC = cfg["tile_co"].size[-1]
      VH = cfg["tile_oh"].size[-1]
      VW = cfg["tile_ow"].size[-1]
  
      kvshape = (CO // VC, CI, KH, KW, VC)
      ovshape = (N, CO // VC, OH // VH, OW // VW, VH, VW, VC)
      oshape = (N, CO, OH, OW)
  
      if dilation_h != 1 or dilation_w != 1:
          # undilate input data
          dvshape = (N, OH // VH, OW // VW, CI, KH, KW, VH, VW)
          data_vec = te.compute(
              dvshape,
              lambda n, h, w, ci, kh, kw, vh, vw: data_pad[n][ci][
                  (h * VH + vh) * HSTR + kh * dilation_h
              ][(w * VW + vw) * WSTR + kw * dilation_w],
              name="data_vec_undilated",
          )
      else:
          dvshape = (
              N,
              OH // VH,
              OW // VW,
              CI,
              VH * HSTR + KH - 1,
              VW * WSTR + KW - 1,
          )
          data_vec = te.compute(
              dvshape,
              lambda n, h, w, ci, vh, vw: data_pad[n][ci][
                  h * VH * HSTR + vh
              ][w * VW * WSTR + vw],
              name="data_vec",
          )
  
      if autotvm.GLOBAL_SCOPE.in_tuning:
          # use "kernel_autotvm" instead of "kernel" to avoid naming conflict with OpenCL keyword
          kernel_vec = tvm.te.placeholder(
              kvshape, kernel.dtype, name="kernel_autotvm"
          )
      else:
          if pre_packed:
              kernel_vec = kernel
          else:
              kernel_vec = te.compute(
                  kvshape,
                  lambda co, ci, kh, kw, vc: kernel[co * VC + vc][ci][
                      kh
                  ][kw],
                  name="kernel_vec",
              )
  
      ci = te.reduce_axis((0, CI), name="ci")
      kh = te.reduce_axis((0, KH), name="kh")
      kw = te.reduce_axis((0, KW), name="kw")
  
      if dilation_h != 1 or dilation_w != 1:
          conv = te.compute(
              ovshape,
              lambda n, co, h, w, vh, vw, vc: te.sum(
                  data_vec[n, h, w, ci, kh, kw, vh, vw].astype(
                      out_dtype
                  )
                  * kernel_vec[co, ci, kh, kw, vc].astype(out_dtype),
                  axis=[ci, kh, kw],
              ),
              name="conv",
          )
      else:
          conv = te.compute(
              ovshape,
              lambda n, co, h, w, vh, vw, vc: te.sum(
                  data_vec[
                      n, h, w, ci, vh * HSTR + kh, vw * WSTR + kw
                  ].astype(out_dtype)
                  * kernel_vec[co, ci, kh, kw, vc].astype(out_dtype),
                  axis=[ci, kh, kw],
              ),
              name="conv",
          )
  
      idxdiv = tvm.tir.indexdiv
      idxmod = tvm.tir.indexmod
  
      output = te.compute(
          oshape,
          lambda n, co, h, w: conv[
              n,
              idxdiv(co, VC),
              idxdiv(h, VH),
              idxdiv(w, VW),
              idxmod(h, VH),
              idxmod(w, VW),
              idxmod(co, VC),
          ],
          name="output_unpack",
          tag="spatial_conv2d_output",
      )
      return output
  
file_path: tvm/python/tvm/topi/arm_cpu/conv2d_spatial_pack.py
imports: from __future__ import absolute_import
  import tvm
  from tvm import te
  from tvm import autotvm
  from tvm.target import Target
  from tvm.autotvm.task.space import SplitEntity, OtherOptionEntity, AnnotateEntity, ReorderEntity
  from ..None import nn
  from ..utils import get_const_tuple
  from ..nn.utils import get_const_int, get_pad_tuple
label_desc: Use image node_id 1 for calling a global function with example usage: conv2d_spatial_pack_nchw(cfg, data, kernel, strides, padding, dilation, out_dtype, num_tile) and returns: output