You are viewing a plain text version of this content. The canonical link for it is here.
Posted to commits@tvm.apache.org by GitBox <gi...@apache.org> on 2020/06/18 10:59:34 UTC
[GitHub] [incubator-tvm-vta] zhanghaohit opened a new pull request #9: Intelfocl support
zhanghaohit opened a new pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9
This PR relates to this [RFC](https://github.com/apache/incubator-tvm/issues/5840).
Main changes are:
- OpenCL driver
- intelfocl implementation for VTA
- add ALU MUL and Load INT_8
This is a basic version as a POC, without much performance optimization. We're still optimizing some parts, and will be submitted as a new PR later.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r456264532
##########
File path: config/intelfocl_sample.json
##########
@@ -0,0 +1,13 @@
+{
+ "TARGET" : "intelfocl",
+ "HW_VER" : "0.0.1",
Review comment:
Agreed. Let bump it up to 0.0.2
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-664791701
> Hi @zhanghaohit , since [apache/incubator-tvm#6092](https://github.com/apache/incubator-tvm/pull/6092) has been merged, kindly retrigger ci testing for this PR?
Hi, we are making some changes now and will push the update today.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] liangfu edited a comment on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
liangfu edited a comment on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-706899033
ci tests failed on tsim seems to be unrelated, @zhanghaohit do you mind retrigger ci to see if it succeeds. (It was [successful previously](https://ci.tlcpack.ai/blue/organizations/jenkins/tvm-vta/detail/PR-12/1/pipeline).)
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r456264177
##########
File path: src/intelfocl/intelfocl_device.cc
##########
@@ -0,0 +1,185 @@
+#include <numeric>
+#include <dmlc/logging.h>
+#include <vta/hw_spec.h>
+#include "intelfocl_device.h"
+#include "aoclutils/aocl_utils.h"
+
+#define CL_STATUS_SUCCESS(x) ((x) == CL_SUCCESS)
+
+void cleanup() {}
Review comment:
This cleanup() function is a callback function from 3rdparty code aoclutils, it gives the user a chance to perform cleanups before calling exit(), if required.
Here our code does not require any special cleanup before exit(), thus the cleanup() was left blank by intention.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r483898612
##########
File path: src/sim/sim_driver.cc
##########
@@ -181,6 +181,51 @@ class SRAM {
}
memset(sram_ptr, 0, kElemBytes * xtotal * op->y_pad_1);
}
+
+ // This is for load 8bits to ACC only
+ void Load_int8(const VTAMemInsn* op,
+ DRAM* dram,
+ uint64_t* load_counter,
+ bool skip_exec) {
+ CHECK_EQ(kBits, VTA_ACC_WIDTH);
+
+ // TODO(zhanghao): extend to other width
+ CHECK_EQ(VTA_ACC_WIDTH, 32);
+ CHECK_EQ(VTA_INP_WIDTH, 8);
Review comment:
Makes total sense thank you
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-663779312
@pasqoc I do 100% agree with you about needing better documentation about what boards are supported, and how the designs get generated (HLS, Chisel, OpenCL etc.), and a white list of working designs. This opens a broader question of hardware CI testing, which would be hugely valuable to catch designs from breaking.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] zhanghaohit commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
zhanghaohit commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-716649049
> ci tests failed on tsim seems to be unrelated, @zhanghaohit do you mind retrigger ci to see if it passes? (It was [successful previously](https://ci.tlcpack.ai/blue/organizations/jenkins/tvm-vta/detail/PR-12/1/pipeline).)
@liangfu Thanks for the comments and sorry for my late reply. I tried to trigger the ci again, but it failed at the same place.
I tried to run `python3 -m pytest -v ${TVM_PATH}/vta/tests/python/unittest` locally. It passed. Any ideas?
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r445322943
##########
File path: config/intelfocl_sample.json
##########
@@ -0,0 +1,13 @@
+{
+ "TARGET" : "intelfocl",
Review comment:
Thank you for the comment, I understand the point here is to differentiate between Chisel-based-soc version and OpenCL-based version.
I think we may need to discuss further about the naming, as intel opencl support both Cloud FPGA and SOC (like Cyclone V)
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego edited a comment on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego edited a comment on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-660622675
> > The code should work for all devices supported by Intel OpenCL for FPGA, namely Intel Arria 10, Stratix V/10 and Cyclone V/10.
>
> Just my humble opinion, given that both "Intel OpenCL for FPGA" and "VTA" requires a large amount of logic utilization, many Cyclone V chips that supports AOCL couldn't get this compiled because of the hardware utilization issue.
> In my understanding of @tmoreau89 's comment, which I would agree, setting a proper target device in vta_config gives the property of being "versatile" in VTA - the Versatile Tensor Accelerator. Specifically, we could take the advantage of the properties in vta_config to define an accelerator that could scale from minimal ones towards data center scale accelerators.
>
> As a side note, the reason we are taking these efforts in building open source projects, in part, we are hoping someone in the community could reproduce what we have done, and could easily start to build something that is even better. With the target being defined too board, an potential grad student could fail to reproduce the result, since not all the student could easily purchase a board with Stratix 10, and low cost Cyclone V boards couldn't get this running. In addition, they're wasting large amount of valuable logic resources, even they could afford a board with Stratix 10. Therefore, we should specify a precise target device for the vta_config.
Thank you for your reply. However, could you explain more on the reason why the design shall not work on Cyclone V FPGAs that supports Intel OpenCL for FPGA?
Precisely as you mentioned, the VTA design is versatile, the user could always change the settings (i.e. LOG_BLOCK and LOG_*_BUF_SIZE) to adjust the resource usage in order to fit their own FPGA boards. Surely a low cost Cyclone V device could not support 64x64 GEMV cores like large Stratix 10 FPGAs do. But the user could always try to use 16x16 or even 4x4 GEMV cores, by setting the LOG_BLOCK lower.
Considering that, we used a relatively small default setting for LOG_BLOCK (4) and Buffers(15, 15, 18, 17). Thus the design should be able to fit into FPGAs comparable to the original Zynq/Zedboard platforms.
We must admit that we don't have those Cyclone V boards on hand, nor has the design been tested on those platforms. However, if there is any issue on compiling the design for a AOCL-compatibale cyclone V board, we will be more than happy to investigate and try to solve the issue together.
In terms of accessibility, we know that high-end could FPGA cards are very expensive. The good news is that nowadays there are many Cloud Service Providers available offering high-end FPGA instances! Those FPGA instances generally only cost few bucks for hour's usage.
In addition, we are also working on porting the design over to Amazon EC2 F1 instances (Xilinx SDAccel). We will update again when we finish testing on the Amazon platforms.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego edited a comment on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego edited a comment on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-660622675
> > The code should work for all devices supported by Intel OpenCL for FPGA, namely Intel Arria 10, Stratix V/10 and Cyclone V/10.
>
> Just my humble opinion, given that both "Intel OpenCL for FPGA" and "VTA" requires a large amount of logic utilization, many Cyclone V chips that supports AOCL couldn't get this compiled because of the hardware utilization issue.
> In my understanding of @tmoreau89 's comment, which I would agree, setting a proper target device in vta_config gives the property of being "versatile" in VTA - the Versatile Tensor Accelerator. Specifically, we could take the advantage of the properties in vta_config to define an accelerator that could scale from minimal ones towards data center scale accelerators.
>
> As a side note, the reason we are taking these efforts in building open source projects, in part, we are hoping someone in the community could reproduce what we have done, and could easily start to build something that is even better. With the target being defined too board, an potential grad student could fail to reproduce the result, since not all the student could easily purchase a board with Stratix 10, and low cost Cyclone V boards couldn't get this running. In addition, they're wasting large amount of valuable logic resources, even they could afford a board with Stratix 10. Therefore, we should specify a precise target device for the vta_config.
Thank you for your reply. However, could you explain more on the reason why the design shall not work on Cyclone V FPGAs that supports Intel OpenCL for FPGA?
Precisely as you mentioned, the VTA design is versatile, the user could always change the settings (i.e. LOG_BLOCK and LOG_*_BUF_SIZE) to adjust the resource usage in order to fit their own FPGA boards. Surely a low cost Cyclone V device could not support 64x64 GEMV cores like large Stratix 10 FPGAs do. But the user could always try to use 16x16 or even 4x4 GEMV cores, by setting the LOG_BLOCK lower.
Considering that, we used a relatively small default setting for LOG_BLOCK (4) and Buffers(15, 15, 18, 17). Thus the design should be able to fit into FPGAs comparable to the original Zynq/Zedboard platforms.
We must admit that we don't have those Cyclone V boards, nor has the design been tested on those platforms. However, if anyone encountered problems when compiling the design for a AOCL-compatibale cyclone V board, we will be more than happy to investigate and try to solve the issue together.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r445324802
##########
File path: src/intelfocl/AOCLUtils/opencl.h
##########
@@ -0,0 +1,122 @@
+// Copyright (C) 2013-2018 Altera Corporation, San Jose, California, USA. All rights reserved.
Review comment:
Thank you for the comment. Sure, we can move it to 3rdparty directory as suggested.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] liangfu edited a comment on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
liangfu edited a comment on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-706899033
ci tests failed on tsim seems to be unrelated, @zhanghaohit do you mind retrigger ci to see if it passes? (It was [successful previously](https://ci.tlcpack.ai/blue/organizations/jenkins/tvm-vta/detail/PR-12/1/pipeline).)
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r477015151
##########
File path: src/sim/sim_driver.cc
##########
@@ -181,6 +181,51 @@ class SRAM {
}
memset(sram_ptr, 0, kElemBytes * xtotal * op->y_pad_1);
}
+
+ // This is for load 8bits to ACC only
+ void Load_int8(const VTAMemInsn* op,
+ DRAM* dram,
+ uint64_t* load_counter,
+ bool skip_exec) {
+ CHECK_EQ(kBits, VTA_ACC_WIDTH);
+
+ // TODO(zhanghao): extend to other width
+ CHECK_EQ(VTA_ACC_WIDTH, 32);
+ CHECK_EQ(VTA_INP_WIDTH, 8);
Review comment:
Should we include an assertion for the WGT width as well?
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r443088598
##########
File path: config/intelfocl_sample.json
##########
@@ -0,0 +1,13 @@
+{
+ "TARGET" : "intelfocl",
Review comment:
Agreed with the above comment
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r456738642
##########
File path: config/intelfocl_sample.json
##########
@@ -0,0 +1,13 @@
+{
+ "TARGET" : "intelfocl",
Review comment:
I believe "Arria 10" is too restrictive here. The code should work for all devices supported by Intel OpenCL for FPGA, namely Arria 10, Stratix V/10 and Cyclone V/10. That's the reason we name it as "intelfocl".
So far we have tested it on both Arria 10 and Stratix 10 boards, and it worked for both boards.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r456699804
##########
File path: config/intelfocl_sample.json
##########
@@ -0,0 +1,13 @@
+{
+ "TARGET" : "intelfocl",
+ "HW_VER" : "0.0.1",
Review comment:
Please apply the 0.0.2 bump thanks
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r477016190
##########
File path: config/vta_cost.py
##########
@@ -0,0 +1,102 @@
+# cost function for intelfocl 32*32 gemm version
+def cal_cost(insn):
+ """
+ Cal the runtime cost statically
+
+ Parameters
+ ------------
+ insn: the insn (json)
+
+ Returns
+ ------------
+ the cost in s
+ """
+ factor = 1000000.0
+ def alu_imm_cost(iter_out, iter_in, uop_bgn, uop_end):
+ x = (uop_end - uop_bgn) * iter_out * iter_in
+ cycles = x + 46
Review comment:
Actually I think it would be good to define constants here, and add a comment that states that these are set to arbitrarily (e.g. 46).
I'm working with a colleague on making easier to perform some of these cycle accurate co-simulation of VTA modules which should help us derive these costs more accurately on a per-VTA variant basis. In the meantime, can we use constant names, e.g. `ALU_LATENCY_CYCLES = 46`
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] liangfu commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
liangfu commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r456884032
##########
File path: config/intelfocl_sample.json
##########
@@ -0,0 +1,13 @@
+{
+ "TARGET" : "intelfocl",
+ "HW_VER" : "0.0.1",
+ "LOG_INP_WIDTH" : 3,
Review comment:
In this case, we would lose the property of being "versatile" in VTA - the Versatile Tensor Accelerator. I think we need to at least warn users that these options are not actually applicable in the proposed context, or we might need to find a way to define the types accordingly with the given vta_config. @tmoreau89 what's your opinion?
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-648436019
That is correct, we'll need to bump the versioning of the bitstream
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-663776880
thanks @remotego @liangfu @pasqoc for the insightful comments. I think that all opinions expressed are very valid!
So far our approach to naming VTA target has been to couple the VTA target with the FPGA board (e.g. pynq, ultra96, de10nano). These targets also contrast with functional simulation (sim), and cycle accurate sim (tsim).
We've been using the VTA target to guide the compilation process to the target device, given that some of the drivers had to be written in a board specific fashion. In the case of the Intel OpenCL FPGA support, I understand that this work captures a variety of hardware backends, including Arria 10, Stratix10 boards etc, and that the driver codebase would remain mostly identical between those boards (unlike Pynq, and Ultra96 that relied on very different ARM SoCs)
However to find a quick resolution to this discussion we can either choose
- to use a specific board name (rather than FPGA family) to indicate that the OCL FPGA design has been tested on this device. This echo @liangfu's concern that we should have concrete targets for the community to reproduce work on.
- keep the naming open as @remotego and @pasqoc are advocating, and classify this target as intelfocl_pcie to indicate that this applies only to PCIE-based OpenCL Intel FPGA devices.
@remotego let us know what you think
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego edited a comment on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego edited a comment on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-664791701
> Hi @zhanghaohit , since [apache/incubator-tvm#6092](https://github.com/apache/incubator-tvm/pull/6092) has been merged, kindly retrigger ci testing for this PR?
Hi, we are making some changes now and will push the update today. Thanks!
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] pasqoc edited a comment on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
pasqoc edited a comment on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-661202272
Hi @liangfu, have you been able to test on the de10nano (with chisel back-end not opencl ... :-))?
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-695611214
Yes. Sorry for the delay. We are working on it and will update Monday or Tuesday.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r491484630
##########
File path: src/oclfpga/oclfpga_device.cc
##########
@@ -0,0 +1,251 @@
+/*
+ * Licensed to the Apache Software Foundation (ASF) under one
+ * or more contributor license agreements. See the NOTICE file
+ * distributed with this work for additional information
+ * regarding copyright ownership. The ASF licenses this file
+ * to you under the Apache License, Version 2.0 (the
+ * "License"); you may not use this file except in compliance
+ * with the License. You may obtain a copy of the License at
+ *
+ * http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing,
+ * software distributed under the License is distributed on an
+ * "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+ * KIND, either express or implied. See the License for the
+ * specific language governing permissions and limitations
+ * under the License.
+ */
+
+#include "oclfpga_device.h"
+#include <dmlc/logging.h>
+#include <vta/hw_spec.h>
+#include <cstring>
+#include <numeric>
+
+#define CL_STATUS_SUCCESS(x) ((x) == CL_SUCCESS)
+
+static const char *kernel_names[] = {"vta_core"};
+
+static cl_platform_id *find_platform(std::vector<cl_platform_id> *platforms,
+ const std::vector<std::string> &supported_platforms) {
+ cl_int status;
+ size_t size;
+ std::vector<char> name;
+ for (auto &id : *platforms) {
+ status = clGetPlatformInfo(id, CL_PLATFORM_NAME, 0, NULL, &size);
+ if (!CL_STATUS_SUCCESS(status)) continue;
+ name.resize(size);
+ status = clGetPlatformInfo(id, CL_PLATFORM_NAME, name.size(), name.data(), NULL);
+ if (!CL_STATUS_SUCCESS(status)) continue;
+ for (auto &p : supported_platforms) {
+ if (strstr(name.data(), p.c_str()) != NULL) {
+ return &id;
+ }
+ }
+ }
+ return NULL;
+}
+
+OCLFPGADevice::OCLFPGADevice() {
+ std::vector<std::string> supported_platforms = {"Intel(R) FPGA SDK for OpenCL(TM)", "Xilinx"};
+ init(supported_platforms);
+}
+
+void OCLFPGADevice::init(const std::vector<std::string> &supported_platforms) {
+ cl_int status;
+ cl_device_id *device;
+ cl_platform_id *platform;
+ cl_uint n;
+ size_t size;
+ std::vector<char> name;
+ std::vector<cl_platform_id> platforms;
+ std::vector<cl_device_id> devices;
+
+ status = clGetPlatformIDs(0, NULL, &n);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query number of OpenCL platforms";
+ platforms.resize(n);
+ CHECK(platforms.size() > 0) << "No OpenCL platform available";
+ status = clGetPlatformIDs(platforms.size(), platforms.data(), NULL);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query OpenCL platform IDs";
+
+ platform = find_platform(&platforms, supported_platforms);
+ CHECK(platform) << "Unable to find supported OpenCL platform";
+
+ status = clGetDeviceIDs(*platform, CL_DEVICE_TYPE_ALL, 0, NULL, &n);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query number of OpenCL devices";
+ devices.resize(n);
+ CHECK(devices.size() > 0) << "No OpenCL device found";
+ status = clGetDeviceIDs(*platform, CL_DEVICE_TYPE_ALL, devices.size(), devices.data(), NULL);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query OpenCL devices IDs";
+
+ device = NULL;
+ for (auto &id : devices) {
+ _context = clCreateContext(NULL, 1, &id, NULL, NULL, &status);
+ if (CL_STATUS_SUCCESS(status)) {
+ status = clGetDeviceInfo(id, CL_DEVICE_NAME, 0, NULL, &size);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query OpenCL device info";
+ name.resize(size);
+ status = clGetDeviceInfo(id, CL_DEVICE_NAME, name.size(), name.data(), NULL);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query OpenCL device name";
+ LOG(INFO) << "Using FPGA device: " << name.data();
+ device = &id;
+ break;
+ } else {
+ LOG(INFO) << "This FPGA Device is not available. Skipped.";
+ }
+ }
+ CHECK(device) << "No FPGA device available";
+ _device = *device;
+}
+
+int OCLFPGADevice::setup(size_t mem_size, std::string aocx_file) {
+ cl_int status;
+ unsigned int argi;
+ size_t size;
+ FILE *binary_file;
+ unsigned char *binary;
+
+ LOG(INFO) << "Using AOCX: " << aocx_file;
Review comment:
We'd need to use the Intel toolchains
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r456257771
##########
File path: hardware/intelfocl/src/vta.cl
##########
@@ -0,0 +1,341 @@
+#pragma OPENCL EXTENSION cl_intel_channels: enable
Review comment:
The code that work for ultra96 is still inside hardware/xilinx
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-660622675
> > The code should work for all devices supported by Intel OpenCL for FPGA, namely Intel Arria 10, Stratix V/10 and Cyclone V/10.
>
> Just my humble opinion, given that both "Intel OpenCL for FPGA" and "VTA" requires a large amount of logic utilization, many Cyclone V chips that supports AOCL couldn't get this compiled because of the hardware utilization issue.
> In my understanding of @tmoreau89 's comment, which I would agree, setting a proper target device in vta_config gives the property of being "versatile" in VTA - the Versatile Tensor Accelerator. Specifically, we could take the advantage of the properties in vta_config to define an accelerator that could scale from minimal ones towards data center scale accelerators.
>
> As a side note, the reason we are taking these efforts in building open source projects, in part, we are hoping someone in the community could reproduce what we have done, and could easily start to build something that is even better. With the target being defined too board, an potential grad student could fail to reproduce the result, since not all the student could easily purchase a board with Stratix 10, and low cost Cyclone V boards couldn't get this running. In addition, they're wasting large amount of valuable logic resources, even they could afford a board with Stratix 10. Therefore, we should specify a precise target device for the vta_config.
Thank you for your reply. However, could you explain more on the reason why the design will not work on Cyclone V FPGAs?
Precisely as you mentioned, the VTA design is versatile, the user could always change the settings (i.e. LOG_BLOCK and LOG_*_BUF_SIZE) to adjust the resource usage in order to fit their own FPGA boards. Surely a low cost Cyclone V device could not support 64x64 GEMV cores like large Stratix 10 FPGAs do. But the user could always try to use 16x16 or even 4x4 GEMV cores, by setting the LOG_BLOCK lower.
Considering that, we used a relatively small default setting for LOG_BLOCK (4) and Buffers(15, 15, 18, 17). Thus the design should be able to fit into FPGAs comparable to the original Zynq/Zedboard platforms.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r456738642
##########
File path: config/intelfocl_sample.json
##########
@@ -0,0 +1,13 @@
+{
+ "TARGET" : "intelfocl",
Review comment:
I believe "Arria 10" is too restrictive here. The code should work for all devices supported by Intel OpenCL for FPGA, namely Arria 10, Stratix V/10 and Cyclone V/10. That's the reason we name it as "intelfocl".
So far we have tested it on both Arria 10 and Stratix 10 boards, and it worked for both boards.
The correct bitstream will be generated automatically according to the card/BSP/toolchain installed on the system. For example, if you have a Stratix 10 card and toolchain(BSP) installed on your system, the bitstream generated will work on your Stratix 10 card.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-646427173
CC-ing @vegaluisjose @huajsj @pasqoc
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r460334828
##########
File path: config/intelfocl_sample.json
##########
@@ -0,0 +1,13 @@
+{
+ "TARGET" : "intelfocl",
+ "HW_VER" : "0.0.1",
+ "LOG_INP_WIDTH" : 3,
Review comment:
Thanks as a resolution I think we can include assertion checks. Thanks for the catch @liangfu
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] zhanghaohit commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
zhanghaohit commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-647311749
> Thanks @zhanghaohit for the PR! I left a couple questions, but will need to review in more depth. One question: have you tested it on the Pynq board to see if some changes may affect the existing design? Given that the ISA has changed slightly (some fields are now wider, like VTA_MEMOP_ID_BIT_WIDTH), it may affect some parameterizations of VTA.
Thanks @tmoreau89 for the comments. Yes. We have tested on the ultra96 board (I only have ultra96 board on hand). It works. The only thing we have to change is to re-compile the bitstream with the new ISA. I think we have to update the pre-compiled bitstream here [https://github.com/uwsampl/vta-distro/tree/master/bitstreams](https://github.com/uwsampl/vta-distro/tree/master/bitstreams), right?
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] zhanghaohit commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
zhanghaohit commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-660502710
> Thanks for the changes. Please apply the 0.0.2 and rename the vta target to something more specific, e.g. "arria10". Also there are some CI errors related to linting that could be addressed. Thanks!
Thanks. The linting error here is due to filetype checking. I think we have to add the opencl filetype to the lint script. I've created a PR here: https://github.com/apache/incubator-tvm/pull/6092
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r456267226
##########
File path: src/intelfocl/intelfocl_device.cc
##########
@@ -0,0 +1,185 @@
+#include <numeric>
+#include <dmlc/logging.h>
+#include <vta/hw_spec.h>
+#include "intelfocl_device.h"
+#include "aoclutils/aocl_utils.h"
+
+#define CL_STATUS_SUCCESS(x) ((x) == CL_SUCCESS)
+
+void cleanup() {}
+
+int IntelFOCLDevice::init(size_t mem_size, std::string aocx_file)
+{
+ cl_int status;
Review comment:
Sure. Thanks. We will update that.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 removed a comment on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 removed a comment on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-646427173
CC-ing @vegaluisjose @huajsj @pasqoc
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r479842626
##########
File path: src/oclfpga/oclfpga_driver.cc
##########
@@ -0,0 +1,94 @@
+/*
+ * Licensed to the Apache Software Foundation (ASF) under one
+ * or more contributor license agreements. See the NOTICE file
+ * distributed with this work for additional information
+ * regarding copyright ownership. The ASF licenses this file
+ * to you under the Apache License, Version 2.0 (the
+ * "License"); you may not use this file except in compliance
+ * with the License. You may obtain a copy of the License at
+ *
+ * http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing,
+ * software distributed under the License is distributed on an
+ * "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+ * KIND, either express or implied. See the License for the
+ * specific language governing permissions and limitations
+ * under the License.
+ */
+
+#include "oclfpga_device.h"
+#include <vta/driver.h>
+#include <tvm/runtime/registry.h>
+#include <string>
+#include <iostream>
+
+#define MEM_ADDR_IDENTIFIER (0x18000000)
Review comment:
Thank you for the comments. This is an arbitrary value created to differentiate virtual and "physical" offset. We convert offset to virtual address by adding MEM_ADDR_IDENTIFIER to the offset value.
In OpenCL, offset of a memory starts at 0. However, by convention, a valid virtual addresses should not be zero. An identifier could help to avoid returning zero as a virtual address.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 edited a comment on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 edited a comment on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-695346293
Hi @zhanghaohit, please let me know if you can address the following changes so we can get this PR merged!
(1) Adding a README file (ideally with instructions on how to compile / run on the FPGA) before we migrate this to official TVM documentation.
(2) Using constants in the vta cost model
(3) Adding comments to MEM_ADDR_IDENTIFIER initialization
(4) 2-space indentation as recommended by @liangfu
(5) Apply CamelCase for function names and comment as much as possible
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] liangfu commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
liangfu commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r483940899
##########
File path: hardware/intelfocl/src/vta.cl
##########
@@ -0,0 +1,360 @@
+/*
+ * Licensed to the Apache Software Foundation (ASF) under one
+ * or more contributor license agreements. See the NOTICE file
+ * distributed with this work for additional information
+ * regarding copyright ownership. The ASF licenses this file
+ * to you under the Apache License, Version 2.0 (the
+ * "License"); you may not use this file except in compliance
+ * with the License. You may obtain a copy of the License at
+ *
+ * http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing,
+ * software distributed under the License is distributed on an
+ * "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+ * KIND, either express or implied. See the License for the
+ * specific language governing permissions and limitations
+ * under the License.
+ */
+
+#pragma OPENCL EXTENSION cl_intel_channels: enable
+
+#include "vta.h"
+
+__attribute__((reqd_work_group_size(1,1,1)))
+__kernel void vta_core(unsigned int insn_count,
+ unsigned int insns_offset,
+ __global insn_T* restrict insns,
+ __global uop_T* restrict uops,
+ __global acc_T* restrict biases,
+ __global inp_T* restrict inputs,
+ __global wgt_T* restrict weights,
+ __global out_T* restrict outputs)
+{
+ /* Local Memories */
+ uop_T uop_mem[VTA_UOP_BUFF_DEPTH];
Review comment:
Since most of C-like code in TVM follows Google C/C++ style, and to be consistent with other part of the code base, I think we should apply 2 spaces as indentation for OpenCL based hardware designs as well.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-687523850
Thanks again @remotego for addressing the comments. If we could address:
(1) Adding a README file (ideally with instructions on how to compile / run on the FPGA) before we migrate this to official TVM documentation.
(2) Using constants in the vta cost model
(3) Adding comments to `MEM_ADDR_IDENTIFIER` initialization
Then the PR is looking in good shape.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r483898387
##########
File path: src/oclfpga/oclfpga_driver.cc
##########
@@ -0,0 +1,94 @@
+/*
+ * Licensed to the Apache Software Foundation (ASF) under one
+ * or more contributor license agreements. See the NOTICE file
+ * distributed with this work for additional information
+ * regarding copyright ownership. The ASF licenses this file
+ * to you under the Apache License, Version 2.0 (the
+ * "License"); you may not use this file except in compliance
+ * with the License. You may obtain a copy of the License at
+ *
+ * http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing,
+ * software distributed under the License is distributed on an
+ * "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+ * KIND, either express or implied. See the License for the
+ * specific language governing permissions and limitations
+ * under the License.
+ */
+
+#include "oclfpga_device.h"
+#include <vta/driver.h>
+#include <tvm/runtime/registry.h>
+#include <string>
+#include <iostream>
+
+#define MEM_ADDR_IDENTIFIER (0x18000000)
Review comment:
Thanks. If you could add a comment or two above that would be great.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-646427144
CC-ing @vegaluisjose @huajsj @pasqoc
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] liangfu commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
liangfu commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r483942608
##########
File path: src/oclfpga/oclfpga_device.h
##########
@@ -0,0 +1,78 @@
+/*
+ * Licensed to the Apache Software Foundation (ASF) under one
+ * or more contributor license agreements. See the NOTICE file
+ * distributed with this work for additional information
+ * regarding copyright ownership. The ASF licenses this file
+ * to you under the Apache License, Version 2.0 (the
+ * "License"); you may not use this file except in compliance
+ * with the License. You may obtain a copy of the License at
+ *
+ * http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing,
+ * software distributed under the License is distributed on an
+ * "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+ * KIND, either express or implied. See the License for the
+ * specific language governing permissions and limitations
+ * under the License.
+ */
+
+#ifndef _3RDPARTY_VTA_HW_SRC_OCLFPGA_OCLFPGA_DEVICE_H_
+#define _3RDPARTY_VTA_HW_SRC_OCLFPGA_OCLFPGA_DEVICE_H_
+
+#define CL_TARGET_OPENCL_VERSION 120
+#define CL_USE_DEPRECATED_OPENCL_1_2_APIS
+#include <CL/opencl.h>
+#include <list>
+#include <vector>
+#include <string>
+
+#define FOCL_MEM_OFF_ERR (SIZE_MAX)
+
+enum kernel_index {
+ KERNEL_VTA_CORE,
+ NUM_OCL_KERNELS
+};
+
+typedef size_t focl_mem_off_t;
+
+typedef struct {
+ focl_mem_off_t offset;
+ size_t size;
+ bool occupied;
+} mem_chunk_t;
+
+class OCLFPGADevice {
+ private:
+ cl_context _context = NULL;
+ cl_device_id _device = NULL;
+ cl_program _program = NULL;
+ cl_mem _mem = NULL;
+ cl_kernel _kernels[NUM_OCL_KERNELS] = {NULL};
+ cl_command_queue _queues[NUM_OCL_KERNELS] = {NULL};
+ std::list<mem_chunk_t> _mem_chunks;
+ size_t _alignment;
+
+ public:
+ OCLFPGADevice();
+
+ void init(const std::vector<std::string> &supported_platforms);
+
+ int setup(size_t mem_size, std::string aocx_file);
+
+ focl_mem_off_t alloc(size_t size);
+
+ void free(focl_mem_off_t offset);
+
+ void write_mem(focl_mem_off_t offset, const void *buf, size_t nbyte);
+
+ void read_mem(focl_mem_off_t offset, void *buf, size_t nbyte);
+
+ int execute_instructions(focl_mem_off_t offset, size_t count);
Review comment:
Please apply `CamelCase` for the function names and kindly add comments for them as well.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-660364642
Thanks for the changes. Please apply the 0.0.2 and rename the vta target to something more specific, e.g. "arria10". Also there are some CI errors related to linting that could be addressed. Thanks!
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r460335117
##########
File path: src/intelfocl/intelfocl_device.cc
##########
@@ -0,0 +1,185 @@
+#include <numeric>
+#include <dmlc/logging.h>
+#include <vta/hw_spec.h>
+#include "intelfocl_device.h"
+#include "aoclutils/aocl_utils.h"
+
+#define CL_STATUS_SUCCESS(x) ((x) == CL_SUCCESS)
+
+void cleanup() {}
Review comment:
@remotego could you add a comment in the code explaining what you just mentioned? thank you
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-646916505
Thanks @zhanghaohit for the PR! I left a couple questions, but will need to review in more depth. One question: have you tested it on the Pynq board to see if some changes may affect the existing design? Given that the ISA has changed slightly (some fields are now wider, like VTA_MEMOP_ID_BIT_WIDTH), it may affect some parameterizations of VTA.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] liangfu commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
liangfu commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r456229024
##########
File path: hardware/intelfocl/src/vta.cl
##########
@@ -0,0 +1,341 @@
+#pragma OPENCL EXTENSION cl_intel_channels: enable
Review comment:
Just curious, as I remember this PR is going to work for ultra96, does this extension work for Xilinx toolchain to compile the hardware design? I might misunderstood somewhere.
##########
File path: src/intelfocl/intelfocl_device.cc
##########
@@ -0,0 +1,185 @@
+#include <numeric>
+#include <dmlc/logging.h>
+#include <vta/hw_spec.h>
+#include "intelfocl_device.h"
+#include "aoclutils/aocl_utils.h"
+
+#define CL_STATUS_SUCCESS(x) ((x) == CL_SUCCESS)
+
+void cleanup() {}
+
+int IntelFOCLDevice::init(size_t mem_size, std::string aocx_file)
+{
+ cl_int status;
Review comment:
Please ensure the source code pass cpplint rules. Specifically, we use 2 spaces for indentation.
##########
File path: config/intelfocl_sample.json
##########
@@ -0,0 +1,13 @@
+{
+ "TARGET" : "intelfocl",
+ "HW_VER" : "0.0.1",
Review comment:
We might need to update this version number, since the HW ISA has changed?
##########
File path: config/intelfocl_sample.json
##########
@@ -0,0 +1,13 @@
+{
+ "TARGET" : "intelfocl",
+ "HW_VER" : "0.0.1",
+ "LOG_INP_WIDTH" : 3,
Review comment:
Since the types are hard coded as
```c
typedef char inp_T;
typedef char wgt_T;
```
, these config variables (e.g. `LOG_INP_WIDTH`) wouldn't actually reflect the hardware design, right?
##########
File path: src/intelfocl/intelfocl_device.cc
##########
@@ -0,0 +1,185 @@
+#include <numeric>
+#include <dmlc/logging.h>
+#include <vta/hw_spec.h>
+#include "intelfocl_device.h"
+#include "aoclutils/aocl_utils.h"
+
+#define CL_STATUS_SUCCESS(x) ((x) == CL_SUCCESS)
+
+void cleanup() {}
Review comment:
actually clean up?
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] liangfu commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
liangfu commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r483942451
##########
File path: src/oclfpga/oclfpga_device.cc
##########
@@ -0,0 +1,251 @@
+/*
+ * Licensed to the Apache Software Foundation (ASF) under one
+ * or more contributor license agreements. See the NOTICE file
+ * distributed with this work for additional information
+ * regarding copyright ownership. The ASF licenses this file
+ * to you under the Apache License, Version 2.0 (the
+ * "License"); you may not use this file except in compliance
+ * with the License. You may obtain a copy of the License at
+ *
+ * http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing,
+ * software distributed under the License is distributed on an
+ * "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+ * KIND, either express or implied. See the License for the
+ * specific language governing permissions and limitations
+ * under the License.
+ */
+
+#include "oclfpga_device.h"
+#include <dmlc/logging.h>
+#include <vta/hw_spec.h>
+#include <cstring>
+#include <numeric>
+
+#define CL_STATUS_SUCCESS(x) ((x) == CL_SUCCESS)
+
+static const char *kernel_names[] = {"vta_core"};
+
+static cl_platform_id *find_platform(std::vector<cl_platform_id> *platforms,
+ const std::vector<std::string> &supported_platforms) {
+ cl_int status;
+ size_t size;
+ std::vector<char> name;
+ for (auto &id : *platforms) {
+ status = clGetPlatformInfo(id, CL_PLATFORM_NAME, 0, NULL, &size);
+ if (!CL_STATUS_SUCCESS(status)) continue;
+ name.resize(size);
+ status = clGetPlatformInfo(id, CL_PLATFORM_NAME, name.size(), name.data(), NULL);
+ if (!CL_STATUS_SUCCESS(status)) continue;
+ for (auto &p : supported_platforms) {
+ if (strstr(name.data(), p.c_str()) != NULL) {
+ return &id;
+ }
+ }
+ }
+ return NULL;
+}
+
+OCLFPGADevice::OCLFPGADevice() {
+ std::vector<std::string> supported_platforms = {"Intel(R) FPGA SDK for OpenCL(TM)", "Xilinx"};
+ init(supported_platforms);
+}
+
+void OCLFPGADevice::init(const std::vector<std::string> &supported_platforms) {
+ cl_int status;
+ cl_device_id *device;
+ cl_platform_id *platform;
+ cl_uint n;
+ size_t size;
+ std::vector<char> name;
+ std::vector<cl_platform_id> platforms;
+ std::vector<cl_device_id> devices;
+
+ status = clGetPlatformIDs(0, NULL, &n);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query number of OpenCL platforms";
+ platforms.resize(n);
+ CHECK(platforms.size() > 0) << "No OpenCL platform available";
+ status = clGetPlatformIDs(platforms.size(), platforms.data(), NULL);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query OpenCL platform IDs";
+
+ platform = find_platform(&platforms, supported_platforms);
+ CHECK(platform) << "Unable to find supported OpenCL platform";
+
+ status = clGetDeviceIDs(*platform, CL_DEVICE_TYPE_ALL, 0, NULL, &n);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query number of OpenCL devices";
+ devices.resize(n);
+ CHECK(devices.size() > 0) << "No OpenCL device found";
+ status = clGetDeviceIDs(*platform, CL_DEVICE_TYPE_ALL, devices.size(), devices.data(), NULL);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query OpenCL devices IDs";
+
+ device = NULL;
+ for (auto &id : devices) {
+ _context = clCreateContext(NULL, 1, &id, NULL, NULL, &status);
+ if (CL_STATUS_SUCCESS(status)) {
+ status = clGetDeviceInfo(id, CL_DEVICE_NAME, 0, NULL, &size);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query OpenCL device info";
+ name.resize(size);
+ status = clGetDeviceInfo(id, CL_DEVICE_NAME, name.size(), name.data(), NULL);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query OpenCL device name";
+ LOG(INFO) << "Using FPGA device: " << name.data();
+ device = &id;
+ break;
+ } else {
+ LOG(INFO) << "This FPGA Device is not available. Skipped.";
+ }
+ }
+ CHECK(device) << "No FPGA device available";
+ _device = *device;
+}
+
+int OCLFPGADevice::setup(size_t mem_size, std::string aocx_file) {
+ cl_int status;
+ unsigned int argi;
+ size_t size;
+ FILE *binary_file;
+ unsigned char *binary;
+
+ LOG(INFO) << "Using AOCX: " << aocx_file;
Review comment:
Just curious, as I've seem `Xilinx` is among part of the `supported_platforms`, does Xilinx toolchain support loading aocx binary files?
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r443088686
##########
File path: config/vta_cost.py
##########
@@ -0,0 +1,102 @@
+# cost function for intelfocl 32*32 gemm version
+def cal_cost(insn):
+ """
+ Cal the runtime cost statically
+
+ Parameters
+ ------------
+ insn: the insn (json)
+
+ Returns
+ ------------
+ the cost in s
+ """
+ factor = 1000000.0
+ def alu_imm_cost(iter_out, iter_in, uop_bgn, uop_end):
+ x = (uop_end - uop_bgn) * iter_out * iter_in
+ cycles = x + 46
Review comment:
How are these costs generally derived? Are they FPGA device specific? If we change the compilation parameters in the OpenCL compiler, could it affect these cost models?
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r456264532
##########
File path: config/intelfocl_sample.json
##########
@@ -0,0 +1,13 @@
+{
+ "TARGET" : "intelfocl",
+ "HW_VER" : "0.0.1",
Review comment:
Agreed. Let's bump it up to 0.0.2
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-683503850
> I've made other minor request, but overall the PR is shaping up well. We can stick to `intelfocl` as a target name. For documentation's stake, can you add a README file under `hardware/intelfocl` to mention what FPGA this has been tested on?
Thank you @tmoreau89! We will add an README listing all the boards we have tested so far.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] liangfu commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
liangfu commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r483942451
##########
File path: src/oclfpga/oclfpga_device.cc
##########
@@ -0,0 +1,251 @@
+/*
+ * Licensed to the Apache Software Foundation (ASF) under one
+ * or more contributor license agreements. See the NOTICE file
+ * distributed with this work for additional information
+ * regarding copyright ownership. The ASF licenses this file
+ * to you under the Apache License, Version 2.0 (the
+ * "License"); you may not use this file except in compliance
+ * with the License. You may obtain a copy of the License at
+ *
+ * http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing,
+ * software distributed under the License is distributed on an
+ * "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+ * KIND, either express or implied. See the License for the
+ * specific language governing permissions and limitations
+ * under the License.
+ */
+
+#include "oclfpga_device.h"
+#include <dmlc/logging.h>
+#include <vta/hw_spec.h>
+#include <cstring>
+#include <numeric>
+
+#define CL_STATUS_SUCCESS(x) ((x) == CL_SUCCESS)
+
+static const char *kernel_names[] = {"vta_core"};
+
+static cl_platform_id *find_platform(std::vector<cl_platform_id> *platforms,
+ const std::vector<std::string> &supported_platforms) {
+ cl_int status;
+ size_t size;
+ std::vector<char> name;
+ for (auto &id : *platforms) {
+ status = clGetPlatformInfo(id, CL_PLATFORM_NAME, 0, NULL, &size);
+ if (!CL_STATUS_SUCCESS(status)) continue;
+ name.resize(size);
+ status = clGetPlatformInfo(id, CL_PLATFORM_NAME, name.size(), name.data(), NULL);
+ if (!CL_STATUS_SUCCESS(status)) continue;
+ for (auto &p : supported_platforms) {
+ if (strstr(name.data(), p.c_str()) != NULL) {
+ return &id;
+ }
+ }
+ }
+ return NULL;
+}
+
+OCLFPGADevice::OCLFPGADevice() {
+ std::vector<std::string> supported_platforms = {"Intel(R) FPGA SDK for OpenCL(TM)", "Xilinx"};
+ init(supported_platforms);
+}
+
+void OCLFPGADevice::init(const std::vector<std::string> &supported_platforms) {
+ cl_int status;
+ cl_device_id *device;
+ cl_platform_id *platform;
+ cl_uint n;
+ size_t size;
+ std::vector<char> name;
+ std::vector<cl_platform_id> platforms;
+ std::vector<cl_device_id> devices;
+
+ status = clGetPlatformIDs(0, NULL, &n);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query number of OpenCL platforms";
+ platforms.resize(n);
+ CHECK(platforms.size() > 0) << "No OpenCL platform available";
+ status = clGetPlatformIDs(platforms.size(), platforms.data(), NULL);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query OpenCL platform IDs";
+
+ platform = find_platform(&platforms, supported_platforms);
+ CHECK(platform) << "Unable to find supported OpenCL platform";
+
+ status = clGetDeviceIDs(*platform, CL_DEVICE_TYPE_ALL, 0, NULL, &n);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query number of OpenCL devices";
+ devices.resize(n);
+ CHECK(devices.size() > 0) << "No OpenCL device found";
+ status = clGetDeviceIDs(*platform, CL_DEVICE_TYPE_ALL, devices.size(), devices.data(), NULL);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query OpenCL devices IDs";
+
+ device = NULL;
+ for (auto &id : devices) {
+ _context = clCreateContext(NULL, 1, &id, NULL, NULL, &status);
+ if (CL_STATUS_SUCCESS(status)) {
+ status = clGetDeviceInfo(id, CL_DEVICE_NAME, 0, NULL, &size);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query OpenCL device info";
+ name.resize(size);
+ status = clGetDeviceInfo(id, CL_DEVICE_NAME, name.size(), name.data(), NULL);
+ CHECK(CL_STATUS_SUCCESS(status)) << "Failed to query OpenCL device name";
+ LOG(INFO) << "Using FPGA device: " << name.data();
+ device = &id;
+ break;
+ } else {
+ LOG(INFO) << "This FPGA Device is not available. Skipped.";
+ }
+ }
+ CHECK(device) << "No FPGA device available";
+ _device = *device;
+}
+
+int OCLFPGADevice::setup(size_t mem_size, std::string aocx_file) {
+ cl_int status;
+ unsigned int argi;
+ size_t size;
+ FILE *binary_file;
+ unsigned char *binary;
+
+ LOG(INFO) << "Using AOCX: " << aocx_file;
Review comment:
Just curious, as I've seem `Xilinx` in among part of the `supported_platforms`, does Xilinx toolchain support loading aocx binary files?
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r445321853
##########
File path: config/vta_cost.py
##########
@@ -0,0 +1,102 @@
+# cost function for intelfocl 32*32 gemm version
+def cal_cost(insn):
+ """
+ Cal the runtime cost statically
+
+ Parameters
+ ------------
+ insn: the insn (json)
+
+ Returns
+ ------------
+ the cost in s
+ """
+ factor = 1000000.0
+ def alu_imm_cost(iter_out, iter_in, uop_bgn, uop_end):
+ x = (uop_end - uop_bgn) * iter_out * iter_in
+ cycles = x + 46
Review comment:
Yes. Those costs are FPGA and configuration specific, and we need to profile/measure the values from the actual hardware.
I believe the values are provided here as an example.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-680545019
I've made other minor request, but overall the PR is shaping up well. We can stick to `intelfocl` as a target name. For documentation's stake, can you add a README file under `hardware/intelfocl` to mention what FPGA this has been tested on?
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] liangfu commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
liangfu commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-664790981
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] liangfu commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
liangfu commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-706899033
ci tests failed on tsim seems to be unrelated,
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego edited a comment on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego edited a comment on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-665845575
> thanks @remotego @liangfu @pasqoc for the insightful comments. I think that all opinions expressed are very valid!
>
> So far our approach to naming VTA target has been to couple the VTA target with the FPGA board (e.g. pynq, ultra96, de10nano). These targets also contrast with functional simulation (sim), and cycle accurate sim (tsim).
>
> We've been using the VTA target to guide the compilation process to the target device, given that some of the drivers had to be written in a board specific fashion. In the case of the Intel OpenCL FPGA support, I understand that this work captures a variety of hardware backends, including Arria 10, Stratix10 boards etc, and that the driver codebase would remain mostly identical between those boards (unlike Pynq, and Ultra96 that relied on very different ARM SoCs)
>
> However to find a quick resolution to this discussion we can either choose
>
> * to use a specific board name (rather than FPGA family) to indicate that the OCL FPGA design has been tested on this device. This echo @liangfu's concern that we should have concrete targets for the community to reproduce work on.
> * keep the naming open as @remotego and @pasqoc are advocating, and classify this target as intelfocl_pcie to indicate that this applies only to PCIE-based OpenCL Intel FPGA devices.
>
> @remotego let us know what you think
Thank you for the reply. I am okay with either approach but I would like to suggest the decoupling between target and back-end. As we have three different back-end right now, we will definitely run into the situation that one particular board will be supported by more than one backend. I don't think we need to choose a backend for the user and it will be great if the user is willing to explore and experience the pros/cons of each backend.
Agreed with you and @liangfu, I would like to keep the current target definition ("specific board") so that the user could have a concrete platform to test the design. However, using board name to imply backend is confusing as I see it. For example, as I mentioned before, the Intel OpenCL code is not restricted to PCIe based boards, it should support SoC based boards as well, which includes Altera DE10 boards. In addition, as Intel OpenCL platform supports software emulation and cycle-accurate simulation as well, it could also have targets of "fsim" and "tsim".
I propose we could place an additional backend option inside vta_config.json, so that user could explicitly spell out the choice of backend for VTA. We are also actively extending our support for Xilinx SDAccel/SDSoc, so that we could support a lot more FPGA cards in the near future.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r479844168
##########
File path: src/sim/sim_driver.cc
##########
@@ -181,6 +181,51 @@ class SRAM {
}
memset(sram_ptr, 0, kElemBytes * xtotal * op->y_pad_1);
}
+
+ // This is for load 8bits to ACC only
+ void Load_int8(const VTAMemInsn* op,
+ DRAM* dram,
+ uint64_t* load_counter,
+ bool skip_exec) {
+ CHECK_EQ(kBits, VTA_ACC_WIDTH);
+
+ // TODO(zhanghao): extend to other width
+ CHECK_EQ(VTA_ACC_WIDTH, 32);
+ CHECK_EQ(VTA_INP_WIDTH, 8);
Review comment:
This code section is for the newly added load_int8 instructions in fsim. The instruction will load and cast 8-bit inputs from input memory and copy onto the 32-bit acc memory. As of now, this instruction only support 8-bit source and 32-bit acc mem as destination, although we may expend the instruction further.
As this instruction does not operate on wgt_mem, no requirement on wgt width is specified here.
On the other hand, for our current intelfocl/vitis implementation, we do have a restriction on the width of wgt mem, and we have already added an assertion there.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] pasqoc commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
pasqoc commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-661202272
Hi @liangfu, have you been able to test on the de10nano?
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] liangfu edited a comment on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
liangfu edited a comment on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-660618430
> The code should work for all devices supported by Intel OpenCL for FPGA, namely Intel Arria 10, Stratix V/10 and Cyclone V/10.
Just my humble opinion, given that both "Intel OpenCL for FPGA" and "VTA" requires a large amount of logic utilization, many Cyclone V chips that supports AOCL couldn't get this compiled because of the hardware utilization issue.
In my understanding of @tmoreau89 's comment, which I would agree, setting a proper target device in vta_config gives the property of being "versatile" in VTA - the Versatile Tensor Accelerator. Specifically, we could take the advantage of the properties in vta_config to define an accelerator that could scale from minimal ones towards data center scale accelerators.
As a side note, the reason we are taking these efforts in building open source projects, in part, we are hoping someone in the community could reproduce what we have done, and could easily start to build something that is even better. With the target being defined too board, an potential grad student could fail to reproduce the result, since not all the student could easily purchase a board with Stratix 10, and low cost Cyclone V boards couldn't get this running. In addition, they're wasting large amount of valuable logic resources, even they could afford a board with Stratix 10. Therefore, we should specify a precise target device for the vta_config.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r456699307
##########
File path: config/intelfocl_sample.json
##########
@@ -0,0 +1,13 @@
+{
+ "TARGET" : "intelfocl",
Review comment:
SO I believe that the solution was to go with "arria10" in this context?
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] pasqoc commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
pasqoc commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-661193715
> > Thanks for the changes. Please apply the 0.0.2 and rename the vta target to something more specific, e.g. "arria10". Also there are some CI errors related to linting that could be addressed. Thanks!
>
> Thank you very much! Sure. We will apply the 0.0.2 and address the linting errors.
> However. I believe "arria10" is too restrictive here. The code should work for all devices supported by Intel OpenCL for FPGA, namely Intel Arria 10, Stratix V/10 and Cyclone V/10. So far we have tested it on both Arria 10 and Stratix 10 boards, and it worked.
I agree with @remotego, arria10 seems too restrictive and would not do justice with what the code actually support, i.e. Intel OpenCL supported boards.
That said, if the VTA implicit rule to add new supported (and working) target devices is that of using "boards" by name, like "pynq", "de10nano", "ultra96", etc. then we should continue doing so.
But in this case I would suggest to add target entries for all devices that are known to work, that is arria10, stratix10, etc. and document the fact that these have the restriction to only work with OpenCL and not VHLS or Chisel like pynq and de10nano.
To avoid explosion, duplication of configurations, we should also probably be thinking of separating target parameters (target, hw_ver) in vta_config.json into a different file vta_target.json so that we can more flexibly specify different architecture configurations and target devices.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego edited a comment on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego edited a comment on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-660622675
> > The code should work for all devices supported by Intel OpenCL for FPGA, namely Intel Arria 10, Stratix V/10 and Cyclone V/10.
>
> Just my humble opinion, given that both "Intel OpenCL for FPGA" and "VTA" requires a large amount of logic utilization, many Cyclone V chips that supports AOCL couldn't get this compiled because of the hardware utilization issue.
> In my understanding of @tmoreau89 's comment, which I would agree, setting a proper target device in vta_config gives the property of being "versatile" in VTA - the Versatile Tensor Accelerator. Specifically, we could take the advantage of the properties in vta_config to define an accelerator that could scale from minimal ones towards data center scale accelerators.
>
> As a side note, the reason we are taking these efforts in building open source projects, in part, we are hoping someone in the community could reproduce what we have done, and could easily start to build something that is even better. With the target being defined too board, an potential grad student could fail to reproduce the result, since not all the student could easily purchase a board with Stratix 10, and low cost Cyclone V boards couldn't get this running. In addition, they're wasting large amount of valuable logic resources, even they could afford a board with Stratix 10. Therefore, we should specify a precise target device for the vta_config.
Thank you for your reply. However, could you explain more on the reason why the design shall not work on Cyclone V FPGAs that supports Intel OpenCL for FPGA?
Precisely as you mentioned, the VTA design is versatile, the user could always change the settings (i.e. LOG_BLOCK and LOG_*_BUF_SIZE) to adjust the resource usage in order to fit their own FPGA boards. Surely a low cost Cyclone V device could not support 64x64 GEMV cores like large Stratix 10 FPGAs do. But the user could always try to use 16x16 or even 4x4 GEMV cores, by setting the LOG_BLOCK lower.
Considering that, we used a relatively small default setting for LOG_BLOCK (4) and Buffers(15, 15, 18, 17). Thus the design should be able to fit into FPGAs comparable to the original Zynq/Zedboard platforms.
We must admit that we don't have those Cyclone V boards on hand, nor has the design been tested on those platforms. However, if there is any issue on compiling the design for a AOCL-compatibale cyclone V board, we will be more than happy to investigate and try to solve the issue together.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] zhanghaohit commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
zhanghaohit commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-659792207
@tmoreau89 @liangfu any other comments? Thanks.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] liangfu commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
liangfu commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r442601905
##########
File path: config/intelfocl_sample.json
##########
@@ -0,0 +1,13 @@
+{
+ "TARGET" : "intelfocl",
Review comment:
I think this should be the specific cloud fpga, instead of intelfocl or aocl.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] liangfu commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
liangfu commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-660618430
> The code should work for all devices supported by Intel OpenCL for FPGA, namely Intel Arria 10, Stratix V/10 and Cyclone V/10.
Just my humble opinion, given that both "Intel OpenCL for FPGA" and "VTA" requires a large amount of logic utilization, many Cyclone V chips that supports AOCL couldn't get this compiled because of the hardware utilization issue.
In my understanding of @tmoreau89 's comment, which I would agree, setting a proper target device in vta_config gives the property of being "versatile" in VTA - the Versatile Tensor Accelerator. Specifically, we could take the advantage of the properties in vta_config to define an accelerator that could scale from minimal ones towards data center scale accelerators.
As a side note, the reason we are taking these efforts in building open source projects, in part, we are hoping someone in the community could reproduce what we have done, and could easily start to build something that is even better. With the target being defined too board, an potential grad student could fail to reproduce the result, since not all the student could easily purchase a board with Stratix 10, and low cost Cyclone V boards couldn't get this running. In addition, they're wasting large amount of valuable logic resources, even they could afford a board with Stratix 10. Therefore, we should specify a precise target device for the vta_config.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego edited a comment on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego edited a comment on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-660411735
> Thanks for the changes. Please apply the 0.0.2 and rename the vta target to something more specific, e.g. "arria10". Also there are some CI errors related to linting that could be addressed. Thanks!
Thank you very much! Sure. We will apply the 0.0.2 and address the linting errors.
However. I believe "arria10" is too restrictive here. The code should work for all devices supported by Intel OpenCL for FPGA, namely Intel Arria 10, Stratix V/10 and Cyclone V/10. So far we have tested it on both Arria 10 and Stratix 10 boards, and it worked.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r443089629
##########
File path: include/vta/hw_spec.h
##########
@@ -404,7 +259,7 @@ typedef struct {
/*! \brief Destination index (indexes accum buffer) */
uint32_t dst_idx : VTA_LOG_ACC_BUFF_DEPTH;
/*! \brief Source index (indexes input buffer for GEMM or accum buffer for ALU) */
- uint32_t src_idx : VTA_LOG_INP_BUFF_DEPTH;
+ uint32_t src_idx : VTA_LOG_ACC_BUFF_DEPTH;
Review comment:
We may want to set it to max(VTA_LOG_ACC_BUFF_DEPTH, VTA_LOG_INP_BUFF_DEPTH)
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego edited a comment on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego edited a comment on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-695611214
Yes. We will address those issues.
Sorry for the delay. We are working on it and will update Monday or Tuesday.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r460334614
##########
File path: config/intelfocl_sample.json
##########
@@ -0,0 +1,13 @@
+{
+ "TARGET" : "intelfocl",
Review comment:
thank you for clarifying, see main conversation thread for resolution
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] liangfu commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
liangfu commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-687598082
Hi @zhanghaohit thanks for the update, I agree with @tmoreau89 that this PR is in good shape, and I left a few comments on the codestyle / document part, otherwise this is good to go.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] liangfu commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
liangfu commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r443953640
##########
File path: src/intelfocl/AOCLUtils/opencl.h
##########
@@ -0,0 +1,122 @@
+// Copyright (C) 2013-2018 Altera Corporation, San Jose, California, USA. All rights reserved.
Review comment:
I think a better place for source code files in this directory should be the `3rdparty` directory.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-695346293
Hi @zhanghaohit, please let me know if you can address the following changes so we can get this PR merged!
(1) Adding a README file (ideally with instructions on how to compile / run on the FPGA) before we migrate this to official TVM documentation.
(2) Using constants in the vta cost model
(3) Adding comments to MEM_ADDR_IDENTIFIER initialization
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-699643211
Hi @tmoreau89,
We have completed the changes listed, and we have also included an README.rst file as a preliminary installation guide.
Please let us know if more changes are required.
Thank you very much!
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] tmoreau89 commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
tmoreau89 commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r477014763
##########
File path: src/oclfpga/oclfpga_driver.cc
##########
@@ -0,0 +1,94 @@
+/*
+ * Licensed to the Apache Software Foundation (ASF) under one
+ * or more contributor license agreements. See the NOTICE file
+ * distributed with this work for additional information
+ * regarding copyright ownership. The ASF licenses this file
+ * to you under the Apache License, Version 2.0 (the
+ * "License"); you may not use this file except in compliance
+ * with the License. You may obtain a copy of the License at
+ *
+ * http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing,
+ * software distributed under the License is distributed on an
+ * "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY
+ * KIND, either express or implied. See the License for the
+ * specific language governing permissions and limitations
+ * under the License.
+ */
+
+#include "oclfpga_device.h"
+#include <vta/driver.h>
+#include <tvm/runtime/registry.h>
+#include <string>
+#include <iostream>
+
+#define MEM_ADDR_IDENTIFIER (0x18000000)
Review comment:
One more question -- where is this derived from? Is this address common to all Intel FPGA PCI-e cards, or is it specific to the given board that is being used?
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r456259647
##########
File path: config/intelfocl_sample.json
##########
@@ -0,0 +1,13 @@
+{
+ "TARGET" : "intelfocl",
+ "HW_VER" : "0.0.1",
+ "LOG_INP_WIDTH" : 3,
Review comment:
Yes. LOG_INP_WIDTH and LOG_WGT_WIDTH are fixed at 3 in the current version.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on a change in pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on a change in pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#discussion_r456895729
##########
File path: config/intelfocl_sample.json
##########
@@ -0,0 +1,13 @@
+{
+ "TARGET" : "intelfocl",
+ "HW_VER" : "0.0.1",
+ "LOG_INP_WIDTH" : 3,
Review comment:
Agreed! I think for this version we will try to assert an Error if user set unsupported LOG_INP_WIDTH and LOG_WGT_WIDTH settings. And in future updates, we will enhance the code to support 16-bit and 32-bit inputs and width.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-660411735
> Thanks for the changes. Please apply the 0.0.2 and rename the vta target to something more specific, e.g. "arria10". Also there are some CI errors related to linting that could be addressed. Thanks!
Sure. We will apply the 0.0.2 and address the linting errors.
However. I believe "arria10" is too restrictive here. The code should work for all devices supported by Intel OpenCL for FPGA, namely Intel Arria 10, Stratix V/10 and Cyclone V/10. So far we have tested it on both Arria 10 and Stratix 10 boards, and it worked.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org
[GitHub] [incubator-tvm-vta] remotego commented on pull request #9: [Hardware][OpenCL] Intelfocl support
Posted by GitBox <gi...@apache.org>.
remotego commented on pull request #9:
URL: https://github.com/apache/incubator-tvm-vta/pull/9#issuecomment-665845575
> thanks @remotego @liangfu @pasqoc for the insightful comments. I think that all opinions expressed are very valid!
>
> So far our approach to naming VTA target has been to couple the VTA target with the FPGA board (e.g. pynq, ultra96, de10nano). These targets also contrast with functional simulation (sim), and cycle accurate sim (tsim).
>
> We've been using the VTA target to guide the compilation process to the target device, given that some of the drivers had to be written in a board specific fashion. In the case of the Intel OpenCL FPGA support, I understand that this work captures a variety of hardware backends, including Arria 10, Stratix10 boards etc, and that the driver codebase would remain mostly identical between those boards (unlike Pynq, and Ultra96 that relied on very different ARM SoCs)
>
> However to find a quick resolution to this discussion we can either choose
>
> * to use a specific board name (rather than FPGA family) to indicate that the OCL FPGA design has been tested on this device. This echo @liangfu's concern that we should have concrete targets for the community to reproduce work on.
> * keep the naming open as @remotego and @pasqoc are advocating, and classify this target as intelfocl_pcie to indicate that this applies only to PCIE-based OpenCL Intel FPGA devices.
>
> @remotego let us know what you think
Thank you for your reply. I am okay with either approach but I would like to suggest the decoupling between target and back-end. As we have three different back-end right now, we will definitely run into the situation that one particular board will be supported by more than one backend. I don't think we need to choose a backend for the user and it will be great if the user is willing to explore and experience the pros/cons of each backend.
Agreed with you and @liangfu, I would like to keep the current target definition ("specific board") so that the user could have a concrete platform to test the design. However, using board name to imply backend is confusing as I see it. For example, as I mentioned before, the Intel OpenCL code is not restricted to PCIe based boards, it should support SoC based boards as well, which includes Altera DE10 boards. In addition, as Intel OpenCL platform supports software emulation and cycle-accurate simulation as well, it could also have targets of "fsim" and "tsim".
I propose we could place an additional backend option inside vta_config.json, so that user could explicitly spell out the choice of backend for VTA.
----------------------------------------------------------------
This is an automated message from the Apache Git Service.
To respond to the message, please log on to GitHub and use the
URL above to go to the specific comment.
For queries about this service, please contact Infrastructure at:
users@infra.apache.org