Skip to content

Commit

Permalink
Fix lint
Browse files Browse the repository at this point in the history
  • Loading branch information
echuraev committed Jun 8, 2022
1 parent edd94cb commit 9f19d71
Show file tree
Hide file tree
Showing 3 changed files with 13 additions and 38 deletions.
18 changes: 3 additions & 15 deletions python/tvm/topi/adreno/conv2d_nchw_winograd.py
Original file line number Diff line number Diff line change
Expand Up @@ -18,21 +18,7 @@
"""Winograd NCHW template for Adreno backend"""

import logging
import tvm
from tvm import te
from tvm import autotvm

from tvm.topi import nn
from tvm.topi.utils import get_const_int, get_const_tuple, traverse_inline
from ..nn.winograd_util import winograd_transform_matrices
from .utils import (
split_to_chunks,
pack_input,
pack_filter,
bind_data_copy,
get_texture_storage,
infer_tile_size,
)
from .conv2d_winograd_common import conv2d_winograd_comp, schedule_conv2d_winograd_impl


Expand Down Expand Up @@ -137,4 +123,6 @@ def conv2d_nchw_winograd_comp(
output: tvm.te.Tensor
4-D or 5-D with shape NCHW or NCHW4c
"""
return conv2d_winograd_comp(cfg, data, kernel, strides, padding, dilation, out_dtype, args, pre_computed, "NCHW")
return conv2d_winograd_comp(
cfg, data, kernel, strides, padding, dilation, out_dtype, args, pre_computed, "NCHW"
)
18 changes: 3 additions & 15 deletions python/tvm/topi/adreno/conv2d_nhwc_winograd.py
Original file line number Diff line number Diff line change
Expand Up @@ -18,21 +18,7 @@
"""Winograd NHWC template for Adreno backend"""

import logging
import tvm
from tvm import te
from tvm import autotvm

from tvm.topi import nn
from tvm.topi.utils import get_const_int, get_const_tuple, traverse_inline
from ..nn.winograd_util import winograd_transform_matrices
from .utils import (
split_to_chunks,
pack_input,
pack_filter,
bind_data_copy,
get_texture_storage,
infer_tile_size,
)
from .conv2d_winograd_common import conv2d_winograd_comp, schedule_conv2d_winograd_impl


Expand Down Expand Up @@ -137,4 +123,6 @@ def conv2d_nhwc_winograd_comp(
output: tvm.te.Tensor
4-D or 5-D with shape NCHW or NCHW4c
"""
return conv2d_winograd_comp(cfg, data, kernel, strides, padding, dilation, out_dtype, args, pre_computed, "NHWC")
return conv2d_winograd_comp(
cfg, data, kernel, strides, padding, dilation, out_dtype, args, pre_computed, "NHWC"
)
15 changes: 7 additions & 8 deletions python/tvm/topi/adreno/conv2d_winograd_common.py
Original file line number Diff line number Diff line change
Expand Up @@ -17,7 +17,6 @@
# pylint: disable=invalid-name,unused-variable,unused-argument
"""Common Winograd implementation for Adreno backend"""

import logging
import tvm
from tvm import te
from tvm import autotvm
Expand Down Expand Up @@ -80,7 +79,7 @@ def conv2d_winograd_comp(
output: tvm.te.Tensor
4-D or 5-D with shape NCHW or NCHW4c
"""
assert layout == "NCHW" or layout == "NHWC"
assert layout in ("NCHW", "NHWC")
tile_size = infer_tile_size(data, layout)

if isinstance(dilation, int):
Expand Down Expand Up @@ -301,9 +300,9 @@ def conv2d_winograd_comp(
else:
output = te.compute(
(N, CO, H, W, COB),
lambda n, co, h, w, cob: inverse[co][n * nH * nW + idxdiv(h, m) * nW + idxdiv(w, m)][
idxmod(h, m)
][idxmod(w, m)][cob].astype(out_dtype),
lambda n, co, h, w, cob: inverse[co][
n * nH * nW + idxdiv(h, m) * nW + idxdiv(w, m)
][idxmod(h, m)][idxmod(w, m)][cob].astype(out_dtype),
name="output",
tag="cast_from_acc" + args["accumulator"][-2:],
)
Expand All @@ -320,9 +319,9 @@ def conv2d_winograd_comp(
else:
output = te.compute(
(N, H, W, CO, COB),
lambda n, h, w, co, cob: inverse[co][n * nH * nW + idxdiv(h, m) * nW + idxdiv(w, m)][
idxmod(h, m)
][idxmod(w, m)][cob].astype(out_dtype),
lambda n, h, w, co, cob: inverse[co][
n * nH * nW + idxdiv(h, m) * nW + idxdiv(w, m)
][idxmod(h, m)][idxmod(w, m)][cob].astype(out_dtype),
name="output",
tag="cast_from_acc" + args["accumulator"][-2:],
)
Expand Down

0 comments on commit 9f19d71

Please sign in to comment.