CN107111663B - 目标为可编程集成电路的异构型多处理器程序编译 - Google Patents

Abstract

OpenCL程序编译可以包括使用处理器生成对异构多处理器设计的第一内核的寄存器传输级(RTL)描述(705)并将所述第一内核的RTL描述与基础平台电路设计集成(710)。所述基础平台电路设计为异构多处理器设计的主机提供可编程集成电路内的静态接口。可以使用所述处理器从所述的第一内核的RTL描述生成第一配置比特流(715)。所述第一配置比特流指定所述第一内核的硬件实现和所述配置比特流的支持数据。所述第一配置比特流和所述支持数据可以被包括在二进制容器内(720)。

Description

目标为可编程集成电路的异构型多处理器程序编译

技术领域

本公开涉及集成电路(IC)，并且更具体涉及将可编程IC并入到异构型多处理器设计中。

背景技术

异构型多处理器框架提供跨平台的标准，并支持现代处理器、服务器、手持/嵌入式设备等的并行编程。开放计算语言(Open Computing Language，简称OpenCL)是用于编写可跨异构型计算平台执行的程序的异构型多处理器框架的示例。异构型计算平台可以包括中央处理单元(CPU)、图形处理单元(GPU)、或数字信号处理器(DSP)等。

异构型多处理器程序(例如，OpenCL程序)包括在主机系统上执行的部分和在设备上执行的一个或多个其他部分。通常，主机系统包括CPU，而设备可以为GPU、DSP等。在设备上执行的可以被称为内核的部分可以是用OpenCL，OpenCL C或另一适用于异构型多处理器框架或OpenCL语言的高级编程语言编写的。在主机上执行的部分可以用例如C或C++编程，并且控制跨各种设备的异构型多处理器环境。

虽然上述环境本质上是异构型的，但是每个特定设备(无论是DSP还是GPU)都具有静态架构。相比之下，诸如现场可编程门阵列(FPGA)的可编程IC具有非常灵活的硬件架构，其可用于硬件加速。然而，为了使用可编程IC作为设备，在可编程IC内实现的电路必须能够与主机交互并在异构型多处理器环境的场景中操作。

发明内容

一种方法包括：使用处理器来生成异构型多处理器设计的第一内核的寄存器传输级(RTL)描述，将所述的第一内核的RTL描述与在可编程集成电路(IC)内提供静态区域的基础平台电路设计集成，所述可编程集成电路向所述异构型多处理器设计的主机提供接口，并且使用处理器并从所述的第一内核的RTL描述生成指定第一内核的硬件实现的第一配置比特流和所述配置比特流的支持数据。所述方法还包括在二进制容器内包含第一配置比特流和支持数据。

一种方法包括：使用处理器来生成异构型多处理器设计的第一内核的RTL描述，将所述的第一内核的RTL描述与基础平台电路设计集成，所述基础平台电路设计在可编程IC内提供静态接口给异构型多处理器设计的主机，并且使用处理器并从所述的第一内核的RTL描述生成所述第一内核的RTL描述的支持数据。该方法还包括在二进制容器内包含所述第一内核的RTL描述和支持数据。

一种系统可以包括被编程以启动可执行操作的处理器。所述可执行操作包括：生成异构型多处理器设计的第一内核的RTL描述，将所述的第一内核的RTL描述与在可编程IC内提供静态区域的基础平台电路设计集成，所述可编程IC向所述异构型多处理器设计的主机提供接口，从所述第一内核的RTL描述生成指定第一内核的硬件实现的第一配置比特流以及用于所述配置比特流的支持数据。该方法还可以包括将所述第一配置比特流和支持数据包含在二进制容器内。

本发明内容部分仅用于介绍某些概念，而不是确认所要求保护的主题的任何关键或基本特征。本发明的布置的其他特征将从附图和下面的详细描述中被认识。

附图说明

本发明的装置在附图中以举例的方式示出。然而，附图不应被解释为将本发明的装置限制在仅仅是所示的具体实施方式中。通过阅读以下详细描述并参照附图，各方面和优点将变得显而易见。

图1是示出用于集成电路(IC)的示例性架构的框图；

图2是示出示例性数据处理系统(系统)的框图；

图3是示出图2的目标平台的示例性架构的框图；

图4是示出包括图2和图3的目标平台的异构型多处理器运行时系统的示例性层的框图；

图5是示出在图3的IC内实现的示例性电路的框图；

图6是示出内核区域的实施例的框图；

图7是示出实现异构型多处理器设计的内核的示例性方法的流程图；

图8是示出用于在可编程IC内实现异构型多处理器设计的内核的示例性过程的块流程图；

图9是示出处理在可编程IC内实现的用异构型多处理器计算语言指定的内核的示例性方法的流程图；

图10是示出以异构型多处理器计算语言指定的内核的示例性处理的框图；

图11是示出以异构型多处理器计算语言指定的内核的示例性处理的框图；

图12是示出以异构型多处理器计算语言指定的内核的示例性处理的框图；

图13-1和13-2总体上示出了以异构型多处理器计算语言指定的内核的处理；

图14是示出在异构型多处理器计算语言中指定的内核的示例性处理的框图；

图15是示例性目录结构；

图16是示出内核执行的示例性方法的流程图。

具体实施方式

尽管本公开用定义新颖的特征的权利要求进行了归纳，但是从结合附图的描述来考虑将更好地理解所述各种特征。本文所述的方法、机器、制造和其任何变型仅仅用于示例。本说明书中公开的具体结构和功能细节不应被解释为限制，而仅仅是作为权利要求的基础，以及作为教导本领域技术人员使用实际上任何适当详细的结构的典型基础。此外，本文中使用的术语和短语不旨在限制，而是提供对所描述的特征的可理解的描述。

本公开涉及集成电路(IC)，更具体地，涉及将可编程IC并入异构型多处理器系统。根据本公开内容中描述的本发明的布置，可以在异构型多处理器设计中使用可编程IC来实现一个或多个内核。异构型多处理器的一个或多个内核可以被编译并转换成使用可编程IC的可编程电路来实现的硬件。在这方面，使用可编程IC实现的内核是硬件加速的，这是因为该内核使用电路来实现，而不是被实现为用于除中央处理单元(CPU)之外的处理器执行的被卸载的可执行程序代码。被综合为硬件的异构型多处理器设计的内核部分，与异构型多处理器设计的主机协同工作。

在一个方面，可编程IC可以提供基础平台。内核可以在基础平台上和/或与所述基础平台协同实现。所述基础平台提供了所述内核与可编程IC耦合到的目标平台和主机进行通信所需的基础架构。例如，所述基础平台可以由目标平台的供应商来实现或决定。因此，所述基础平台可以根据所使用的可编程IC的特定型号或类型以及与可编程IC一起使用的目标平台的型号或类型而变化。

本文描述的本发明的布置可以被实现为由数据处理系统执行的方法或过程。在一个示例中，该方法可以针对异构型多处理器设计的实现，其中在可编程IC的可编程电路中实现了一个或多个内核。在另一示例中，该方法可以针对包括使用可编程IC实现的内核的异构型多处理器系统的操作，例如运行时操作(runtime operation)。

在另一个方面，本发明的布置可以被实现为具有CPU的数据处理系统。所述数据处理系统可以执行针对异构型多处理器设计(例如编译时间方法(compile time method))的实施的方法，其中在可编程IC的可编程电路中实现一个或多个内核。所述数据处理系统还可以包括可编程IC。在这种情况下，所述数据处理系统可以执行包括使用可编程IC实现的内核的异构型多处理器设计的操作(例如，运行时操作)的方法。

在另一个方面，本发明的布置可以被实现为IC。所述IC可以包括基础平台。所述IC还可以被配置为包括在其中实现的一个或多个与基础平台协作操作的内核。所述IC可以实现涉及在所述IC内实现的内核和/或各种主机交互的操作的运行时方法。

在另一个方面，本发明的布置可以被实现为存储程序代码的非暂态计算机可读存储介质，所述程序代码在执行时使处理器和/或系统执行和/或启动在此描述的各种方法和/或流程。

为了说明的简单和清楚，附图中所示的元件不一定按比例绘制。例如，为了清楚起见，相对于其它元件，某些元件的尺寸可能被放大。此外，在认为适当的情况下，在附图中重复参考数字以指示相应的，可比的或类似的特征。

图1是示出用于IC的示例性架构100的框图。在一个方面，架构100在现场可编程门阵列(FPGA)类型的IC中实现。在架构100包括处理器的情况下，架构100也表示SOC类型的IC。SOC是包括执行程序代码的处理器和一个或多个其他电路系统的IC。所述电路系统在与处理器在同一衬底中实现。所述电路系统可以彼此协同工作以及与处理器协同工作。

如图所示，架构100包括几种不同类型的可编程电路，例如逻辑，块。例如，架构100可以包括大量不同的可编程单元块(tile)，包括多千兆比特收发器(MGT)101、可配置逻辑块(CLB)102、随机存取存储器块(BRAM)103、输入/输出块(IOB)104、配置和时钟逻辑(CONFIG/CLOCKS)105、数字信号处理块(DSP)106、专用I/O块107(例如，配置端口和时钟端口)以及其他可编程逻辑108，诸如数字时钟管理器、模拟数字转换器、系统监控逻辑等。

在一些IC中，每个可编程单元块包括可编程互连元件(INT)111，其具有去到和来自每个相邻单元块中对应的INT 111的标准化连接。因此，INT 111合起来实现了所示IC的可编程互连结构。每个INT 111还包括去到和来自同一单元块内的可编程逻辑元件的连接，如图1顶部的示例所示。

例如，CLB 102可以包括可被编程为实现用户逻辑加上单个INT 111的可配置逻辑元件(CLE)112。除了一个或多个INT之外，BRAM 103还可以包括BRAM逻辑元件(BRL)113。通常，包含在单元块中的INT 111的数量取决于单元块的高度。如图所示，BRAM单元块具有与五个CLB相同的高度，但是也可以使用其他数字(例如，四个)。除了适当数量的INT 111之外，DSP单元块106可以包括DSP逻辑元件(DSPL)114。除了INT 111的一个实例，IOB 104还可以包括例如I/O逻辑元件(IOL)115的两个实例。对于本领域技术人员将清楚，连接到例如IOL 115的实际I/O焊盘通常不限于IOL 115的区域。

在图1所示的例子中，例如由区域105，107和108形成的管芯中心附近的柱状区域可用于配置、时钟和其他控制逻辑。从该列延伸的水平区域109被用来分布跨在可编程IC的宽度上的时钟和配置信号。

利用图1所示的架构的一些IC包含了附加逻辑块，其打断了构成IC大部分的常规柱状结构的附加逻辑块。附加逻辑块可以是可编程块和/或专用电路。例如，描绘为PROC110的可选处理器块跨越多列CLB和BRAM。

在一个方面，PROC 110被实现为专用电路，例如作为硬接线处理器，其被制造为实现IC的可编程电路的管芯(die)的一部分。PROC 110可以表示各种不同处理器类型和/或系统,在复杂程度上包括从单个处理器(例如，能够执行程序代码的单个内核)到具有一个或多个内核、块、协处理器、接口等的整个处理器系统的范围。

在另一个方面，在架构100中省略了PROC 110，并且其被所描述的可编程块的其他变体中的一个或多个替代。此外，这种块可以用于形成“软处理器”，因为可以使用各种可编程电路的块来形成如PROC 110的情况一样的可以执行程序代码的处理器。

短语“可编程电路”是指IC内的可编程电路元件，例如本文所述的各种可编程或可配置电路块或单元块，以及互连电路，其根据加载到IC中的配置数据选择性地耦合各种电路块，单元块和/或元件。例如，在图1中所示的诸如CLB 102和BRAM 103之类的在PROC 110之外的部分被认为是IC的可编程电路。

通常，直到配置数据被加载到IC中之后才可以确定可编程电路的功能。一组配置位可用于对诸如FPGA的IC的可编程电路进行编程。所述配置位通常被称为“配置比特流”。通常，可编程电路在没有先将配置比特流加载到IC中的情况下是不可操作的或可用的。所述配置比特流有效地实现或实例化可编程电路内的特定电路设计。所述电路设计指定，例如，可编程电路块的功能方面和各种可编程电路块之间的物理连接。

“硬连线”或“硬化”(即不可编程)的电路被制造为IC的一部分。与可编程电路不同，硬连线电路或电路块不是在IC制造之后通过加载配置比特流实现的。硬连线电路通常被认为具有专用电路块和互连，例如其在没有先将配置比特流加载到IC中的情况下是可用的，例如PROC 110。

在一些情况下，硬连线电路可以具有一个或多个操作模式，其可以根据存储在IC内的一个或多个存储器元件中的寄存器设置或值来设置或选择。所述操作模式可以，例如，通过将配置比特流加载到IC来设置。尽管有这种能力，但是硬连线电路不被认为是可编程电路，因为硬接线电路在作为IC的一部分被制造时是可操作的并且具有特定的功能。

在SOC的情况下，所述配置比特流可以指定要在可编程电路内实现的电路和要由PROC 110或软处理器执行的程序代码。在一些情况下，架构100包括将配置比特流加载到适当的配置存储器和/或处理器存储器的专用配置处理器。与可能包含的PROC 110不同，配置处理器不执行用户程序代码。在其他情况下，架构100可以利用PROC 110来接收配置比特流，将配置比特流加载到适当的配置存储器中，和/或提取用于执行的程序代码。

图1旨在说明可用于实现包括可编程电路(例如可编程结构(programmablefabric))的IC的示例性架构。例如，列中的逻辑块的数量，列的相对宽度，列的数量和顺序，列中包括的逻辑块的类型，逻辑块的相对大小以及包括在图1的顶部的互连/逻辑实现纯粹是示范性的。在实际的IC中，例如，通常在CLB出现的任何地方包括多于一个相邻列的CLB，以便有效地实现用户电路设计。然而，相邻CLB列的数量可随IC的总体尺寸而变化。此外，IC内的诸如PROC 110的块的大小和/或位置仅仅是为了说明的目的，并不意图作为限制。

图2是示出示例性数据处理系统(系统)200的框图。在一个方面，系统200可以表示编译时(compile-time)系统，其可以作为可编程IC内的电路被编程为实现异构型多处理器设计(例如程序)的内核。如本文所定义的，“异构型多处理器设计”是包括在主机系统上执行的部分和在不同设备或处理器上执行的称为内核的至少一个附加部分的程序。异构型多处理器设计的一个例子是OpenCL程序或设计。在一个示例中，在主机上执行的部分可以以与在不同设备或处理器上执行的部分不同的编程语言来指定。可编程IC可以具有参考图1所描述的架构。

在另一个方面，系统200可以表示运行时异构型多处理器系统，其中处理器用作主机，并且可编程IC实现一个或多个内核。如本文所定义的，“异构型多处理器系统”是包括两个或多个处理器的计算系统。两个或多个处理器可以是不同类型的处理器。例如，异构型多处理器系统可以包括中央处理单元(CPU)，图形处理单元(GPU)，数字信号处理器(DSP)，诸如FPGA的可编程IC等。异构型多处理器系统可以是OpenCL系统。

如图所示，系统200包括通过系统总线215或其他合适的电路耦合到存储器元件210的至少一个处理器205，例如，中央处理单元(CPU)。系统200将程序代码存储在存储器元件210内。处理器205执行经由系统总线215从存储器元件210取得的程序代码。在一个方面，系统200被实现为计算机或其他适于存储和/或执行程序代码的数据处理系统。然而，应当理解，系统200可以以包括能够执行本公开所描述的功能的处理器和存储器的任何系统的形式来实现。此外，系统200可以被实现为一个或多个联网的数据处理系统，例如服务器。

存储器元件210包括一个或多个物理存储器设备，例如本地存储器220和一个或多个大容量存储设备225。本地存储器220指的是随机存取存储器(RAM)或其他非持久存储器件，一般在实际执行程序代码时使用。大容量存储设备225可以被实现为硬盘驱动器(HDD)，固态驱动器(SSD)或其他持久数据存储设备。系统200还可以包括一个或多个高速缓存存储器(未示出)，其提供至少一些程序代码的临时存储，以便减少在执行期间必须从大容量存储设备225获取程序代码的次数。

输入/输出(I/O)设备(例如键盘230，显示设备235和点击设备240)可选地可以耦合到系统200。I/O设备可以直接或通过中间的I/O控制器耦合到系统200。还可以耦合网络适配器245到系统200，以使系统200能够通过中间的私有或公共网络与其他系统、计算机系统、远程打印机、远程存储设备和/或目标平台260耦合。调制解调器、电缆调制解调器、以太网卡和无线收发器是可以与系统200一起使用的不同类型的网络适配器245的示例。通信端口250(例如通用串行总线端口，火线端口，外围组件互连(PCI)和/或PCI Express(PCIe)端口等)也可以耦合到系统200以允许系统200耦合到另一个系统，例如包括目标平台260的上述系统中的任何一个。

在一个方面，存储器元件210存储电子设计自动化(EDA)应用程序255。EDA应用程序255可以被存储在例如系统200表示编译时系统的实现中。EDA应用程序255可以包括一个或多个不同的组件或模块。以可执行程序代码的形式实现的EDA应用程序255由系统200执行。因此，EDA应用程序255被认为是系统200的集成部分。EDA应用程序255以及当执行EDA应用程序255时系统200使用、生成和/或操作的任何数据项均属于功能数据结构，其当被用作系统200的一部分时赋予功能性。作为编译时系统，主机应用程序258可以从系统200中排除。

在编译时系统的情况下，用户通过系统200操作、执行EDA应用程序255。系统200可以接收作为输入的异构型多处理器设计275并将异构型多处理器设计275的一个或多个内核综合为可以在IC 270内实现的电路。系统200可以生成并输出二进制容器(binarycontainer)280。在一个方面，二进制容器280可以包括其中的内容的描述和一个或多个配置比特流，不管是部分的还是完整的。在另一个方面，二进制容器280可以包括其中的内容的描述、一个或多个可执行模拟文件和/或一个或多个寄存器传输级(RTL)文件，其可以在RTL或硬件描述语言模拟器中进行模拟。在这种情况下，二进制容器280除了可执行模拟文件和/或RTL文件之外还可以包括一个或多个配置比特流，不管是部分的还是完整的。二进制容器280可以被存储在存储器元件210中和/或通过网络适配器245和/或通信端口250提供给另一个系统。

在另一个方面，存储器元件210存储主机应用程序258。主机应用程序258可以被存储在例如系统200表示异构型多处理器运行时系统的实现中。主机应用程序258可以包括一个或多个不同的组件或模块。以可执行程序代码的形式实现的主机应用程序258由系统200执行。因此，主机应用程序258被认为是系统200的集成部分。主机应用程序258以及当执行主机应用程序258时系统200使用、生成和/或操作的任何数据项属于功能数据结构，其当被用作系统200的一部分时赋予功能性。作为运行时系统，可以从系统200中排除EDA应用程序255。

系统200可以通过通信链接265耦合到目标平台260。在实施运行时系统的情况下，目标平台260耦合到系统200或被认为是系统200的一部分。因此，应当理解，在编译时系统情况下可以排除目标平台260。从目标平台260延续，通信链接265可以被实现为可操作为耦合到通信端口250和/或网络适配器245的各种不同的有线和/或无线连接中的任何一种。

目标平台260可以被实现为电路板，例如具有在其上实施的电路的印刷电路板。目标平台可以被实现为卡，其可以被插入用于通信端口250的机械连接器(例如，在系统200内或系统200外部)。目标平台260可以包括耦合到通信链接265的连接器。所述连接器可以使用目标平台260的电路耦合到IC 270。

IC 270可以使用插孔、插座、另一种安装技术(例如将IC 270直接焊接到目标平台260)或诸如此类而被耦合到目标平台260。IC 270通过目标平台260耦合到通信链接265。在一个方面，IC 270是可编程IC。例如，可以使用参考图1描述的架构来实现IC 270。在另一个方面，IC 270可以被实现为SOC。IC 270可以实现异构型多处理器设计的一个或多个内核作为电路。异构型多处理器设计可以是OpenCL设计。

在运行时系统的情况下，处理器205可以作为主机操作。异构型多处理器设计的一个或多个内核可以在IC 270内实现。在操作期间，根据需要可以在IC 270内实现新的和/或不同的内核，因为IC 270可以在运行中被配置或视情况而动态地重新配置，而不会导致未被配置或视情况而重新配置的IC 270的其它部分中断。

图3是示出图2的目标平台260的示例性架构的框图。如图所示，IC 270和RAM 345耦合到目标平台260。目标平台260还包括耦合到IC 270的连接器350。尽管被示为卡边缘类型的连接器，但是应当理解，连接器350可以实现为各种不同的连接器类型。此外，目标平台260可以包括一个或多个附加部件(未示出)。所述附加组件例如可以耦合在连接器350和IC270之间。

IC 270包括静态区域335和内核区域340。在一个方面，静态区域335包括支持异构型多处理器编程模型所需的基础设施IP。在一个示例中，异构型多处理器编程模型是OpenCL模型。静态区域335，例如，在运行时期间将内核区域340与位于目标平台260上的其他组件(例如RAM 345和/或其他系统，例如主机，例如处理器205)进行通信地链接。静态区域335，例如，可以实现用于与主机通信的软件接口。在一个方面，静态区域335可以是由目标平台260的供应商和/或制造商提供的电路实现的。

内核区域340表示实现IC 330内核的部分。在一个方面，内核区域340可以具有与静态区域335的存储器映射接口。与静态区域335不同，内核区域340可以被动态生成并与静态区域335集成。例如，不同的内核和不同内核的组合可以在运行期间的不同时间实现于内核区域340内。

图4是示出包括目标平台260的异构型多处理器运行时系统的示例性层的框图。在一个示例中，异构型多处理器运行时系统是OpenCL系统。如图所示，该主机执行在主机应用程序内实现的运行时层405。如所讨论的，所述主机可以被实现为参考图1描述的系统200的处理器205。目标平台软件层415在目标平台电路内实现。运行时层405通过公共低级驱动器接口410与目标平台软件层415进行通信。例如，运行时层405使用在公共低级驱动器410中定义的，标准的文件记录的应用编程接口(API)来与目标平台软件层415进行通信。目标平台软件层415，例如，可以被实现为内核驱动程序。

目标平台软件层415，在目标平台的电路260中执行，通过目标平台专用编程接口420(例如，硬件编程接口)与静态区域335通信。静态区域335为内核区域340提供时钟和复位信号430。静态区域335还通过耦合到控制寄存器(未示出)的存储器映射从接口440向内核区域340提供信息。内核区域340通过耦合到RAM 345的存储器映射总线主接口435向静态区域335提供信息。

图5是示出在图3的IC 270内实现的示例性电路的框图。更具体地，图5示出了可以用于实现静态区域335的示例性架构。框505、510、515、520和525中的每一个表示电路块。作为静态区域335的一部分的块505-525和内核区域340中的每一个可以在IC 270的可编程电路内实现。

如图所示，静态区域335可以包括耦合到总线直接存储器访问(DMA)控制器510的总线端点505。总线DMA控制器510耦合到互连515。互连515耦合到互连520和内核区域340。互连520耦合到内核区域340和存储器控制器525。存储器控制器525耦合到实现在芯片IC270外的RAM 345。

总线端点505被配置为通过总线与异构型多处理器设计的主机进行通信。可以包括总线DMA控制器510以支持主机RAM(例如本地存储器220)和目标平台260上的RAM 345之间的DMA功能。在一个方面，总线DMA控制器510包括主接口530。互连515可以包括从接口535和主接口540和545。如图所示，从接口535耦合到主接口530。内核区域340包括从接口550和主接口555。互连515的主接口545耦合到内核区域340的从接口550。

互连520包括从接口560和565以及主接口570。存储器控制器525包括从接口575。如图所示，互连515的主接口540耦合到互连520的从接口560。内核区域340的主接口555耦合到互连520的从接口565。

互连515和520是被配置为将两个或更多个其他电路块耦合在一起的电路块。在一个方面，互连515和520可以被实现为将一个或多个存储器映射的主设备与一个或多个存储器映射的从设备耦合的电路块。互连电路块实现的示例是符合英国剑桥的

有限公司的

AXI第4版规范的实例。然而，应当理解，可以使用其他互连类型和/或技术来实现互连515和520。本公开不被提供的示例性互连电路块限制。

如图5的架构所示，总线DMA控制器510和内核区域340用作存储器控制器525的主控。互连515允许主机，例如，通过总线读取和写入RAM 345。互连520支持为存储器控制器525创建两个总线主控，即总线DMA控制器510和内核区域340。

内核区域340最初可以被实现为用于编译的异构型多处理器设计内核的容器。在一个方面，内核区域340可以被实现为用于编译内核的具有占位符的分级IP。内核区域340中可以包括一个或多个内核。在一个示例中，在内核区域340中可以包括多达16个内核。来自主机的命令可以通过从接口550接收。内核区域340可以通过主接口555向存储器控制器525提供命令。时钟和复位信号通过主接口545和从接口550之间的连接提供给内核区域340和在其中实现的任何内核。

图6是示出内核区域340的实施例的框图。如图所示，内核区域340包括附加互连605和615。互连605包括耦合到互连515的主接口545的从接口550。互连605还包括耦合到内核电路610-1的从接口630的主接口625。主接口625还可以耦合到一个或多个其他内核电路610-N，其中N是整数值。

统一标示为内核电路610-1至610-N的内核电路610可以表示相同内核电路的多个实例，并且因此表示相同内核的多个实例。在另一示例中，内核电路610可以表示两个或多个不同的内核电路。在另一个示例中，内核电路610可以表示第一内核电路和一个或多个附加的不同内核电路的一个或多个实例。互连615具有耦合到每个内核电路610的主接口640和645的从接口650。互连615包括耦合到互连520的从接口565的主接口555。

在一个方面，互连605和615可以支持多达16个不同的内核电路实例，16个不同的内核电路或不超过16个的它们的组合。如所讨论的，所提供的可以在内核区域340内实现的内核和/或内核实例的特定数量的目的是为了说明而不是限制。

在异构型多处理器框架(如OpenCL)中，并行内核调用可以被描述为称为NDRange的1，2或3维索引空间。NDRange被细分为工作组。工作组包括多个工作项。例如，NDRange中的每个点都称为工作项。

异构型多处理器设计的内核被编译成一个或多个计算单元。系统设计者，例如用户，为给定内核确定要并行实现的计算单元的数量。在一个方面，内核的计算单元的数量表示在内核区域340内实现并且并行操作的内核电路的实例数。每个计算单元能够处理由主机确定和提供的一个工作组。

在图6的示例中，每个内核电路块610-1至610-N表示工作单元的电路。内核电路块610-1至610-N可以表示相同内核的工作单元(例如并行操作的多个实例)或并行操作的两个或多个内核的工作单元，其中所述内核中的一个或多个在具有多个实例的内核区域340内实施。

图7是示出实施异构型多处理器设计的内核的示例性方法700的流程图。在一个示例中，异构型多处理器设计可以是OpenCL设计。方法700可以由编译时系统执行，例如参考图2描述的系统。方法700可以开始于其中异构型多处理器设计包括内核的状态，该内核在C，C++，OpenCL，OpenCL C，OpenCL兼容的高级编程语言或其他高级编程语言中指定。在一个方面，可以使用各种高级编程语言中的任一种来指定内核。在另一个方面，用于指定内核的高级编程语言可以是支持并行或并行操作的显式规范或符号的高级编程语言。系统可以访问内核。

在框705中，系统生成内核的RTL描述。所述RTL描述可以使用硬件描述语言(HDL)来指定。如本文所定义的，术语“硬件描述语言”或“HDL”是便于诸如集成电路的数字系统的文档化、设计和制造的计算机语言。HDL将程序验证技术与专家系统设计方法相结合。例如，使用HDL，用户可以设计和指定电子电路，描述电路的操作，并创建测试以验证电路的操作。HDL包括正在建模的电子系统的空间和时间结构以及行为的标准的基于文本的表达。HDL语法和语义包括用于表示并发的显式符号。与大多数高级编程语言相反，HDL还包括明确的时间概念，这是数字系统的主要属性。

在框710中，系统将内核的RTL描述与基础平台集成。在一个方面，基础平台可以与在静态区域335内实现的并且参考图4和/或5描述的电路设计类似或相同。

在框715中，系统生成配置比特流和支持数据。所述配置比特流指定内核的硬件实现，例如参考图6所描述的计算单元。在一个方面，例如，所述配置比特流可以是部分比特流，例如其仅指定内核或一个或多个内核。在另一个方面，所述配置比特流可以是完整比特流，其指定内核或者根据情况可以是多内核，以及基础平台。

所述支持数据描述所述配置比特流和/或所述配置比特流的内容。在一个方面，所述支持数据可以指定内核实现中包括的IP块和/或内核的列表。在另一个方面，所述支持数据可以指定可编程IC内的二维坐标位置，当指定为部分配置比特流时，该内核的硬件实现将在所述位置被实现。

在框720中，所述系统包括二进制容器内的所述配置比特流和所述支持数据。在一个方面，所述二进制容器可以包括多个单独的文件。例如，二进制容器可以包括一个或多个配置比特流和一个或多个支持数据文件。

在另一个方面，内核的RTL描述可以被包括在二进制容器内。然后可以将RTL描述与RTL模拟器一起使用，以便将内核实现作为整体异构型多处理器设计仿真的一部分进行测试。例如，在异构型多处理器设计的运行时模拟期间，主机可以向RTL模拟器提供包含RTL描述的二进制容器。RTL模拟器可以访问二进制容器中的RTL描述。在另一个方面，内核的可执行版本可以被包含在可以使用用于测试和/或模拟目的的处理器来执行的二进制容器中。例如，在异构型多处理器设计的运行时模拟期间，主机可以将包含可执行版本的内核的二进制容器提供给模拟器。应当理解，内核的可执行版本可以是内核的硬件实现的可执行模型。模拟器可以从二进制容器访问内核的可执行版本。因此，二进制容器支持多个不同的内核实现，无论是用于具有可编程IC的运行时间的配置比特流，还是用于在数据处理系统上进行仿真的可执行版本和/或用于使用RTL模拟器进行仿真的RTL版本。

所述二进制容器可以包括仅有配置比特流；仅有内核的可执行版本；仅有内核的RTL版本；配置比特流和内核的RTL版本；配置比特流和内核的可执行版本；内核的可执行和RTL版本；或配置比特流，内核的可执行版本和内核的RTL版本。对于在二进制容器内实现的内核版本的任何前述组合，也可以包括支持数据。CPU和/或GPU供应商在异构型多处理器设计中使用的现有容器，特别是OpenCL设计，与“内存(in memory)”和地址映射对象一起使用。这样的容器不支持同一容器内的内核或多个不同内核类型的仿真版本。

虽然所使用的二进制容器可以支持多种不同类型的内核实现，但是在一个方面，第一容器可以包括第一类型的内核实现，例如配置比特流，RTL描述或可执行的，而第二二进制容器可以包括不同类型的内核实现。在另一个方面，第一容器可以包括指定第一内核的部分配置比特流，而第二容器可以包括指定第二和不同内核的部分配置比特流。

用于异构型多处理器计算语言(例如，用于主机和内核之间的通信的OpenCL)的标准API仅支持二进制对象文件。由系统生成的二进制容器符合这个二进制要求，即所有预编译的内核都可以通过自包含的对象访问。在框720中生成的二进制容器可以被主机用来在运行时间内在可编程IC内实现内核电路，例如计算单元。

图8是示出用于在可编程IC内实现异构型多处理器设计的内核的示例性过程800的块流程图。在一个示例中，异构型多处理器设计可以是OpenCL设计。过程800可以由诸如参考图2描述的执行EDA应用程序255的系统执行。在一个方面，EDA应用程序255可以被实现为OpenCL编译器工具。过程800示出了示例性的编译时(compile-time)系统实现。

如图所示，异构型多处理器设计275(例如也称为异构型多处理器应用，OpenCL设计或OpenCL应用)可以包括主机代码805和内核815。应当理解，虽然图示为一个内核，但是异构型多处理器设计275也可以包括可以通过过程800实现的多于一个的内核。主机代码805是在主机中执行的异构型多处理器设计275的一部分。主机代码805可以用诸如C，C++等的高级编程语言来指定。

如本文所定义，术语“高级编程语言”是指用于对数据处理系统进行编程的编程语言或指令集，其中指令具有来自数据处理系统的细节(例如机器语言)的强抽象。例如，高级编程语言可以自动化或隐藏诸如存储器管理的数据处理系统的操作的方面。抽象的数量通常定义了编程语言有多么的“高级”。当使用高级编程语言时，用户不需要面对高级编程语言将执行的数据处理系统的寄存器、存储器地址等困难。在这方面，高级编程语言包括很少或根本没有在一对一的基础上直接翻译成数据处理系统的本地操作码的指令。高级编程语言的示例包括但不限于C，C++，SystemC等。

主机代码805被提供给C编译器840或其他高级语言编译器。C编译器840生成主机代码805的目标代码版本(图示为App.o 860)。链接器885接收异构型多处理器运行时库875App.o 860，并且生成主机应用程序894。异构型多处理器运行时库875可以包括用于与目标平台通信的公共低级驱动程序。主机应用程序894由运行时异构型多处理器系统的CPU执行。

异构型多处理器高级综合块890接收内核815并生成kernel.hdl 892。Kernel.hdl892是内核815的RTL版本。系统汇编器850接收Kernel.hdl 892和基础平台描述825。在一个方面，基础平台描述825可以是描述实际基础平台的多方面的元数据文件。如上所述，所述基础平台是在可编程IC 270的静态区域335内实现的电路。

例如，由基础平台描述825，系统汇编器850确定目标平台和用于内核实现的可编程IC的特定类型。例如，系统汇编器850可以标识目录，该目录指定关于基础平台的实施细节以及用于主机与目标平台和基础平台通信所需的低级驱动程序。所识别的目录可以包括基础平台的一个或多个打包的IP。系统汇编器850可以检索基础平台的打包的IP，包括将基础平台与内核耦合的互连IP。例如，互连IP可以指定将Kernel.hdl 892集成或并入基础平台的封装IP所需的各种互连电路块。系统汇编器850生成二进制容器280。系统汇编器850可以生成指定与包含在二进制容器280中的基础平台集成的内核815的硬件实现的配置比特流。

例如，包含在二进制容器280中的每个配置比特流可以实现从内核815或kernel.Hdl 892确定的一个或多个计算单元(视具体情况而定)。如所讨论的，系统设计者确定要为给定内核并行实现的计算单元的数量。

根据用户偏好，系统汇编器850可以在如之前所描述的二进制容器280内包括Kernel.hdl 892，例如，用于RTL仿真的内核815的RTL版本和/或用于模拟的内核815的可执行的例如目标代码版本280。系统汇编器850还包括二进制容器280内的支持数据(未示出)。

在一个方面，系统汇编器850将内核815与基础平台集成。例如，具有在基础平台描述825和kernel.Hdl 892中指定的信息的系统汇编器850可以通过执行诸如技术映射、布局、布线等功能来将内核815与基础平台集成，从而产生配置比特流。所述配置比特流可以是指定基础平台和内核两者的完整配置比特流或仅指定内核的部分配置比特流。在任何情况下，系统汇编器850使用指定的互连IP将基础平台与内核耦合。

在另一个方面，系统汇编器850可以生成二进制容器280，以包括除配置比特流之外的文件。例如，如上所述，内核815可以被提供给异构型多处理器编译器，其产生内核815的处理器可执行的版本(例如目标代码)。内核815的可执行版本，例如内核815的硬件实现的可执行模型，可以被提供给系统汇编器850。系统汇编器850可以包括二进制容器280内的内核815的可执行版本代替配置比特流。在另一示例中，系统汇编器850可以包括二进制容器280内的kernel.hdl 892代替配置比特流。

图9是示出处理在IC 270内实现的用异构型多处理器计算语言指定的内核的示例性方法900的流程图。在一个示例中，异构型多处理器计算语言可以是OpenCL。方法900可以由编译时系统执行，例如参考图2描述的系统。在一个方面，方法900示出了在内核的RTL描述的生成期间执行的各种操作，所述内核初始时用OpenCL C，C，C++，另一高级编程语言，或者在本公开内容中指出的一种语言的衍生物和/或变体所指定。

在框905中，该系统识别并且映射所述内核的存储器访问。所述异构型多处理器全局存储器可以被映射到主存储器总线。例如，OpenCL全局存储器可能被映射到AXI主存储器总线。内核参数可以映射到从控制总线。例如，内核参数可以映射到AXI从控制总线。

在框910中，该系统识别由内核使用的参数，并且包括用于IC内核的硬件实现的存储器映射内的参数。在框915中，该系统将变量与IC的存储器结构相关联，作为内核的专用存储器。在框920中，该系统将本地存储器指令与IC的存储器结构相关联，作为内核的本地存储器。

在框925中，该系统为所述内核生成控制流程图。在一个方面，系该统将内核转换为LLVM中间表示(IR)格式。由LLVM IR格式，该系统通过识别其中的数据流来生成控制流程图。在框930中，该系统使用所述控制流程图来识别内核的并行区域。该并行区域可以在控制流程图中被隔离。例如，对于控制流程图中的每个并行区域，该区域将具有进入区域的一个控制边缘和离开该区域的一个控制边缘。

在框935中，该系统可选地围绕每个并行区域构建“for”循环。识别并行区域并将每个区域表示为“for”循环允许将作为数据并行实现的内核表示为诸如C，C++等的连续高级编程语言。在框940中，该系统使用流水线生成电路描述。例如，通过将并行区域表示为“for”循环，系统可以将该区域综合为将被综合的高级编程语言，例如C，C++等。

图10是示出异构型多处理器系统的示例性存储器架构1000的框图。在一个示例中，异构型多处理器系统是OpenCL系统。如图所示，主机1005包括主机存储器1010。主机1005可以被实现为处理器205，而主机存储器1010可被实现为存储器元件210。主机1005耦合到目标平台260以及全局存储器和常量存储器(constant memory)1015。如所讨论的，可以由存储器控制器(未示出)提供对全局存储器和常量存储器1015的访问。全局存储器和常量存储器1015可以被实现为RAM 345，其中存储器控制器被实现在IC 270内。然而应当理解，存储器控制器可被实现为在IC 270外部的目标平台260上的存储器控制器，但其被配置为与IC 270通信。

IC 270包括计算单元1020和1025。虽然在IC 270中示出了两个计算单元，但是应当理解，IC 270可以包括少于或多于两个计算单元。此外，特定计算单元和在IC 270内实现的计算单元的特定数量可以在运行时期间改变。计算单元1020和1025被实现为内核区域340的一部分。出于说明的目的，未示出静态区域335。

如图所示，计算单元1020包括本地存储器1030，处理元件1040和1045以及专用存储器1060和1065。本地存储器1030由处理元件1040和1045共享。处理单元1040和1045中的每一个均耦合到专用存储器1060和1065的单独的未共享的一个。计算单元1025包括本地存储器1035，处理元件1050和1055以及专用存储器1070和1075。本地存储器1035由处理元件1050和1055共享。处理单元1050和1055中的每一个均耦合到专用存储器1070和1075中的单独的未共享的一个。计算单元1020和1025都可以访问全局存储器和常量存储器1020。

在一个实施例中，可以使用目标平台上的RAM、主机RAM和/或主机的一个或多个高速缓存存储器来实现主机存储器1010和全局存储器和经常存储器1020。例如，本地存储器1030和1035可以使用一个或多个BRAM 103在IC 270内实现。可以使用CLB 102内包括的查找表RAM来实现专用存储器1060、1056、1070和1075。

提供将IC 270的存储器结构分配到图10的存储器架构1000的存储器仅仅为了说明的目的。应当理解，在综合期间，根据IC 270的存储器结构的可用性和所需的存储器的量，可以使用IC 270的一个或多个其他存储器结构来实现专用存储器和/或本地存储器。

图11是示出以异构型多处理器计算语言指定的内核的示例性处理的框图。尽管图11以OpenCL为示例，但是应当理解，可以以除OpenCL之外的高级编程语言来指定内核，并且本公开内容中描述的本发明的布置不限于所提供的示例。在一个方面，图11示出了在图9的框905中执行的处理。如图所示，该系统识别内核中的“global int”指令。“global int”指令指示从主机传递到内核的特定缓冲区。该系统将存储器访问映射为互连上的事务(transaction)。

在另一个方面，图11示出了在图9的框910中执行的处理。在图11的例子中，该系统确定所使用的标识符(id)，如由“get_local_id(0)”函数所示。例如，从主机传入内核的指针被指定在实现于内核中的寄存器映射。诸如所述id之类的数据被主机写入内核，例如由在运行时期间在主机中执行的主机应用程序写入。例如，主机可以将任何必需的数据(诸如id之类)写入到内核电路610的寄存器映射的适当的寄存器中。

该系统进一步从对内核程序代码的分析识别内核使用的任何隐式参数。可能需要从主机提供给内核的隐式参数的示例包括但不限于ND范围的大小、工作组的大小等。在某些情况下，隐式参数可能不会通过主机和内核之间的接口传递。然而，这些参数可以通过寄存器映射传递。

图12是示出以异构型多处理器计算语言指定内核的示例性处理的框图。虽然图12以OpenCL为示例，应当理解，可以以除OpenCL之外的高级编程语言来指定内核，并且在本公开内描述的本发明的布置不限于所提供的示例。在一个方面，图12示出了在图9的框920和925中执行的处理。在图12的示例中，参考框920，诸如“id”之类的变量与实现内核的专用存储器1060的存储器结构相关联。专用存储器的示例可以包括流水线寄存器，小阵列，BRAM，查找表RAM等。参考框925，该系统将每个“local int”存储器指令与内核中的本地存储器1030(诸如BRAM)相关联。

图13-1和13-2(作为集体统称为图13)展示了用异构型多处理器计算语言指定内核的处理。在一个方面，图13示出了在图9的框930和935中执行的处理。参考图13-1，其示出了用于内核1300的示例性OpenCL C源代码。尽管图13以OpenCL为示例，但是应当理解，可以以OpenCL之外的高级编程语言来指定内核，并且本公开内描述的本发明的布置不限于所提供的示例。

该系统将内核1300的并行区域识别为区域1305、1310和1315。作为在框930中识别并行性的一部分，该系统可以识别控制并行性的特定指令和/或构造。该系统可以识别内核1300内的“屏障(barrier)”指令的实例。例如，“屏障”指令指示在任何工作项目可以超过屏障指令之前，所有工作项目必须到达屏障指令。所述屏障指令可以用作存储器栅栏或同步机制。该系统可以标识“async_work_group_copy()”指令(在本文中称为“async”)的实例。所述“async”指令指定所有工作项目必须使用相同的参数到达副本。因此，在一个方面，该系统通过识别内核1300内的控制并行性的指令来识别内核1300的并行区域。

所述异构型多处理器执行和存储器模型保证区域1305、1310和1315中的每一个可以完全并行、完全依次或以不同的组合来实现。必须遵守的序列化伴随着直接影响并行性的指令和/或构造(例如屏障和/或async指令)而出现。

图13-2示出了用于内核1300的框925的数据流程图的生成。指示了并行区域1305、1310和1315。对应于包含在区域1305中的“for”部分的结束或返回，包括了第四并行区域1305-1。

图14是示出用异构型多处理器计算语言指定内核的示例性处理的框图。在一个示例中，异构型多处理器计算语言可以是OpenCL。在一个方面，图14示出了在图9的框935和940中执行的处理。在图14的例子中，示出了对应于灰度转换的区域1310的处理。每个循环迭代处理一个工作项。整个循环处理一个工作组。所述循环可以被实现为流水线，每个时钟周期将一个新工作项引入该流水线。如图所示，该系统在并行区域1310附近创建“for”循环结构。所述电路描述使用如图所示的流水线生成，其中每列工作项对应于内核的一个流水线阶段。每行工作项对应于一个循环。

内核的示例性寄存器映射展示如下。

//0x00:Control signals

//bit 0-ap_start(Read/Write/COH)

//bit 1-ap_done(Read/COR)

//bit 2-ap_idle(Read)

//bit 3-ap_ready(Read)

//bit 7-auto_restart(Read/Write)

//others-reserved

//0x04:Global Interrupt Enable Register

//bit 0-Global Interrupt Enable(Read/Write)

//others-reserved

//0x08:IP Interrupt Enable Register(Read/Write)

//bit 0-Channel 0(ap_done)

//bit 1-Channel 1(ap_ready)

//others-reserved

//0x0c:IP Interrupt Status Register(Read/TOW)

//bit 0-Channel 0(ap_done)

//bit 1-Channel 1(ap_ready)

//others-reserved

//0x10:Data signal of group_id_x

//bit 31～0-group_id_x[31:0](Read/Write)

//0x14:reserved

//0x18:Data signal of group_id_y

//bit 31～0-group_id_y[31:0](Read/Write)

//0x1c:reserved

//0x20:Data signal of group_id_z

//bit 31～0-group_id_z[31:0](Read/Write)

//0x24:reserved

//0x28:Data signal of global_offset_x

//bit 31～0-global_offset_x[31:0](Read/Write)

//0x2c:reserved

//0x30:Data signal of global_offset_y

//bit 31～0-global_offset_y[31:0](Read/Write)

//0x34:reserved

//0x38:Data signal of global_offset_z

//bit 31～0-global_offset_z[31:0](Read/Write)

//0x3c:reserved

//0x40:Data signal of matrix

//bit 31～0-matrix[31:0](Read/Write)

//0x44:reserved

//0x48:Data signal of maxIndex

//bit 31～0-maxIndex[31:0](Read/Write)

//0x4c:reserved

//0x50:Data signal of s1

//bit 31～0-s1[31:0](Read/Write)

//0x54:reserved

//0x58:Data signal of s2

//bit 31～0-s2[31:0](Read/Write)

//0x5c:reserved

//(SC＝Self Clear,COR＝Clear on Read,TOW＝Toggle on Write,

COH＝Clear on Handshake)

异构型多处理器HLS 890，如参考图8所述，为每个编译成RTL的内核生成如上所示的自定义寄存器映射。所述主机可以使用寄存器映射将位于目标平台上的设备存储器中的缓冲器(例如，OpenCL缓冲器)的地址、标量参数传递到内核，以及控制信号以控制内核。所述主机也可以使用寄存器映射将OpenCL规范要求的组id和组偏移量传递给内核。在一个方面，寄存器映射可以被包括在生成的二进制容器中。例如，寄存器映射可以是并入先前描述的二进制容器中的支持数据的一部分。

示例性平台元数据文件展示如下。

在一个方面，上面所示的平台元数据文件是参考图8所描述并且提供给系统链接器830的基础平台描述825的实现。如图所示，所述平台元数据文件指定了可编程IC耦合到的目标平台或板的类型。此外，所述平台元数据文件指示了板上可编程IC的特定特征，例如模型和/或类型以及特定区域(例如，静态区域)的时钟频率。链接器830可以从所述平台元数据文件识别目标平台，并访问为平台元数据文件中指定的为目标平台命名的目录结构。

图15是示例性目录结构1500。所指示的顶级目录使用与可由系统链接器830从平台元数据文件读取的目标平台相同的名称。在这个例子中，所述顶层目录叫做“BoardName”。然而，参考上面提供的示例性平台元数据文件，所述顶级目录可以被指定为“VC690”或其派生词。在任何情况下，系统链接器830使用图15的目录结构来获得平台FPGA 845。为展示的目的，图15中的目录为粗体字。例如，“Board Name”，“driver”，“ipi”和“local_lib”是目录。列出的其余项是文件和/或软件包。

通过正确打包并添加到系统的指定平台目录中的文件，该系统可以自动定位任何新添加的目标平台。在所示的示例中，“driver_file.so”是主机通过总线与目标平台通信的低级驱动程序。如图所示，driver_file.so位于“driver”目录中。所述的平台元数据文件，在图15的示例中被称为“platform.xml”，被放在根目录下。如参考图3、4和5所述的在IC的静态区域中使用的任何封装的IP，可以存储在“local_lib”目录中。名为“bp.tcl”的基础平台框图表TCL文件和静态区域电路设计中的任何顶级设计约束文件(例如时序和/或物理约束)包含在“ipi”目录中。

在图15中描绘为“driver_file.so”的公共低级驱动程序可以包括具有多个功能的API。所述的公共低级驱动程序API(以下简称“驱动程序API”)允许在主机中执行异构型多处理器运行时程序代码以与目标平台进行通信。例如，驱动程序API支持将配置比特流下载到可编程IC中，分配和/或解除分配缓冲区，将缓冲区从主机存储器迁移到目标平台存储器，将目标平台存储器迁移到主机存储器，以及通过内核的控制端口和如在可编程IC内实现的内核进行通信。

所述的驱动程序API还支持地址空间。所述地址空间可用于访问目标平台的外围设备。例如，目标平台的每个外围设备可能具有地址空间的其自己的存储器映射范围。目标平台可选择地具有可用于寻址目标平台的所有外围设备的平面存储器空间(flat memoryspace)。

所述的driver_file.so可以支持各种数量，例如可以在目标平台上读取或写入的缓冲器(例如DMA缓冲器)的最小尺寸。此外，可以支持称为“enums”的一个或多个枚举的地址空间。存储器操作可以使用平面寻址或相对寻址。示例性的枚举可以包括但不限于，XCL_ADDR_SPACE_DEVICE_FLAT，XCL_ADDR_SPACE_DEVICE_RAM，XCL_ADDR_KERNEL_CTRL和XCL_ADDR_SPACE_MAX。

所述驱动程序API支持多种设备访问操作，包括但不限于：

●xclDeviceHandle xclOpen(const char*deviceName)

●void xclClose(xclDeviceHandle handle)

●int xclGetDeviceInfo(xclDeviceHandle handle，xclDeviceInfo*info)(xclDeviceHandle handle)

所述驱动程序API通过操作“int xclLoadBitstream(xclDeviceHandle handle，const char*fileName)”来支持配置比特流的加载操作。因此，主机可以启动将配置比特流(无论是完整的还是部分的)加载到IC中，以在运行时期间根据需要在硬件中实现一个或多个不同的内核。

所述驱动程序API提供用于管理目标平台存储器的各种操作。例如，目标平台的供应商需要使用以下API来提供内存管理：

●uint64_t xclAllocDeviceBuffer(xclDeviceHandle handle，size_t size)

操作“xclAllocDeviceBuffer”在目标平台上分配指定大小的缓冲区，并返回目标平台RAM中分配的缓冲区的偏移量作为返回值。该偏移作为缓冲区句柄。OpenCL运行时(runtime)随后将返回的句柄传递给OpenCL内核。OpenCL内核将使用返回的句柄对目标平台RAM中分配的缓冲区执行总线主控读取和/或写入操作。主机不直接对目标平台RAM写入。在没有空闲块的情况下，该函数应返回-1。

●void xclFreeDeviceBuffer(xclDeviceHandle handle，uint64_t buf)

操作“xclFreeDeviceBuffer”释放先前由xclAllocDeviceBuffer分配的内存。释放的内存稍后可能会因为对xclAllocDeviceBuffer的另一次调用而被重用。传递以前未被xclAllocDeviceBuffer分配的缓冲区句柄将导致错误状况。

●size_t xclCopyBufferHost2Device(xclDeviceHandle handle，uint64_tdest，const void*src，size_t size，size_t seek)

操作“xclCopyBufferHost2Device”将主机缓冲区的内容复制到驻留在目标平台上的目标缓冲区中。src单元是指主机缓冲区指针，dest是指设备缓冲区句柄。传递先前未由xclAllocDeviceBuffer分配的dest句柄将导致错误。seek单元指定dest句柄中的偏移量。在size加上seek大于先前分配的设备缓冲区的尺寸时传递size将导致错误。在提供的示例中，PCIe DMA用于迁移缓冲区。

●size_t xclCopyBufferDevice2Host(xclDeviceHandle handle，void*dest，uint64_t src，size_t size，size_t skip)

操作xclCopyBufferDevice2Host将内容从目标平台驻留缓冲区复制到主机缓冲区。srs单元指的是设备缓冲区句柄，dest单元指的是主机缓冲区指针。错误将导致传递先前未由xclAllocDeviceBuffer分配的src句柄。skip单元指定src句柄中的偏移量。在size加上skip大于先前分配的设备缓冲区的尺寸时传递size将导致错误。在提供的示例中，PCIe DMA用于迁移缓冲区。

●size_t xclWrite(xclDeviceHandle handle，xclAddressSpace space，uint64_t offset，const void*hostBuf，size_t size)

操作xclWrite将主机缓冲区hostBuf的内容复制到目标平台地址映射中的特定位置。所述的hostBuf的内容用于对目标平台的外设进行编程。例如，在主机中执行的OpenCL运行时使用此操作将参数发送到可编程IC内的内核。所述偏移量与地址空间相关。

●size_t xclRead(xclDeviceHandle handle，xclAddressSpace space，uint64_t offset，void*hostbuf，size_t size)

操作xclRead将数据从目标平台地址映射中的特定位置复制到主机缓冲区hostBuf。该操作用于读取目标平台外围设备的状态。例如，OpenCL运行时程序库使用此操作来确定内核是否已完成运行。所述偏移量与地址空间相关。

主机的操作系统需要内核DMA驱动程序以与目标平台进行通信。在一个方面，公共低级驱动程序API可以分层在内核DMA驱动器的顶部，以使OpenCL运行时与驱动程序的细节隔离。所述驱动程序应是多线程安全的。OpenCL运行时可选地使用多于一个线程同时读取和写入设备。

在另一个方面，驱动程序API可以包括中断功能。例如，内核的寄存器映射可以包括一个或多个存储器位置，其中内核可能存储标志在所述位置。在寄存器映射的指定存储器位置检测到标志可能会导致静态区域通过作为驱动程序API一部分提供的功能来触发对主机的中断。

尽管上面示出的示例针对OpenCL实现，但是应当理解，可以使用任何异构型多处理器计算语言，并且可以相应地调整作为API的一部分描述的各种操作。

图16是示出内核执行的示例性方法1600的流程图。方法1600开始于内核在可编程IC内实现并且目标平台与主机通信地链接的状态。例如，主机可以包括一个或多个二进制容器或者可以访问一个或多个二进制容器。主机访问二进制容器，并将配置比特流文件从二进制容器提供给IC，例如到RAM 345。主机可以启动作为驱动程序API的部分而被描述的配置比特流加载操作，从而使IC加载配置比特流并实现由配置比特流指定的内核。如所讨论的，主机可以导致一个或多个不同的配置比特流(无论是部分的还是完整的)在运行时的不同时间被加载以实现IC内的一个或多个不同的内核。

在框1605中，主机应用程序被初始化。所述主机应用程序包括如图8所示的异构型多处理器运行时程序库。在框1610中，主机应用程序在主机存储器中分配缓冲区。在框1615中，主机应用程序发起传送，将缓冲器内容从主机存储器发送到目标平台存储器。

在框1620中，主机应用程序通过从接口向内核发信号开始运行。在框1625中，主机应用程序可选地开始轮询目标平台以监视完成的信号(done signal)。在框1630中，所述内核(即内核的硬件实现)执行或开始运行。所述内核加载并存储来自目标平台存储器的数据。在框1635中，内核响应于完成处理而将存储器映射寄存器中的状态改变为完成，或产生中断。在框1640中，主机应用程序将目标平台存储器的更新的缓冲器内容(即结果)传送到主机存储器。在框1645中，异构型多处理器运行时(例如在主机或主机应用程序中执行的)从主机存储器读取缓冲器。

为了说明的目的，阐述了具体的术语的含义以提供对本文公开的各种发明构思的透彻理解。然而，这里使用的术语仅用于描述本发明的具体方面，而不是限制性的。

如在本公开内容中所定义的，术语“一(个)”和“一(种)”表示一个或多于一个。如本文所定义的术语“多个”是指两个或多于两个。如本文所定义的术语“另一个”是指至少第二个或更多个。除非另有说明，本文所定义的术语“耦合”是指连接，无论是直接地没有任何中间元件，还是间接地连接一个或多个中间元件。两个元件也可以通过通信信道、路径、网络或系统机械地、电气地或通信地耦合。

如本文所定义，术语“自动”意味着没有用户干预。如本文所定义，术语“用户”是指人。本文定义的术语“和/或”是指一个或多个相关列出的项目的任何和所有可能的组合。当在本公开中使用时，术语“包括”和/或“包含”是指所述特征，整数，步骤，操作，元件和/或组件的存在，但不排除存在或添加一个或多个其他特征，整数，步骤，操作，元件，组件和/或其组合。尽管这里可以使用术语“第一”，“第二”等来描述各种元素，但是这些元素不应该被这些术语限制，因为这些术语仅用于将一个元素与另一个元素区分开，除非上下文另有说明。

如本文所定义，术语“如果”是指“当的时候”，“一旦”，“响应于确定”，“响应于检测”，“应答于确定”或“应答于检测”，取决于上下文语境。类似地，短语“如果被确定”或短语“如果(所述情况或事件)被检测到”，如本文所定义的，是指“一旦确定”，“响应于确定”，“应答于确定”，“一旦检测到(所述情况或事件)”，“响应于检测到(所述情况或事件)”或“应答于检测到(所述情况或事件)”，根据上下文语境。

在本公开内容中，相同的附图标记用于表示终端、信号线、导线及其对应的信号。在这方面，在本公开内容中，术语“信号”、“线”、“连接”、“端子”和“引脚”可能有时互换使用。还应当理解，术语“信号”、“线”等可以表示一个或多个信号，例如通过单根导线传送单个位或通过多个并行导线传送多个并行位。此外，每个线或信号可以表示由信号或线连接的两个或更多个组件之间的双向通信(视情况而定)。

本公开内容描述的一个或多个方面可以硬件或硬件和软件的组合来实现。一个或多个方面可以在一个系统中以集中方式实现，或者以分布式方式实现，其中不同的元素分布在多个互连的系统中。任何种类的数据处理系统或适用于执行本文描述的方法的至少一部分的其他设备都是适合的。

一个或多个方面还可以嵌入到计算机程序产品中，该计算机程序产品包括能够实现本文描述的方法的所有特征。计算机程序产品包括计算机可读数据存储介质。如本文所定义，短语“计算机可读存储介质”是指包含或存储由指令执行系统，装置或设备使用或与指令执行系统，设备或设备结合使用的程序代码的存储介质。如本文所定义的，“计算机可读存储介质”是非暂态的，因此本身不是暂时传播信号。计算机可读存储介质的示例可以包括但不限于光介质，磁介质，磁光介质，诸如RAM的计算机存储器，大容量存储设备(例如，硬盘)等。

附图中的流程图和框图示出了根据本文公开的本发明设备的各个方面的系统，方法和计算机程序产品的可能实现的架构，功能和操作。在这方面，流程图或框图中的每个块可以表示代码的模块、段或部分，其包括用于实现指定功能的一个或多个可执行指令。还将注意到，框图和/或流程图的每个块和方框图和/或流程图说明中的块的组合可以由执行指定功能或动作的专用的基于硬件的系统、或专用硬件和计算机指令的组合来实现。

在一个方面，流程图示例中的块可以以对应于各个块中数字的增加的数字顺序来执行。在其他方面，块可以以与块中的数字不同或变化的顺序执行。例如，可以基本上同时执行连续示出的两个或更多个块。在其他情况下，有时可能以相反的顺序执行两个或多个块，这取决于所涉及的功能。在其他情况下，可以以不同的顺序执行一个或多个块，结果被存储并用于后续或其他不立即跟随的块中。

在本文中，术语“计算机程序”，“软件”，“应用程序”，“计算机可用程序代码”，“程序代码”，“可执行代码”，衍生物和/或其组合意味着旨在使数据处理系统直接或以下面的任一或两者执行特定功能的一组指令的任何语言，代码或符号：a)转换为另一种语言，代码或符号；b)以不同的物质形式复制。例如，程序代码可以包括但不限于子例程，函数，过程，对象方法，对象实现，可执行应用，小应用程序，小服务程序，源代码，目标代码，共享库/动态加载库和/或设计用于在计算机系统上执行的其他指令序列。

因此，在本公开中，使用诸如“处理”或“运算”或“计算”或“确定”或“显示”等术语的陈述指的是数据处理系统的动作和处理，例如计算机系统或类似的电子计算设备，其将表示为计算机系统的寄存器和/或存储器内的物理(电子)量的数据操作和变换为类似地表示为计算机系统存储器和/或寄存器或其他此类信息存储，传输或显示设备中作为物理量的其他数据。

所附权利要求书中的所有方法或步骤加功能元件的相应结构，物质，作用和等同物旨在包括与特别要求保护的其它要求保护的元件相组合地执行该功能的任何结构，物质或作用。

一种方法包括：使用处理器来生成异构型多处理器设计的第一内核的RTL描述；将第一内核的RTL描述与提供可编程IC内的静态区域的基础平台电路设计集成，所述可编程IC向所述异构型多处理器设计的主机提供接口；从第一内核的RTL描述并使用所述处理器生成指定第一内核的硬件实现的第一配置比特流以及支持配置比特流的数据。该方法还包括在二进制容器内包含所述第一配置比特流和支持数据。

在一个示例中，所述异构型多处理器设计是OpenCL设计，所述第一内核用OpenCL指定。

在一个方面，所述支持数据包括在所述可编程IC内实现的第一内核的硬件实现的二维位置。

该方法可以包括加载所述第一内核的配置比特流，该配置比特流在异构型多处理器设计的运行时在可编程IC内创建第一内核的硬件实现的一个实例。

该方法还可以包括加载所述第一内核的配置比特流，该配置比特流在异构型多处理器设计的运行时在可编程IC内创建第一内核的硬件实现的多个实例。

该方法还可以包括将指定异构型多处理器设计的第二内核的硬件实现的第二配置比特流包括在第二二进制容器内。所述第二内核的硬件实现的至少一个实例可以在可编程IC内创建。

在一个方面，生成第一配置比特流可以包括生成第一配置比特流作为指定内核电路的部分配置比特流。在另一个方面，生成第一配置比特流可以包括生成第一配置比特流作为指定内核电路和基础平台电路的完整配置比特流。

一种方法包括：使用处理器来生成异构型多处理器设计的第一内核的RTL描述；将第一内核的RTL描述与基础平台电路设计集成，该基础平台电路设计提供可编程IC内的静态区域给异构型多处理器设计的主机；使用所述处理器并从第一内核的RTL描述生成第一内核的RTL描述的支持数据。该方法还包括在二进制容器内包含所述第一内核的RTL描述和支持数据。

在一个示例中，所述异构型多处理器设计是OpenCL设计，第一内核用OpenCL指定。

该方法可以包括生成内核的可执行版本并且在二进制容器内包括所述内核的可执行版本。

该方法还可以包括，主机在运行时将RTL描述从二进制容器提供给RTL模拟器并且在RTL模拟器中模拟内核的RTL描述。

该方法还可以包括在第二二进制容器内包括异构型多处理器设计的第二内核的RTL描述。

一种系统，可以包括被编程以启动可执行操作的处理器。所述可执行操作包括，生成异构型多处理器设计的第一内核的RTL描述；将所述第一内核的RTL描述与提供可编程IC内的静态区域的基础平台电路设计集成，所述可编程IC向所述异构型多处理器设计的主机提供接口；使用所述处理器并从第一内核的RTL描述生成指定第一内核的硬件实现的第一配置比特流和配置比特流的支持数据。该方法还包括在二进制容器内包含所述第一配置比特流和支持数据。

在一个示例中，所述异构型多处理器设计是OpenCL设计，第一内核在OpenCL中指定。

所述支持数据可以包括在可编程IC内实现的第一内核的硬件实现的二维位置。

所述可执行操作可以包括加载所述第一内核的配置比特流，该配置比特流在异构型多处理器设计的运行期间创建在可编程IC内的第一内核的硬件实现的一个实例。

所述可执行操作还可以包括加载所述第一内核的配置比特流，该配置比特流在异构型多处理器设计的运行期间创建在可编程IC内的第一内核的硬件实现的多个实例。

所述可执行操作还可以包括将指定异构型多处理器设计的第二内核的硬件实现的第二配置比特流包括在第二二进制容器内。所述可执行操作可以包括在可编程IC内创建第二内核的硬件实现的至少一个实例。

在一个方面，生成第一配置比特流可以包括生成第一配置比特流作为指定内核电路的部分配置比特流。在另一个方面，生成第一配置比特流可以包括生成第一配置比特流作为指定内核电路和基础平台电路的完整配置比特流。

在另一示例中，IC包括静态的并且在IC和主机处理器之间提供接口的第一区域。所述第一区域包括具有第一主接口的第一互连电路块和具有第一从接口的第二互连电路块。所述IC包括耦合到所述第一区域的第二区域。所述第二区域实现异构型多处理器设计的内核，并且包括耦合到第一互连电路块的第一主接口并被配置为从主机处理器接收命令的从接口。所述第二区域还包括耦合第二互连电路块的第一从接口的主接口，其中所述第二区域的所述主接口是用于存储器控制器的主控。

在一个示例中，异构型多处理器设计是OpenCL设计。

在一个方面，第二区域可以在运行时动态地重新配置以在主机处理器的控制下实现不同的内核。在另一个方面，第二区域可以在运行时期间被动态地重新配置，以在主机处理器的控制下实现不同的内核，同时保持第一区域完整。

第一区域可以包括总线端点和耦合到总线端点的DMA控制器。所述第一区域可以包括耦合到第一互连电路块的从接口的主接口。所述第一互连电路块可以包括第二主接口。第二互连电路块可以包括耦合到第一互连电路块的第二主接口的第二从接口。

第一区域还可以包括存储器控制器。所述存储器控制器可以包括耦合到第二互连电路块的主接口的从接口。

第一互连电路块和第二互连电路块可以被实现为AXI互连电路块。

IC还可以耦合到主机处理器，用异构型多处理器设计的主机程序代码进行编程。

DMA控制器可以被配置为存储器控制器的主控。

第一区域可以被配置为通过第一互连电路块向第二区域提供时钟信号和复位信号。

第二区域可以包括耦合到第一互连电路块的存储器映射寄存器。

在一个方面，第二区域包括第三互连电路块，其具有耦合到第一互连电路块的第一主接口的第二区域的从接口和耦合到第一内核电路块610-1的输入的主接口。第二区域还可以包括第四互连电路块，其具有耦合到第一内核电路块610-1的输出的从接口和耦合到第二接口电路块的第一从接口的主接口。

第二区域还可以包括具有耦合到第三互连电路块的主接口的输入和耦合到第四互连电路块的从接口的输出的第二内核电路块。

在另一示例中，一种方法包括在IC内提供静态的实现IC和主机处理器之间的接口的第一区域，在第一区域内包含具有第一主接口的第一互连电路块和具有第一从接口的第二互连电路块，并且提供耦合到第一区域的第二区域。该方法还可以包括在第二区域内实现异构型多处理器设计的内核，并且在第二区域内包含耦合到第一互连电路块的第一主接口的从接口。所述内核被配置为从主机处理器接收命令。该方法还包括在第二区域内包含耦合第二互连电路块的第一从接口的主接口，其中第二区域的主接口是用于存储器控制器的主控。

在一个示例中，所述异构型多处理器设计是OpenCL设计。

在一个方面，所述方法可以包括在IC的运行时期间动态地重新配置第二区域，以在主机处理器的控制下实现不同的内核。在另一个方面，该方法可以包括在运行时动态地重新配置第二区域，以在主机处理器的控制下实现不同的内核，同时保持第一区域完整。

该方法可以包括在第一区域内提供总线端点并且在第一区域内提供耦合到总线端点的DMA控制器，并且包括耦合到第一互连电路块的从接口的主接口。第一互连电路块可以包括第二主接口。第二互连电路块可以包括耦合到第一互连电路块的第二主接口的第二从接口。

该方法还可以包括在第一区域内提供具有耦合到第二互连电路块的主接口的从接口的存储器控制器。

该方法还可以包括向主机处理器提供异构型多处理器设计的主机程序代码。

该方法还可以包括在第二区域内提供第三互连电路块，其具有耦合到第一互连电路块的第一主接口的第二区域的从接口和耦合到第一内核电路块的输入的主接口。可以在第二区域内提供第四互连电路块。第四互连电路块可以包括耦合到第一内核电路块的输出的从接口和耦合到第二接口电路块的第一从接口的主接口。

该方法还可以包括在第二区域内提供第二内核电路块，其具有耦合到第三互连电路块的主接口的输入和耦合到第四互连电路块的从接口的输出。

在不脱离本公开的精神或基本属性的情况下，本公开内容中描述的特征可以以其他形式实施。因此，应参考所附的权利要求而不是上述公开内容，来确定这些特征和实施方式的范围。

Claims (15)

1.一种方法，其特征在于，包括：

使用处理器从第一内核的高级编程语言描述生成所述第一内核的寄存器传输级描述，其中所述第一内核是异构型多处理器设计的一部分；

将所述第一内核的寄存器传输级描述与在可编程集成电路内提供静态区域的基础平台电路设计集成，所述可编程集成电路向所述异构型多处理器设计的主机提供接口，并向所述可编程集成电路内的可动态重新配置的内核区域提供接口；

使用所述处理器并从所述第一内核的寄存器传输级描述，生成指定所述第一内核的硬件实现的第一配置比特流和包含所述第一内核的寄存器映射的所述配置比特流的支持数据，其中所述第一内核的硬件实现适于在所述内核区域内实现；并且

生成多个文件，其中所述多个文件包括包含所述第一配置比特流的第一文件和包含所述支持数据的第二文件，其中所述第二文件的寄存器映射可由所述主机用来与在所述可编程集成电路内实现的第一内核进行通信。

2.根据权利要求1所述的方法，其特征在于，所述内核区域在运行时期间可动态地重新配置以实现不同内核的硬件实现，同时保持静态区域完整。

3.根据权利要求1所述的方法，其特征在于，还包括：

加载所述第一内核的配置比特流，所述第一内核的配置比特流在所述异构型多处理器设计的运行时期间在所述可编程集成电路内创建所述第一内核的硬件实现的实例，其中所述主机将所述配置比特流从所述多个文件提供给所述可编程集成电路。

4.根据权利要求2所述的方法，其特征在于，所述静态区域包括：

耦合到所述内核区域并且被配置为向所述可编程集成电路的内核区域提供时钟信号的第一互连电路块；

耦合到所述第一互连电路块的总线直接存储器存取控制器；和

耦合到所述总线直接存储器存取控制器的总线端点。

5.根据权利要求4所述的方法，其特征在于，所述静态区域包括将所述内核区域耦合到存储器控制器的第二互连电路块。

6.根据权利要求1所述的方法，其特征在于，生成所述第一配置比特流包括：

生成所述第一配置比特流作为指定内核电路的部分配置比特流。

7.根据权利要求1所述的方法，其特征在于，生成所述第一配置比特流包括：

生成所述第一配置比特流作为指定内核电路和基础平台电路的完整配置比特流。

8.一种系统，其特征在于，包括：

被编程以启动可执行操作的处理器，所述可执行操作包括：

从第一内核的高级编程语言描述生成所述第一内核的寄存器传输级描述，其中所述第一内核是异构型多处理器设计的一部分；

将所述第一内核的寄存器传输级描述与在可编程集成电路内提供静态区域的基础平台电路设计集成，所述可编程集成电路向所述异构型多处理器设计的主机提供接口，并向所述可编程集成电路内的可动态重新配置的内核区域提供接口；

从所述第一内核的寄存器传输级描述，生成指定所述第一内核的硬件实现的第一配置比特流和包含第一内核的寄存器映射的所述配置比特流的支持数据，其中所述第一内核的硬件实现适于在所述内核区域内实现；以及

生成多个文件，其中所述多个文件包括包含所述第一配置比特流的第一文件和包含所述支持数据的第二文件，其中所述第二文件的寄存器映射可由所述主机用来与所述可编程集成电路内实现的第一内核进行通信。

9.根据权利要求8所述的系统，其特征在于，所述内核区域在运行时期间可动态地重新配置以实现不同内核的硬件实现，同时保持静态区域完整。

10.根据权利要求8所述的系统，其特征在于，所述可执行操作还包括：

加载所述第一内核的配置比特流，所述第一内核的配置比特流在所述异构型多处理器设计的运行时期间在所述可编程集成电路内创建所述第一内核的硬件实现的实例，其中所述主机将所述配置比特流从所述多个文件提供给所述可编程集成电路。

11.根据权利要求9所述的系统，其特征在于，所述可执行操作还包括：

加载所述第一内核的配置比特流，所述第一内核的配置比特流在所述异构型多处理器设计的运行时期间在所述可编程集成电路内创建所述第一内核的硬件实现的多个实例，其中所述主机将所述配置比特流从所述多个文件提供给所述可编程集成电路。

12.根据权利要求9所述的系统，其特征在于，所述静态区域包括：

耦合到所述内核区域并且被配置为给所述可编程集成电路的内核区域提供时钟信号的第一互连电路块；

耦合到所述第一互连电路块的总线直接存储器存取控制器；和

耦合到所述总线直接存储器存取控制器的总线端点。

13.根据权利要求12所述的系统，其特征在于，所述静态区域包括将所述内核区域耦合到存储器控制器的第二互连电路块。

14.根据权利要求8所述的系统，其特征在于，生成所述第一配置比特流包括：

生成所述第一配置比特流作为指定内核电路的部分配置比特流。

15.根据权利要求8所述的系统，其特征在于，生成所述第一配置比特流包括：

生成所述第一配置比特流作为指定内核电路和基础平台电路的完整配置比特流。

Images