大侠好，欢迎来到FPGA技术江湖，江湖偌大，相见即是缘分。大侠可以关注FPGA技术江湖，在“闯荡江湖”、"行侠仗义"栏里获取其他感兴趣的资源，或者一起煮酒言欢。“煮酒言欢”进入IC技术圈，这里有近50个IC技术公众号。
大侠好，欢迎来到“艮林子”专栏，本次为艮林子首次和大侠见面，新春佳节之际，略备薄礼，不成敬意，给大侠带来“Xilinx Vitis 系列连载”，给大侠提供参考学习的资料，如有不足之处，还请多多指教。
由于内容过多，首篇给大侠列出目录，后续有时间分篇连载发出，欢迎各位大侠关注“FPGA技术江湖”，我们是真正做事的团队，希望大家多多支持，如果想获取更多资源，可以加大辉哥微信，进交流群，可以永久交流学习，共同进步。话不多说，咱们就进入正题。
今天带来第二篇，介绍开发应用程序和构建和运行程序。后续还会带来详细的开发教程，敬请关注。话不多说，咱们就进入正题。
第一篇放个超链接。
Vitis指南 | Xilinx Vitis 系列（一）
VITIS简介
作者：艮林子 校对：陆辉
连载目录
一、Vitis 2019.2软件平台发行说明
1.1 Vitis软件平台中的内容
1.1.1 硬件加速的应用程序开发流程
1.2 改变行为
1.2.1 从SDAccel迁移
1.2.2 从SDSoC迁移
1.3 支持平台
1.3.1 数据中心加速卡
1.3.2 嵌入式平台
二、安装
三、Vitis加速环境简介
3.1 简介与概述
3.2 FPGA加速
3.3 执行模型
3.4 建立过程
3.4.1 主机程序构建过程
3.4.2 FPGA二进制构建过程
3.5 建立目标
3.5.1 软件仿真
3.5.2 硬件仿真
3.5.3 系统
3.6 使用Vitis软件平台加速应用程序的方法论
3.6.1 文件受众和范围
3.6.2 加速：工业类比
3.6.3 设计设备加速应用程序的方法
3.6.4 开发C / C 内核的方法论
四、开发应用程序
4.1 程序设计模型
4.2 主机申请
4.2.1 设置OpenCL环境
4.2.2 在FPGA中执行命令
4.2.3 后处理和FPGA清理
4.2.4 摘要
4.3 RTL内核
4.3.1 RTL内核的要求
4.3.2 RTL内核开发流程
4.3.2.1 将RTL代码打包为Vivado IP
4.3.2.2 创建内核描述XML文件
4.3.2.3 从RTL内核创建XO文件
4.3.3 RTL内核向导
4.3.4 摘要
五、构建和运行程序
5.1 设置Vitis环境
5.2 建立目标
5.2.1 软件仿真
5.2.2 硬件仿真
5.2.3 系统硬件目标
5.3 构建主机程序
5.4 构建FPGA二进制文件
5.4.1 在Vivado HLS中创建内核
六、对应用程序进行性能分析，优化和调试
七、使用Vitis分析仪
7.1 处理报告
7.2 了解葡萄分析仪
7.3 创建运行配置
7.4 配置葡萄分析仪
八、使用Vitis IDE
8.1 Vitis命令选项
8.2 创建Vitis IDE项目
8.2.1 启动Vitis IDE工作区
8.2.2 创建一个应用程序项目
8.2.3 管理平台和存储库
8.2.4 了解Vitis IDE
8.2.5 添加源
8.2.5.1 添加源文件
8.2.5.2 创建和编辑新的源文件
8.2.5.3 在项目编辑器视图中工作
8.2.5.4 在助理视图中工作
8.3 建立系统
8.3.1 Vitis IDE指导视图
8.3.2 从Vitis IDE使用Vivado工具
8.4 Vitis IDE调试流程
8.5 配置Vitis IDE
8.5.1 Vitis项目设置
8.5.2 Vitis Build配置设置
8.5.3 Vitis运行配置设置
8.5.4 Vitis二进制容器设置
8.5.5 Vitis硬件功能设置
8.5.6 Vitis工具链设置
8.5.6.1 Vitis内核编译器和链接器选项
8.5.6.2 emconfigutil设置
8.5.6.3 G 主机编译器和链接器设置
8.6 项目进出口
8.6.1 导出Vitis项目
8.6.2 导入葡萄项目
正文
四、开发应用程序
4.1 程序设计模型
Vitis心开发套件支持使用行业标准异构计算的OpenCL™框架（https://www.khronos.org/opencl/）。主机程序在处理器（x86或Arm®）上执行，并通过Xilinx运行时（XRT）卸载计算密集型任务，从而使用OpenCL编程范例在运行于可编程逻辑（PL）的硬件内核上执行。
在Vitis核心开发套件中，目标设备可以包括Xilinx®MPSoC或UltraScale ™ FPGA，这些处理器连接到处理器，例如通过PCIe总线连接到x86主机，或通过AXI4接口连接到Arm处理器。FPGA包含一个实现和执行硬件内核的可编程区域。
FPGA平台包含一个或多个全局存储库。从主机到内核以及从内核到主机的数据传输是通过这些全局内存库进行的。FPGA中运行的内核可以具有一个或多个存储器接口。从全局内存库到这些内存接口的连接是可配置的，因为它们的功能由内核编译选项确定。
可以在Xilinx器件的PL中实现多个内核，从而大大加快了应用程序的速度。单个内核也可以被实例化多次。内核的实例数量是可编程的，并且由构建FPGA二进制文件时指定的链接选项确定。有关指定这些选项的更多信息，请参阅链接内核。
4.2 主机申请
在Vitis核心开发套件中，主机代码使用行业标准OpenCL API 用C或C 语言编写。所述葡萄 核心开发套件提供的OpenCL 1.2嵌入的配置文件符合的运行时API。
通常，主机代码的结构可以分为三个部分：
- 1.设置环境。
- 2.核心命令执行，包括执行一个或多个内核。
- 3.后处理和资源释放。
4.2.1 设置OpenCL环境
Vitis核心开发套件中的主机代码遵循OpenCL编程范例。为了正确设置环境，主机应用程序需要初始化标准的OpenCL结构：目标平台，设备，上下文，命令队列和程序。
- 1.平台：初始化后，主机应用程序应需要识别一个由一个或多个Xilinx设备组成的平台。对应的OpenCL API调用clGetPlatformIDs来发现一组可用的OpenCL平台上给定系统。之后，通过与字符串匹配，clGetPlatformInfo用于识别基于Xilinx设备的平台 。cl_platform_vendor"Xilinx".
- 2.设备：找到Xilinx平台后 ，应用程序需要识别相应的Xilinx设备。
- 3.子设备：在Vitis核心开发工具包中，有时设备包含单个内核或不同内核的多个内核实例。尽管OpenCL API clCreateSubDevices允许主机代码将设备划分为多个子设备，但Vitis核心开发套件支持均分的子设备（使用CL_DEVICE_PARTITION_EQUALLY），每个子设备 包含一个内核实例。
- 4.语境：该clCreateContextAPI用于创建包含Xilinx设备的上下文，该设备将与主机进行通信。
- 5.命令队列 ：该clCreateCommandQueueAPI会为每个设备的一个或多个命令队列。FPGA可以包含多个内核，可以是相同或不同的内核。开发主机应用程序时，有两种主要的编程方法可以在设备上执行内核：
- 1.单个乱序命令队列：可以通过同一命令队列请求多个内核执行。XRT会以任何顺序尽快分配内核，从而允许在FPGA上并发执行内核。
- 2.多个有序命令队列：每个内核执行都将从不同的有序命令队列中请求。在这种情况下，XRT从不同的命令队列分派内核，通过在设备上同时运行它们来提高性能。
4.2.2 在FPGA中执行命令
一旦OpenCL的环境初始化，主机应用程序是准备发出命令到设备和交互与内核。这些命令包括：
- 1.设置内核。
- 2.到/从FPGA的缓冲区传输。
- 3.在FPGA上执行内核。
- 4.事件同步
4.2.3 后处理和FPGA清理
在主机代码的末尾，应使用适当的释放功能来释放所有分配的资源。如果没有正确释放资源，则Vitis 核心开发工具包可能无法生成与性能相关的正确配置文件和分析报告。
4.2.4 摘要
如先前主题中所述，Vitis核心开发工具包中建议的主机程序编码风格包括以下几点：
- 1.如果需要，在每个OpenCL API调用之后添加错误检查以进行调试。
- 2.在Vitis核心开发套件中，一个或多个内核被分别编译/链接以构建.xclbin文件。该API clCreateProgramWithBinary用于cl_program从内核二进制文件构建 对象。
- 3.clSetKernelArg在对缓冲区执行任何入队操作之前，请使用buffer设置内核参数（）。
- 4.使用clEnqueueMigrateMemObjects或将数据从主机代码来回传输到内核clEnqueueMapBuffer。
- 5.使用posix_memalign在4K边界对齐主机内存的指针（适用于的PCIe为基础的平台）。
- 6.最好将乱序命令队列用于FPGA上的并发命令执行。
- 7.执行整个工作负载clEnqueueTask，而不是使用来拆分工作负载clEnqueueNDRangeKernel。
- 8.使用事件同步命令clFinish 和和clWaitForEvents来解决异步OpenCL API调用的依赖关系。
- 9.完成后释放所有OpenCL分配的资源。
4.3 RTL内核
如FPGA二进制构建过程中所述，Vitis核心开发套件中的每个硬件内核都独立编译为Xilinx对象（.xo）文件。这些文件可以组合到应用程序项目中，以链接到FPGA可执行文件（xclbin）。这包括从Vivado Design Suite打包现有RTL IP的功能，以用于Vitis 应用程序加速开发流程。
许多硬件工程师拥有现有的RTL IP（包括基于Vivado®IP集成商的设计），或者更喜欢在RTL中实现内核并使用Vivado工具进行开发。尽管Vitis核心开发套件支持使用打包的RTL设计，但它们必须遵守在加速的应用程序开发流程和运行时库中使用的软件和硬件要求。
4.3.1 RTL内核的要求
RTL设计必须同时满足接口和软件要求，才能在Vitis IDE中用作RTL内核。为了满足这些要求，可能有必要添加或修改原始RTL设计。
- 内核接口要求 为了满足Vitis核心开发套件执行模型，RTL内核必须遵守“ 内核属性”中描述的要求。RTL内核必须至少具有一个时钟接口端口，才能向内核逻辑提供时钟。
- 内核软件要求 RTL内核具有与C / C 和OpenCL内核相同的软件接口模型。在宿主程序中，它们被视为具有无效返回值，指针参数和标量参数的函数。
Vitis的核心开发工具包执行模型使然如下：
- 标量参数通过AXI4-Lite从属接口直接写入内核。
- 指针参数从主机程序传递到内存，或从内存传递，并且RTL内核通过一个或多个AXI4内存映射接口读取/写入内存中的数据。
- 主机程序通过AXI4-Lite从接口通过控制寄存器（如下所示）控制内核。
如果RTL设计具有不同的执行模型，则必须对其进行调整以确保其将以这种方式运行。
- 打断 RTL内核可以选择具有一个包含单个中断的中断端口。端口名称必须被调用interrupt并且为高电平有效。当全局中断允许（GIE）位和中断允许寄存器（IER）位均被置位时，该位被使能。 默认情况下，IER使用内部ap_done信号触发中断。此外，仅当将1写入IP中断状态寄存器的bit-0时，该中断才被清除。 如果将interrupt端口添加到RTL内核，则kernel.xml文件需要包含此信息。所述kernel.xml使用RTL内核向导时自动生成。默认情况下，RTL内核向导会interrupt在Control Register模块中创建一个名为的单个中断端口以及中断逻辑。这反映在为RTL内核以及关联的component.xml和kernel.xml文件生成的Verilog代码中。
4.3.2 RTL内核开发流程
本节说明为Vitis核心开发套件创建RTL内核的三步过程，其中包括：
- 1.将RTL块打包为标准的Vivado IP。
- 2.为RTL内核创建XML描述文件。
- 3.将RTL内核打包到Xilinx对象（.xo）文件中。
打包的RTL内核作为Xilinx目标文件提供，文件扩展名为.xo。该文件是封装Vivado IP对象（包括源文件）和关联的内核XML文件的容器。该.xo文件可以与其他内核相结合，并与目标平台连接，并内置硬件或硬件仿真流。
提示：除非您为内核提供C模型，否则RTL内核不适合软件仿真。
4.3.2.1 将RTL代码打包为Vivado IP
必须将RTL内核打包为适合IP集成商使用的Vivado IP。有关Vivado工具中IP打包的详细信息，请参见《Vivado Design Suite用户指南：创建和打包自定义IP （UG1118）》。
RTL内核所需的接口必须打包如下：
- 该AXI4-精简版接口名称必须打包为S_AXI_CONTROL，但潜在的AXI端口可以以不同的方式命名。
- 的AXI4接口必须打包为AXI4主端点具有64位地址的支持。
- ap_clk并且ap_clk_2必须打包为时钟接口（ap_clk_2仅当RTL内核有两个时钟时才需要）。
- ap_rst_n并且ap_rst_n_2必须打包为低电平有效复位接口（当RTL内核复位时）。
- ap_clk必须打包为与所有AXI4-Lite，AXI4和AXI4-Stream接口关联。
若要测试RTL内核对于IP集成器是否正确打包，请尝试在IP集成器中将打包的内核IP实例化为模块设计。有关该工具的信息。请参考Vivado Design Suite用户指南：使用IP Integrator（UG994）设计IP子系统。
内核IP应该显示上述各种接口。在画布视图中检查IP。可以通过在画布上选择接口来查看AXI接口的属性。然后在“块接口属性”窗口中，选择“ 属性”选项卡并展开“ 配置”表条目。如果接口是只读或只写的，则可以删除未使用的AXI通道，并将其READ_WRITE_MODE设置为只读或只写。
4.3.2.2 创建内核描述XML文件
必须为每个RTL内核创建一个名为kernel.xml的XML内核描述文件，以便可以在Vitis应用程序加速开发流程中使用它。所述 kernel.xml文件指定内核由运行时需要的和类似的寄存器映射属性和端口葡萄流动。以下是kernel.xml文件的示例。
4.3.2.3 从RTL内核创建XO文件
最后一步是将RTL IP及其 kernel.xml文件打包到Xilinx目标文件（.xo）中，以便可以在Vitis核心开发套件中使用该内核。这是通过Vivado Design Suite中的package_xoTcl命令完成的。
4.3.3 RTL内核向导
RTL内核向导会自动执行您需要执行的某些步骤，以确保将RTL IP打包到Vitis编译器可以使用的内核对象（.xo）中。RTL内核向导：
- 指导您完成为RTL内核指定接口要求的过程，并根据提供的信息生成顶级RTL包装器。
- 自动生成AXI4-Lite接口模块，包括顶层包装器中包含的控制逻辑和寄存器文件。
- 在确保RTL IP与包装器之间正确连接之后，可以在顶层包装器中包含示例内核IP模块，您可以将其替换为自己的RTL IP设计。
- 自动生成一个kernel.xml文件以匹配向导中的内核规范。
- 为生成的RTL内核包装器生成一个简单的模拟测试台。
- 生成示例主机程序以运行和调试RTL内核。
可以从Vitis IDE或Vivado IP目录访问RTL内核向导。无论哪种情况，它都会创建一个包含示例设计的Vivado项目，以用作定义自己的RTL内核的模板。
具体示例见《ug1393》117页。
4.3.4 摘要
本节总结了RTL内核IP的VLNV，软件功能原型和从上一页中选择的选项创建的硬件控制寄存器。函数原型传达了如果是C函数的内核调用的样子。请参阅主机代码生成的示例，该示例如何设置内核调用的内核参数。寄存器映射显示主机软件ID，自变量名称，硬件寄存器偏移量，类型和关联的接口之间的关系。在继续生成内核之前，请查看本节的正确性。
- 在Vivado IDE中使用RTL内核项目
如果从Vitis IDE 启动了RTL内核向导，则在“摘要”页面上单击“ 确定 ” 后，将打开带有示例IP项目的Vivado Design Suite，以使您可以完成RTL内核代码。
如果从Vivado IP目录中启动了RTL内核向导，则在“摘要”页面上单击“ 确定 ” 后，RTL内核向导IP将实例化到当前项目中。从那里必须执行以下步骤：
- 1.当出现“生成输出产品”对话框时，单击“ 跳过”以将其关闭。
- 2.右键单击 添加到“源”视图的<kernel_name> .xci文件，然后选择“ 打开IP示例设计”。
- 3.在“打开示例设计”对话框中，指定“ 示例项目”目录，或接受默认值，然后单击“ 确定”
五、构建和运行程序
5.1 设置Vitis环境
这部分略，如有想要了解可参见开发文档，点击阅读原文可以获取原英文开发文档。
5.2 建立目标
Vitis工具的构建目标定义了在编译和链接期间创建的FPGA二进制文件（.xclbin）的性质和内容。共有三种不同的构建目标：用于验证和调试目的的两个仿真目标：软件仿真和硬件仿真，以及用于生成加载到Xilinx器件中的FPGA二进制文件（.xclbin）的默认系统硬件目标。
针对仿真目标的编译比针对实际硬件的编译要快得多。仿真运行是在仿真环境中执行的，该仿真环境提供了增强的调试可见性，并且不需要实际的加速卡。
5.2.1 软件仿真
软件仿真的主要目标是确保主机程序和内核的功能正确性。对于软件仿真，主机代码和内核代码都被编译为在主机x86处理器上运行。该v 编译器内核代码来创建FPGA二进制文件，以便运行主机程序和内核代码在一起的最小变换。软件仿真流程可用于优化算法，调试功能问题，并让开发人员快速遍历代码以进行改进。保留了通过快速编译和运行迭代进行开发的编程模型。
在Vitis统一软件平台的上下文中，CPU上的软件仿真与典型的CPU / GPU编程迭代开发过程相同。在这种类型的开发风格中，程序员在开发过程中不断地编译和运行应用程序
5.2.2 硬件仿真
硬件仿真流程使程序员能够检查从C，C 或OpenCL内核代码合成的FPGA二进制文件的RTL描述的功能正确性。
每个内核都被编译为硬件模型（RTL）。在硬件仿真过程中，内核在Vivado逻辑仿真器中运行，并带有波形查看器以检查内核设计。另外，硬件仿真为硬件实现提供了性能和资源估计。
在硬件仿真中，编译和执行时间比软件仿真要长，但是它提供了详细的，周期精确的内核活动视图。Xilinx建议您在硬件仿真过程中使用小的数据集进行验证，以使运行时间易于管理。
5.2.3 系统硬件目标
当构建目标是系统硬件时，通过在设计上运行综合和实现v 为Xilinx器件构建FPGA二进制文件。因此，与在Vitis IDE中生成软件或硬件仿真目标相比，此构建目标通常需要更长的时间。但是，最终的FPGA二进制文件可以加载到加速卡或嵌入式处理器平台的硬件中，并且该应用程序可以在其实际操作环境中运行。
5.3 构建主机程序
这部分略，如有想要了解可参见开发文档。
5.4 构建FPGA二进制文件
内核代码是用C，C ，OpenCL C或RTL编写的，通过将内核代码编译成Xilinx目标文件（.xo），然后将.xo文件链接 到FPGA二进制文件（.xclbin）中来构建，如下图所示。
如上所述，该过程分为两个步骤：
- 1.从内核源代码构建Xilinx目标文件。 对于C，C 或OpenCL内核，该v -c命令将源代码编译为Xilinx对象（.xo）文件。多个内核被编译成单独的.xo 文件。 对于RTL内核，该package_xo命令将生成.xo文件以用于链接。 有关更多信息，请参考RTL内核。 您还可以创建直接在Vivado HLS工具中工作的内核目标文件（.xo）。有关更多信息，请参考在Vivado HLS中直接编译内核。
- 2.编译后，该v -l 命令将一个或多个内核对象（.xo）与硬件平台（.xsa）链接在一起，以生成FPGA二进制文件（.xclbin）。 提示：该v 命令可以从命令行，脚本或类似的构建系统中使用make，也可以通过使用Vitis IDE中讨论的GUI通过GUI 使用。
5.4.1 在Vivado HLS中创建内核
这部分略，如有想要了解可参见开发文档《ug1393》160页，点击阅读原文可以获取原英文开发文档。
六、对应用程序进行性能分析，优化和调试
这部分略，如有想要了解可参见开发文档，点击阅读原文可以获取原英文开发文档。
本篇到此结束，下篇还会继续连载，欢迎大侠关注！
END
后续会持续更新，带来Vivado、 ISE、Quartus II 、candence等安装相关设计教程，学习资源、项目资源、好文推荐等，希望大侠持续关注。
大侠们，江湖偌大，继续闯荡，愿一切安好，有缘再见！
