Introduce Apple BNNS backend #7299

apeskov · 2021-01-16T16:01:38Z

This is simple JSON based runtime which offload execution of some operation into Accelerate frameworks via BNNS api.

Works only for next Apple platforms:

macOS 11.0 and later
iOS 14.0 and later

List of supported primitives:

conv2d and fusing with bias and relu
dense and fusing with bias and relu/gelu
batch_matmul

Thanks for contributing to TVM! Please refer to guideline https://tvm.apache.org/docs/contribute/ for useful information and tips. After the pull request is submitted, please request code reviews from Reviewers by @ them in the pull request thread.

leandron

Reviewed mostly the Python sources.

python/tvm/relay/op/contrib/bnns.py

leandron · 2021-01-18T10:27:33Z

python/tvm/relay/op/contrib/bnns.py

+        if len(b_shape) == 4:
+            return bool(b_shape[0] == 1 and b_shape[2] == 1 and b_shape[3] == 1)
+        elif len(b_shape) == 3:
+            return bool(b_shape[1] == 1 and b_shape[2] == 1)


Why do we need these bool() calls here after the expression?

b_shape[0] == 1 is an instance of tvm.tir.expr.IntImm. External consumer of this function expect bool value and further "bool like" processing lead to exception.

This explicit cast was a shorter way to avoid this.

Understood, please mark this one a closed.

python/tvm/relay/op/contrib/bnns.py

leandron · 2021-01-18T10:37:57Z

tests/python/contrib/test_bnns/infrastructure.py

+    Parameters
+    ----------
+    space: List[List[Any]]
+        A list of different options with varying values to test.
+    r_factor: (optional) int
+        The repeat factor.
+


There is a mix of numpy-style docstrings and other formats (in other Python sources in this same PR as well). I think it would be better to make it consistent, as this is quite a big chunk of work.

Once one is picked, please check the other source files in this PR.

FrozenGene · 2021-01-18T10:46:21Z

Hi @apeskov Thanks for your contribution. Could you add more introduction for Apple BNNS backend in your description? I think this will help others know more about it. Thanks!

This is simple JSON based runtime which offload execution of some operation into Accelerate frameworks via BNNS api. Works only for: * macOS 11.0 and later * iOS 14.0 and later Supported primitives: * conv2d and fusing with bias and relu * dense and fusing with bias and relu/gelu * batch_matmul Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

Also fix some pylint issues Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

Supports `nn.max_pool2d`, `nn.avg_pool2d`, `nn.global_max_pool2d` and `nn.global_avg_pool2d` operations

jwfromm · 2021-03-01T17:08:59Z

@FrozenGene @leandron can you guys take another look at this PR? It seems to have come a long way in the last month.

leandron

In general it looks OK to me. A few comments on the example snippets, plus kindly asking for TVMC support.

leandron · 2021-03-01T17:39:27Z

docs/deploy/bnns.rst

+    with tvm.transform.PassContext(opt_level=3):
+        model = partition_for_bnns(model, params=params)


Does this need to be inside the PassContext? Other targets don't required that.

No, it is not required, thank you.

leandron · 2021-03-01T17:42:57Z

tests/python/contrib/test_bnns/test_pooling.py

+    if skip_codegen_test():
+        return


the use of this idiom here, will make the test be marked as "passed" when BNNS is not available. Please check how to use the @pytest.mark.skipif decorator.

This applies to all tests using this if statement.

leandron · 2021-03-01T17:48:29Z

docs/deploy/bnns.rst

+    target = "llvm -mtriple=x86_64-apple-darwin20.2.0"
+
+    with tvm.transform.PassContext(opt_level=3):
+        model = partition_for_bnns(model, params=params)  # to markup operations to be offloaded to BNNS


This probably doesn't need to be nested into the PassContext, can you check?

You are right, thanks

leandron · 2021-03-01T17:50:40Z

python/tvm/relay/op/contrib/bnns.py

+from ...dataflow_pattern import wildcard, is_op, is_expr
+
+
+def partition_for_bnns(mod, params=None):


Do you mind also checking how to integrate this in TVMC, by binding this function into:

tvm/python/tvm/driver/tvmc/composite_target.py

Lines 34 to 43 in 51dc332

REGISTERED_CODEGEN = {

"compute-library": {

"config_key": None,

"pass_pipeline": partition_for_arm_compute_lib,

},

"ethos-n77": {

"config_key": "relay.ext.ethos-n.options",

"pass_pipeline": partition_for_ethosn,

},

}

echuraev · 2021-03-05T14:19:27Z

@FrozenGene @leandron could you please take a look on this PR one again?

FrozenGene · 2021-03-06T02:45:01Z

docs/deploy/bnns.rst

+----------------------------------------
+
+Operations to be offloaded on BNNS execution must be annotated before passing of module for compilation.
+All opps annotated by `partition_for_bnns` will be offloaded for BNNS execution. The rest of the ops


opps -> ops

Thank you, I fixed it.

FrozenGene · 2021-03-06T02:46:37Z

docs/deploy/bnns.rst

+
+Important note: BNNS support primitives only with constant weights. To satisfy this requirements we have
+to map constants to related tensor abstraction in relay representation. To freeze tensors and operate
+with them as with constants you may need to call ONNX importer with special flag "freeze_params=True"


as with -> as

FrozenGene · 2021-03-06T02:47:56Z

docs/deploy/bnns.rst

+
+
+Input data layout for operations to be offloaded to BNNS execution
+----------------------------------------


I wonder whether we have problem if we don't align -------- with words, other places have same problem

FrozenGene · 2021-03-06T03:37:07Z

python/tvm/relay/op/contrib/bnns.py

+            transform.AnnotateTarget("bnns"),
+            #   If you no need in per layer performance statistic you can
+            #   uncomment next line
+            # transform.MergeCompilerRegions(),


Does it have effect on end to end network performance?

No, it doesn't have effect on performance. But if we use this transformation, then it will be more difficult to compare per layer statistic with initial topology due to some kernels will be merged.

leandron

One minor nit. Apart from that LGTM.

leandron · 2021-03-09T14:48:57Z

tests/python/contrib/test_bnns/infrastructure.py

+    Common device configuration for python tests.
+
+    Check tests/python/contrib/arm_compute_lib/ for the presence of an test_config.json file.
+    This file can be used to override the default configuration here which will attempt to run the Arm
+    Compute Library runtime tests locally if the runtime is available. Changing the configuration
+    will allow these runtime tests to be offloaded to a remote Arm device via a tracker for example.


Does this need some updating to refer to BNNS instead?

Thank you. I updated this comment and double checked that we don't have any refers to ARC in BNNS tests.

jwfromm · 2021-03-12T01:54:38Z

Thanks @apeskov, @FrozenGene, @leandron, and @echuraev. This is now merged and will be an awesome addition to TVM!

* Introduce Apple BNNS backend This is simple JSON based runtime which offload execution of some operation into Accelerate frameworks via BNNS api. Works only for: * macOS 11.0 and later * iOS 14.0 and later Supported primitives: * conv2d and fusing with bias and relu * dense and fusing with bias and relu/gelu * batch_matmul Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * [BNNS] Add conv2d DW test Also fix some pylint issues Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * [BNNS] Fix clang-format issues Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * [BNNS] Refactoring. Add TView abstraction Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * [BNNS] Add several more onnx topologies into tests Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * [BNNS] Avoid redundant tensor allocation Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * [BNNS] Fix conv_splitter issue Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * [BNNS] Fix isse with bias {1,1,1,1} Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * [BNNS] Min. Rename file Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * Fix review comments. Initial Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * [BNNS] test refactoring Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * [BNNS] Fix cpplint issues Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * [BNNS] Fix clang-format issues Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * Fix python format Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * Fix pylint issues Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * [BNNS] Fix pylint. Second attempt Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * [BNNS] Add integration documentation * Check onnx import before use Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * [BNNS] Add instance normalization operator * Add fusing sigmoid activation after conv2d * min changes Signed-off-by: Alexander Peskov <peskovnn@gmail.com> * Add pooling operations to BNNS runtime Supports `nn.max_pool2d`, `nn.avg_pool2d`, `nn.global_max_pool2d` and `nn.global_avg_pool2d` operations * Fix lint * Fix lint * Apply comments * Fix documentation * Fix comment to refer to BNNS Co-authored-by: dlexplorer <elvin.nnov@gmail.com> Co-authored-by: Egor Churaev <egor.churaev@gmail.com>

leandron requested changes Jan 18, 2021

View reviewed changes

ZihengJiang added the status: WIP label Jan 20, 2021

apeskov force-pushed the ap/bnns-runtime branch from 19a75c1 to 47dd249 Compare January 27, 2021 18:33

echuraev force-pushed the ap/bnns-runtime branch 4 times, most recently from cf50a8f to 89a4bee Compare February 25, 2021 09:18

apeskov and others added 22 commits February 26, 2021 00:30

[BNNS] Add conv2d DW test

3552b8a

Also fix some pylint issues Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

[BNNS] Fix clang-format issues

12e6f3e

Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

[BNNS] Refactoring. Add TView abstraction

a9a9f10

Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

[BNNS] Add several more onnx topologies into tests

ca72944

Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

[BNNS] Avoid redundant tensor allocation

c3bf919

Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

[BNNS] Fix conv_splitter issue

d54576d

Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

[BNNS] Fix isse with bias {1,1,1,1}

331d8e0

Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

[BNNS] Min. Rename file

d31063e

Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

Fix review comments. Initial

d67f6b7

Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

[BNNS] test refactoring

7da6a26

Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

[BNNS] Fix cpplint issues

6aecc42

Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

[BNNS] Fix clang-format issues

4cdacf1

Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

Fix python format

1060b98

Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

Fix pylint issues

7203493

Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

[BNNS] Fix pylint. Second attempt

83b7be3

Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

[BNNS] Add integration documentation

7ede113

Check onnx import before use

1747896

Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

[BNNS] Add instance normalization operator

1a39265

Add fusing sigmoid activation after conv2d

b70a89b

min changes

9d5945a

Signed-off-by: Alexander Peskov <peskovnn@gmail.com>

Add pooling operations to BNNS runtime

30e2c0f

Supports `nn.max_pool2d`, `nn.avg_pool2d`, `nn.global_max_pool2d` and `nn.global_avg_pool2d` operations

echuraev added 2 commits February 26, 2021 00:31

Fix lint

c7f3705

Fix lint

d444846

Fix lint

d444846

echuraev force-pushed the ap/bnns-runtime branch from 89a4bee to d444846 Compare February 25, 2021 21:31

apeskov marked this pull request as ready for review February 26, 2021 19:07

leandron requested changes Mar 1, 2021

View reviewed changes

Apply comments

81479db

FrozenGene reviewed Mar 6, 2021

View reviewed changes

Fix documentation

eaffeab

leandron reviewed Mar 9, 2021

View reviewed changes

Fix comment to refer to BNNS

0103e70

FrozenGene approved these changes Mar 10, 2021

View reviewed changes

leandron approved these changes Mar 10, 2021

View reviewed changes

echuraev approved these changes Mar 10, 2021

View reviewed changes

jwfromm merged commit 68b81ad into apache:main Mar 12, 2021

leandron mentioned this pull request Mar 12, 2021

[TVMC][VitisAI] Enable Vitis AI target through TVMC #7577

Merged

Provide feedback

Saved searches

Use saved searches to filter your results more quickly

Uh oh!

Introduce Apple BNNS backend #7299

Introduce Apple BNNS backend #7299

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

		with tvm.transform.PassContext(opt_level=3):
		model = partition_for_bnns(model, params=params)

		from ...dataflow_pattern import wildcard, is_op, is_expr


		def partition_for_bnns(mod, params=None):

	REGISTERED_CODEGEN = {
	"compute-library": {
	"config_key": None,
	"pass_pipeline": partition_for_arm_compute_lib,
	},
	"ethos-n77": {
	"config_key": "relay.ext.ethos-n.options",
	"pass_pipeline": partition_for_ethosn,
	},
	}



		Input data layout for operations to be offloaded to BNNS execution
		----------------------------------------

Introduce Apple BNNS backend #7299

Introduce Apple BNNS backend #7299

Uh oh!

Conversation

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Uh oh!

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Uh oh!

Uh oh!