DE102020127705A1 - Techniken für einen effizienten fabric-attached-speicher - Google Patents

Techniken für einen effizienten fabric-attached-speicher Download PDF

Info

Publication number
DE102020127705A1
DE102020127705A1 DE102020127705.9A DE102020127705A DE102020127705A1 DE 102020127705 A1 DE102020127705 A1 DE 102020127705A1 DE 102020127705 A DE102020127705 A DE 102020127705A DE 102020127705 A1 DE102020127705 A1 DE 102020127705A1
Authority
DE
Germany
Prior art keywords
fabric
memory
gpu
address
attached
Prior art date
Legal status (The legal status is an assumption and is not a legal conclusion. Google has not performed a legal analysis and makes no representation as to the accuracy of the status listed.)
Pending
Application number
DE102020127705.9A
Other languages
English (en)
Inventor
John Feehrer
Denis Foley
Mark Hummel
Vyas Venkataraman
Ram Gummadi
Samuel H. Duncan
Glenn Dearth
Brian Kellehrer
Current Assignee (The listed assignees may be inaccurate. Google has not performed a legal analysis and makes no representation or warranty as to the accuracy of the list.)
Nvidia Corp
Original Assignee
Nvidia Corp
Priority date (The priority date is an assumption and is not a legal conclusion. Google has not performed a legal analysis and makes no representation as to the accuracy of the date listed.)
Filing date
Publication date
Application filed by Nvidia Corp filed Critical Nvidia Corp
Publication of DE102020127705A1 publication Critical patent/DE102020127705A1/de
Pending legal-status Critical Current

Links

Images

Classifications

    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06T—IMAGE DATA PROCESSING OR GENERATION, IN GENERAL
    • G06T1/00—General purpose image data processing
    • G06T1/20—Processor architectures; Processor configuration, e.g. pipelining
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06F—ELECTRIC DIGITAL DATA PROCESSING
    • G06F13/00—Interconnection of, or transfer of information or other signals between, memories, input/output devices or central processing units
    • G06F13/14—Handling requests for interconnection or transfer
    • G06F13/16—Handling requests for interconnection or transfer for access to memory bus
    • G06F13/1668—Details of memory controller
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06F—ELECTRIC DIGITAL DATA PROCESSING
    • G06F12/00—Accessing, addressing or allocating within memory systems or architectures
    • G06F12/02—Addressing or allocation; Relocation
    • G06F12/06—Addressing a physical block of locations, e.g. base addressing, module addressing, memory dedication
    • G06F12/0607—Interleaved addressing
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06F—ELECTRIC DIGITAL DATA PROCESSING
    • G06F13/00—Interconnection of, or transfer of information or other signals between, memories, input/output devices or central processing units
    • G06F13/14—Handling requests for interconnection or transfer
    • G06F13/16—Handling requests for interconnection or transfer for access to memory bus
    • G06F13/1605—Handling requests for interconnection or transfer for access to memory bus based on arbitration
    • G06F13/1652—Handling requests for interconnection or transfer for access to memory bus based on arbitration in a multiprocessor architecture
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06F—ELECTRIC DIGITAL DATA PROCESSING
    • G06F13/00—Interconnection of, or transfer of information or other signals between, memories, input/output devices or central processing units
    • G06F13/38—Information transfer, e.g. on bus
    • G06F13/40—Bus structure
    • G06F13/4004—Coupling between buses
    • G06F13/4022—Coupling between buses using switching circuits, e.g. switching matrix, connection or expansion network
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06F—ELECTRIC DIGITAL DATA PROCESSING
    • G06F17/00—Digital computing or data processing equipment or methods, specially adapted for specific functions
    • G06F17/10—Complex mathematical operations
    • G06F17/16—Matrix or vector computation, e.g. matrix-matrix or matrix-vector multiplication, matrix factorization
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06F—ELECTRIC DIGITAL DATA PROCESSING
    • G06F9/00—Arrangements for program control, e.g. control units
    • G06F9/06—Arrangements for program control, e.g. control units using stored programs, i.e. using an internal store of processing equipment to receive or retain programs
    • G06F9/44—Arrangements for executing specific programs
    • G06F9/455—Emulation; Interpretation; Software simulation, e.g. virtualisation or emulation of application or operating system execution engines
    • G06F9/45533—Hypervisors; Virtual machine monitors
    • G06F9/45558—Hypervisor-specific management and integration aspects
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06N—COMPUTING ARRANGEMENTS BASED ON SPECIFIC COMPUTATIONAL MODELS
    • G06N20/00—Machine learning
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06T—IMAGE DATA PROCESSING OR GENERATION, IN GENERAL
    • G06T1/00—General purpose image data processing
    • G06T1/60—Memory management
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06F—ELECTRIC DIGITAL DATA PROCESSING
    • G06F12/00—Accessing, addressing or allocating within memory systems or architectures
    • G06F12/02—Addressing or allocation; Relocation
    • G06F12/08—Addressing or allocation; Relocation in hierarchically structured memory systems, e.g. virtual memory systems
    • G06F12/10—Address translation
    • G06F12/1027—Address translation using associative or pseudo-associative address translation means, e.g. translation look-aside buffer [TLB]
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06F—ELECTRIC DIGITAL DATA PROCESSING
    • G06F12/00—Accessing, addressing or allocating within memory systems or architectures
    • G06F12/02—Addressing or allocation; Relocation
    • G06F12/08—Addressing or allocation; Relocation in hierarchically structured memory systems, e.g. virtual memory systems
    • G06F12/10—Address translation
    • G06F12/1027—Address translation using associative or pseudo-associative address translation means, e.g. translation look-aside buffer [TLB]
    • G06F12/1036—Address translation using associative or pseudo-associative address translation means, e.g. translation look-aside buffer [TLB] for multiple virtual address spaces, e.g. segmentation
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06F—ELECTRIC DIGITAL DATA PROCESSING
    • G06F12/00—Accessing, addressing or allocating within memory systems or architectures
    • G06F12/14—Protection against unauthorised use of memory or access to memory
    • G06F12/1408—Protection against unauthorised use of memory or access to memory by using cryptography
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06F—ELECTRIC DIGITAL DATA PROCESSING
    • G06F9/00—Arrangements for program control, e.g. control units
    • G06F9/06—Arrangements for program control, e.g. control units using stored programs, i.e. using an internal store of processing equipment to receive or retain programs
    • G06F9/44—Arrangements for executing specific programs
    • G06F9/455—Emulation; Interpretation; Software simulation, e.g. virtualisation or emulation of application or operating system execution engines
    • G06F9/45533—Hypervisors; Virtual machine monitors
    • G06F9/45558—Hypervisor-specific management and integration aspects
    • G06F2009/45583—Memory management, e.g. access or allocation
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06F—ELECTRIC DIGITAL DATA PROCESSING
    • G06F2212/00—Indexing scheme relating to accessing, addressing or allocation within memory systems or architectures
    • G06F2212/15—Use in a specific computing environment
    • G06F2212/152—Virtualized environment, e.g. logically partitioned system
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06F—ELECTRIC DIGITAL DATA PROCESSING
    • G06F2212/00—Indexing scheme relating to accessing, addressing or allocation within memory systems or architectures
    • G06F2212/30—Providing cache or TLB in specific location of a processing system
    • G06F2212/302—In image processor or graphics adapter
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06F—ELECTRIC DIGITAL DATA PROCESSING
    • G06F2212/00—Indexing scheme relating to accessing, addressing or allocation within memory systems or architectures
    • G06F2212/45—Caching of specific data in cache memory
    • G06F2212/454—Vector or matrix data
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06F—ELECTRIC DIGITAL DATA PROCESSING
    • G06F2212/00—Indexing scheme relating to accessing, addressing or allocation within memory systems or architectures
    • G06F2212/65—Details of virtual memory and virtual address translation
    • G06F2212/651—Multi-level translation tables
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06F—ELECTRIC DIGITAL DATA PROCESSING
    • G06F2212/00—Indexing scheme relating to accessing, addressing or allocation within memory systems or architectures
    • G06F2212/65—Details of virtual memory and virtual address translation
    • G06F2212/657—Virtual address space management
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06N—COMPUTING ARRANGEMENTS BASED ON SPECIFIC COMPUTATIONAL MODELS
    • G06N3/00—Computing arrangements based on biological models
    • G06N3/02—Neural networks
    • G06N3/04—Architecture, e.g. interconnection topology
    • G06N3/045—Combinations of networks
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06N—COMPUTING ARRANGEMENTS BASED ON SPECIFIC COMPUTATIONAL MODELS
    • G06N3/00—Computing arrangements based on biological models
    • G06N3/02—Neural networks
    • G06N3/06—Physical realisation, i.e. hardware implementation of neural networks, neurons or parts of neurons
    • G06N3/063—Physical realisation, i.e. hardware implementation of neural networks, neurons or parts of neurons using electronic means
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06N—COMPUTING ARRANGEMENTS BASED ON SPECIFIC COMPUTATIONAL MODELS
    • G06N3/00—Computing arrangements based on biological models
    • G06N3/02—Neural networks
    • G06N3/08—Learning methods
    • G—PHYSICS
    • G06—COMPUTING; CALCULATING OR COUNTING
    • G06T—IMAGE DATA PROCESSING OR GENERATION, IN GENERAL
    • G06T2200/00—Indexing scheme for image data processing or generation, in general
    • G06T2200/28—Indexing scheme for image data processing or generation, in general involving image processing hardware

Landscapes

  • Engineering & Computer Science (AREA)
  • Theoretical Computer Science (AREA)
  • Physics & Mathematics (AREA)
  • General Physics & Mathematics (AREA)
  • General Engineering & Computer Science (AREA)
  • Software Systems (AREA)
  • Mathematical Physics (AREA)
  • Data Mining & Analysis (AREA)
  • Mathematical Optimization (AREA)
  • Computational Mathematics (AREA)
  • Pure & Applied Mathematics (AREA)
  • Mathematical Analysis (AREA)
  • Computing Systems (AREA)
  • Computer Hardware Design (AREA)
  • Medical Informatics (AREA)
  • Evolutionary Computation (AREA)
  • Computer Vision & Pattern Recognition (AREA)
  • Artificial Intelligence (AREA)
  • Algebra (AREA)
  • Databases & Information Systems (AREA)
  • Multi Processors (AREA)
  • Memory System Of A Hierarchy Structure (AREA)
  • Storage Device Security (AREA)
  • Techniques For Improving Reliability Of Storages (AREA)

Abstract

Fabric-Attached-Speicher (FAM) stellt einen Speicherpool bereit, auf den ein oder mehrere Prozessoren, wie beispielsweise eine oder mehrere Grafikverarbeitungseinheiten (GPUs), über ein Netzwerk-Fabric zugreifen können. In einem Beispiel wird eine Technik offenbart zur Verwendung von unvollkommenen Prozessoren als Speichercontroller, um einen Zugriff auf Speicher, der für die unvollkommenen Prozessoren lokal ist, durch andere Prozessoren als Fabric-Attached-Speicher zu ermöglichen. In einem anderen Beispiel wird eine Speicheradressenverdichtung innerhalb der Fabric-Elemente verwendet, um den verfügbaren Speicherplatz voll auszunutzen.

Description

  • GEBIET
  • Diese Technologie bezieht sich auf Fabric-Attached-Speicher („FAM“) und insbesondere auf Fabric-Attached-Speicher, der Adressverdichtung über schnelle Daten-Interconnects verwendet. Ein weiterer Aspekt dieser Technologie bezieht sich auf die Verwendung und/oder Wiederverwendung von Grafikverarbeitungseinheiten (GPUs) mit reduzierten Fähigkeiten als kostengünstige Fabric-Attached-Speicher-Controller, die in der Lage sind, atomische Funktionen und/oder andere Speicherbefehle nativ zu verarbeiten.
  • HINTERGRUND & KURZBESCHREIBUNG
  • Die Menge an Daten, die Computer verwalten und verarbeiten müssen, ist explosionsartig angestiegen. Soziale Medien, künstliche Intelligenz und das Internet der Dinge haben alle einen Bedarf geschaffen, riesige Datenmengen zu speichern und schnell zu verarbeiten.
  • Der Trend in der modernen Datenverarbeitung war es, hochleistungsfähige, massiv parallel verarbeitende Systeme einzusetzen und so große Berechnungsaufgaben in viele kleinere Aufgaben aufzubrechen, die gleichzeitig ausgeführt werden können. Da solche parallelen Verarbeitungsarchitekturen weit verbreitet geworden sind, hat dies wiederum einen Bedarf an Hochleistungsspeicher mit großer Kapazität und geringer Latenz geschaffen, der große Datenmengen speichern und Parallelprozessoren einen schnellen Zugriff bereitstellen kann.
  • Speicher mit hoher Bandbreite (high bandwidth memory, HBM), der direkt mit GPUs oder anderen Parallelprozessoren verbunden ist, stellt eine hohe Zugriffsbandbreite bei geringer Latenz bereit, aber seine Kapazität kann relativ begrenzt sein und daher für massiv parallele Arbeitslasten mit sehr hohen Anforderungen an die Speicherkapazität und/oder die Bandbreite nicht ausreichen. Wenn ein Kunde in der Vergangenheit Speicherkapazität mit hoher Leistung und niedriger Latenz erhöhen wollte, musste der Kunde mehr GPUs kaufen, um den typischerweise mit jeder GPU gebündelten Hochleistungsspeicher in die GPU-Systemstruktur bzw. das GPU-System-Fabric zu integrieren. Die Bereitstellung von mehr GPUs als für die Rechenfunktionen notwendig kann jedoch kostspielig sein und erhöht den Energiebedarf. Einige Kunden sind daher möglicherweise an einem größeren Speichergrundriss mit weniger GPUs interessiert.
  • Eine Alternative bestand darin, Systemspeicher (SYSMEM) zu verwenden - d. h. Speicher, der an die zentrale(n) Verarbeitungseinheit(en) (CPUs) des Systems angeschlossen ist. Moderne Rechenarchitekturen können auch GPUs mit Zugriff auf große Mengen an nichtflüchtigem Systemspeicher über z.B. NVMe (Non-Volatile Memory express)-Laufwerke und PCIe (Peripheral Component Interconnect express)-Peer-to-Peer-Zugriff bereitstellen. Ein Problem bei der Verwendung von Systemspeicher oder nichtflüchtigem Speicher, der als PCIe-Gerät angeschlossen ist, ist jedoch, dass die Zugriffsbandbreite in vielen Architekturen durch solche PCIe- oder andere relativ langsame Daten-Interconnects begrenzt ist. Je nach Interconnect zwischen CPU und GPU ist die Speichermodell-Semantik der GPU möglicherweise nicht mit den gleichen Leistungsmerkmalen über die Verbindung abbildbar. Infolgedessen müssen Anwendungen möglicherweise ein alternatives Programmiermodell verwenden, anstatt den Speicher mit der GPU-Speicher-Semantik zu behandeln. Diese Art des Zugriffs kann auch ein Block-Eingabe/Ausgabe (E/A)-Programmiermodell erzwingen (im Gegensatz zu z.B. Wortadressierbarkeit), mit den damit verbundenen Overheads und Latenzverlusten.
  • Obwohl moderne Systemspeicherkapazität relativ reichlich zu sein scheint, stoßen darüber hinaus einige massiv-parallel verarbeitende Systeme in Bezug auf Speicherkapazität an ihre Grenzen. Die Systemspeicherkapazität ist im Allgemeinen auf der Grundlage des maximalen Adressraums der verwendeten CPU(s) begrenzt. Viele moderne CPUs können zum Beispiel nicht auf mehr als etwa drei Terabyte (TB) zugreifen. Diese Kapazität (drei Millionen Millionen Bytes) mag nach viel klingen, reicht aber für bestimmte massiv parallele GPU-Operationen wie beispielsweise Deep Learning, Datenanalyse, medizinische Bildgebung und Grafikverarbeitung möglicherweise nicht aus.
  • Aus der Software-Perspektive werden GPUs immer schneller und ermöglichen es Systemen, mehr Rechenoperationen in kürzerer Zeit durchzuführen. Höhere Rechenkapazitäten erfordern mehr Daten, was wiederum impliziert, dass es sinnvoll wäre, einen schnellen Zugriff auf mehr gespeicherte Daten bereitzustellen. Die Speicherbandbreite hat jedoch nicht so schnell skaliert wie die GPU-Rechenkapazitäten. Das bedeutet, dass es immer wichtiger wird, die GPUs - welche Datenverbraucher sind - vollständig mit Daten zu versorgen, mit denen sie arbeiten können.
  • Um dieses Problem zu lösen, hat NVIDIA ein Hochgeschwindigkeits-Datenverbindungs-Fabric namens NVLINK™ entwickelt, welches eine höhere Datenübertragungsgeschwindigkeit zwischen GPU-Rechenkomponenten bereitstellt. Fabric-Interconnect-Anordnungen wie NVLINK™und NVSWITCH™ ermöglichen es GPUs, als Peers über schnelle, hoch skalierbare Multiprozessor-Interconnects miteinander zu kommunizieren, die den Bandbreiten-Engpass bzw. Flaschenhals langsamerer Arten von Datenverbindungen vermeiden. Dadurch kann eine GPU auf den lokalen Speicher einer anderen GPU zugreifen, als wäre es ihr eigener, welches es dem Entwickler ermöglicht, die Speicherressourcen mehrerer GPUs zu bündeln. Siehe zum Beispiel USP7275123 , USP7,627,723 und USP7,451,259 . Das NVLINK™-Konstrukt ist langsamer als die Bandbreite lokalen On-Chip-Speichers, aber immer noch viel schneller als PCIe oder andere derartige Datenverbindungen, die häufig für den Zugriff auf Hauptsystemspeicher oder andere an das PCIe-Fabric angeschlossene Speichervorrichtungen verwendet werden.
  • Fabric-Attached-Speicher bzw. Fabric Attached Memory („FAM“) wurde bereits als ein Konzept zur Disaggregierung bzw. Entkopplung von Speicher von Rechenressourcen definiert, wodurch die Speicherkapazität unabhängig von der Rechenkapazität wachsen kann. FAM wurde zum Beispiel von Anbietern von Rechenzentrumsinfrastrukturen wie beispielsweise Hewlett Packard Enterprise (HPE) über Industriestandards wie beispielsweise Gen-Z eingesetzt. So hat HPE zum Beispiel kürzlich eine speicherzentrierte „Maschine“ angekündigt, die das Gen-Z Open Standard Memory Interconnect Fabric nutzt. Siehe zum Beispiel https://genzconsortium.org/wpcontent/uploads/2018/05/20170303_Gen-Z-DRAM-and-SCM-Overview.pdf. Achermann et al, „Separating Translation from Protection in Address Spaces with Dynamic Remapping", Proceedings of the 16th Workshop on Hot Topics in Operating Systems Pages 118-124 (Whistler, BC, Canada, May 07 - 10, 2017); und Chen, Fei et al, „Billion node graph inference: iterative processing on The Machine" (Hewlett Packard Labs HPE-2016-101, 2016). Trotz solcher früheren Arbeiten bleiben viele Herausforderungen in Bezug auf effiziente, kostengünstige FAM-Implementierungen mit hoher Kapazität bestehen.
  • Die Technologie hierin löst das Problem, wie GPU-Speicherkapazität auf sehr hohe Mengen (z.B. einige 10 bis einige 100 TB) und Bandbreiten (z.B. mehrere TB/s) für Multi-GPU-Systeme erhöht werden kann, ohne dass die Anzahl der GPUs und/oder CPUs erhöht werden muss. Fabric-Attached-Speicher bzw. Fabric Attached Memory ist eine Möglichkeit, die Stärke und den Wert einer Hochgeschwindigkeits-Datenverbindung mit hoher Bandbreite zwischen GPUs, wie z.B., aber nicht beschränkt auf, NVIDIA NVLINK™zu nutzen, um es einem Benutzer zu ermöglichen, die der GPU zugängliche Speicherkapazität zu erhöhen, ohne auch die GPU-Rechenkapazität erhöhen zu müssen.
  • Die beispielhaften, nicht beschränkenden Ausführungsformen ermöglichen es einem Benutzer, die Speicherkapazität und die GPU-Bandbreite zu erhöhen, ohne die GPU-Speicherberechnungsressourcen erhöhen zu müssen. Der Effekt eines solchen Fabric-Attached-Speichers ist die Entkopplung des Speichers in solchen Systemen von GPU-Rechenressourcen, wodurch die Speicherkapazität unabhängig von der GPU-Rechenkapazität wachsen kann. Einige GPU-Arbeitslasten haben sehr hohe Anforderungen an die Speicherkapazität und/oder die Bandbreite. Daher können manche Anwendungen von einem größeren Speichergrundriss, aber relativ gesehen weniger GPUs profitieren. Wie nachstehend im Einzelnen erläutert wird, ist es jedoch trotz einer solchen Entkopplung in vielen Anwendungen höchst wünschenswert, den Fabric-Attached-Speicher mit einigen GPU-ähnlichen Schnittstellenfähigkeiten auf kostengünstige Weise auszustatten - z.B. damit Fabric-Attached-Speicher GPU-basierte hardwarebeschleunigte Speicherzugriffsfunktionen wie beispielsweise „atomische“ Speicherzugriffsanforderungen implementieren kann und damit das Interconnect-Fabric anderweitig auf den Fabric-Attached-Speicher auf die gleiche Weise und unter Verwendung derselben Mechanismen zugreifen kann, die für den Zugriff auf direkt an die GPU angeschlossenen lokalen Speicher verfügbar sind. Wie nachstehend im Einzelnen beschrieben wird, stellt die beispielhafte, nicht beschränkende Technologie hierin diese und andere Fähigkeiten bereit.
  • Die hierin beispielhaft beschriebenen, nicht beschränkenden Technologien ermöglichen eine variable Größe des Fabric-Attached-Speichers und stellen Techniken für die Adresszuordnung und die Verteilung von Speicherzugriffsanforderungen bereit, um sicherzustellen, dass die Kapazität des Fabric-Attached-Speichers vollständig genutzt wird. Beispielsweise kann eine Anwendung, die auf einer „Source-GPU“ (d. h. einer Computervorrichtung, die auf den Fabric-Attached-Speicher zugreifen möchte) läuft, Adressen generieren, die einen potenziell großen Adressraum definieren, z.B. Hunderte von Terabyte (TB). In einigen nicht beschränkenden Ausführungsformen kann dieser Adressraum den eigenen lokal angeschlossenen Speicher der Source-GPU, die lokal angeschlossenen Speicher anderer GPUs und den Fabric-Attached-Speicher beinhalten oder darauf abgebildet werden. Indessen wird jedoch jede einzelne Fabric-Attached-Speichervorrichtung (d. h. eine Steuereinrichtung bzw. ein Controller wie beispielsweise eine GPU mit reduzierter Rechenkapazität oder ein kundenspezifisches ASIC und zugeordneter gebündelter flüchtiger oder nichtflüchtiger Halbleiter-Hochleistungsspeicher wie beispielsweise DIMM, welcher z.B. jede beliebigen interessierenden Speichertechnologien einschließlich z.B. DDR, GDDR, HBM, NVRAM, NVMe usw. beinhalten kann) im Allgemeinen einen Adressraum bereitstellen, der viel kleiner ist (z.B. in der Größenordnung von etwa 1, 2 oder 4 TB als einige Beispiele). Im Allgemeinen kann eine beliebige Anzahl solcher individueller Fabric-Attached-Speichervorrichtungen oder -module vorhanden sein, die an das Interconnect-Fabric angeschlossen sind, und kann der Endbenutzer weiteren Fabric-Attached-Speicher hinzufügen, soweit dies mit Kosten-Leistungs-Kompromissen und der Skalierbarkeit des Fabric bzw. der Verbindungsstruktur (d. h. der Anzahl der Links und Switches) vereinbar ist.
  • Ein Vorteil der beispielhaften, nicht beschränkenden Technologie ist, dass Endbenutzer die Kapazität des Fabric-Attached-Speichers bequem erweitern können, um eine bessere Leistung zu erzielen und Thrashing bzw. Überlastung zu reduzieren, ohne dass Softwareanwendungen neu geschrieben oder modifiziert werden müssen. Dementsprechend stellt die hier beschriebene beispielhafte, nicht beschränkende Technologie automatische Mechanismen bereit zur Verwendung von Entropie, um Speicherzugriffsanforderungen automatisch auf verfügbare Interconnect-Verbindungen und zugeordnete Fabric-Attached-Speichervorrichtungen zu verteilen, um Kommunikation und Speicher-/Zugriffslast auszubalancieren. Ferner ist es in beispielhaften nicht beschränkenden Ausführungsformen nicht erforderlich, dass jede Fabric-Attached-Speichervorrichtung mit allen verfügbaren GPU-Interconnect-Verbindungen verbunden ist - im Gegenteil, eine bestimmte Fabric-Attached-Speichervorrichtung kann mit einer relativ kleinen Teilmenge von Interconnect-Verbindungen verbunden sein - obwohl in manchen Anwendungen vorzugsweise ausreichend Fabric-Attached-Speicher bereitgestellt ist, damit die Source-GPU auf einen Fabric-Attached-Speicher über alle oder viele ihrer Verbindungen zugreifen kann. Dieses strukturelle Merkmal, das es einer Fabric-Attached-Speichervorrichtung ermöglicht, sich mit im beispielsweise Vergleich zu einer Rechen-GPU einem reduzierten Satz von Verbindungen mit dem Interconnect-Fabric zu verbinden, ist nützlich bei der Bereitstellung kostengünstiger Fabric-Attached-Speichermodule, schafft aber auch einige Adressierungs-, Routing- und Kapazitätsauslastungsmöglichkeiten, die die vorliegende beispielhafte Technologie ausnutzt.
  • Insbesondere stellen die beispielhaften, nicht beschränkenden Ausführungsformen Techniken und Mechanismen für die automatische Handhabung der Adresszuordnung und des Anforderungsroutings zwischen den von der GPU erzeugten physischen Adressen und den Adresspositionen des Fabric-Attached-Speichers bereit, so dass die Kapazität des Fabric-Attached-Speichers vollständig genutzt werden kann, obwohl die Source-GPU physische Adressen erzeugen kann, die Adressräume definieren, die viel größer sind als diejenigen jeder bestimmten Fabric-Attached-Speichervorrichtung, und obwohl die Source-GPU solche physischen Adressen über entropie-selektierte Interconnect-Verbindungen senden kann, während sie effizient und flexibel Daten-Striping über ein Array solcher Fabric-Attached-Speichervorrichtungen unterstützt.
  • Durch direkten Anschluss von Speicher an ein skalierbares Hochgeschwindigkeits-Fabric, das aus Hochgeschwindigkeits-Interprozess-Kommunikationsverbindungen wie beispielsweise NVLINK™und NVSWITCH™ von NVIDIA besteht, kann die Technologie hierin eine viel höhere Kapazität und Bandbreite als CPU-Speicher, auf den über PCIe zugegriffen wird, mehr Flexibilität und eine kostengünstigere Plattform für die Ausführung speicherintensiver Arbeitslasten bereitstellen. Der Speichergrundriss und die Leistung können daher von Rechenfähigkeiten „disaggregiert“ (entkoppelt) werden, und dieser FAM-Ansatz ermöglicht es GPUs, ihr Speichermodell durch Ausgeben von Loads, Stores und Atomics mit Adressierbarkeit auf Wortebene direkt an Fabric-Attached-Speicher mit entsprechenden Sichtbarkeits- und Ordnungsgarantien so zu erweitern, dass es FAM abdeckt. Dies ist besonders wertvoll für GPUs oder spezialisierte ASICs für Deep-Learning-Anwendungen.
  • Die Technologie hierin stellt ferner Verbesserungen für FAM bereit, die kostengünstige FAM-Module („FAMMs“) auf Basis von „Floor Swept“- und/oder GPUs mit geringeren Fähigkeiten bereitstellen. Wie vorstehend erörtert wurde, ist es in vielen Implementierungen wünschenswert, kostengünstig einen GPU-ähnlichen Peer-to-Peer-Zugriff auf Fabric-Attached-Speicher bereitzustellen. Ein nicht beschränkender Aspekt bestimmter Ausführungsformen der vorliegenden Technologie ist der Einsatz von GPUs der unteren Leistungsklasse, die andernfalls aufgrund einer Abfallquote des Produktionsertrags ausgemustert würden, als relativ einfache und stromsparende Speichercontroller, die als FAMM-Vorrichtungen arbeiten. Manche GPU-Architekturen beinhalten einen hochentwickelten Hochleistungs-Speichercontroller für den Zugriff auf seinen lokalen Frame-Pufferspeicher, der typischerweise GDDR- und/oder HBM-Technologie verwendet. Anstatt sich auf die mechanischen, elektrischen und protokolltechnischen Beschränkungen von Industriestandard-Speicherformfaktoren (d.h. JEDEC-DIMMs) verlassen zu müssen und an Produkt-Roadmaps von Drittanbietern gebunden zu sein, kann ein Systemdesigner „native“ GPU-Teile nutzen, um die Gesamtsystemleistung, die Kosten und die Resilienz bzw. Ausfallsicherheit enger zu optimieren.
  • Einfache Erweiterungen der CUDA® Speichermanagement-APIs von NVIDIA (oder anderen Anbietern) erlauben es, Anwendungsspeicher an FAM anzuheften und als GPU-Peer-Speicher zu betrachten. Alternativ oder zusätzlich kann sich der Benutzer für Unified Virtual Memory (UVM) und Seitenmigration entscheiden, um transparent zwischen lokalem Videospeicher einer GPU und FAM auf einer bedarfsweisen bzw. On-Demand-Basis zu wechseln. Siehe zum Beispiel USP 9,767,036 ; USP 9,830,210 ; USP 9,424,201 ; USP 9,639,474 & USP 10,133,677.
  • Die beispielhafte, nicht beschränkende Technologie hierin unterstützt verschiedene Programmierparadigmen: eine gegebene FAM-Region kann beispielsweise von mehreren GPUs, die zusammen an einem großen, hohe Leistung erfordernden Rechen (high performance computing, HPC)-Problem arbeiten, gemeinsam genutzt werden oder für eine einzelne GPU in einer Cloud Service Provider (CSP)-Umgebung dediziert sein, in der jede GPU eine virtuelle Maschine (VM) eines anderen Kunden ausführt. Falls eine Leistungs- oder Fehlerisolierung zwischen den verschiedenen GPUs, die auf unterschiedliche FAM-Regionen zugreifen, erwünscht ist, kann dies durch eine Fabric-Topologie-Konstruktion oder durch eine Programmierung von Engpasssteuerfunktionen in den Interconnect-Fabric-Switches erreicht werden. Zusätzlich kann eine Teilmenge von FAM-Gebern bestimmten GPUs, Benutzern und/oder VMs zugewiesen werden, um durch Richtlinien definierte Quality-of Service-Garantien zwischen GPUs oder Tenants zu ermöglichen.
  • Ein beispielhaftes, nicht beschränkendes System verbindet somit eine oder einen Satz von „Source-GPUs“ mit einem oder einem Satz von Fabric-Attached-Speichermodulen (FAMMs) über ein NVLINK™-Interconnect-Fabric, das mit NVLINK™Switches aufgebaut ist. Die Source-GPUs verschachteln („sprühen“) Speicheranforderungen über einen programmierbaren Satz von NVLINK™s und diese Anforderungen werden durch das Fabric an den Satz von FAMM-Vorrichtungen weitergeleitet. In einigen nicht beschränkenden Implementierungen sind eine „Geber“-GPU (welche wie hierin beschrieben reduzierte Fähigkeit haben kann) und diskrete DRAM-Chips, mit denen sie über ihre Frame-Puffer (FB)-Schnittstelle verbunden ist, zusammen auf einer Leiterplatte platziert, die als FAM-Grund- bzw. Basisplatte bezeichnet wird. Ein Gesamtsystem kann eine beliebige Anzahl dieser FAM-Basisplatten haben - keine, eine, zwei, drei oder n, wobei n eine beliebige ganze Zahl ist.
  • In einer nicht beschränkenden Ausführungsform verbindet jedes FAMM über eine kleine Anzahl von NVLINK™-Verbindungen (z.B. 2 oder 4) mit dem Fabric, im Vergleich zu einer größeren Anzahl von Verbindungen, die der Source-GPU zur Verfügung stehen. In einigen nicht beschränkenden Ausführungsformen ist die Spender-GPU innerhalb eines FAMM so strukturiert, nicht als eine vollwertige GPU verwendet werden kann, weil einige Teile ihrer Engines und/oder ihres Caches fehlerhaft sind, dauerhaft deaktiviert sind oder nicht existieren; aber zumindest einige ihrer NVLINK™-Verbindungen und ihrer Speicherschnittstellenbereiche voll funktionsfähig sind. Die Spender-GPU des FAMM benötigt nur eine minimale Anzahl funktionsfähiger Engines, um Speicherinitialisierungs- und Diagnoseoperationen durchzuführen, die beim Einschalten oder beim Wechsel der dem FAMM zugewiesenen Gast-VM durch den Cloud Service Provider (CSP) ausgeführt werden. In beispielhaften nicht beschränkenden Ausführungsformen kann eine abgespeckte Version des GPU-Treibers oder anderer Software diese Funktionen ebenso wie die Interrupt-Behandlung für Speicher- und GPU-interne Fehler übernehmen.
  • Zusätzliche nicht-beschränkende Merkmale und Vorteile beinhalten:
    • • In einigen nicht beschränkenden Ausführungsformen die Verwendung von Floor-Swept-GPUs als FAM-Speichercontroller („FAM-Spender“) anstelle von Industriestandard-DIMMs mit Speichercontrollern von Drittanbietern. Dies stellt höhere Kompatibilität bereit, reduziert die Abhängigkeit von Formfaktoren und Standards von Drittanbietern, senkt die Gesamtsystemkosten, nutzt die Ausgereiftheit und die bekannten Funktionen des hauseigenen GPU-Speichercontrollers (sowohl für Leistung als auch für Ausfallsicherheit) und erlaubt eine engere Integration von Rechen- und Speichersystemelementen.
    • • Weil die Source-GPUs und die FAM-Spender-GPUs in einigen Ausführungsformen dasselbe Protokoll verwenden, kann die Source-GPU den vollständigen Satz von Transaktionen ausgeben, die von dem Fabric-Protokoll unterstützt werden, einschließlich „atomischer“ Operationen sowie des Satzes von Speicher-Lese- und -Schreib-Transaktionen. Solche atomischen Operationen können zum Beispiel arithmetische Funktionen wie beispielsweise atomicAdd(), atomicSub(), atomicExch(), atomicMin(), atomicMax(), atomicInc(), atomicDec(), atomicCAS(); bitweise Funktionen wie atomicAnd(), atomicOr(), atomicXor(); und andere Funktionen beinhalten. Die Fähigkeit einiger nicht beschränkender Ausführungsformen, native „Atomics“ durchzuführen, ist besonders wertvoll, da viele Arbeitslasten Atomics für Synchronisationsoperationen verwenden. Eine Alternative zu nativen „Atomics“ ist die Nachahmung von „Atomics“ unter Verwendung von Lesen-Ändern-Schreiben bzw. „read-modify-write“ (RMW)-Operationen, was höhere Latenzzeiten und potenzielle Belastungen für die Fabric-Switches mit sich bringt, um die notwendige Umwandlung zwischen RMWs und „Atomics“ durchzuführen.
    • • Die Fähigkeit, physische Seiten, die auf FAM abgebildet sind, derart über mehrere FAM-Spender zu verschachteln, dass die Bandbreite der Source-GPU zu FAM auf die Gesamtbandbreite für alle Spender in dem „Stripe“ skaliert werden kann, die zusammen eine bestimmte Speicherseite bilden. Mit „Stripe“ ist einer der logischen Sätze von FAM-Vorrichtungen gemeint, die dazu organisiert sind, sich so an das Fabric anzuschließen, dass die Speicherleistung, die Zuverlässigkeit oder beides erhöht wird.
    • • Aufgaben der Speicherinitialisierung und -diagnose werden lokal in den Spender-GPUs und ihren abgespeckten Treibern durchgeführt und nicht von einer Host-CPU oder von Hardware-Engines in dem Fabric gesteuert. Bei einem Wechsel des Eigentümers einer FAM-Region kann deren Inhalt aus Sicherheitsgründen gelöscht werden, und in einigen Fällen werden zu diesem Zeitpunkt einfache Diagnosen ausgeführt. Das Auslagern dieser Aufgaben von einer zentralen Ressource bedeutet, dass lokale Komponenten eine FAM-Region schneller von einer virtuellen Maschine (VM) auf eine andere übertragen können, wenn Gast-Arbeitslasten innerhalb der Cloud migrieren; es gibt weniger Ausfallzeiten für neue VMs und keine Auswirkungen auf laufende VMs, deren Ressourcen nicht verschoben werden.
    • • Bereitstellen einer skalierbaren Hardware-/Software-Plattform für Kunden, die eine Vielzahl von Arbeitslasten ausführen, die die Rechenkapazität mehrerer High-End-GPUs erfordern - z.B. Deep Learning, Graph-Analysen, Recommender Engines, HPC, medizinische Bildgebung, Bildrendering, Datenbank- und Transaktionsverarbeitung usw. Für viele dieser Anwendungen wachsen die Anforderungen an die Speicherbandbreite und/oder -kapazität schneller als die GPU- oder CPU-Rechenanforderungen. Die Technologie hierin erweitert das Portfolio von Rechenzentren und anderen Infrastrukturen, indem sie mehr Flexibilität bei der Mischung von Rechen- und Speicherressourcen ermöglicht.
    • • o In einigen Ausführungsformen kann die Software Verbindungen und/oder FAM-Spender virtuell deaktivieren, um den fortgesetzten Betrieb des Systems mit verminderter Kapazität oder Bandbreite auf eine vom Administrator kontrollierte Weise zu ermöglichen. Anwendungen, die FAM verwenden, müssten nicht weiter modifiziert werden, um mit der reduzierten Kapazität oder Bandbreite umzugehen.
    • • In einigen Ausführungsformen können einzelne defekte Seiten auf einem gegebenen FAM-Spender neu zugeordnet und in Software so gesteuert werden, dass ein nachfolgender Job ECC-Doppelbitfehler oder das Steckenbleiben bei Fehlern im Speicher vermeiden kann, ohne dass ein gesamtes FAM-Chassis neu gestartet werden muss.
    • • Die Technologie nutzt die Leistungs- und Skalierbarkeitseigenschaften von Hochgeschwindigkeits-Interconnect-Fabrics wie beispielsweise NVLINK™ und NVSWITCH™.
    • • Die Möglichkeit, Silizium wiederzuverwenden, das sonst aufgrund von Fehlern in den für normale Anwendungen erforderlichen Einheiten verschrottet werden müsste.
    • • Die architektonischen Konzepte sind allgemein genug, auf beliebige Multi-GPU-Systeme und auf zukünftige größere Plattformen, die mehrere Chassis in einem Einschubschrank umfassen, angewendet werden können.
    • • In Kombination mit Software-Erweiterungen für die Allozierung/Verwaltung von FAM als Peer-Speicher und Erweiterungen, die eine Migration zwischen Videospeicher und FAM ermöglichen, baut dieses Hardware-Konzept auf bestehenden Multi-GPU-Systemen auf und ermöglicht eine Roadmap, die in die Zukunft reicht.
  • Figurenliste
  • Die folgende detaillierte Beschreibung beispielhafter, nicht beschränkender, illustrativer Ausführungsformen ist in Verbindung mit den Zeichnungen zu lesen, in welchen:
    • 1 ein nicht beschränkendes Beispiel eines Fabric-Attached-Speichersystems zeigt;
    • 2 eine High-Level-Software-Ansicht der Fabric-Attached-Speicherarchitektur der 1 zeigt;
    • 3 eine beispielhafte GPU mit reduzierten Fähigkeiten zur Verwendung in einem Fabric-Attached-Speicher zeigt;
    • 4 ein Beispiel für Fabric-Attached-Speicher-Striping zeigt;
    • 5 eine beispielhafte Unterteilung von Stripes in 4 in logische Stripes so, dass jedes FAMM mehrere Stripes bereitstellt, zeigt;
    • 6, 7 und 8 beispielhafte, nicht beschränkende Server- oder andere Chassiskonfigurationen zeigen;
    • 9 eine beispielhafte nicht beschränkende Adresszuordnung zeigt;
    • 10 ein beispielhaftes nicht beschränkendes Source-GPU-„Sprühen“ (Verschachtelung) mit einer entropiebasierten Verbindungsauswahl zeigt;
    • 11 und 12 eine beispielhafte Adressübersetzung zeigen;
    • 13, 14A und 14B eine beispielhafte Map-Slot-Programmierung zeigen;
    • 15 eine detailliertere nicht beschränkende Map-Slot-Programmierung zeigt;
    • 16 ein Beispiel für eine Map-Slot-Programmierung zeigt, die Ziel-FAMM-IDs zuweist;
    • 17 ein Beispiel für eine nicht-begrenzende Kennungszuweisung für FAMMs und deren Spiegelung in Interconnect-Fabric- Routing-Map - Slots;
    • 18 eine beispielhafte GPU veranschaulicht;
    • 19 einen beispielhaften Universalverarbeitungscluster innerhalb der GPU veranschaulicht;
    • 20 ein konzeptionelles Diagramm einer beispielhaften Grafikverarbeitungs-Pipeline, die von der GPU implementiert wird, ist;
    • 21 eine beispielhafte Speicherpartitionseinheit der GPU veranschaulicht;
    • 22 einen beispielhaften Streaming-Multiprozessor veranschaulicht;
    • 23 ein beispielhaftes Konzeptdiagramm eines unter Verwendung der GPU implementierten Verarbeitungssystems ist; und
    • 24 ein Blockdiagramm eines beispielhaften Verarbeitungssystems mit zusätzlichen Eingabevorrichtungen und Ausgabevorrichtungen ist.
  • DETAILLIERTE BESCHREIBUNG BEVORZUGTER
  • AUSFÜHRUNGSFORMEN
  • Beispielhaftes nicht beschränkendes System 100
  • 1 ist ein Blockdiagramm eines beispielhaften, nicht beschränkenden Systems 100, das Fabric-Attached-Speicher (FAM) unterstützt. In dem in 1 gezeigten System 100 kommuniziert eine Vielzahl (N) von GPUs 102(0), 102(1), ... 102(N) über ein Hochleistungs-Interconnect-Fabric mit hoher Bandbreite, wie z.B. NVLINK™von NVIDIA, miteinander. Andere Systeme können eine einzelne GPU 102(0) bereitstellen, die mit NVLINK™verbunden ist.
  • Das NVLINK™-Interconnect-Fabric bzw. die NVLINK™-Verbindungsstruktur (welche Verbindungen bzw. Links 108, 110 und Switch(es) 104 beinhaltet) stellt mehrere Hochgeschwindigkeitsverbindungen NVL(0)-NVL(k) bereit, die die GPUs 102 verbinden. In dem gezeigten Beispiel ist jede GPU 102 über k Hochgeschwindigkeitsverbindungen 108(0)-108(k) mit dem Switch 104 verbunden. Somit ist die GPU 102(0) über Verbindungen 108(00)-108(0k) mit dem Switch 104 verbunden, ist die GPU 102(1) über Verbindungen 108(10)-108(lk) mit dem Switch verbunden und so weiter. In einigen Ausführungsformen ist k=12. In anderen Ausführungsformen können die verschiedenen GPUs 102 jedoch über eine unterschiedliche Anzahl von Verbindungen 108 mit dem Switch 104 verbunden sein, oder können einige GPUs direkt mit anderen GPUs verbunden sein, ohne über den Switch 104 verbunden zu sein (siehe z.B. 23).
  • In der gezeigten Ausführungsform kann jede GPU 102 Hochgeschwindigkeitsverbindungen 108 und den Switch 104 verwenden, um mit dem Speicher zu kommunizieren, der von einer oder allen der anderen GPUs 102 bereitgestellt wird. Beispielsweise kann es Fälle und Anwendungen geben, in welchen jede GPU 102 mehr Speicher benötigt, als von ihrem eigenen lokal angeschlossenen Speicher bereitgestellt wird. Als einige nicht beschränkende Anwendungsfälle kann dann, wenn das System 100 ein Deep Learning-Training großer Modelle unter Verwendung von Netzwerkaktivierungs-Auslagerung, ein Analysieren von Massendaten bzw. „Big Data“ (z.B. RAPIDS-Analysen (ETL) in Speicherdatenbankanalysen, Graph-Analysen usw.), eine computergestützte Pathologie unter Verwendung von Deep Learning, eine medizinische Bildgebung, ein Grafik-Rendering oder dergleichen durchführt, es mehr Speicher benötigen, als als Teil jeder GPU 102 verfügbar ist.
  • Als eine mögliche Lösung kann jede GPU 102 von 1 die Verbindungen 108 und den Switch 104 verwenden, um auf lokalen Speicher jeder beliebigen anderen GPU zuzugreifen, als ob es eigener lokale Speicher der GPU wäre. Somit kann jede GPU 102 mit ihrem eigenen lokal angeschlossenen Speicher ausgestattet sein, auf den sie zugreifen kann, ohne Transaktionen über das Interconnect-Fabric zu initiieren, sie kann aber auch das Interconnect-Fabric verwenden, um einzelne Wörter des lokalen Speichers anderer GPUs, die mit dem Fabric bzw. der Struktur verbunden sind, zu adressieren/auf diese zuzugreifen. In einigen nicht beschränkenden Ausführungsformen ist jede GPU 102 in der Lage, auf einen solchen lokalen Speicher anderer GPUs unter Verwendung hardwarebeschleunigter atomischer MMU-Funktionen, die einen Speicherplatz lesen, den gelesenen Wert modifizieren und die Ergebnisse zurück in den Speicherplatz schreiben, ohne Befehle zum Laden in Register und Speichern aus dem Register heraus zu benötigen (siehe oben), zuzugreifen.
  • Ein solcher Zugriff durch eine GPU auf den lokalen Speicher einer anderen GPU kann aus der Perspektive einer Anwendung, die auf der den Zugriff veranlassenden GPU ausgeführt wird, „derselbe“ (wenn auch nicht ganz so schnell) sein, als ob die GPU auf ihren eigenen lokal angeschlossenen Speicher zugreifen würde. Hardware innerhalb jeder GPU 102 und Hardware innerhalb des Switches 104 stellen notwendige Adressübersetzungen bereit, um von der laufenden Anwendung verwendete virtuelle Adressen in physische Speicheradressen des eigenen lokalen Speichers der GPU und des lokalen Speichers einer oder mehrerer anderer GPUs abzubilden. Wie hierin erklärt wurde, wird ein solcher Peer-to-Peer-Zugriff auf Fabric-Attached-Speicher ausgedehnt, ohne dass damit verbundene Kosten für das Hinzufügen weiterer rechenfähiger GPUs anfallen.
  • 1 (und siehe auch 26 für eine andere Ansicht) zeigt auch, dass jede GPU 102 auf ein Hauptspeichersystem 114 innerhalb des Adressraums/der Adressräume einer oder mehrerer CPUs 116/150 zugreifen kann. Da jedoch die Verbindung zwischen dem Switch 104 und dem Hauptspeichersystem 114 über einen relativ langsamen PCIe-Bus bzw. relativ langsame PCIe-Busse 112 erfolgt, kann der Zugriff von GPUs 102 auf das Hauptspeichersystem 114 eine relativ hohe Latenz und damit eine langsame Leistung involvieren.
  • Um für die GPUs 102 Zugriff auf zusätzlichen Hochleistungsspeicher mit niedriger Latenz bereitzustellen, stellt das System nach 1 eine neue Art von GPU-Peer - Fabric-Attached-Speichermodulen (FAMMs) 106 bereit, die jeweils einen spezialisierten Speichercontroller und zugeordneten Hochleistungsspeicher umfassen. Die GPUs 102 kommunizieren mit den FAMMs 106 über die gleiche Hochgeschwindigkeitsverbindung 108, 110, 104, die die GPUs zur Kommunikation untereinander verwenden. Folglich ist jedes der FAMMs 106 mit dem Switch 104 über eine oder mehrere Hochgeschwindigkeitsverbindungen 110 verbunden, die, in einer beispielhaften, nicht beschränkenden Ausführungsform, die gleiche Bandbreite haben können wie die Verbindungen 108, die die GPUs 102 zur Kommunikation mit dem Switch 104 verwenden. Jedes der FAMMs 106 kann mit dem Switch 104 über eine beliebige Anzahl von Verbindungen 110 kommunizieren, obwohl in einigen nicht beschränkenden Fällen die Anzahl von Verbindungen 110, die jedes FAMM 106 zur Kommunikation mit dem Switch verwendet, geringer ist als die Anzahl (k) der Verbindungen, die jede GPU 102 zur Kommunikation mit dem Switch verwendet.
  • Bisher befanden sich auf der anderen Seite des NVLINK™-Interconnect-Fabric 108, 110, 104 aus der Perspektive einer GPU 102 oder einer CPU andere (.z.B. Peer-) Rechen-GPUs. Die vorliegende nicht-beschränkende Technologie stellt den GPUs 102 Peer-to-Peer-Zugriff auf eine andere Art von Vorrichtung bereit - auf zusätzlichen FAM-Speicher 106, der viel schneller ist als der Systemspeicher 114 und der (kollektiv) Kapazitäten bietet, die (potenziell) viel größer sind als der eigene, lokal angeschlossene Speicher der GPUs und der Pool an lokalem Speicher, der mit allen Rechen-GPUs in dem System verbunden ist. Daher sieht dieser zusätzliche FAM-Speicher 106 unter Verwendung der nicht beschränkenden Technologie hierin wie lokal verbundener Speicher oder Peer-Speicher in dem Sinne aus, dass bestehende Anwendungen auf den FAM-Speicher auf die gleiche Weise zugreifen können wie auf Peer-GPU-Speicher (d. h. zusätzlichen Speicher, der lokal zu anderen GPUs 102 gehört). Eine GPU-Anwendung kann leicht den zusätzlichen Fabric-Attached-Speicher 106, auf den über NVLINK™ 108, 110, 104 zugegriffen werden kann, ohne oder mit nur wenigen Änderungen nutzen und die Möglichkeit erhalten, ihre Arbeitsausführung in zusätzlichen Hochleistungsspeicher zu speichern. Die beispielhafte, nicht beschränkende Technologie ermöglicht es somit einer GPU 102, viel höhere Speicherzugriffsbandbreite zu erhalten, als sie es durch den Zugriff auf Hauptsystemspeicher 114 mit Kapazitäten, die zumindest so groß sind wie (und in einigen Ausführungsformen viel größer als) die Speicherkapazitäten des für die CPU 116 verfügbaren Speichers 114, könnte.
  • Ferner unterstützt in einer beispielhaften, nicht beschränkenden Ausführungsform die beispielhafte, nicht beschränkende Technologie das gesamte GPU-Speichermodell - was bedeutet, dass alle der Operationen, die in die Anwendung integriert sind, alle nativ ausgeführt werden und keinerlei Emulation oder andere langsamere Pfadanpassungen erfordern, wie z.B. für atomische GPU-Operationen (die sich von einem oder dem Satz von Atomics unterscheiden können, die auf der CPU 116 vorhanden sind). Solche Schnittstellen zwischen GPU-Atomics und CPU-Atomics könnten langsamere, softwarevermittelte Operationen oder in manchen Fällen einen Hardware-Übersetzer oder einen anderen Vermittler erfordern - welches immer noch langsamer ist als die Möglichkeit, GPU-Atomics nativ auszuführen.
  • Beispielhafte FAM-Implementierung
  • 2 zeigt eine beispielhafte Implementierung des Systems nach 1 mit 8 GPUs 102(0)-107(7). Die NVLINK™-Switches 104 können auf einer GPU-Basisplatte und einer Mittenebene in einem Multi-GPU-System angeordnet sein. 2 zeigt, dass der/die Switch(es) 104 über mehrere funktionale Switch-Module 104A0-104A5, 104B0-104B5 verteilt sein können, die von einem Service-Prozessor 152 überwacht werden. 2 zeigt ferner mehrere FAM-Platten oder Backplanes (die horizontalen Blöcke im unteren Teil der Zeichnung), die jeweils eine Vielzahl von FAMMs 106 implementieren und jeweils von einer FAM-Service-Prozessor(en) (SP)-CPU 154 überwacht werden. Es kann eine beliebige Anzahl von FAM-Platten oder Backplanes geben. Eine FAM-Serviceprozessor-CPU 154 befindet sich in einer beispielhaften Ausführungsform auf oder in der Nähe jeder FAM-Basisplatte und wird zur Verwaltung der Vorrichtungen (FAMMs und Switches, falls vorhanden) auf der Basisplatte verwendet. Die FAM SP-CPU 154 kann in einer Implementierung ein anderes Betriebssystem als die Host-CPUs 150 und ein anderes Betriebssystem als der Service-Prozessor 152, der die Source-GPUs 102 und die Switches 104 auf einer GPU-Basisplatte und der Mittenebene (falls vorhanden) verwaltet, ausführen. Die FAM SP-CPU 154 kann zum Beispiel alle der FAMMs 106 auf der Basisplatte über eine oder mehrere Verbindungen wie beispielsweise PCIe verwalten. Die FAM SP-CPU 154 führt in einer Ausführungsform Anweisungen aus, die in einem zusätzlichen nichtflüchtigen Speicher, der mit ihr verbunden ist, gespeichert sind, um einige oder alle der folgenden Verwaltungsfunktionen auszuführen:
    • • Initialisierung der Spender
      • o Konfiguration der Speichercontroller-Register
      • o Nullsetzen des Inhalts des DRAM
    • • Fehlerüberwachung und -behandlung
      • o DRAM SBEs und DBEs
      • o Speichercontroller-interne Fehler
    • • Leistungsüberwachung
      • o Konfiguration und Abfrage von Leistungsmonitoren
      • o Verarbeitung der von Monitoren gelesenen Werte zur Berechnung von Statistiken über den Durchsatz usw.
    • • Zeilen-Remapper-Funktionen
      • o Ansprechen auf Interrupts, die SBE- und DBE-Ereignisse anzeigen
      • o Verwalten einer Tabelle pro FBPA (Frame-Puffer-Partitionsadresse), die Adress-Remapping durchführt
    • • Umgebungsüberwachung (Strom, Wärme) (dies kann auch von einer nicht gezeigten Basisplatten-Verwaltungssteuereinrichtung bzw. einem Baseboard Management Controller (BMC) gehandhabt werden; oder statt eines separaten BMC auf der FAM-Basisplatte wird der vorhandene Chassis-BMC, der die Source-GPUs 102 überwacht, seinen Aufgabenbereich auf die Überwachung der FAMMs 106 erweitern).
  • „Floor Swept“-GPUs als disaggregierte Fabric-Attached-Speichercontroller
  • Beispielhafte, nicht beschränkende Ausführungsformen stellen eine Disaggregation zwischen GPUs 102 und Speicher bereit durch Implementieren des FAMM 106 unter Verwendung von relativ preiswerter Low-End-Speichercontroller-Hardware, die in einigen Fällen im Vergleich zu einer vollwertigen GPU viel weniger kostspielig und weniger energieintensiv ist, aber dennoch einen voll funktionsfähigen Peer-to-Peer-Zugriff bieten kann. Solche Speichercontroller-Hardware wird hauptsächlich oder ausschließlich für die Kommunikation mit DRAM oder anderen Halbleiterspeichern verwendet und muss keine Aufgaben ausführen, die nicht für den Speicherzugriff und die Steuerung benötigt werden, wie z.B. Rechen- oder Kopierfunktionen.
  • Eine nicht beschränkende Möglichkeit ist die Implementierung von FAMMs 106 unter Verwendung so genannter „Floor Swept“-GPUs, die andernfalls aufgrund von Herstellungsfehlern, die verhindern, für Rechenanwendungen richtig funktionieren, nicht in Produkten verkauft werden würden oder könnten. Falls die Defekte solcher „Floor Swept“-GPU-Komponenten die Fähigkeit der Komponente, mit anderen GPUs zu kommunizieren, an der Interconnect-Fabric teilzunehmen und auf gebündelten Speicher zuzugreifen, nicht beeinträchtigen, kann die Komponente als ein Fabric-Attached-Speichercontroller verwendet werden, und können andere Funktionen dauerhaft deaktiviert oder abgeschaltet werden, um Energie zu sparen.
  • In einigen nicht beschränkenden Ausführungsformen arbeitet die Spender-GPU innerhalb des FAMM 106 als reine Slave-Vorrichtung, d. h. sie reagiert nur auf Anforderungen, die von der Verbindung 108 empfangen werden; sie initiiert keine Anforderungen auf dem Fabric (aber andere Typen von FAM-Spender-GPUs könnten solche Anforderungen initiieren). Die dazu konfigurierte Spender-GPU wird als ein „Floor Sweep“-Teil bezeichnet, bei dem die nicht funktionsfähigen Einheiten abgeschmolzen oder anderweitig absichtlich deaktiviert sind, so weniger Strom verbrauchen (z.B. in einigen Fällen nur Leckstrom). Siehe z.B. 3. In beispielhaften Ausführungsformen wird ein Mechanismus bereitgestellt, damit auf dem System ausgeführte Software in der Lage ist, eine solche FAM-Spender-GPU zu identifizieren und sie von einer Rechen-GPU zu unterscheiden.
  • Ein Vorteil der Verwendung einer Teilmenge einer „normalen“ GPU als FAM-Speichercontroller ist, dass ein Speichercontroller mit einer solchen Teilmenge von Fähigkeiten in der Lage ist, mit anderen GPUs 102 zu kommunizieren und dabei einen vollständigen Satz von Funktionalitäten zu verwenden, einschließlich beispielsweise Lese-, Schreib- und „atomische“ Speicherzugriffsfunktionen. Allgemein führt, wie vorstehend erörtert wurde, eine atomische Funktion eine atomische Lese-Änderung-Schreib-Operation an einem (z.B. 32-Bit- oder 64-Bit-) Wort durch, das sich in globalem oder gemeinsam genutztem Speicher befindet, wobei Hardwarebeschleunigung verwendet wird. Zum Beispiel liest atomicAdd() ein Wort an einer Adresse im globalen oder gemeinsam genutzten Speicher, addiert eine Zahl dazu und schreibt das Ergebnis an dieselbe Adresse zurück. Die Operation ist „atomisch“ in dem Sinne, garantiert ohne Beeinträchtigung durch andere Threads ausgeführt wird. Mit anderen Worten führt die Speichercontroller-Hardware typischerweise die atomische Operation aus, und kein anderer Thread kann auf diese Adresse zugreifen, bis die Operation abgeschlossen ist.
  • Weil atomische Inter-GPU-Befehle in den Fabric-Attached-Speichern 106 verfügbar sind, die von einigen nicht beschränkenden Ausführungsformen hierin bereitgestellt werden, kann eine „Quell“-GPU 102, die versucht, über einen „Spender“-GPU-basierten Speichercontroller 106 auf Speicher zuzugreifen, einen vollständigen Satz von Inter-GPU-Kommunikationsprotokolltransaktionen einschließlich solcher atomischer Funktionen verwenden, wodurch die Anwendung eine bessere Leistung erhalten kann. Die Leistung wird erhöht, weil die Atomics bzw. atomischen Funktionen nativ in Hardware ausgeführt werden können, was Geschwindigkeitsvorteile mit sich bringt. Ferner wird die Kompatibilität aufrechterhalten, so dass dieselben Threads, die für die Kommunikation mit anderen GPUs 102 ausgelegt sind, auch auf den Fabric-Attached-Speicher 106 zugreifen können, auch wenn auf diesen FAM nicht notwendigerweise über eine vollwertige GPU zugegriffen wird. Während atomische Funktionen unter Verwendung grundlegenderer Lese-Änderung-Schreib-Befehle und anderer Techniken emuliert werden können, ist es in einigen nicht beschränkenden Beispielen sehr effizient, Spender-GPUs mit Fähigkeiten nativ implementierter atomischer Funktionen auszustatten.
  • Einige beispielhafte, nicht beschränkende Implementierungen unterstützen atomische Funktionen möglicherweise nicht nativ. Die Unfähigkeit, atomische Funktionen nativ zu unterstützen, kann Anwendungen auf der Source-GPU unterstützen, die umgeschrieben oder ursprünglich entwickelt wurden, um die Operationen nativer atomischer Funktionen durch Lese-/Änderung-/Schreib-Anweisungen zu ersetzen, oder die Emulation von atomischen Funktionen auf den Spender-GPUs erfordern. Dies würde die Leistung verringern, könnte in bestimmten Anwendungen aber dennoch gut funktionieren.
  • In einer beispielhaften, nicht beschränkenden Ausführungsform kann es möglich sein, ein spezialisiertes Stück Hardware zu entwerfen oder zu konstruieren, wie z.B. einen spezialisierten Speichercontroller, der keine GPU ist, aber dennoch ausreichende Funktionalität bereitstellt, um an der hierin beschriebenen Fabric-Attached-Speicherarchitektur zu partizipieren. Eine solche Implementierung könnte eine sehr einfache GPU-ähnliche Vorrichtung sein, die mit einem Speichercontroller ausgestattet ist. Eine solche Vorrichtung könnte über minimale Funktionalität verfügen, die für die Verarbeitung von NVLINK™-Befehlen erforderlich ist, einschließlich atomischer Funktionen sowie einiger primitiver Engines bzw. Maschinen für Stammfunktionen, die eine Initialisierung und ein Löschen von Speicher durchführen können. Eine beispielhafte minimale GPU-Konfiguration, die zur Implementierung des FAMM 106 erforderlich ist, könnte eine logischphysische Verbindungsabbildungsfunktion bzw. Link-Mapping-Funktion, zwei NVLINK™-Ports (welche zwei, könnte von Spender zu Spender variieren) und bestimmte andere Funktionen, z.B. für die Verarbeitung von atomischen Funktionen, Übersetzung eingehender Adressen und andere Funktionen, umfassen. Wie das Blockdiagramm in 3 zeigt, könnten solche Mindestfunktionen z.B. Folgendes beinhalten:
    • • Kapazität. Die nominale Kapazität geht von 3DS (3D-Stacking) und 4H (4-high) 16 GB x8 Halbleiterspeicherteilen in einigen Beispielausführungen aus.
    • • Bandbreite. Die DRAM-Schnittstelle ist in einigen Ausführungsformen auf die bidirektionale NVLINK™-Bandbreite für die beiden an das FAM-DIMM (Dual-Inline-Speichermodul(e)) angeschlossenen Hochgeschwindigkeitsverbindungen abgestimmt. Es gibt zwei Szenarien, in denen die erzielbare DRAM-Bandbreite geringer ist: (1) Strom von kleinen Schreibvorgängen, z.B. weniger als 32B. Bei Schreibvorgängen dieser Größe muss der GPU-Speichercontroller ein Read-Modify-Write bzw. Lesen-Ändern-Schreiben durchführen. Das Gleiche gilt für atomische NVLINK™-Operationen. Die meisten interessierenden FAM-Arbeitslasten beinhalten keine langen Sequenzen kleiner Schreibvorgänge oder atomischer Operationen, d. h. sie sind eher sporadisch. (2) Zufällige Abfolge von Adressen über das DIMM oder anderen Hochgeschwindigkeits-Speicheradressraum hinweg. Zufällige Zugriffe werden zu einer höheren Häufigkeit von L2-Cache-Fehlzugriffen führen und werden generell eine schlechte DRAM-Effizienz erzeugen, weil mehr Zugriffe auf geschlossene Bänke erfolgen werden. Eine geschlossene Bank kann vor dem DRAM-Lesezyklus geöffnet (aktiviert) werden und der daraus resultierende Overhead raubt verfügbare Bandbreite. Diese Art von Muster wird für viele FAM-Arbeitslasten nicht erwartet, ist aber möglich. Diese spezifischen Beschränkungen sind nur Beispiele und nicht beschränkend, obwohl viele Speichercontroller eine gewisse Zugriffsgröße haben werden, bei welcher die Schreibleistung abfällt, weil sie Lesen-Ändern-Schreiben durchführen müssen, und auch das Verhalten von offenen gegenüber geschlossenen Bänken aufweisen werden.
    • • Zeilen-Remapper. Zusätzliche Zuverlässigkeits/Zugänglichkeits/ Bedienbarkeits (Reliability/Accessibility/Serviceability; RAS)-Funktionen in der Spender-GPU können, zusammen mit Software zu deren Verwaltung, problemlos für FAM eingesetzt werden. Diese Funktionen werden wichtiger mit Kapazitätsniveaus von FAM, die in den Bereich von einigen 10 oder möglicherweise einigen 100 TB reichen. Ein Beispiel ist die GPU-Zeilen-Remapper-Funktion, welche eine Reihe von freien DRAM-Speicherplätzen pro Bank reserviert. Die Zeilen-Remapping-Funktion(en) ist/sind hilfreich als ein Ausfallsicherheitsmerkmal in FAM. Wenn ein unkorrigierbarer Fehler (z.B. ein Doppelbitfehler (DBE)) erfasst wird, kann das System heruntergefahren werden, damit die Systemsoftware die DRAM-Seite, bei der der DBE aufgetreten ist, neu auf eine Ersatzseite zuordnen kann. Software, die die Spender-GPU verwaltet, kann die Neuzuordnungszugriffe der Zeilen-Remapper-Tabelle auf die fehlerhafte Zeile auf eine der reservierten Zeilen konfigurieren. In einigen Ausführungsformen wird die Neuzuordnung aufgrund von Sicherheitsbedenken nicht on-the-fly durchgeführt.
    • • Verbindungs- bzw. Link L1 Cache Translation Lookaside Buffer (TLB)-Abdeckung der gesamten DRAM-Kapazität. Software kann die Fabric Linear Address (FLA)-Fähigkeit (siehe unten) verwenden, um eine Seite neu zuzuordnen, bei der ein DBE zwischen dem Zeitpunkt der Erfassung des DBE und dem Herunterfahren des Systems zur Durchführung der Neuzuordnung aufgetreten ist.
    • • Unterstützung für eingehendes NVLINK™-Lesen-Ändern-Schreiben. Dies dient der Interoperabilität, falls neue atomische NVLINK™-Operationen, die von der GPU nicht nativ unterstützt werden, hinzugefügt werden.
    • • Fähigkeit zum Selbsttest und zur Initialisierung des DRAM. Um diese Funktionen durchzuführen, muss ein minimaler Satz von Engines verfügbar und eingeschaltet sein.
    • • Die Spender-GPU kann je nach ihren „Floor Swept“-Fähigkeiten in die Lage versetzt werden, bestimmte Housekeeping- und Verwaltungsaufgaben von dem zentralisierten Systemverwaltungsprozessor oder der Host-CPU auszulagern, wobei sie Operationen wie beispielsweise Speicherdiagnosen zum Zeitpunkt der Systeminitialisierung oder Sicherheitsmaßnahmen (z.B. Löschen des Speicherinhalts des Spenders, wenn sein Besitz von einer VM zu einer anderen wechselt) durchführt.
  • In einer Ausführungsform hat der FAMM 106-Speichercontroller also keine GPU-Rechenfähigkeiten, umfasst aber:
    • • ein Boot-ROM;
    • • einen DDR-Speichercontroller, der in der Lage ist, die atomischen Funktionen ohne Emulation in Hardware zu beschleunigen;
    • • einen DRAM-Zeilen-Remapper
    • • einen Daten-Cache;
    • • eine Crossbar-Interconnection;
    • • eine Fabric-Interconnect-Schnittstelle, die zu Peer-to-Peer-Kommunikation über das Interconnect-Fabric mit GPUs fähig ist; und
    • • DRAM-Schnittstellenschaltungen.
  • Hierin bedeutet „GPU-Rechenfähigkeit“ parallele Mehrfachkern- bzw. Multi-Core, Multi-Thread-Ausführungsberechnungsfähigkeiten, welche Hardware-beschleunigtes, Grafik-Pipeline-basiertes Shading, Echtzeit-Raytracing, Deep-Learning-Beschleunigung und/oder Echtzeit-Computer-Vision umfassen, die durch Streaming-Multiprozessor-Kerne, Raytracing-Kerne, Tensor-Kerne und Textur-Einheiten implementiert werden, wie sie zum Beispiel durch VOLTA-, TURING- und/oder AMPERE-GPU-Architekturen von NVIDIA beispielhaft dargestellt werden.
  • Eine nochmals weitere Option wäre es (wie in 3 gezeigt ist), voll funktionsfähige GPUs zu nehmen und die meisten der wie vorstehend beschriebenen Rechen- und Kopier-Engine-Fähigkeiten solcher GPUs abzuschmelzen, um sie für die Verwendung (nur) als FAM-Spender-GPUs geeignet zu machen. In einem nicht beschränkenden Beispiel können die folgenden Einheiten für FAM abgeschmolzen werden, um Strom zu sparen: NVENC, NVDEC, Anzeige, 2 NVL3-Ports und alle Syspipes außer Syspipe0. So können z.B. alle Grafik-Engines und Pipelines und alle Anzeigefunktionen; alle Mikrocontroller; alle Rechen-Engines und alle Kopier-Engines in den als FAM-Speichercontroller verwendeten GPUs nicht-reversibel abgeschmolzen oder deaktiviert (oder anderweitig nicht vorgesehen) werden. Eine solche absichtliche Verschlechterung (welche auch durch eine Umgestaltung bzw. ein Redesign erreicht werden könnte) könnte beispielsweise bestimmte Rechenfunktionen deaktivieren, die typischerweise in jeder, den meisten oder allen der voll funktionsfähigen GPUs in dem System vorhanden sein würden, wie z.B. einige oder alle der folgenden: (a) atomische Addition, die mit 32-Bit-Gleitkommawerten im globalen und gemeinsamen Speicher arbeitet (atomicAdd()); (b) atomische Addition, die mit 64-Bit-Gleitkommawerten im globalen und gemeinsamen Speicher arbeitet (atomicAdd()); (c) Warp-Vote- und Ballot- bzw. Abstimmungsfunktionen; (d) Memory-Fence-Funktionen; (e) Synchronisations-funktionen; (f) Oberflächenfunktionen; (g) 3D-Gitter von Thread-Blöcken; (h) Unified-Memory-Programmierung, sofern hierin nicht anders erläutert; (i) Trichterverschiebung; (j) dynamische Parallelität; (k) halbgenaue Gleitkommaoperationen: (1) Addition, Subtraktion, Multiplikation, Vergleich, Warp-Misch-Funktionen, Konvertierung; und (m) Tensorkern. Solche absichtlich verschlechterten (und/oder umgestalteten) Vorrichtungen wären folglich für Rechenfunktionalität nicht brauchbar (und könnten auch nicht einfach von Endanwendern modifiziert werden, um verlorene Rechenfunktionalität wiederherzustellen), würden aber immer noch genügend GPU-ähnliche Funktionalität bereitstellen, um die hierin beschriebenen Fabric-Attached-Speicherarchitekturen zu unterstützen. Ein solches Abschmelzen oder andere Modifikationen hätten den zusätzlichen Vorteil, den Stromverbrauch reduzieren, was in Anwendungen von Vorteil sein könnte, in denen es viele Spender-GPUs gibt und/oder in denen die Leistungsanforderungen besonders wichtig sind (z.B. in autonomen Fahrzeugen, tragbaren Computern, Raumfahrzeugen, Tauchbooten oder anderen Anwendungen, in denen der Stromverbrauch und/oder die Wärmeentwicklung zu minimieren ist).
  • Beispielhafte, nicht beschränkende Daten-Stripes
  • Daten-Striping ist eine Technik, die es einem Prozessor, wie beispielsweise einer GPU, ermöglicht, seinen Speicher über eine Reihe verschiedener FAMM-Speichervorrichtungen zu verteilen. Unter Verwendung von Data Striping kann eine Source-GPU 102 Daten parallel in N verschiedene Speichervorrichtungen wie beispielsweise FAMMs 106 schreiben. Die N Speichervorrichtungen können die Zugriffe parallel in 1/N-tel der Zeit durchführen, die eine Speichervorrichtung für die sequenzielle Durchführung derselben Zugriffe benötigt hätte.
  • Das FAM-System 100 hierin unterstützt solches softwaremäßig alloziertes Speicher-Striping. In einer beispielhaften Ausführungsform, wie in 4 gezeigt, weist zum Beispiel Software einer Anwendung FAMMs 106 in der Granularität von Streifen bzw. „Stripes“ zu, wobei ein Stripe über eine Sammlung von FAMMs 106 gespeichert wird. 4 zeigt ein Beispiel mit 8 FAMMs 106 pro Switch 104, wodurch Software acht 6-breite Streifen von DIMMs oder anderem Speicher erstellen kann. In einem System mit 8 GPUs 102 stellt dies einen einzelnen Stripe pro GPU und zwei zusätzliche Stripes, die von GPUs verwendet werden können, die zusätzliche Kapazität benötigen, bereit. Allgemein kann Software auf einer bestimmten FAM Basisplatte Stripes unterschiedlicher Breite allozieren, obwohl die üblichere Konfiguration aus einheitliche Stripe-Breiten besteht, wie hier gezeigt. Streifen können „horizontal“ sein, wo ein gegebener Switch 104 einen einzelnen FAMM 106 zu dem Streifen beiträgt, oder „vertikal“, wo ein gegebener Switch mehrere FAMMs 106 zu dem Streifen beiträgt.
  • Sammlungen von FAMMs können somit als ein „Stripe“ an das Fabric angeschlossen werden, um der Source-GPU mehr Gesamtkapazität oder Speicherbandbreite zur Verfügung zu stellen, wobei die Anzahl von FAMMs, die einen Stripe umfassen, und die Anzahl von Verbindungen, über welche auf den Stripe zugegriffen wird, von Speicherzuweisungssoftware in Abhängigkeit von den Kapazitäts- und Bandbreitenanforderungen der Anwendung konfiguriert werden kann.
  • GPUs 102 können in einigen Anwendungen den Speicher auf einem gegebenen Stripe gemeinsam nutzen, anstatt exklusiven Zugriff auf den Stripe zu haben, und der Switch 104 kann dies durch geeignete Programmierung von Routingtabellen (wie nachstehend erläutert) unterstützen. Falls die GPUs 102 zu separaten virtuellen Maschinen gehören (z.B. in einem Cloud-Rechenzentrum, in dem das System von mehreren Tenants genutzt wird), dann kann die Nichtinterferenz-Eigenschaft dazu beitragen, Leistung und Fehlerisolierung zwischen den VMs/Benutzern zu gewährleisten. Insbesondere kann ein gegebener Stripe durch das Design des Switching-Fabric und/oder durch die Programmierung von Switch-Routingtabellen (siehe unten) so aufgebaut werden, dass der Stripe einer einzelnen Source-GPU und/oder einer einzelnen VM vorbehalten ist; Zugriffe von anderen GPUs oder VMs werden durch Sicherheitsüberprüfungen in den Switches verhindert. Ein gegebener Stripe kann auch von mehreren Source-GPUs, die unter derselben VM laufen, oder von GPUs, die unter verschiedenen VMs laufen, gemeinsam genutzt werden, abhängig von dem Modell der gemeinsamen Datennutzung für die Anwendung. Für beide Modelle - dedizierte oder gemeinsam genutzte Stripes - können Techniken zur Engpasskontrolle, wie z.B. die Begrenzung der Injektionsrate, in den Source-GPUs oder Switches eingesetzt werden, um sicherzustellen, dass die Bandbreite für den Satz von FAMM-Stripes gleichmäßig unter den Source-GPUs aufgeteilt wird.
  • Wie in 5 gezeigt ist, kann ein auf einem FAMM 106 basierender Stripe-Adressraum selbst unterteilt oder partitioniert sein, um mehrere „logische Stripes“ innerhalb desselben physischen FAM-Speichers zu erstellen. Dies ist z.B. in einem Mehrknotensystem hilfreich, wenn die Anzahl von Source-GPUs 102 die Anzahl von Stripes übersteigt. In dieser Art von System ist es hilfreich, jeder Source-GPU 102 eine Teilmenge der Kapazität eines Stripes zuzuweisen. Die Unterteilung von Stripes wird durch Programmierung der Source-GPU 102 und der Routingtabellen des NVSWITCH™ 104 erreicht (siehe unten in Verbindung mit den 13-15) und hat keinen Einfluss auf die Hardwarefunktion. In 5 beinhaltet eine Platte acht 6-breite Stripes, die jeweils in 3 logische Stripes unterteilt sind. 24 Source-GPUs 102 erhalten jeweils einen logischen Stripe. Natürlich ist das gezeigte Striping nur ein nicht-beschränkendes Beispiel und sind andere Striping-Muster und oder Verteilungen möglich.
  • Die Fähigkeit der Verschachtelung über mehrere Spender 106, die einen „Stripe“ von FAM erzeugen, hinweg ist für die Leistung wertvoll, weil die Bandbreite einer Source-GPU 102 zu FAM nicht durch die Bandbreite eines einzelnen FAMM 106 zu dem Fabric begrenzt ist. Abhängig davon, wie viele Source-GPUs sich eine FAM-Basisplatte teilen, kann eine beliebige gegebene Source-GPU 102 potenziell bis hin zu allen ihrer Verbindungen zum Fabric beim Zugriff auf FAM sättigen.
  • Es wird angemerkt, dass das vorstehend beschriebene Konzept der Datenstreifen unabhängig von dem nachstehend diskutierten hardwarebasierten „Sprüh“-Konzept ist. Insbesondere werden Datenstreifen durch Software (z.B. die Anwendung(en), die auf einer Source-GPU 102 läuft/laufen) ausgewählt und programmiert und durch Routingtabellen gehandhabt, wohingegen sich das „Sprühen“ (z.B. wie in Verbindung mit 10 unten beschrieben) darauf bezieht, wie die beispielhafte, nicht beschränkende Ausführungsform Speicherzugriffsanforderungen effizient über das Interconnect-Fabric kommuniziert. In den beispielhaften Ausführungsformen steuern dieselben Routingtabellen (siehe 14A, 14B, 15), die das Daten-Striping verwalten (basierend auf der Zuordnung von physischen Interconnect-Adressen zu bestimmten FAMM 106-Regionen), auch zusätzliche Datentransformationen, die das „Sprühen“ sowie Ungleichheiten zwischen der globalen Adressraumgröße und dem Adressraum einzelner FAMMs 106 berücksichtigen.
  • Beispiel für einen nicht-begrenzenden Formfaktor
  • Das System 100 kann unter Verwendung einer beliebigen Anzahl von verschiedenen Formfaktoren implementiert sein. Einige Implementierungen können jedoch Vorteile in Bezug auf Kosten und Zweckmäßigkeit bieten. Zum Beispiel können in einigen nicht-beschränkenden Ausführungsformen mehrere FAMMs 106 auf einer gemeinsamen Leiterplatte angeordnet sein, wodurch eine signifikante Speichererweiterung durch einfaches Hinzufügen einer weiteren Einzelplatine zu einem System ermöglicht wird. Genauer gesagt können in einer nicht beschränkenden Ausführungsform mehrere FAMMs 106 zusammen auf einer FAM-Basisplatte („Einschub“) angeordnet sein, welche denselben Formfaktor wie die Source-GPU-Basisplatte hat, die GPU 102-Rechenressourcen bereitstellt. Ein Rechenzentrums-Einschubschrank kann beispielsweise abhängig von den Anforderungen an Rechenleistung und Speicher für die ausgeführten Kunden-Arbeitslasten mit einer unterschiedlichen Mischung aus Source-GPU- und FAM-Basisplatten bestückt sein.
  • 6 zeigt ein beispielhaftes FAM-Chassis, das Kunden, die Big-Data-Anwendungen mit Multi-GPU-Beschleunigung ausführen, die Option eines größeren Speichergrundrisses bereitstellt. In dieser Ausführungsform wird eine dedizierte FAM-Basisplatte (Einsatz bzw. „Tray“) über den GPU- und CPU-Subsystemen hinzugefügt, um ein speichererweitertes System zu schaffen. Die dedizierte FAM-Basisplatte kann eine Reihe von FAMM-Vorrichtungen und zugeordneten gebündelten Hochleistungsspeicher bereitstellen. In diesem Beispiel sind die FAM- und GPU-Einsätze austauschbar ausgelegt, so dass es möglich ist, Rechenleistung gegen mehr Speicher zu tauschen oder umgekehrt.
  • Als ein weiteres Beispiel sei ein Multi-GPU-System der in 6 gezeigten Art mit einer Reihe von Rechen-GPU-Basisplatten betrachtet. Die 7 und 8 zeigen, dass viele andere Konfigurationen (z.B. Mischungen von GPU-Basisplatten und FAM-Basisplatten) möglich sind. So ist es z.B. möglich, ein oder mehrere GPU-Basisplatten auszulagern und durch ein oder mehrere FAM-Basisplatten zu ersetzen. Die FAM-Speicherkapazität ist viel größer als es mit CPU-DRAM oder HBM (Videospeicher) erreicht werden kann, und die FAM-Speicherbandbreite ist viel größer als es mit PCIe zu sysmem bzw. Hauptspeicher möglich ist. Das Wertversprechen ist höhere Kapazität und höhere Bandbreite als herkömmlicher Systemspeicher bieten kann. Der FAM nutzt die hohe Bandbreite, die niedrige Latenz und die hohe Skalierbarkeit von NVLINK™+NVSWITCH™ oder anderen Interconnect-Fabrics mit hoher Bandbreite. FAM liefert mehr Speicher, ohne den Kunden dazu zu zwingen, mehr GPUs oder CPUs zu kaufen. FAM kann darüber hinaus im Einsatz virtualisiert werden - so ist es beispielsweise möglich, pro virtueller Maschine eine „Scheibe“ bzw. einen „Slice“ von FAM zuzuweisen (jede Rechen-GPU kann eine oder mehrere virtuelle Maschinen unterstützen).
  • Beispielhafte nicht beschränkende Adresszuordnungen/ Adresstransformationen
  • In der aktuellen GPU-Architektur wird Hardware bereitgestellt, um zwischen der virtuellen Speicheradresse einer Anwendung und einer physischen Speicheradresse zu übersetzen. Genauer gesagt wird in einer nicht beschränkenden Ausführungsform eine lineare Fabric-Adresse (Fabric Linear Address, FLA) über die Fabric-Interconnect und somit innerhalb eines Adressraums bereitgestellt, der von GPUs in bzw. auf verschiedenen Basisplatten (Knoten) verwendet wird, die über Lesevorgänge/Schreibvorgänge/atomische Funktionen miteinander kommunizieren. Siehe zum Beispiel die US-Anmeldung Nr. 16/198,649 , eingereicht am 21.11.18, mit dem Titel „Distributed Address Translation In A Multi-Node Interconnect Fabric“, die die Implementierung eines linearen Fabric-Adress (FLA)-Raums offenbart, um einen globalen virtuellen Adressraum bereitzustellen, in welchen verschiedene Verarbeitungsknoten einen oder mehrere Bereiche lokalen physischen Speichers eindeutig abbilden können (siehe die nachstehende Diskussion der Adresszuordnung). Auf diese Weise kann auf den gemeinsam genutzten lokalen physischen Speicher an einem bestimmten Verarbeitungsknotens von jedem oder mehreren anderen Verarbeitungsknoten über eindeutige und verwaltbare Adressbereiche innerhalb des FLA-Raums zugegriffen werden. Beispielhafte Ausführungsformen hierin nutzen FLA, um es den GPUs 102 zu ermöglichen, über das Interconnect-Fabric hinweg auf durch FAMMs 106 bereitgestellten Speicher zuzugreifen.
  • Wie in 9 gezeigt ist, übersetzt die Source-GPU 102 ihre Adresse von einer Form in eine andere, führt der Switch 104 eine zusätzliche Adressübersetzung durch, und führt die Spender-GPU 106 eine nochmals zusätzliche Adressübersetzung durch. In anderen Ausführungsformen könnte die Adressübersetzung eher von der Source-GPU 102 als von dem Switch 104 und der Spender-GPU 106; eher von dem Switch 104 als von der Source-GPU 102 und der Spender-GPU 106; oder eher von der Spender-GPU 106 als von der Source-GPU 102 und dem Switch 104 durchgeführt werden. Allgemein könnte folglich die Adressübersetzung abhängig von Anwendung und Kontext von der einen oder der anderen GPU und/oder von dem Interconnect-Fabric selbst durchgeführt werden.
  • Wie nachstehend erläutert wird, stellen beispielhafte Ausführungsformen des Interconnect-Fabric und/oder Schnittstellen dazu Hardware bereit, die mehrere verschiedene Arten von Adresstransformationen durchführt:
    • (1) eine als „Durchmischen“ bzw. „Swizzle“ bezeichnete Transformation verwendet Entropie, um auszuwählen, welche NV-Links des Interconnect-Fabric eine Source-GPU 102 verwendet, um eine Speicherzugriffsanforderung über das Interconnect-Fabric zu kommunizieren oder zu „sprühen“ (die „Durchmischung“ bestimmt das Sprühmuster) - wodurch sichergestellt wird, dass die Source-GPU nicht auf einer bestimmten Verbindung „kampiert“, sondern ihre Zugriffsanforderungen über alle verfügbaren Verbindungen verteilt; und
    • (2) eine als „Verdichtung“ bezeichnete Transformation, welche die Löcher in dem Speicherraum, die durch die Adressraum-Verschachtelung erzeugt wurden, verdichtet, wodurch das FAMM effizienter genutzt wird. Die Verdichtung berücksichtigt Größenunterschiede zwischen dem Adressraum einer Source-GPU 102 und dem Adressraum eines Fabric-Attached-Speichers, indem sie die Adresse, die die Source-GPU erzeugt, in einen Bereich von Adresswerten, die das FAMM 106 aufnehmen kann, teilt oder „quetscht“ (oder in anderen Ausführungsformen multipliziert/erweitert).
  • Die vorstehenden Transformationen sind theoretisch unabhängig (eine könnte ohne die andere verwendet werden), aber falls „Durchmischung“ verwendet wird, um die Adressen der Source-GPU 102 für die Zwecke der Verbindungsauswahl zu transformieren, muss dieselbe oder eine andere Komponente (z.B. der Switch 104 und/oder das FAMM 106) in einer nicht beschränkenden Ausführungsform die Adresse unter Verwendung desselben Algorithmus wie die Source-GPU vor der Adressverdichtung durchmischen, um eine Eins-zu-Eins-Entsprechung zwischen Adressen und eindeutigen Speicherplätzen in dem FAM-Adressraum zu erhalten. In nicht beschränkenden Beispielen führt der Fabric-Switch dieselbe Durchmischung durch wie die Source-GPU, und erfolgt die Verdichtung auf einer nicht durchmischten Adresse. Das in der Source-GPU durchgeführte Durchmischen randomisiert die Verbindungsauswahl für eine bestimmte Adresse, ändert aber nicht die tatsächliche auf NVLINK™gesendete Adresse, die der Switch-Port sieht.
  • Sprüh- und Durchmischungs-Entropie-basierte Adressumwandlung
  • In Übereinstimmung mit einem weiteren beispielhaften, nicht beschränkenden vorteilhaften Merkmal kann eine Source-GPU 102 die volle Inter-GPU-Kommunikationsbandbreite für den Zugriff auf Fabric-Attached-Speicher nutzen, indem die Fabric-Attached-Speicher-Zugriffe über mehrere Spender-Fabric-Attached-Speicher hinweg verschachtelt werden. Die Source-GPU ist somit in der Lage, Speicherzugriffe über mehrere Verbindungen/Interconnects des an sie angeschlossenen Fabric zu „sprühen“ (zu verschachteln), um auf einen angeschlossenen Speicherpool über eine Vielzahl von Spender-Speichercontroller-Hardwareeinheiten zuzugreifen.
  • 10 zeigt, wie eine Source-GPU 102 ihre Speicherzugriffsanforderungen über mehrere (in diesem Fall 12) verschiedene Interconnect-Verbindungen 106 verschachteln oder „sprühen“ kann. In beispielhaften, nicht beschränkenden Ausführungsformen transportiert jede solche Verbindung/Interconnect eine Teilmenge des Adressraums, die als eine „Adressenebene“ bezeichnet wird. Diese Adressen leiten die Daten an verschiedene FAM-Spender-Hardware 106 auf den Fabric-Attached-Speicher-Verbindungsschaltkreisen. Dadurch kann die Source-GPU ihre volle Bandbreite nutzen, während der Zugriff auf den gesamten FAM-Adressraum erhalten bleibt. Während das Sprühen zu einer Leistungsverbesserung im Vergleich zur Verwendung einer festen Schrittweite über eine Anzahl N verschiedener Verbindungen beitragen kann, könnten andere beispielhafte, nicht beschränkende Implementierungen feste oder variable Schrittweiten wählen, je nach Anwendung und damit verbundenem Bedarf.
  • Im Einzelnen können Zugriffsmuster einer gegebenen Source-GPU 102 potenziell sehr regelmäßig oder sehr unregelmäßig sein, abhängig von der Arbeit, die die GPU ausführt. Falls das Zugriffsmuster regelmäßig ist, dann könnten, je nachdem, wie die Schrittweite der Speicherzugriffe ist, alle diese Zugriffe am Ende über dieselbe Verbindung 108 ausgehen. Falls keine Vorkehrungen getroffen werden, könnte die Source-GPU 102 am Ende auf bestimmten Verbindungen 108 „Hotspots“ ausbilden, welches einige Verbindungen überlasten könnte, während andere Verbindungen im Leerlauf belassen werden. Wie 10 zeigt, kann zur Lösung dieses Problems der Source-Peer dazu programmiert werden, Daten über eine programmierbare Anzahl von Verbindungen 108 zu „sprühen“. Somit kann die Anordnung zum Beispiel mit nur einem einzigen FAMM 106-Peer arbeiten, aber falls mehr FAMM 106-Peers verfügbar sind, kann die Source-GPU ihre Daten zufällig oder pseudozufällig über irgendeines oder alle dieser FAMMs 106 und zugeordnete Verbindungen sprühen.
  • Das Sprühen hat die Wirkung eines Lastausgleichs von Speicherverkehr über die verschiedenen Verbindungen hinweg, so dass keine überlastet und keine sehr wenig ausgelastet ist. Es gibt verschiedene Wege, dieses Sprühen durchzuführen. Eine Technik besteht zum Beispiel darin, die Adresse zu nehmen und sie zu herumzumischen oder zu „durchmischen“ (siehe 11, Block 210), um Speicherschrittweitenmuster zu eliminieren, die immer wieder auf dieselbe Verbindung oder dieselben Verbindungen 108 treffen, während sie andere Verbindungen selten oder nie benutzen. Während ein solches „Durchmischen“ an sich bekannt ist, implementiert es die vorliegende Technologie in Kombination mit anderen Techniken, um einzigartige Vorteile bereitzustellen.
  • 9 zeigt, dass die Source-GPU 102 eine virtuelle Adresse (VA) generiert, die sie an ihre Speicherverwaltungseinheit (MMU) anlegt. Falls die MMU feststellt, dass auf die virtuelle Adresse über das Interconnect-Fabric zugegriffen werden muss, wird die zugeordnete physische Adresse generiert und „durchmischt“, um eine durchmischte Adresse („swizaddr“) zu erzeugen. Eine Modulo-L-Operation, die auf die durchmischte Adresse anspricht (mit L = Anzahl der Interconnect-Verbindungen, die der Source-GPU 102 zur Verfügung stehen), bestimmt eine Interconnect-Verbindungs-ID („Nvlink_ID“) der Verbindung 108, die verwendet werden wird, um die Zugriffsanforderung auszusenden (in dem gezeigten Beispiel gibt es keine Entsprechung zwischen Adresse und Verbindung - jede beliebige Verbindung kann verwendet werden, um jede beliebige Adresse auszusenden). Ein Multiplexer wird dann verwendet, um die Verbindung 108 im Ansprechen auf die bestimmte Verbindungs-ID auszuwählen, und ein Datensender sendet die Adresse über die ausgewählte Verbindung aus.
  • Als ein praktisches Beispiel sei angenommen, dass die Spender-GPU 102 eine 2-MB-Seite des Speichers durchläuft. In beispielhaften, nicht beschränkenden Anordnungen würde die Source-GPU ihre zugeordneten Speicheranforderungen über ihre miteinander verbundenen Verbindungen bzw. Links verschachteln. Indessen gibt es in beispielhaften, nicht beschränkenden Ausführungsformen Hardwarekomponenten innerhalb der Source-GPU 102, die ein „Kampieren“ auf einer bestimmten Verbindung verhindern, und eine „Durchmischungs“-Funktion, die Adressbits zufällig verteilt, damit die Source-GPU sich nicht auf eine bestimmte Verbindung „konzentriert“ - alles mit dem Ziel, die Nutzung der Verbindungsressourcen durch Verhindern einer Über- oder Unterauslastung einer bestimmten Verbindung des Interconnects zu maximieren. In einer nicht beschränkenden Ausführungsform, die auf Galois-Mathematik basiert, erzeugt eine „Durchmischung“ „Entropie“ durch Heranziehen eines Bereichs von Adressbits, Multiplizieren jedes derselben mit einer Zahl in einem vordefinierten Galois-„String“, Akkumulieren der Produkte über XOR und dann ODER-Verknüpfen des Ergebnisses in einen Bereich von niedrigeren Adressbits, um die durchmischte Adresse zu erzeugen.
  • 10 zeigt ein beispielhaftes Ergebnis eines solchen „Durchmischens“. Es sei angenommen, dass die Source-GPU 102 eine Sequenz (0-23) von aufsteigenden Adressen durch den Adressraum erzeugt, wobei jede Adresse ein 256B großes Segment (oder Segment einer anderen Größe) bzw. einen solchen „Chunk“ oder Block des Speichers adressiert. Die „Durchmischungs“-Operation bewirkt, dass die Speicherzugriffsanforderungen aus mehreren (in diesem Fall zwölf) verschiedenen Verbindungen 108 heraus „gesprüht“ werden. In der beispielhaften Ausführungsform wird eine nicht umgewandelte physische Adresse über die Verbindung gesendet - keine „durchmischte“ oder anderweitig umgewandelte Adresse. So wird z.B. die Anforderung für die Adresse 0 über die Verbindung 0 ausgesendet und wird die Anforderung für die Adresse 1 über die Verbindung 1 gesendet, aber dann wird auch die Anforderung für die Adresse 2 über die Verbindung 1 gesendet. Die Anforderung für die Adresse 3 wird über die Verbindung 11 gesendet, die Anforderung für die Adresse 4 wird über die Verbindung 4 gesendet, die Anforderung für die Adresse 5 wird über die Verbindung 1 gesendet, die Anforderung für die Adresse 6 wird über Verbindung 6 gesendet, die Anforderung für die Adresse 7 wird über die Verbindung 7 gesendet, und die Anforderung für die Adresse 8 wird ebenfalls über die Verbindung 7 gesendet. Und so weiter. Die Anforderungen werden in einer zufälligen oder pseudozufälligen Weise über alle verfügbaren Verbindungen 108(0-11) „gesprüht“, wobei die Anforderungen so auf die verschiedenen Verbindungen verteilt werden, dass keine Verbindung unterausgelastet und keine Verbindung überlastet wird. Durch dieses Durchmischen wird ein schnellerer Zugriff erreicht, weil die Zugriffsanforderungen über die Verbindungen hinweg lastbalanciert werden.
  • In früheren NVIDIA-Architekturen wie beispielsweise VOLTA und TURING wurde ein solches Sprühen auch dann durchgeführt, wenn zwei GPUs 102 peer-to-peer miteinander kommunizierten. In dieser Situation waren jedoch alle Verbindungen von einer GPU 102(a) mit der Peer-GPU 102(b) verbunden. In der beispielhaften, nicht beschränkenden FAM-Ausführungsform hierin, wie sie in 9 dargestellt ist, ist dagegen typischerweise nur eine Teilmenge der Verbindungen von einer Source-GPU 102 mit einem bestimmten FAMM 106 verbunden. Zum Beispiel kann es 12 Verbindungen 108 geben, die von der Source-GPU kommen, aber nur zwei Verbindungen 110, die ein bestimmtes einer Vielzahl von FAMMs 106 verbinden. Die Source-GPU 102 kann daher ihre Anforderungen über mehrere FAMMs 106 hinweg sprühen bzw. verteilen, z.B. auf sechs FAMMs mit jeweils zwei Verbindungen 110. Da jede einzelne Verbindung 110 (zumindest in einigen nicht beschränkenden Ausführungsformen) in der Bandbreite an die GPU-Verbindungen 108 angepasst ist, kommunizieren die 12 Source-GPU-Verbindungen 108 mit 12 FAM-Verbindungen 108, die mit insgesamt sechs verschiedenen FAMMs 108 verbunden sind - wobei in der beispielhaften Ausführungsform alle Verbindungen in der Bandbreite angepasst sind.
  • 11 zeigt die entsprechenden unterschiedlichen Adressräume, wobei der virtuelle Adress (VA)-Raum der Source-GPU 102 als Block 202 dargestellt ist. Die MMU-Übersetzungsoperation (Block 204) übersetzt von der virtuellen Adresse (VA) (Block 202), die von der Anwendung angegeben wurde, in eine Fabric-Lineare (physische) Adresse (FLA) (oder in diesem Fall den FAMLA-Raum, d. h. die lineare Adresse des Fabric-Attached-Speichers) (Block 208), die von dem NVLINK™-Switch-Fabric an das Ziel-FAMM 106 zu transportieren ist.
  • In beispielhaften Ausführungsformen wird eine Speicherseite über FAM-DIMMs hinweg in einem Streifen verteilt, abhängig davon, wie eine „Peer Aperture“ (nachstehend erläutert) programmiert ist und wie das Interconnect Fabric aufgebaut ist. Das Mapping bzw. die Zuordnung kann pro Betriebssysteminstanz physische Volumen mit Funktions-/Leistungs-Isolierung bereitstellen, die jeweils in logische Volumen unterteilt sind. Die Anwendungsschichten können beliebige von verschiedenen Speicherzuweisungsmodellen verwenden. Wie hierin beschrieben wurde, kann die virtuelle Adresse in der MMU (Speicherverwaltungseinheit) einer Source-GPU 102 in die Seite des physischen Adressraums übersetzt werden, die über FAM gestriped wird. Implizit wird der Speicherpool zur Seitenmigration über UVM-Oversubscription (nachstehend beschrieben) erweitert, z.B. unter Verwendung eines Befehls wie beispielsweise cudaMallocManaged(). Wenn Speicher unter Verwendung der cudaMallocManaged()-API alloziert wird, kann er entweder bei Bedarf oder durch die Systemsoftware im Ansprechen auf Richtlinien in den/aus dem in FAM abgebildeten Speicher migriert und evoziert werden. Die Benutzeranwendung bräuchte keine Modifikation, um auf einem physischen System mit FAM zu laufen, und würde lediglich eine höhere Leistung bei GPU-Zugriffen auf einen arbeitenden Satz beobachten, der größer ist als die Kapazität des Speichers der Source-GPU. Explizit können also Befehle wie beispielsweise cudaMallocO und eine neue CUDA-Treiber-API verwendet werden, um FAM als Pinned Memory zu allozieren/deallozieren. Resource Manager (RM)-Programme können die FAM-Parameter der GPU pro Peer-Aperture beziehen. Der Fabric Manager + RM können NVSWITCH™-Routentabellen programmieren. Die Software kann auch dazu verwendet werden, das Ausscheiden von Speicherseiten aufgrund von nicht korrigierbaren Fehlern zu ermöglichen. Die FAM-Spender-Fehlersignalisierung für fatale Fehler in dem Geber oder in dem Speicher selbst kann so gestaltet sein, genügend Informationen liefert, um eine bestimmte Source-GPU und/oder VM zu indizieren, so dass Software „chirurgische“ Aktionen durchführen kann, wobei nur die von den FAM-Fehlern betroffene GPU oder VM heruntergefahren wird, während andere GPUs oder VMs von diesen Aktionen isoliert sind.
  • Die beispielhafte, nicht beschränkende Technologie hierin verwendet das Konstrukt einer „Peer-Apertur“, um es einer Source-GPU 102 zu erlauben, auf den Fabric-Attached-Speicher 106 zuzugreifen. In einigen nicht beschränkenden Beispielen ist „Peer“ eine Sammlung von GPUs mit reduzierter Fähigkeit oder anderen Speichercontrollern, die an eine Fabric-Attached-Speicher-Basisplatte angeschlossen sind. In beispielhaften, nicht beschränkenden Ausführungsformen ist die physische Speicheradresse in der NVLINK™-Architektur mit etwas assoziiert, das als eine „Apertur“ bezeichnet wird. Diese Apertur gibt der GPU 102 ein Fenster (siehe die Angabe „N Bytes in FAM-Slice“ in 11 Block 202) entweder in den Systemspeicher (an die CPU angeschlossen), in den Speicher eines Peers (d.h. Speicher, der an eine andere GPU, d.h. einen Peer, angeschlossen und mit dem NVLINK™-Fabric verbunden ist) oder - im vorliegenden Fall - in den Speicher eines FAMM 106. Die beispielhafte, nicht beschränkende Technologie erweitert somit das Konzept einer Peer-Apertur bzw. Öffnung für den Zugriff auf ein oder eine Sammlung von FAMMs 106. Wie 11 zeigt, identifiziert der PTE, dass das Ziel ein Peer ist, und zeigt auf eine bestimmte Peer-Apertur, z.B. eine von 8. Ein Peer-Apertur-Register kann verwendet werden, um die Anzahl interner Hubs der GPU 102, die zum Sprühen des Datenverkehrs zu verwenden sind, sowie die Anzahl der NV-Links 108, die auf jedem solchen internen Hub zu verwenden sind, zu speichern/zu identifizieren. Ein Pro-Peer-Apertur-Register steuert somit die „Sprüh“-Breite.
  • Beispielhafte, nicht-begrenzende Durchmischung & Verdichtung
  • Wie vorstehend erläutert wurde, verwenden beispielhafte, nicht beschränkende Ausführungsformen Entropie, um Speicherzugriffe über mehrere FAMMs 106 und zugeordnete Verbindungen 110 hinweg zu verschachteln. In Abwesenheit jeglicher Funktionalität, die von dem Switch 104 bereitgestellt wird, würde das Ergebnis mehrere FAMM-Ströme sein, von welchen jeder auf 1/N des Adressraums zugreift, wobei N die Anzahl der FAMMs ist. Dies würde implizieren, dass ein bestimmtes eines der FAMMs 106 jede N-te Adresse erhalten wird. Ohne Maßnahmen zur Modulation des Adressstroms, der an ein bestimmtes FAMM gerichtet ist, könnte dies zu einer geringen Auslastung auf den FAMMs 106 führen, d.h. die Auslastung könnte 1/N-tel der Kapazität betragen, zu der das einzelne FAMM in der Lage ist. Dies wäre eine verschwenderische Nutzung der FAM-Speicherkapazität.
  • Genauer gesagt, führt das Sprühen, das eine Eins-zu-Eins-Wiederzuordnung bzw. Remapping von einer ursprünglichen globalen FAM-Adresse zu einer Interconnect-Adresse einschließlich einer Verbindungs-ID bereitstellt, dazu, dass die ursprünglichen Adressen in (einem) schrittlosen/unregelmäßigen Intervall(en) in verschiedene „Eimer“ bzw. „Buckets“ fallen. Falls die ursprüngliche Adresse in dem Bereich von 1 ... X liegt und die Interconnect-Adresse ebenfalls in dem Bereich von 1 ... X liegt, dann kann der Interconnect-Adressraum in Segmente unterteilt werden, die auf den lokalen Adressraum jedes FAMM 106 abgebildet werden.
  • Es sei beispielsweise angenommen, dass das Fabric die ursprüngliche Adresse (z.B. im Bereich von 0-12 GB liegend) erhält und der lokale Adressraum eines FAMM 106 viel kleiner ist (z.B. im Bereich von 0-2 GB liegt). Aufgrund der Durchmischung der Originaladresse und der Auswahl eines FAMM 106 auf der Grundlage der durchmischten Adressen würde das Ergebnis sein, dass ursprüngliche Adressen im Umfang von 2 GB an einen einzigen FAMM 106 gesendet werden, wobei die Adressen unregelmäßig verteilt sind. Zum Beispiel könnte ein FAMM 106 die ursprünglichen Adressen bzw. Originaladressen 0KB, 256KB, 320KB, 448KB usw. erhalten, aber niemals die Originaladressen 64KB, 128KB, 192KB, 384KB usw., unter der Annahme, dass die Adressen auf 64KB-Grenzen fallen.
  • Um diese ineffiziente Speichernutzung zu verhindern, werden in einigen beispielhaften, nicht beschränkenden Ausführungsformen an dem FAMM 106 oder in einem Switch 104 oder einem anderen Element, das Teil des Interconnection Fabric ist, die Originaladressen auf den lokalen Adressraum des FAMM als 0KB, 64KB (entsprechend der Originaladresse 256KB), 128K (entsprechend der Originaladresse 320KB), 194K (ursprünglich 448KB) usw. neu abgebildet (verdichtet). Oder es wird eine andere Art der Neuzuordnung von Originaladressraum zu FAMM-Adressraum verwendet, um sicherzustellen, dass auf den gesamten verfügbaren FAMM-Speicheradressraum unter Verwendung einer globalen Originaladresse zugegriffen werden kann.
  • In einigen beispielhaften, nicht beschränkenden Ausführungsformen, wie in 9 gezeigt, teilt der Switch 104 die Adresse auf und entfernt „Löcher“, um einen linearisierten Adressraum bereitzustellen, der mit dem Adressraum eines Fabric-Attached-Speichers 106 übereinstimmt oder zumindest „hineinpasst“. Der Switch 104 in einigen nicht beschränkenden Ausführungsformen ist folglich in der Lage, einen verschachtelten Strom heranzuziehen, der nur eine von L Adressen hat, und ihn durch die Anzahl von Verbindungen L („divL“) herunterzuteilen, so dass der herauskommende Adressstrom linear ist, d. h. aus aufeinanderfolgenden Adressen besteht. Dadurch wird die volle Ausnutzung der FAM-Speicherkapazität erreicht. Eine beispielhafte, nicht beschränkende Ausführungsform verwendet folglich einen Prozess des Heranziehens einer Adresse und des Manipulierens derselben auf der Grundlage programmierter FAM-Parameter derart, dass das Ziel-FAM-Modul einen linearen Adressraum ohne Löcher sieht.
  • 12 veranschaulicht die Wirkung der Durchmischungs-Funktion auf Adressen und insbesondere, wie die Durchmischungs-Funktion in der Source-GPU 102 effektiv den Strom von Adressen moduliert, die an einem gegebenen Switch-Port erscheinen. 12 zeigt, dass die Durchmischung der Source-GPU 102, der in dem Switch 104 oder einem anderen Teil des Interconnects repliziert wird, den Raum, der abgebildet werden muss, effektiv „aufbläht“, um die durch die Durchmischung eingeführte Zufälligkeit zu berücksichtigen, und dann die Verdichtung verwendet wird, um die Adresse in dem linearen FAM-Adressraum (FAMLA) durch L (die Anzahl der Verbindungen in dem Sprühen) zu teilen, um Adressen in den verdichteten FAM-Adress (FAMCA)-Raum zu transformieren. Mit anderen Worten ist die Verdichtungsoperation, die an jeder durchmischten Adresse durchgeführt wird:
    • FAMLA/L.
  • Zur weiteren Erläuterung führt in beispielhaften Ausführungsformen wie in 9 gezeigt der Switch 104 ein Durchmischen aus, um dasselbe Durchmischen zu erzeugen, das in der Source-GPU vor der Adressverdichtung durchgeführt wird. In beispielhaften Ausführungsformen muss der Switch 104 vor der Verdichtung mit der GPU-Durchmischung übereinstimmen. Wie vorstehend erläutert wurde, basiert die in dem Interconnect-Fabric durchgeführte Durchmischung auf der Durchmischungsfunktion der Source-GPU - d.h. einer Funktion, die in der Source-GPU 102 durchgeführt wird, um Adressen nach dem Zufallsprinzip über einen Satz von Pfaden hinweg zu verteilen, um eine ungleiche Nutzung eines bestimmten Pfads oder einer Teilmenge von Pfaden (Kampieren) aufgrund von Merkmalen des Adressschrittmusters zu reduzieren. In nicht beschränkenden Ausführungsformen hierin kehrt das Interconnect-Fabric (oder, in anderen Ausführungsformen, das FAMM) die Durchmischung um, die die Source-GPU 102 vor der Verdichtung durchführt, um Adresskonflikte zu verhindern.
  • In beispielhaften, nicht beschränkenden Ausführungsformen erzeugt die anfängliche Durchmischung, die von der Source-GPU (absichtlich) durchgeführt wird, eine nicht lineare Verteilung von Adressen über die verschiedenen Verbindungen hinweg. Die Adresse, die auf irgendeine bestimmten Verbindung platziert wird, ist jedoch die ursprüngliche Adresse. Falls die bereits durchmischten Adressen einfach verdichtet wurden, ohne zu berücksichtigen, bereits zufällig oder anderweitig ungleichmäßig über den Adressraum hinweg verteilt wurden, werden zumindest einige Arten der Verdichtung Kollisionen verursachen.
  • In einer beispielhaften Ausführungsform ist die von dem Switch 104 empfangene Adresse die unverarbeitete (nicht durchmischte) Adresse. Vor der Verdichtung muss der Switch 104 die Adresse mit der GPU übereinstimmend transformieren (durchmischen), um die Adresse in die korrekte Form zum Erzeugen einer bijektiven Zuordnung zu erzeugen.
  • Als eine beispielhafte Adresstransformation und Adressverdichtung sei angenommen, dass es nur eine Source-GPU 102 gibt, die Zugriffe auf eine Vielzahl von FAM-Gebern 106 generiert, z.B. über 12 Verbindungen 108 zu sechs FAMMs 106, von welchen jeder mit einem Paar (2) von Verbindungen 110 verbunden ist. Es sei angenommen, dass die Source-GPU 102 eine Adresse von z.B. zwischen 0 und 12 GB erzeugt. Es sei angenommen, dass jedes FAMM 106 über 2 GB Speicher verfügt. In diesem Fall wird die von der Source-GPU 102 erzeugte Adresse im Bereich von 0-12 GB liegen, während der Adressbereich jedes der Spender 106 im Bereich von 0-2 GB liegt. In einigen beispielhaften, nicht beschränkenden Ausführungsformen wird die Source-GPU 102 die Verteilung von Anforderungsübertragungen über ihre 12 Verbindungen 108(0)-108(11) hinweg randomisieren, um die Auslastung der 12 Verbindungen auszugleichen. Unter der Annahme, dass die Anforderung ein Speicherlesevorgang oder Speicherschreibvorgang ist, wird sie die Speicheradresse auf die ausgewählte Verbindung gelegt, wobei diese Speicheradresse eine Adresse innerhalb des 0-12 GB-Adressraums spezifiziert. Diese spezielle ausgewählte Verbindung ist jedoch nur mit einem FAMM 106 verbunden, das einen Adressraum von 0-2 GB hat.
  • Daher greift in einer beispielhaften, nicht beschränkenden Ausführungsform, wie in 9 gezeigt, ein zwischengeschalteter Switch 104 auf mit ihm verbundene FAMMs 106 und darauf, welche Adressbereichserwartungen diese FAMMs haben, zu. Ein FAMM 106 hat in diesem speziellen Fall die Erwartung, eine Adresse mit 0-2 GB zu erhalten, so dass der Switch 104 die ursprüngliche Adresse so umwandelt, innerhalb des Adressraums des FAMM liegt. Falls z.B. die Source-GPU 102 eine Adresse in dem Bereich von 6 GB erzeugt, ist es wünschenswert, dass der Switch 104 sie so transformiert, innerhalb des Bereichs von 0-2 GB liegt, den ein FAMM erwartet. Der Switch 104 stellt also anhand der von der Source-GPU 102 empfangenen Adresse und der Verbindung 108, über die sie empfangen wird, fest, dass die Anforderung für ein FAMM 106(i) bestimmt ist und dass das FAMM 106(i) eine Adresse innerhalb eines Bereichs von 0-2 GB erwartet. Der Switch 104 ändert daher die Adresse so, in das Speicheradressfenster des FAMM 106(i) passt, das die Zugriffsanforderung bearbeiten wird. Siehe die 11 & 12.
  • Wie in 9 gezeigt ist, „macht“ der Switch 104 die randomisierte Verbindungsauswahl/Durchmischung „rückgängig“, die von der Source-GPU 102 durchgeführt wurde, bevor diese Verdichtung (Teilung) stattfindet. Andernfalls können die Nichtlinearitäten in der Verbindungsauswahl der ursprünglichen Source-GPU 102 zu Speicheradressenkollisionen führen. Dies liegt daran, dass es auf der Seite der Source-GPU 102 keine lineare Aufteilung von Speicheradressen gibt, so dass alle Speicherzugriffe innerhalb eines ersten Speicherbereichs über eine erste Verbindung 108(00) gesendet werden, alle Speicherzugriffe innerhalb eines zweiten Speicheradressbereichs über eine zweite Verbindung 108(01) gesendet werden, und so weiter. Demgegenüber verhindert in einigen Ausführungsformen die Source-GPU 102 absichtlich, dass irgendeine solche 1-zu-1-Entsprechung auftritt, um die Auslastung der verschiedenen Verbindungen 108 auszugleichen. Weil die Source-GPU 102 die Verbindungsauswahl zufällig vornimmt, kann jede ihrer verbundenen Verbindungen potenziell eine Adresse innerhalb des Bereichs von 0-12 GB sehen. In der beispielhaften, nicht beschränkenden Ausführungsform geht jedoch nicht jede Verbindung 108 zu jedem Spender 106, da in einigen beispielhaften, nicht beschränkenden Ausführungsformen die Nichtsymmetrie zwischen der (z.B. in einigen Ausführungsformen größeren) Anzahl von Verbindungen 108, die mit der Source-GPU 102 verbunden sind, und der (z.B. in solchen Ausführungsformen kleineren) Anzahl von Verbindungen 110, die mit jedem FAMM 106 verbunden sind, besteht.
  • In einer beispielhaften, nicht beschränkenden Ausführungsform kann der Switch 104 eine Adressdurchmischung und dann eine Verdichtung in einem Eingangsmodul des Switches 104 für Zugriffsports (die mit der Source-GPU 102 verbunden sind) durchführen. Ein solches Eingangsmodul kann Routingtabellen (siehe 13) beinhalten, die mit den Parametern des FAM-Ziels und des Fabrics zwischen der Source-GPU 102 und dem Ziel-FAMM 106 programmiert sind. Der Port des Switches 104 an dem ersten Hop (der mit der Source-GPU 102 verbunden ist) verwendet diese Informationen, um die Interconnect-Adresse, die er empfängt, so zu manipulieren, dass jedem FAMM 106 ein linearer Adressstrom präsentiert wird. Zusätzlich zu der „Verdichtung“ der Adresse (z.B. durch Division durch die Anzahl der Verbindungen, über welche die Source-GPU Anforderungen verschachtelt) kann der Switch 104 auch ein Versetzen bzw. Offsetting (Hinzufügen eines festen Versatzes zu der verdichteten Adresse) und maskierte Rewrite-bzw. Neuschreibe-Operationen auf Teile der Adresse anwenden. Diese Operationen können zur Relokation nützlich sein, wenn der FAM von mehreren Gästen in einem virtualisierten System gemeinsam genutzt wird. Das FAMM 106 kann auch dazu konfiguriert sein, dass es eine Adressvalidierung und Adressübersetzung der eingehenden Adresse durchführt, falls eine Fabric Linear Address (FLA)-Funktion in dem Fabric aktiviert ist.
  • Beispielhaftes Interconnect-Fabric-Routing
  • In beispielhaften, nicht beschränkenden Ausführungsformen stellt der Switch 104 Routingtabellen bereit, die verwendet werden, um die physische Adresse, die die Source-GPUs 102 bereitstellen, über das Interconnect-Fabric abzubilden. Diese Routingtabellen stellen ein Routing zu Bestimmungs-FAMM-106-Zielen, die durch software-spezifische „TgtID“-Informationen bezeichnet werden, bereit sowie wie die Verdichtung durchzuführen ist. Solche Routingtabellen basieren in einer beispielhaften Ausführungsform auf „Map-Slot“-Einträgen - und zwar auf solchen Map-Slot-Einträgen in einer Eingangsportrouten-Neuzuordnungstabelle 1302 eines Ebene-1-Switches 104. 13 ist ein Blockdiagramm, das solche beispielhaften Routingtabellen zeigt, welche verwendet werden können, um ein Daten-Striping wie vorstehend beschrieben zu ermöglichen und auch um sicherzustellen, dass Adresstransformationen durch das Interconnect-Fabric geeignet gehandhabt werden. In einigen nicht beschränkenden Ausführungsformen wird die Neuzuordnung von Routingtabellenadressen verwendet, um konvergente Ebenen zu disambiguieren.
  • In dem Beispiel von 13 unterhält der Switch 104 eine Eingangsport-Routingtabelle 1514 für jede eingehende Verbindung 108. Die Routingtabelle 1514 enthält Informationen, die verwendet werden, um eingehende physische Adressen auf FAMMs 106 abzubilden - einschließlich Steuerinformationen für die selektive Durchführung einer Durchmischung/Verdichtung. Solche Informationen werden in der Eingangsroutingtabelle 1514 bereitgestellt, weil der Switch 104 in einer Ausführungsform nicht für den Zugriff auf ein FAMM 106 vorgesehen ist und daher dazu gesteuert wird, selektiv die vorstehend beschriebenen Transformationen durchzuführen oder bestimmte Transformationen nicht durchzuführen, abhängig davon, ob eine Speicherzugriffsanweisung für ein FAMM oder für einen anderen Peer als ein FAMM bestimmt ist. Zusätzlich kann in beispielhaften, nicht beschränkenden Implementierungen die Routingtabelle 1514 des Switches 104 verwendet werden, um festzulegen, ob Basis- und/oder Grenzwertprüfungen an der eingehenden Adresse durchgeführt werden sollen (dies ist als ein Sicherheitsmerkmal in Fällen von FAMM 106-Partitionen nützlich, die unregelmäßig große Speicherkapazitäten bereitstellen, z.B. 46 GB im Gegensatz zu 64 GB, falls 64 GB der Bereich ist, der durch den Map-Slot abgebildet wird, um sicherzustellen, dass es keinen unberechtigten Zugriff auf FAM gibt). Wie vorstehend erörtert wurde, durchläuft die durchmischte Adresse eine modL-Funktion, worin L die Anzahl von Verbindungen ist, über welche ein Sprühen erfolgt. In einem bestimmten, nicht beschränkenden Beispiel kann das Durchmischen daher den Adressbereich, der von einem gegebenen Port gesehen wird, um bis zu 2^(N+1) - 1 Bytes (siehe 12), über den Bereich hinaus vergrößern, den eine reguläre (nicht durchmischte) Verschachtelung über die Ports erzeugen würde, worin N der Index des Bits höchster Ordnung ist, das von der Durchmischung manipuliert wird, um einen Port auszuwählen. In einem beispielhaften, nicht-beschränkenden Beispiel hängt N von L ab und überschreitet nie 11, so dass die maximale Erhöhung der Map-Slot-Grenze bzw. des Map-Slot Limits 2^ 12 = 4 KB beträgt. Dieser „Slop“ bzw. „Schlick“ im Grenzwert kann bei der Software-Speicherzuweisung berücksichtigt werden. Praktisch gesehen impliziert jedoch dann, wenn in dieser speziellen, nicht beschränkenden beispielhaften Ausführungsform die Felder für Map-Slot-Basis und Map-Slot-Limit eine Granularität von 2^20 = 1 MB haben, bedeutet dies, dass der „Slop“ wirklich 1 MB beträgt.
  • Außerdem zusätzlich verwendet eine beispielhafte, nicht beschränkende Ausführungsform die Routingtabellen 1514, um einen „Shuffle“ bzw. „Mischen“-Modus zu programmieren, um ein perfektes Mischen von verdichteten Adressen von mehreren Verbindung 108-Ports durchzuführen, die unterschiedliche (z.B. gerade und ungerade) Speicherebenen bedienen und deren Verkehr auf demselben FAMM 106 konvergiert, um Kollisionen in Adressen von den mehreren Ports zu verhindern. Die Verwendung von „Mischen“ kann die Anzahl offener Seiten in DRAM reduzieren. Eine alternative oder zusätzliche Technik besteht darin, den programmierbaren Versatz in den Routingtabellen 1514 zu verwenden, der selektiv angewendet (z.B. hinzugefügt) werden kann, um verschiedene feste Partitionen in dem Adressraum eines FAMM 106 zu erstellen, wobei die verschiedenen Partitionen verschiedenen Verbindung 108-Ports entsprechen.
  • Wie in den 14A & 14B gezeigt ist, können die Routingtabellen innerhalb des Switches 104, die auf aufeinanderfolgenden Map-Slots basieren, verwendet werden, um den gesamten Adressraum des DIMMs mit nutzbarer DIMM-Kapazität innerhalb eines FAMM 106 abzubilden. In dem in 14A gezeigten Beispiel verwendet der Switch 104 die Routingtabelleninformationen, um Adressen, die Anforderungen zugeordnet sind, die an zwei verschiedenen eingehenden Verbindung 108-Ports empfangen werden, dem DIMM-Adressraum desselben FAMM 106 zuzuordnen. Dieses Diagramm zeigt der Einfachheit in der Darstellung halber die Programmierung von Map-Slots ohne Durchmischung oder Verdichtung. Die Basis- und/oder Limit- bzw. Grenzwertprüfung kann jedoch für MS0_C und MS1_Z durchgeführt werden, da diese Zuordnungen in diesem speziellen Beispiel für weniger als einen „vollen“ (in diesem Fall 64 GB) Bereich gelten. Daher kann die Basis-/Grenzwert-Prüfung Map-Slot-weise deaktiviert werden, und wird für FAM erwartet, für alle Map-Slots, die das FAM-Ziel vollständig abbilden, deaktiviert wird; sie wird in einigen Ausführungsformen nur für Map-Slots aktiviert, für welche der 64 GB-Bereich durch das zugrunde liegende FAMM nicht vollständig abgebildet wird.
  • Hingewiesen wird auf die beispielhaften Map-Slot-Versätze (welche zu den physischen Adressen addiert werden können) für die durch MS1 X, MS1 Y und MS1_Z in den gezeigten Beispielen spezifizierten Zuordnungen, damit die Zuordnung den maximalen DIMM-Bereich überspannen kann (in einem bestimmten Beispiel 1 TB mit 16 MB Granularität). Es könnte eine effizientere Packung des Adressraums vorgenommen werden - dies ist nur ein Beispiel.
  • 15 veranschaulicht ein einfaches Beispiel dafür, wie Map-Slots für eine gegebene Eingangsroutingtabelle des Switches 104 aus dem linearen FAM-Adressraum (FAMLA), der von der Source-GPU gesehen wird, und dem verdichteten FAM-Adressraum (FAMCA), der an dem Verbindung 110-Eingang zu dem FAMM 106 gesehen wird, abgebildet werden könnten. In diesem Beispiel findet keine Verdichtung von FAMLA zu FAMCA statt, weil die Source-GPU 102 nur eine einzige Verbindung zur Kommunikation mit diesem speziellen FAMM 106 verwendet.
  • In beispielhaften, nicht beschränkenden Ausführungsformen können die Switch-Routingtabellen ferner softwareprogrammierbare Bestimmungs-Ziel-ID („TgtID“)-Felder enthalten, die Bestimmungs-FAMMs 106 für bestimmte Adressbereiche spezifizieren/zuweisen. 16 zeigt ein Beispiel, in dem die Source-GPU 102 Verkehr über 12 Verbindungen 108 sprüht bzw. verteilt, was bedeutet, dass der Switch 104 verdichten muss, um den linearen FAMLA-Adressraum in den verdichteten FAMCA-Adressraum zu transformieren. In diesem Beispiel können aufeinanderfolgende Map-Slots für jeden Ebene-1-Switch 104 auf der GPU-Basisplatte programmiert werden, wobei jeder Ebene-1-Switch Verkehr über zwei seiner Ausgangsports 108 (gerichtet durch eine „TgtID“-Programmierung in dem Map-Slot) an eine gegebene Spalte eines 6-breiten Slice von FAM, das der Source-GPU zugewiesen ist, ausgibt.
  • 17 zeigt ein Beispiel dafür, wie die „TgtID“-Map-Slot-Programmierung den FAMMs 106 auf einer FAM-Basisplatte zugewiesen werden könnte, wobei (in diesem speziellen Beispiel) 48 FAMMs106 angenommen werden, bei denen jedem FAMM 106 ein eindeutiger TgtID-Wert zugewiesen wird, der in die Map-Slot-Routingtabelle des L1-Switches programmiert ist.
  • Beispielhafte, nicht beschränkende Parallelverarbeitungs-GPU-Architektur zur Durchführung der vorstehend beschriebenen Operationen und Verarbeitung
  • Nachstehend wird eine beispielhafte veranschaulichende Architektur beschrieben, die von Fabric-Attached-Speicher profitieren kann und in welcher die vorstehenden Techniken und Strukturen implementiert sein können. Die folgenden Informationen dienen der Veranschaulichung und sollten als in keiner Weise beschränkend verstanden werden. Jedes der folgenden Merkmale kann optional mit oder ohne Ausschluss anderer beschriebener Merkmale integriert sein.
  • 18 veranschaulicht, dass die in 1 gezeigte GPU 102 als ein Multi-Threaded-Multi-Core-Prozessor implementiert sein kann, der auf einer oder mehreren integrierten Schaltungsvorrichtungen implementiert ist. Die GPU 102 ist eine latenzverbergende Architektur, die dazu ausgelegt ist, viele Threads parallel zu verarbeiten. Ein Thread (z.B. ein Ausführungsthread) ist eine Instanziierung eines Satzes von Anweisungen, die zur Ausführung durch die GPU 102 konfiguriert sind. In einer Ausführungsform ist die GPU 102 dazu konfiguriert, eine Grafik-Rendering-Pipeline zur Verarbeitung von dreidimensionalen (3D) Grafikdaten zu implementieren, um zweidimensionale (2D) Bilddaten für die Anzeige auf einer Anzeigevorrichtung, wie z.B. einer Flüssigkristallanzeige (LCD)-Vorrichtung, zu erzeugen. In anderen Ausführungsformen kann die GPU 102 für die Durchführung von Universalberechnungen verwendet werden.
  • Wie vorstehend erörtert wurde, können ein oder mehrere GPUs 102, wie gezeigt, dazu konfiguriert sein, Tausende von High Performance Computing (HPC)-Anwendungen, Rechenzentrumsanwendungen und Anwendungen maschinellen Lernens zu beschleunigen. Die GPU 102 kann dazu konfiguriert sein, zahlreiche Deep-Learning-Systeme und -Anwendungen zu beschleunigen, darunter autonome Fahrzeugplattformen, Deep Learning, hochpräzise Sprach-, Bild- und Texterkennungssysteme, intelligente Videoanalyse, molekulare Simulationen, Arzneimittelentdeckung, Krankheitsdiagnose, Wettervorhersage, Big-Data-Analyse, Astronomie, Molekulardynamiksimulation, Finanzmodellierung, Robotik, Fabrikautomatisierung, Echtzeit-Sprachübersetzung, Online-Suchoptimierung und personalisierte Benutzerempfehlungen und dergleichen.
  • Wie in 18 gezeigt ist, beinhaltet die GPU 102 eine Eingabe/Ausgabe- bzw. Input/Output (I/O)-Einheit 305, eine Frontend-Einheit 315, eine Scheduler-Einheit 320, eine Arbeitsverteilungseinheit 325, einen Verteiler bzw. Hub 330, eine Kreuz- bzw. Querschiene (Crossbar; Xbar) 370, einen oder mehrere Universalverarbeitungscluster bzw. General Processing Cluster (GPC) 350 und eine oder mehrere Partitionseinheiten 380. Die GPU 102 kann über eine oder mehrere schnelle NVLINK™310-Verbindungen, die ein Interconnect-Fabric mit Fabric-Attached-Speicher wie vorstehend erörtert ausbilden, mit einem Host-Prozessor oder anderen PPUs 300 verbunden sein. Die GPU 102 kann über eine oder mehrere weitere Interconnect-Verbindung(en) 302 mit einem Host-Prozessor oder anderen Peripheriegeräten verbunden sein (siehe 2). Die GPU 102 kann darüber hinaus mit einem lokalen Hochleistungsspeicher verbunden sein, der eine Anzahl von Speichervorrichtungen 304 umfasst. In einer Ausführungsform kann der lokale Speicher eine Anzahl von dynamischen DRAM (Dynamic Random Access Memory)-Vorrichtungen umfassen. Die DRAM-Vorrichtungen können als ein HBM (High-Bandwidth Memory)-Subsystem konfiguriert sein, wobei mehrere DRAM-Dies innerhalb jeder Vorrichtung gestapelt sind. Dieselben oder vergleichbare solche Speichervorrichtungen sind in jedem FAMM 106 enthalten.
  • Die NVLINK™ 108-Verbindung ermöglicht es Systemen, zu skalieren, und umfasst eine oder mehrere PPUs 300, die mit einer oder mehreren CPUs 150 kombiniert sind, unterstützt Cache-Kohärenz zwischen den PPUs 300 und CPUs sowie CPU-Mastering. Daten und/oder Befehle können durch den NVLINK™ 108 über den Hub 330 zu/von anderen Einheiten der GPU 102, wie z.B. einer oder mehreren Kopier-Engines, einem Video-Encoder, einem Video-Decoder, einer Energieverwaltungseinheit usw. (nicht explizit gezeigt), übertragen werden. Der NVLINK™ 108 wird in Verbindung mit 22 näher beschrieben.
  • Die E/A-Einheit 305 ist dazu konfiguriert, Kommunikationen (z.B. Befehle, Daten usw.) von einem Host-Prozessor 150 über die Verbindung 302 zu senden und zu empfangen. Die E/A-Einheit 305 kann mit dem Host-Prozessor 150 direkt über die Verbindung 302 oder über ein oder mehrere Zwischenvorrichtungen wie z.B. eine Speicherbrücke kommunizieren. In einer Ausführungsform kann die E/A-Einheit 305 mit einem oder mehreren anderen Prozessoren, wie beispielsweise einer oder mehreren der PPUs 300, über die Verbindung 302 kommunizieren. In einer Ausführungsform implementiert die E/A-Einheit 305 eine Peripheral Component Interconnect Express (PCIe)-Schnittstelle für die Kommunikation über einen PCIe-Bus und ist die Verbindung 302 ein PCIe-Bus. In alternativen Ausführungsformen kann die E/A-Einheit 305 andere Arten von gut bekannten Schnittstellen zur Kommunikation mit externen Vorrichtungen implementieren.
  • Die E/A-Einheit 305 dekodiert Pakete, die über die Verbindung 302 empfangen wurden. In einer Ausführungsform stellen die Pakete Befehle dar, die dazu konfiguriert sind, die GPU 102 zu veranlassen, verschiedene Operationen durchzuführen. Die E/A-Einheit 305 überträgt die dekodierten Befehle an verschiedene andere Einheiten der GPU 102, wie es die Befehle vorgeben können. Zum Beispiel können einige Befehle an die Frontend-Einheit 315 übertragen werden. Andere Befehle können an den Hub 330 oder andere Einheiten der GPU 102, wie z.B. eine oder mehrere Kopier-Engines, einen Video-Encoder, einen Video-Decoder, eine Energieverwaltungseinheit usw. (nicht explizit gezeigt), übertragen werden. Mit anderen Worten ist die E/A-Einheit 305 dazu konfiguriert, die Kommunikation zwischen und unter den verschiedenen logischen Einheiten der GPU 102 zu routen.
  • In einer Ausführungsform kodiert ein von dem Host-Prozessor 150 ausgeführtes Programm einen Befehlsstrom in einem Puffer, der der GPU 102 Arbeitslasten zur Verarbeitung bereitstellt. Eine Arbeitslast kann mehrere Anweisungen und Daten umfassen, die von diesen Anweisungen zu verarbeiten sind. Der Puffer ist ein Bereich in einem Speicher, auf den sowohl der Host-Prozessor 150 als auch die GPU 102 zugreifen (z.B. lesen/schreiben) können. Beispielsweise kann die E/A-Einheit 305 dazu konfiguriert sein, über Speicheranforderungen, die über die Verbindung 302 übertragen werden, auf den Puffer in einem Systemspeicher zuzugreifen, der mit der Verbindung 302 verbunden ist. In einer Ausführungsform schreibt der Host-Prozessor 150 den Befehlsstrom in den Puffer und überträgt dann einen Zeiger auf den Anfang des Befehlsstroms an die GPU 102. Die Frontend-Einheit 315 empfängt Zeiger auf einen oder mehrere Befehlsströme. Die Frontend-Einheit 315 verwaltet den einen oder die mehreren Ströme, liest Befehle aus den Strömen und leitet Befehle an die verschiedenen Einheiten der GPU 102 weiter.
  • Die Frontend-Einheit 315 ist mit einer Scheduler-Einheit 320 gekoppelt, die die verschiedenen GPCs 350 für die Verarbeitung von Aufgaben konfiguriert, die durch den einen oder die mehreren Ströme bzw. Streams definiert sind. Die Scheduler-Einheit 320 dazu konfiguriert, Zustandsinformationen in Bezug auf die verschiedenen, von der Scheduler-Einheit 320 verwalteten Aufgaben nachzuverfolgen. Der Zustand kann anzeigen, welchem GPC 350 eine Aufgabe zugewiesen ist, ob die Aufgabe aktiv oder inaktiv ist, eine Prioritätsstufe, die mit der Aufgabe verbunden ist, und so weiter. Die Scheduler-Einheit 320 verwaltet die Ausführung einer Vielzahl von Tasks auf dem einen oder mehreren GPCs 350.
  • Die Scheduler-Einheit 320 ist mit einer Arbeitsverteilungseinheit 325 gekoppelt, die dazu konfiguriert ist, Aufgaben zur Ausführung auf den GPCs 350 zu verteilen. Die Arbeitsverteilungseinheit 325 kann eine Anzahl von geplanten Aufgaben nachverfolgen, die von der Scheduler-Einheit 320 empfangen wurden. In einer Ausführungsform verwaltet die Arbeitsverteilungseinheit 325 einen Pool anstehender Aufgaben und einen Pool aktiver Aufgaben für jeden der GPCs 350. Der Pool anstehender Aufgaben kann eine Anzahl von Slots (z.B. 32 Slots) umfassen, die Aufgaben enthalten, die zur Verarbeitung durch einen bestimmten GPC 350 zugewiesen sind. Der Pool aktiver Aufgaben kann eine Anzahl von Slots (z.B. 4 Slots) für Aufgaben enthalten, die aktiv von den GPCs 350 bearbeitet werden. Wenn ein GPC 350 die Ausführung einer Aufgabe beendet, wird diese Aufgabe aus dem Pool aktiver Aufgaben für den GPC 350 entfernt und wird eine der anderen Aufgaben aus dem Pool anstehendender Aufgaben ausgewählt und zur Ausführung auf dem GPC 350 eingeplant. Falls eine aktive Aufgabe auf dem GPC 350 im Leerlauf war, z.B. während des Wartens auf die Auflösung einer Datenabhängigkeit, dann kann die aktive Aufgabe aus dem GPC 350 verdrängt und in den Pool anstehender Aufgaben zurückgeführt werden, während eine andere Aufgabe in dem Pool anstehender Aufgaben ausgewählt und zur Ausführung auf dem GPC 350 eingeplant wird.
  • Die Arbeitsverteilungseinheit 325 kommuniziert mit dem einen oder mehreren GPCs 350 über die Crossbar bzw. XBar 370. Die XBar 370 ist ein Verbindungsnetzwerk, das viele der Einheiten der GPU 102 mit anderen Einheiten der GPU 102 koppelt. Zum Beispiel kann die XBar 370 dazu konfiguriert sein, die Arbeitsverteilungseinheit 325 mit einem bestimmten GPC 350 zu koppeln. Obwohl dies nicht explizit dargestellt ist, können auch eine oder mehrere andere Einheiten der GPU 102 über den Hub 330 mit der XBar 370 verbunden sein.
  • Die Aufgaben werden von der Scheduler-Einheit 320 verwaltet und von der Arbeitsverteilungseinheit 325 an einen GPC 350 weitergeleitet. Der GPC 350 ist dazu konfiguriert, die Aufgabe zu verarbeiten und Ergebnisse zu erzeugen. Die Ergebnisse können von anderen Tasks innerhalb des GPC 350 verbraucht, über die XBar 370 an einen anderen GPC 350 weitergeleitet oder im Speicher 304 abgelegt werden. Die Ergebnisse können über die Partitionseinheiten 380, die eine Speicherschnittstelle zum Lesen und Schreiben von Daten in/aus dem Speicher 304 implementieren, in den Speicher 304 geschrieben werden. Die Ergebnisse können über den NVLINK™ 108 an eine andere PPU 304 oder CPU übertragen werden. In einer Ausführungsform beinhaltet die GPU 102 eine Anzahl U von Partitionseinheiten 380, die gleich der Anzahl von separaten und unterschiedlichen Speichervorrichtungen 304 ist, die mit der GPU 102 gekoppelt sind. Eine Partitionseinheit 380 wird nachstehend in Verbindung mit 20 näher beschrieben.
  • In einer Ausführungsform führt ein Host-Prozessor 150 einen Treiberkern aus, der eine Anwendungsprogrammierschnittstelle (API) implementiert, die es einer oder mehreren auf dem Host-Prozessor ausgeführten Anwendungen ermöglicht, Operationen zur Ausführung auf der GPU 102 zu planen. In einer Ausführungsform werden mehrere Rechenanwendungen gleichzeitig von der GPU 102 ausgeführt und stellt die GPU 102 Isolierung, Dienstgüte (QoS) und unabhängige Adressräume für die mehreren Rechenanwendungen bereit. Eine Anwendung kann Anweisungen (z.B. API-Aufrufe) generieren, die den Treiberkern veranlassen, eine oder mehrere Aufgaben zur Ausführung durch die GPU 102 zu generieren. Der Treiberkern gibt Aufgaben an einen oder mehrere Streams aus, die von der GPU 102 verarbeitet werden. Jede Aufgabe kann eine oder mehrere Gruppen von zusammenhängenden Threads umfassen, die hier als ein Warp bezeichnet werden. In einer Ausführungsform umfasst ein Warp mehrere (z.B. 32) zusammengehörige Threads, die parallel ausgeführt werden können. Zusammenhängende Threads können sich auf eine Vielzahl von Threads beziehen, die Anweisungen zur Ausführung der Aufgabe enthalten und die Daten über einen gemeinsamen Speicher austauschen können.
  • 19 veranschaulicht einen GPC 350 der GPU 102 von 18 in Übereinstimmung mit einer Ausführungsform. Wie in 19 gezeigt ist, beinhaltet jeder GPC 350 eine Anzahl von Hardwareeinheiten zur Verarbeitung von Aufgaben. In einer Ausführungsform beinhaltet jeder GPC 350 einen Pipelineverwalter bzw. Pipeline-Manager 410, eine Vorrasterbetriebseinheit bzw. Pre-Raster Operations Unit (PROP) 415, eine Raster-Engine 425, eine Arbeitsverteilungs- bzw. Work Distribution Crossbar (WDX) 480, eine Speicherverwaltungseinheit bzw. Memory Management Unit (MMU) 490 und einen oder mehrere Datenverarbeitungscluster bzw. Data Processing Cluster (DPCs) 420. Es wird deutlich, dass der GPC 350 andere Hardwareeinheiten anstelle der in 20 gezeigten Einheiten oder zusätzlich zu diesen enthalten kann, einschließlich beispielsweise einer Echtzeit-Raytracing-Engine, einer Kopier-Engine, einem Deep-Learning-Beschleuniger, einem Bildverarbeitungsbeschleuniger und anderer Beschleunigungshardware.
  • In einer Ausführungsform wird der Betriebsablauf des GPC 350 durch den Pipeline-Manager 410 gesteuert. Der Pipeline-Manager 410 verwaltet die Konfiguration des einen oder der mehreren DPCs 420 für die Verarbeitung von Aufgaben, die dem GPC 350 zugewiesen sind. In einer Ausführungsform kann der Pipeline-Manager 410 zumindest einen des einen oder der mehreren DPCs 420 konfigurieren, um zumindest einen Teil einer in 20 gezeigten Grafik-Rendering-Pipeline zu implementieren. Zum Beispiel kann ein DPC 420 dazu konfiguriert sein, dass er ein Vertex-Shader-Programm auf dem programmierbaren Streaming-Multiprozessor (SM) 440 ausführt. Der Pipeline-Manager 410 kann auch dazu konfiguriert sein, von der Arbeitsverteilungseinheit 325 empfangene Pakete an die entsprechenden logischen Einheiten innerhalb des GPC 350 weiterzuleiten. Beispielsweise können einige Pakete an Hardwareeinheiten mit fester Funktion im PROP 415 und/oder in der Raster-Engine 425 weitergeleitet werden, während andere Pakete an die DPCs 420 zur Verarbeitung durch die Stammfunktionen- bzw. Primitiv-Engine 435 oder den SM 440 weitergeleitet werden können. In einer Ausführungsform kann der Pipeline-Manager 410 zumindest einen des einen oder der mehreren DPCs 420 konfigurieren, um ein neuronales Netzwerkmodell und/oder eine Rechenpipeline zu implementieren.
  • Die PROP-Einheit 415 ist dazu konfiguriert, die von der Raster-Engine 425 und den DPCs 420 erzeugten Daten an eine Rasterbetriebs- bzw. Raster Operations (ROP)-Einheit weiterzuleiten, die in Verbindung mit 21 näher beschrieben wird. Die PROP-Einheit 415 kann auch dazu konfiguriert sein, Optimierungen für die Farbüberblendung durchzuführen, Pixeldaten zu organisieren, Adressübersetzungen durchzuführen und dergleichen.
  • Grafikverarbeitungs-Pipeline
  • In einer Ausführungsform ist die GPU 102 als eine Grafikverarbeitungseinheit (GPU) konfiguriert. Die GPU 102 ist dazu konfiguriert, Befehle zu empfangen, die Shader-Programme zur Verarbeitung von Grafikdaten spezifizieren. Grafikdaten können als ein Satz von Stammfunktionen bzw. Primitiven wie Punkte, Linien, Dreiecke, Quadrate, Dreiecksstreifen und dergleichen definiert sein. Typischerweise enthält eine Stammfunktion Daten, die eine Anzahl von Scheitelpunkten für die Stammfunktion (z.B. in einem Modellraum-Koordinatensystem) sowie Attribute, die jedem Scheitelpunkt der Stammfunktion zugeordnet sind, spezifizieren. Die GPU 102 kann dazu konfiguriert sein, die Grafik-Stammfunktionen bzw. Grafikprimitive zu verarbeiten, um einen Bildpuffer bzw. Frame Buffer zu erzeugen (z.B. Pixeldaten für jedes der Pixel der Anzeige).
  • Eine Anwendung schreibt Modelldaten für eine Szene (z.B. eine Sammlung von Scheitelpunkten und Attributen) in einen Speicher, wie z.B. einen Systemspeicher oder Speicher 304. Die Modelldaten definieren jedes der Objekte, die auf einer Anzeige sichtbar sein können. Die Anwendung tätigt dann einen API-Aufruf an den Treiberkern, der die Modelldaten zum Rendern und Anzeigen anfordert. Der Treiberkern liest die Modelldaten und schreibt Befehle in den einen oder die mehreren Streams, um Operationen zur Verarbeitung der Modelldaten durchzuführen. Die Befehle können auf verschiedene Shader-Programme verweisen, die auf den SMs 440 der GPU 102 implementiert sind, einschließlich eines oder mehrerer Vertex-Shader, Hull-Shader, Domain-Shader, Geometrie-Shader und eines Pixel-Shaders. Beispielsweise können einer oder mehrere der SMs 440 dazu konfiguriert sein, ein Vertex-Shader-Programm auszuführen, das eine durch die Modelldaten definierte Anzahl von Scheitelpunkten verarbeitet. In einer Ausführungsform können die verschiedenen SMs 440 dazu konfiguriert sein, verschiedene Shader-Programme gleichzeitig ausführen. Beispielsweise kann eine erste Teilmenge von SMs 440 dazu konfiguriert sein, ein Vertex-Shader-Programm auszuführen, während eine zweite Teilmenge von SMs 440 dazu konfiguriert sein kann, ein Pixel-Shader-Programm auszuführen. Die erste Teilmenge von SMs 440 verarbeitet Scheitelpunktdaten, um verarbeitete Scheitelpunktdaten zu erzeugen, und schreibt die verarbeiteten Scheitelpunktdaten in den L2-Cache 460 und/oder den Speicher 304. Nachdem die verarbeiteten Scheitelpunktdaten gerastert (z.B. von dreidimensionalen Daten in zweidimensionale Daten im Bildschirmraum transformiert) wurden, um Fragmentdaten zu erzeugen, führt die zweite Teilmenge von SMs 440 einen Pixel-Shader aus, um verarbeitete Fragmentdaten zu erzeugen, die dann mit anderen verarbeiteten Fragmentdaten gemischt und in den Bildpuffer in dem Speicher 304 geschrieben werden. Das Vertex-Shader-Programm und das Pixel-Shader-Programm können gleichzeitig ausgeführt werden und verschiedene Daten aus derselben Szene in einer Pipeline verarbeiten, bis alle Modelldaten für die Szene in den Bildpuffer gerendert worden sind. Dann wird der Inhalt des Bildpuffers an eine Anzeigesteuervorrichtung bzw. einen Display-Controller zur Anzeige auf einer Anzeigevorrichtung übertragen.
  • 20 ist ein konzeptionelles Diagramm einer Grafikverarbeitungs-Pipeline 600, die von der GPU 102 von 18 in Übereinstimmung mit einer Ausführungsform implementiert wird. Die Grafikverarbeitungs-Pipeline 600 ist ein abstraktes Ablaufdiagramm der Verarbeitungsschritte, die implementiert sind, um 2D-Computer-generierte Bilder aus 3D-Geometriedaten zu erzeugen. Wie gut bekannt ist, können Pipeline-Architekturen Operationen mit langer Latenzzeit effizienter durchführen, indem sie die Operation in eine Vielzahl von Stufen aufteilen, wobei der Ausgang jeder Stufe mit dem Eingang der nächstfolgenden Stufe gekoppelt ist. Somit empfängt die Grafikverarbeitungs-Pipeline 600 Eingangsdaten 601, die von einer Stufe zur nächsten Stufe der Grafikverarbeitungs-Pipeline 600 übertragen werden, um Ausgangsdaten 602 zu erzeugen. In einer Ausführungsform kann die Grafikverarbeitungs-Pipeline 600 eine Grafikverarbeitungs-Pipeline repräsentieren, die durch die OpenGL®-API definiert ist. Optional kann die Grafikverarbeitungs-Pipeline 600 im Kontext der Funktionalität und der Architektur der vorherigen Figuren und/oder jeder nachfolgenden Figur(en) implementiert sein.
  • Wie in 20 gezeigt ist, umfasst die Grafikverarbeitungs-Pipeline 600 eine Pipeline-Architektur, die eine Anzahl von Stufen beinhaltet. Die Stufen beinhalten, sind aber nicht beschränkt auf, eine Datenassemblierungsstufe 610, eine Vertex-Shading-Stufe 620, eine Stammfunktion-Assemblierungsstufe 630, eine Geometrie-Shading-Stufe 640, eine Ansichtsfenster- bzw. Viewport-Skalierungs-, Cull- und Clip-Stufe (VSCC) 650, eine Rasterisierungsstufe 660, eine Fragment-Shading-Stufe 670 und eine Rasteroperationsstufe 680. Wie vorstehend beschrieben wurde, können die Software-Shading-Algorithmen, die in Verbindung mit einer solchen Shading-Hardware arbeiten, optimiert werden, um die Berechnungszeit zu reduzieren.
  • In einer Ausführungsform umfassen die Eingangsdaten 601 Befehle, die die Verarbeitungseinheiten dazu konfigurieren, die Stufen der Grafikverarbeitungs-Pipeline 600 und geometrische Stammfunktionen bzw. Primitive (z.B. Punkte, Linien, Dreiecke, Quadrate, Dreiecksstreifen oder Fächer usw.) zu implementieren, die von den Stufen zu verarbeiten sind. Die Ausgangsdaten 602 können Pixeldaten (z.B. Farbdaten) umfassen, die in einen Bildpuffer oder eine andere Art von Oberflächendatenstruktur in einem Speicher kopiert werden.
  • Die Datenassemblierungsstufe 610 empfängt die Eingangsdaten 601, die Scheitelpunktdaten für Oberflächen höherer Ordnung, Primitive oder ähnliches spezifizieren. Die Datenassemblierungsstufe 610 sammelt die Scheitelpunktdaten in einem temporären Speicher oder einer Warteschlange, z.B. durch Empfangen eines Befehls von dem Host-Prozessor, der einen Zeiger auf einen Puffer im Speicher beinhaltet, und Lesen der Scheitelpunktdaten aus dem Puffer. Die Scheitelpunktdaten werden dann an die Vertex-Shading-Stufe 620 zur Verarbeitung übertragen.
  • Die Vertex-Shading-Stufe 620 verarbeitet Vertex- bzw. Scheitelpunktdaten, indem sie einen Satz von Operationen (z.B. einen Vertex-Shader oder ein Programm) einmal für jeden der Scheitelpunkte ausführt. Scheitelpunkte können z.B. als ein 4-Koordinaten-Vektor (z.B. <x, y, z, w>) angegeben sein, der einem oder mehreren Scheitelpunktattributen (z.B. Farbe, Texturkoordinaten, Oberflächennormale usw.) zugeordnet ist. Die Vertex-Shading-Stufe 620 kann einzelne Scheitelpunktattribute wie beispielsweise Position, Farbe, Texturkoordinaten und dergleichen manipulieren. Mit anderen Worten führt die Vertex-Shading-Stufe 620 Operationen an den Scheitelpunktkoordinaten oder anderen Scheitelpunktattributen durch, die einem Scheitelpunkt zugeordnet sind. Solche Operationen beinhalten üblicherweise Beleuchtungsoperationen (z.B. Ändern von Farbattributen für einen Scheitelpunkt) und Transformationsoperationen (z.B. Ändern des Koordinatenraums für einen Scheitelpunkt). Beispielsweise können Scheitelpunkte unter Verwendung von Koordinaten in einem Objektkoordinatenraum angegeben werden, die durch Multiplikation der Koordinaten mit einer Matrix transformiert werden, die die Koordinaten aus dem Objektkoordinatenraum in einen Weltraum oder einen für eine Vorrichtung normalisierten Koordinatenraum (normalized-device coordinate, NCD) übersetzt. Die Vertex-Shading-Stufe 620 erzeugt transformierte Scheitelpunktdaten, die an die Stammfunktion-Assemblierungsstufe 630 übertragen werden.
  • Die Stammfunktion-Assemblierungsstufe 630 sammelt die von der Vertex-Shading-Stufe 620 ausgegebenen Scheitelpunkte und gruppiert die Scheitelpunkte in geometrische Primitive zur Verarbeitung durch die Geometrie-Shading-Stufe 640. Zum Beispiel kann die Stammfunktion-Assemblierungsstufe 630 dazu konfiguriert sein, alle drei aufeinanderfolgenden Scheitelpunkte als ein geometrisches Primitiv (z.B. ein Dreieck) zur Übertragung an die Geometrie-Shading-Stufe 640 gruppiert. In einigen Ausführungsformen können bestimmte Scheitelpunkte für aufeinanderfolgende geometrische Primitive wiederverwendet werden (z.B. können sich zwei aufeinanderfolgende Dreiecke in einem Dreiecksstreifen zwei Scheitelpunkte teilen). Die Stammfunktion-Assemblierungsstufe 630 überträgt geometrische Primitive (z.B. eine Sammlung zugeordneter Scheitelpunkte) an die Geometrie-Shading-Stufe 640.
  • Die Geometrie-Shading-Stufe 640 verarbeitet geometrische Primitive, indem sie einen Satz von Operationen (z.B. einen Geometrie-Shader oder ein Programm) auf den geometrischen Primitiven durchführt. Tesselationsoperationen können ein oder mehrere geometrische Primitive aus jedem geometrischen Primitiv erzeugen. Mit anderen Worten kann die Geometrie-Shading-Stufe 640 jedes geometrische Element in ein feineres Netz aus zwei oder mehr geometrischen Grundelementen unterteilen, das vom Rest der Grafikverarbeitungs-Pipeline 600 verarbeitet wird. Die Geometrie-Shading-Stufe 640 überträgt geometrische Primitive an die Viewport-SCC-Stufe 650.
  • In einer Ausführungsform kann die Grafikverarbeitungs-Pipeline 600 innerhalb eines Streaming-Multiprozessors arbeiten, und können die Vertex-Shading-Stufe 620, die Stammfunktion-Assemblierungsstufe 630, die Geometrie-Shading-Stufe 640, die Fragment-Shading-Stufe 670 und/oder die damit verbundene Hardware/Software sequenziell Verarbeitungsvorgänge durchführen. Sobald die sequenziellen Verarbeitungsvorgänge abgeschlossen sind, kann in einer Ausführungsform die Viewport-SCC-Stufe 650 die Daten verwenden. In einer Ausführungsform können Stammfunktionsdaten, die von einer oder mehreren der Stufen in der Grafikverarbeitungs-Pipeline 600 verarbeitet wurden, in einen Zwischenspeicher bzw. Cache (z.B. einen L1-Cache, einen Vertex-Cache usw.) geschrieben werden. In diesem Fall kann in einer Ausführungsform die Viewport-SCC-Stufe 650 auf die Daten im Cache zugreifen. In einer Ausführungsform sind die Viewport-SCC-Stufe 650 und die Rasterisierungsstufe 660 als Schaltungen mit fester Funktion implementiert.
  • Die Viewport-SCC-Stufe 650 führt eine Ansichtsfensterskalierung, ein Culling und ein Clipping der geometrischen Primitive durch. Jede Oberfläche, auf die gerendert wird, ist einer abstrakten Kameraposition zugeordnet. Die Kameraposition repräsentiert einen Standort eines Betrachters, der auf die Szene blickt, und definiert einen Sichtkegelstumpf, der die Objekte der Szene umschließt. Der Sichtkegelstumpf kann eine Betrachtungsebene, eine hintere Ebene und vier Clipping-Ebenen beinhalten. Jedes geometrische Primitiv, das vollständig außerhalb des Sichtkegelstumpfes liegt, kann aussortiert (z.B. verworfen) werden, da das geometrische Primitiv nicht zur endgültigen gerenderten Szene beitragen wird. Jedes geometrische Primitiv bzw. Element, das sich teilweise innerhalb und teilweise außerhalb des Sichtkegelstumpfes befindet, kann abgeschnitten werden (z.B. in ein neues geometrisches Element umgewandelt werden, das innerhalb des Sichtkegelstumpfes liegt). Darüber hinaus können geometrische Primitive jeweils auf der Grundlage einer Tiefe des Sichtkegelstumpfs skaliert werden. Alle potenziell sichtbaren geometrischen Primitive werden dann an die Rasterisierungsstufe 660 übertragen.
  • Die Rasterisierungsstufe 660 wandelt die geometrischen 3D-Primitive in 2D-Fragmente um (die z.B. für die Anzeige verwendet werden können, usw.). Die Rasterisierungsstufe 660 kann dazu konfiguriert sein, die Scheitelpunkte der geometrischen Primitive zu verwenden, um einen Satz von Ebenengleichungen aufzustellen, aus denen verschiedene Attribute interpoliert werden können. Die Rasterisierungsstufe 660 kann auch eine Abdeckungsmaske für eine Vielzahl von Pixeln berechnen, die anzeigt, ob eine oder mehrere Abtastpositionen für das Pixel das geometrische Primitiv durchschneiden. In einer Ausführungsform kann auch eine z-Prüfung durchgeführt werden, um festzustellen, ob das geometrische Primitiv von anderen geometrischen Primitiven verdeckt wird, die bereits gerastert worden sind. Die Rasterisierungsstufe 660 erzeugt Fragmentdaten (z.B. interpolierte Scheitelpunktattribute, die einer bestimmten Abtastposition für jedes abgedeckte Pixel zugeordnet sind), die an die Fragment-Shading-Stufe 670 übertragen werden.
  • Die Fragment-Shading-Stufe 670 verarbeitet Fragmentdaten, indem sie einen Satz von Operationen (z.B. einen Fragment-Shader oder ein Programm) auf jedem der Fragmente ausführt. Die Fragment-Shading-Stufe 670 kann Pixeldaten (z.B. Farbwerte) für das Fragment erzeugen, wie z.B. durch Ausführen von Beleuchtungsoperationen oder Abtasten von Texturkarten unter Verwendung interpolierter Texturkoordinaten für das Fragment. Die Fragment-Shading-Stufe 670 erzeugt Pixeldaten, die an die Rasteroperationsstufe 680 übertragen werden.
  • Die Rasteroperationsstufe 680 kann verschiedene Operationen an den Pixeldaten durchführen, wie z.B. Alphatests, Schablonentests und ein Mischen der Pixeldaten mit anderen Pixeldaten, die anderen, dem Pixel zugeordneten Fragmenten entsprechen. Wenn die Rasteroperationsstufe 680 die Verarbeitung der Pixeldaten (z.B. der Ausgangsdaten 602) abgeschlossen hat, können die Pixeldaten in ein Rendering-Ziel geschrieben werden, wie z.B. in einen Bildpuffer, einen Farbpuffer oder dergleichen. Die Raster-Engine 425 beinhaltet eine Reihe von Hardwareeinheiten mit fester Funktion, die für die Durchführung verschiedener Rasteroperationen konfiguriert sind. In einer Ausführungsform beinhaltet die Raster-Engine 425 eine Setup-Engine, eine Grobraster-Engine, eine Culling-Engine, eine Clipping-Engine, eine Feinraster-Engine und eine Kachelvereinigungs- bzw. Tile Coalescing-Engine. Die Setup-Engine empfängt transformierte Scheitelpunkte und erzeugt Ebenengleichungen, die dem durch die Scheitelpunkte definierten geometrischen Primitiv zugeordnet sind. Die Ebenengleichungen werden an die Grobraster-Engine übertragen, um Abdeckungsinformationen (z.B. eine x, y-Abdeckungsmaske für eine Kachel) für das Primitiv zu erzeugen. Die Ausgabe der Grobraster-Engine wird an die Culling-Engine übertragen, in der Fragmente, die mit dem Primitiv verbunden sind und einen z-Test nicht bestehen, aussortiert werden, und nicht aussortierte Fragmente werden an eine Clipping-Engine übertragen, in der Fragmente, die außerhalb eines Sichtkegelstumpfs liegen, abgeschnitten werden. Die Fragmente, die das Beschneiden und Aussortieren überstehen, können an die Feinraster-Engine weitergeleitet werden, um Attribute für die Pixelfragmente auf der Grundlage der von der Setup-Engine erzeugten Ebenengleichungen zu generieren. Die Ausgabe der Raster-Engine 425 umfasst Fragmente, die z.B. von einem Fragment-Shader verarbeitet werden, der in einem DPC 420 implementiert ist.
  • Es wird deutlich, dass eine oder mehrere zusätzliche Stufen in der Grafikverarbeitungs-Pipeline 600 zusätzlich zu oder anstelle von einer oder mehreren der vorstehend beschriebenen Stufen enthalten sein können. Verschiedene Implementierungen der abstrakten Grafikverarbeitungs-Pipeline können unterschiedliche Stufen implementieren. Ferner können eine oder mehrere der vorstehend beschriebenen Stufen in einigen Ausführungsformen aus der Grafikverarbeitungs-Pipeline ausgeschlossen sein (z.B. die Geometrie-Shading-Stufe 640). Andere Arten von Grafikverarbeitungs-Pipelines sind im Rahmen der vorliegenden Offenbarung denkbar. Ferner kann jede der Stufen der Grafikverarbeitungs-Pipeline 600 durch eine oder mehrere dedizierte Hardwareeinheiten in einem Grafikprozessor wie der GPU 102 implementiert sein. Andere Stufen der Grafikverarbeitungs-Pipeline 600 können durch programmierbare Hardwareeinheiten wie den SM 440 der GPU 102 implementiert sein.
  • Die Grafikverarbeitungs-Pipeline 600 kann über eine Anwendung implementiert sein, die von einem Host-Prozessor, wie z.B. einer CPU 150, ausgeführt wird. In einer Ausführungsform kann ein Gerätetreiber eine Anwendungsprogrammierschnittstelle (API) implementieren, die verschiedene Funktionen definiert, die von einer Anwendung verwendet werden können, um grafische Daten für die Anzeige zu erzeugen. Der Gerätetreiber ist ein Softwareprogramm, das eine Vielzahl von Anweisungen enthält, die den Betrieb der GPU 102 steuern. Die API stellt eine Abstraktion für einen Programmierer bereit, die es ihm ermöglicht, spezialisierte Grafikhardware wie beispielsweise die GPU 102 zu verwenden, um die grafischen Daten zu erzeugen, ohne dass der Programmierer den spezifischen Befehlssatz für die GPU 102 verwenden muss. Die Anwendung kann einen API-Aufruf enthalten, der an den Gerätetreiber für die GPU 102 weitergeleitet wird. Der Gerätetreiber interpretiert den API-Aufruf und führt verschiedene Operationen aus, um auf den API-Aufruf zu reagieren. In einigen Fällen kann der Gerätetreiber Operationen durch Ausführen von Anweisungen auf der CPU durchführen. In anderen Fällen kann der Gerätetreiber Operationen durchführen, zumindest teilweise, indem er Operationen auf der GPU 102 unter Verwendung einer Eingabe/Ausgabe-Schnittstelle zwischen der CPU und der GPU 102 startet. In einer Ausführungsform ist der Gerätetreiber dazu konfiguriert, die Grafikverarbeitungs-Pipeline 600 unter Verwendung der Hardware des Grafikprozessors 102 zu implementieren.
  • Verschiedene Programme können innerhalb der GPU 102 ausgeführt werden, um die verschiedenen Stufen der Grafikverarbeitungs-Pipeline 600 zu implementieren. Zum Beispiel kann der Gerätetreiber einen Kern bzw. Kernel auf der GPU 102 starten, um die Vertex-Shading-Stufe 620 auf einem SM 440 (oder mehreren SMs 440) auszuführen. Der Gerätetreiber (oder der anfängliche Kern, der von der PPU 400 ausgeführt wird) kann auch andere Kerne auf der PPU 400 starten, um andere Stufen der Grafikverarbeitungs-Pipeline 600 auszuführen, wie z.B. die Geometrie-Shading-Stufe 640 und die Fragment-Shading-Stufe 670. Darüber hinaus können einige der Stufen der Grafikverarbeitungs-Pipeline 600 auf fester Einheitshardware implementiert sein, wie z.B. ein Rasterisierer oder ein Datenassembler, die innerhalb der PPU 400 implementiert sind. Es wird deutlich, dass Ergebnisse von einem Kern von einer oder mehreren dazwischenliegenden Hardwareeinheiten mit fester Funktion verarbeitet werden können, bevor sie von einem nachfolgenden Kern auf einem SM 440 verarbeitet werden.
  • Wie in 19 gezeigt ist, beinhaltet jeder DPC 420, der in dem GPC 350 enthalten ist, einen M-Pipe-Controller (MPC) 430, eine Stammfunktionen-Engine 435 und einen oder mehrere SMs 440. Der MPC 430 steuert den Betrieb des DPC 420 und leitet die von dem Pipeline-Manager 410 empfangenen Pakete an die entsprechenden Einheiten in dem DPC 420 weiter. Beispielsweise können Pakete, die einem Scheitelpunkt zugeordnet sind, an die Stammfunktionen-Engine 435 weitergeleitet werden, die dazu konfiguriert ist, Scheitelpunktattribute, die dem Scheitelpunkt zugeordnet sind, aus dem Speicher 304 abzurufen. Im Gegensatz dazu können Pakete, die mit einem Shader-Programm assoziiert sind, an den SM 440 übertragen werden.
  • Der SM 440 umfasst einen programmierbaren Streaming-Prozessor, der dazu konfiguriert ist, Aufgaben zu verarbeiten, die durch eine Anzahl von Threads repräsentiert werden. Jeder SM 440 ist multi-threaded und dazu konfiguriert, eine Vielzahl von Threads (z.B. 32 Threads) aus einer bestimmten Gruppe von Threads gleichzeitig auszuführen. In einer Ausführungsform implementiert der SM 440 eine SIMD (Single-Instruction, Multiple-Data)-Architektur, bei der jeder Thread in einer Gruppe von Threads (z.B. einem Warp) dazu konfiguriert ist, einen anderen Satz von Daten auf der Grundlage desselben Satzes von Anweisungen zu verarbeiten. Alle Threads in der Gruppe von Threads führen die gleichen Anweisungen aus. In einer anderen Ausführungsform implementiert der SM 440 eine SIMT (Single-Instruction, Multiple Thread)-Architektur, bei der jeder Thread in einer Gruppe von Threads dazu konfiguriert ist, einen anderen Datensatz auf der Grundlage desselben Befehlssatzes zu verarbeiten, wobei jedoch einzelne Threads in der Gruppe von Threads während der Ausführung divergieren dürfen. In einer Ausführungsform werden ein Programmzähler, ein Aufrufstapel und ein Ausführungsstatus für jeden Warp beibehalten, wodurch Gleichzeitigkeit zwischen Warps und serielle Ausführung innerhalb von Warps ermöglicht wird, wenn Threads innerhalb des Warps divergieren. In einer anderen Ausführungsform werden ein Programmzähler, ein Aufrufstapel und ein Ausführungsstatus für jeden einzelnen Thread beibehalten, wodurch gleiche Gleichzeitigkeit zwischen allen Threads innerhalb und zwischen Warps ermöglicht wird. Wenn der Ausführungsstatus für jeden einzelnen Thread beibehalten wird, können Threads, die dieselben Anweisungen ausführen, zusammengeführt und parallel ausgeführt werden, um maximale Effizienz zu erzielen. Der SM 440 wird nachstehend in Verbindung mit 22 näher beschrieben.
  • Die MMU 490 aus 19 stellt eine Schnittstelle zwischen dem GPC 350 und der Partitionseinheit 380 bereit. Wie vorstehend beschrieben wurde, kann die MMU 490 die Übersetzung von virtuellen Adressen in physische Adressen, den Speicherschutz und die Arbitrierung von Speicheranforderungen bereitstellen. In einer Ausführungsform wie vorstehend beschrieben stellt die MMU 490 einen oder mehrere Translation-Lookaside-Buffer (TLBs) bereit, um die Übersetzung von virtuellen Adressen in physische Adressen in dem Speicher 304 durchzuführen.
  • 21 veranschaulicht eine Speicherpartitionseinheit 380 der GPU 102 von 18 gemäß einer Ausführungsform. Wie in 21 gezeigt ist, beinhaltet die Speicherpartitionseinheit 380 eine Raster Operations (ROP)-Einheit 450, einen Level Zwei (L2)-Cache 460 und eine Speicherschnittstelle 470. Die Speicherschnittstelle 470 ist mit dem Speicher 304 gekoppelt. Die Speicherschnittstelle 470 kann 32-, 64-, 128-, 1024-Bit-Datenbusse oder dergleichen für Hochgeschwindigkeitsdatenübertragung implementieren. In einer Ausführungsform inkorporiert die GPU 102 U Speicherschnittstellen 470, eine Speicherschnittstelle 470 pro Paar von Partitionseinheiten 380, wobei jedes Paar von Partitionseinheiten 380 mit einem entsprechenden Speichergerät 304 verbunden ist. Beispielsweise kann die GPU 102 mit bis zu Y Speichervorrichtungen 304, wie z.B. Speicherstapel mit hoher Bandbreite oder Grafikspeicher mit doppelter Datenrate, Version 5, synchroner dynamischer Direktzugriffsspeicher oder andere Arten von persistenten Speichern, verbunden sein.
  • In einer Ausführungsform implementiert die Speicherschnittstelle 470 eine HBM2-Speicherschnittstelle und ist Y gleich der Hälfte von U. In einer Ausführungsform befinden sich die HBM2-Speicherstapel auf demselben physischen Gehäuse wie die GPU 102, was im Vergleich zu herkömmlichen GDDR5-SDRAM-Systemen erhebliche Energie- und Flächeneinsparungen ermöglicht. In einer Ausführungsform beinhaltet jeder HBM2-Stapel vier Speicherchips und ist Y gleich 4, wobei der HBM2-Stapel zwei 128-Bit-Kanäle pro Chip für insgesamt 8 Kanäle und eine Datenbusbreite von 1024 Bit beinhaltet.
  • In einer Ausführungsform unterstützt der Speicher 304, wie vorstehend beschrieben wurde, Single-Error Correcting Double Error Detecting (SECDED) Error Correction Code (ECC) zum Schutz der Daten. ECC stellt eine höhere Zuverlässigkeit für Rechenanwendungen bereit, die empfindlich auf Datenbeschädigung reagieren. Zuverlässigkeit ist besonders wichtig in großen Cluster-Computing-Umgebungen, in denen PPUs 300 sehr große Datensätze verarbeiten und/oder Anwendungen über längere Zeiträume ausführen.
  • In einer Ausführungsform implementiert die GPU 102 eine mehrstufige Speicherhierarchie. In einer Ausführungsform unterstützt die Speicherpartitionseinheit 380 einen vereinheitlichten Speicher, um einen einzigen vereinheitlichten virtuellen Adressraum für den CPU- und GPU 102-Speicher bereitzustellen, was die gemeinsame Nutzung von Daten zwischen virtuellen Speichersystemen ermöglicht. In einer Ausführungsform wird die Häufigkeit von Zugriffen einer GPU 102 auf Speicher, der sich auf anderen Prozessoren befindet, nachverfolgt, um sicherzustellen, dass Speicherseiten in den physischen Speicher der GPU 102 verschoben werden, die häufiger auf die Seiten zugreift. In einer Ausführungsform unterstützt der NVLINK™ 310 Adressübersetzungsdienste, die es der GPU 102 ermöglichen, direkt auf die Seitentabellen einer CPU zuzugreifen und der GPU 102 vollen Zugriff auf den CPU-Speicher zu ermöglichen.
  • In einer Ausführungsform übertragen Kopier-Engines Daten zwischen mehreren PPUs 300 oder zwischen PPUs 300 und CPUs. Die Kopier-Engines können Seitenfehler für Adressen erzeugen, die nicht in den Seitentabellen abgebildet sind. Die Speicherpartitionseinheit 380 kann dann die Seitenfehler bearbeiten und die Adressen in die Seitentabelle abbilden, woraufhin die Kopier-Engine die Übertragung durchführen kann. In einem herkömmlichen System wird der Speicher für mehrere Kopier-Engine-Operationen zwischen mehreren Prozessoren gepinnt (z.B. nicht auslagerbar), wodurch der verfügbare Speicher erheblich reduziert wird. Mit Hardware Page Faulting können Adressen an die Kopier-Engines weitergegeben werden, ohne dass man sich Gedanken darüber machen muss, ob die Speicherseiten resident sind, und ist der Kopiervorgang transparent.
  • Daten aus dem Speicher 304 oder einem anderen Systemspeicher können von der Speicherpartitionseinheit 380 abgerufen und in dem L2-Cache 460 gespeichert werden, der sich auf dem Chip befindet und von den verschiedenen GPCs 350 gemeinsam genutzt wird. Wie gezeigt ist, beinhaltet jede Speicherpartitionseinheit 380 einen Teil des L2-Cache 460, der einer entsprechenden Speichervorrichtung 304 zugeordnet ist. Caches der unteren Ebene können dann in verschiedenen Einheiten innerhalb der GPCs 350 implementiert sein. Zum Beispiel kann jeder der SMs 440 einen Cache der Ebene eins (L1) implementieren. Der L1-Cache (der ein einheitlicher Cache und gemeinsam genutzter Speicher sein kann) ist privater Speicher, der einem bestimmten oder mehreren SM(s) 440 zugeordnet ist. Daten aus dem L2-Cache 460 können abgerufen und in jedem der L1-Caches zur Verarbeitung in den Funktionseinheiten der SMs 440 gespeichert werden. Der L2-Cache 460 ist mit der Speicherschnittstelle 470 und der XBar 370 gekoppelt.
  • Die ROP-Einheit 450 führt Grafikrasteroperationen durch, die sich auf die Pixelfarbe beziehen, wie z.B. Farbkomprimierung, Pixelüberblendung und dergleichen. Die ROP-Einheit 450 implementiert auch eine Tiefenprüfung in Verbindung mit der Raster-Engine 425, wobei sie eine Tiefe für eine Abtaststelle, die einem Pixelfragment zugeordnet ist, von der Culling-Engine der Raster-Engine 425 empfängt. Die Tiefe wird gegen eine entsprechende Tiefe in einem Tiefenpuffer für eine dem Fragment zugeordnete Abtastposition getestet. Falls das Fragment den Tiefentest für die Abtastposition besteht, aktualisiert die ROP-Einheit 450 den Tiefenpuffer und überträgt ein Ergebnis des Tiefentests an die Raster-Engine 425. Es wird deutlich, dass die Anzahl der Partitionseinheiten 380 von der Anzahl der GPCs 350 abweichen kann und daher jede ROP-Einheit 450 mit jedem der GPCs 350 gekoppelt sein kann. Die ROP-Einheit 450 nachverfolgt die von den verschiedenen GPCs 350 empfangenen Pakete und bestimmt, an welchen GPC 350 ein von der ROP-Einheit 450 erzeugtes Ergebnis über die Xbar 370 weitergeleitet wird. Obwohl die ROP-Einheit 450 in 21 innerhalb der Speicherpartitionseinheit 380 enthalten ist, kann sich die ROP-Einheit 450 in anderen Ausführungsformen auch außerhalb der Speicherpartitionseinheit 380 befinden. Beispielsweise kann sich die ROP-Einheit 450 in dem GPC 350 oder einer anderen Einheit befinden.
  • 22 veranschaulicht den Streaming-Multiprozessor 440 aus 19 in Übereinstimmung mit einer Ausführungsform. Wie in 22 gezeigt ist, beinhaltet der SM 440 einen Befehls-Cache 505, eine oder mehrere Scheduler-Einheiten 510, eine Registerdatei 520, einen oder mehrere Verarbeitungskerne 550, eine oder mehrere Spezialfunktionseinheiten (SFUs) 552, eine oder mehrere Lade-/ Speichereinheiten (LSUs) 554, ein Verbindungsnetzwerk 580, und einen Shared Memory/L1-Cache 570.
  • Wie vorstehend beschrieben wurde, verteilt die Arbeitsverteilungseinheit 325 Aufgaben zur Ausführung auf den GPCs 350 der GPU 102. Die Aufgaben werden einem bestimmten DPC 420 innerhalb eines GPCs 350 zugewiesen, und falls die Aufgabe einem Shader-Programm zugeordnet ist, kann die Aufgabe einem SM 440 zugewiesen werden. Die Scheduler-Einheit 510 empfängt die Aufgaben von der Arbeitsverteilungseinheit 325 und verwaltet die Anweisungsplanung für einen oder mehrere Thread-Blöcke, die dem SM 440 zugewiesen sind. Die Scheduler-Einheit 510 plant Thread-Blöcke für die Ausführung als Warps von parallelen Threads ein, wobei jedem Thread-Block zumindest ein Warp zugeordnet ist. In einer Ausführungsform führt jeder Warp 32 Threads aus. Die Scheduler-Einheit 510 kann eine Vielzahl verschiedener Thread-Blöcke verwalten, indem sie die Warps den verschiedenen Thread-Blöcken zuweist und dann Anweisungen aus der Vielzahl verschiedener kooperativer Gruppen an die verschiedenen Funktionseinheiten (z.B. Kerne 550, SFUs 552 und LSUs 554) während jedes Taktzyklus versendet.
  • Kooperative Gruppen bzw. Cooperative Groups ist ein Programmiermodell zum Organisieren von Gruppen kommunizierender Threads, das es Entwicklern ermöglicht, die Granularität auszudrücken, mit der Threads kommunizieren, was den Ausdruck reichhaltigerer, effizienterer paralleler Dekompositionen ermöglicht. APIs für kooperativen Start unterstützen Synchronisierung zwischen Thread-Blöcken für die Ausführung paralleler Algorithmen. Herkömmliche Programmiermodelle bieten ein einziges, einfaches Konstrukt zum Synchronisieren kooperierender Threads: eine Sperre über alle Threads eines Threadblocks (z.B. die Funktion syncthreads()). Programmierer möchten jedoch oft Gruppen von Threads mit einer kleineren Granularität als Thread-Blöcke definieren und innerhalb der definierten Gruppen synchronisieren, um eine höhere Leistung, Designflexibilität und Software-Wiederverwendung in Form von kollektiven gruppenweiten Funktionsschnittstellen zu ermöglichen.
  • Kooperative Gruppen ermöglichen es Programmierern, Gruppen von Threads explizit auf Sub-Block- (z.B. so klein wie ein einzelner Thread) und Multi-Block-Granularität zu definieren und kollektive Operationen wie beispielsweise Synchronisation auf den Threads in einer kooperativen Gruppe durchzuführen. Das Programmiermodell unterstützt eine saubere Komposition über Software-Grenzen hinweg, so dass Bibliotheken und Utility-Funktionen innerhalb ihres lokalen Kontexts sicher synchronisieren können, ohne Annahmen über Konvergenz treffen zu müssen. Primitive für kooperative Gruppen ermöglichen neue Muster kooperativer Parallelität, einschließlich Producer-Consumer-Parallelität, opportunistischer Parallelität und globaler Synchronisierung über ein ganzes Netz von Thread-Blöcken.
  • Eine Sende-Einheit 515 ist dazu konfiguriert, Anweisungen an eine oder mehrere der Funktionseinheiten zu übermitteln. In der Ausführungsform umfasst die Scheduler-Einheit 510 zwei Sende-Einheiten 515, die es ermöglichen, dass zwei verschiedene Anweisungen aus demselben Warp während jedes Taktzyklus versendet werden. In alternativen Ausführungsformen kann jede Scheduler-Einheit 510 eine einzelne Sende-Einheit 515 oder zusätzliche Sende-Einheiten 515 beinhalten.
  • Jeder SM 440 beinhaltet eine Registerdatei 520, die einen Satz von Registern für die Funktionseinheiten des SM 440 bereitstellt. In einer Ausführungsform ist die Registerdatei 520 zwischen jeder der Funktionseinheiten aufgeteilt, so dass jeder Funktionseinheit ein dedizierter Teil der Registerdatei 520 zugeordnet ist. In einer anderen Ausführungsform ist die Registerdatei 520 zwischen den verschiedenen Warps aufgeteilt, die von dem SM 440 ausgeführt werden. Die Registerdatei 520 stellt einen temporären Speicher für Operanden bereit, die mit den Datenpfaden der Funktionseinheiten verbunden sind.
  • Jeder SM 440 umfasst L Verarbeitungskerne 550. In einer Ausführungsform beinhaltet der SM 440 eine große Anzahl (z.B. 128 usw.) unterschiedlicher Verarbeitungskerne 550. Jeder Kern 550 kann eine vollständig pipelinefähige Verarbeitungseinheit mit einfacher, doppelter und/oder gemischter Genauigkeit beinhalten, die eine arithmetische Gleitkomma-Logikeinheit und eine arithmetische Ganzzahl-Logikeinheit beinhaltet. In einer Ausführungsform implementieren die arithmetischen Gleitkomma-Logikeinheiten den Standard IEEE 754-2008 für Gleitkomma-Arithmetik. In einer Ausführungsform beinhalten die Kerne 550 64 Gleitkomma-Kerne mit einfacher Genauigkeit (32 Bit), 64 Ganzzahl-Kerne, 32 Gleitkomma-Kerne mit doppelter Genauigkeit (64 Bit) und 8 Tensor-Kerne.
  • Tensor-Kerne sind dazu konfiguriert, Matrixoperationen durchzuführen, und in einer Ausführungsform sind ein oder mehrere Tensor-Kerne in den Kernen 550 enthalten. Insbesondere sind die Tensorkerne dazu konfiguriert, Deep-Learning-Matrixarithmetik durchführen, wie z.B. Faltungsoperationen für das Training und die Inferenzierung neuronaler Netze. In einer Ausführungsform arbeitet jeder Tensorkern auf einer 4x4-Matrix und führt eine Matrixmultiplikations- und Akkumulationsoperation D=A*B+C durch, worin A, B, C und D 4x4-Matrizen sind.
  • In einer Ausführungsform sind die Matrixmultiplikationseingänge A und B 16-Bit-Gleitkommamatrizen, während die Akkumulationsmatrizen C und D 16-Bit-Gleitkomma- oder 32-Bit-Gleitkomma-Matrizen sein können. Tensorkerne arbeiten auf 16-Bit-Gleitkomma-Eingangsdaten mit 32-Bit-Gleitkomma-Akkumulation. Die 16-Bit-Gleitkomma-Multiplikation erfordert 64 Operationen und resultiert in einem Produkt mit voller Genauigkeit, das dann unter Verwendung von 32-Bit-Gleitkomma-Addition mit den anderen Zwischenprodukten für eine 4x4x4-Matrixmultiplikation akkumuliert wird. In der Praxis werden Tensorkerne verwendet, um viel größere zweidimensionale oder höherdimensionale Matrixoperationen durchzuführen, die aus diesen kleineren Elementen aufgebaut sind. Eine API, wie z.B. die CUDA 9 C++ API, exponiert spezialisierte Matrix-Lade-, Matrix-Multiplikations- und Akkumulationssowie Matrix-Speicher-Operationen, um Tensorkerne von einem CUDA-C++ Programm aus effizient zu nutzen. Auf der CUDA-Ebene geht die Schnittstelle auf Warp-Ebene von Matrizen der Größe 16x16 aus, die sich über alle 32 Threads des Warps erstrecken.
  • In einigen Ausführungsformen ist Transpositionshardware in den Verarbeitungskernen 550 oder einer anderen Funktionseinheit (z.B. SFUs 552 oder LSUs 554) enthalten und dazu konfiguriert, Matrixdaten zu erzeugen, die von Diagonalen gespeichert werden, und/oder die ursprüngliche Matrix und/oder transponierte Matrix aus den von Diagonalen gespeicherten Matrixdaten zu erzeugen. Die Transpositionshardware kann innerhalb des Ladepfads von dem gemeinsam genutzten Speichers570 zu der Registerdatei 520 des SM 440 bereitgestellt sein.
  • In einem Beispiel können die von Diagonalen gespeicherten Matrixdaten aus dem DRAM geholt und in dem gemeinsam genutzten Speicher 570 gespeichert werden. Wenn die Anweisung zur Durchführung der Verarbeitung unter Verwendung der von Diagonalen gespeicherten Matrixdaten verarbeitet wird, kann die in dem Pfad des gemeinsamen Speichers 570 und der Registerdatei 520 angeordnete Transpositionshardware die ursprüngliche Matrix, die transponierte Matrix, die verdichtete ursprüngliche Matrix und/oder die verdichtete transponierte Matrix bereitstellen. Bis zu der allerletzten Speicherung vor der Anweisung können die Einzelmatrixdaten, die in Diagonalen gespeichert sind, beibehalten werden, und wird der durch die Anweisung bestimmte Matrixtyp nach Bedarf in der Registerdatei 520 erzeugt.
  • Jeder SM 440 umfasst darüber hinaus M SFUs 552, die spezielle Funktionen (z.B. Attributauswertung, reziproke Quadratwurzel und dergleichen) durchführen. In einer Ausführungsform können die SFUs 552 eine Baumdurchlaufeinheit bzw. Tree Traversal Unit enthalten, die zum Durchlaufen einer hierarchischen Baumdatenstruktur konfiguriert ist. In einer Ausführungsform können die SFUs 552 eine Textureinheit beinhalten, die dazu konfiguriert ist, Texturkarten-Filterungsoperationen durchzuführen. In einer Ausführungsform sind die Textureinheiten dazu konfiguriert, Texturkarten (z.B. ein 2D-Array von Texeln) aus dem Speicher 304 zu laden und die Texturkarten abzutasten, um abgetastete Texturwerte zur Verwendung in von dem SM 440 ausgeführten Shader-Programmen zu erzeugen. In einer Ausführungsform werden die Texturkarten in dem Shared Memory/L1-Cache 470 gespeichert. Die Textureinheiten implementieren Texturoperationen, wie z.B. Filteroperationen, unter Verwendung von Mip-Maps (z.B. Texturkarten mit unterschiedlichen Detailstufen). In einer Ausführungsform beinhaltet jeder SM 340 zwei Textureinheiten.
  • Jeder SM 440 umfasst darüber hinaus N LSUs (Load-Store Units) 554, die Lade- und Speicheroperationen zwischen dem Shared Memory/L1-Cache 570 und der Registerdatei 520 implementieren. Jeder SM 440 beinhaltet ein Verbindungsnetzwerk 580, das jede der Funktionseinheiten mit der Registerdatei 520 und die LSU 554 mit der Registerdatei 520 und dem Shared Memory/L1-Cache 570 verbindet. In einer Ausführungsform ist das Verbindungsnetzwerk 580 eine Crossbar, die dazu konfiguriert sein kann, jede der Funktionseinheiten mit jedem der Register in der Registerdatei 520 zu verbinden und die LSUs 554 mit der Registerdatei 520 und Speicherplätzen in dem Shared Memory/L1-Cache 570 zu verbinden.
  • Der Shared Memory/L1-Cache 570 ist ein Array von On-Chip-Speicher, der die Datenspeicherung und Kommunikation zwischen dem SM 440 und der Stammfunktionen-Engine 435 sowie zwischen Threads in dem SM 440 ermöglicht. In einer Ausführungsform umfasst der Shared Memory/L1-Cache 570 128 KB Speicherkapazität und befindet sich in dem Pfad von dem SM 440 zu der Partitionseinheit 380. Der Shared Memory/L1-Cache 570 kann zum Zwischenspeichern von Lese- und Schreibvorgängen verwendet werden. Einer oder mehrere des Shared Memory/L1-Cache 570, des L2-Cache 460 und des Speichers 304 sind Sicherungskopiespeicher.
  • Die Kombination von Daten-Cache und Shared-Memory-Funktionalität in einem einzigen Speicherblock stellt die beste Gesamtleistung für beide Arten von Speicherzugriffen bereit. Die Kapazität ist als ein Zwischenspeicher bzw. Cache durch Programme nutzbar, die keinen gemeinsam genutzten Speicher verwenden. Falls z.B. gemeinsam genutzter Speicher dazu konfiguriert ist, die Hälfte der Kapazität zu nutzen, können Textur- und Lade-/Speicher-Operationen die verbleibende Kapazität nutzen. Durch die Integration in den Shared Memory/L1-Cache 570 kann der Shared Memory/L1-Cache 570 als eine Leitung mit hohem Durchsatz zum Streamen von Daten fungieren und gleichzeitig einen Zugriff mit hoher Bandbreite und niedriger Latenz auf häufig wiederverwendete Daten bereitstellen.
  • Bei einer Konfiguration für allgemeine parallele Berechnungen kann eine einfachere Konfiguration als bei der Grafikverarbeitung verwendet werden. Genauer werden die in 18 gezeigten Grafikverarbeitungseinheiten mit festen Funktionen umgangen, wodurch ein viel einfacheres Programmiermodell entsteht. In der Konfiguration für allgemeine parallele Berechnungen weist die Arbeitsverteilungseinheit 325 Blöcke von Threads zu und verteilt diese direkt an die DPCs 420. Die Threads in einem Block führen dasselbe Programm aus, wobei eine eindeutige Thread-ID in der Berechnung verwendet wird, um sicherzustellen, dass jeder Thread eindeutige Ergebnisse erzeugt, wobei der SM 440 zur Ausführung des Programms und zur Durchführung von Berechnungen, der Shared Memory/L1-Cache 570 zur Kommunikation zwischen Threads und die LSU 554 zum Lesen und Schreiben des globalen Speichers über den Shared Memory/L1-Cache 570 und die Speicherpartitionseinheit 380 verwendet werden. Wenn er für allgemeine parallele Berechnungen konfiguriert ist, kann der SM 440 auch Befehle schreiben, die die Scheduler-Einheit 320 verwenden kann, um neue Arbeit auf den DPCs 420 zu starten.
  • Die GPU 102 kann in einem Desktop-Computer, einem Laptop-Computer, einem Tablet-Computer, Servern, Supercomputern, einem Smartphone (z.B. einem drahtlosen, tragbaren Gerät), einem persönlichen digitalen Assistenten (PDA), einer Digitalkamera, einem Fahrzeug, einer am Kopf getragenen Anzeige bzw. einem Head Mounted Display, einem tragbaren elektronischen Gerät und dergleichen enthalten sein. In einer Ausführungsform ist die GPU 102 auf einem einzigen Halbleitersubstrat verkörpert. In einer anderen Ausführungsform ist die GPU 102 in einem System auf einem Chip bzw. System-on-a-Chip (SoC) zusammen mit einer oder mehreren anderen Vorrichtungen wie beispielsweise zusätzlichen PPUs 300, dem Speicher 304, einer RISC (Reduced Instruction Set Computer)-CPU, einer MMU (Memory Management Unit), einem DAC (Digital-Analog-Wandler) und ähnlichem enthalten.
  • In einer Ausführungsform kann die GPU 102 auf einer Grafikkarte enthalten sein, die eine oder mehrere Speichervorrichtungen 304 beinhaltet. Die Grafikkarte kann dazu konfiguriert sein, mit einem PCIe-Steckplatz auf einer Hauptplatine bzw. einem Motherboard eines Desktop-Computers verbunden zu werden. In einer weiteren Ausführungsform kann die GPU 102 eine integrierte Grafikverarbeitungseinheit (iGPU) oder ein Parallelprozessor sein, der im Chipsatz der Hauptplatine enthalten ist.
  • Beispielhaftes Rechensystem
  • Systeme mit mehreren GPUs, Fabric-Attached-Speicher und CPUs werden in einer Vielzahl von Branchen verwendet, da Entwickler mehr Parallelität in Anwendungen wie beispielsweise Berechnungen künstlicher Intelligenz exponieren und nutzen. Leistungsstarke GPU-beschleunigte Systeme mit zehn bis vielen tausend Rechenknoten werden in Rechenzentren, Forschungseinrichtungen und Supercomputern eingesetzt, um immer größere Probleme zu lösen. Da die Anzahl der Verarbeitungsvorrichtungen innerhalb der Hochleistungssysteme steigt, müssen die Kommunikations- und Datenübertragungsmechanismen skaliert werden, um die erhöhte Bandbreite zu unterstützen.
  • 23 ist ein konzeptionelles Diagramm eines Verarbeitungssystems 500, das unter Verwendung der GPU 102 gemäß einer Ausführungsform implementiert ist. Das beispielhafte System 500 kann dazu konfiguriert sein, die in dieser Anwendung offenbarten Verfahren zu implementieren. Das Verarbeitungssystem 500 beinhaltet eine CPU 530, einen Switch 555 und mehrere PPUs 300 sowie entsprechende Speicher 304. Das NVLINK™ 108-Interconnect-Fabric stellt Hochgeschwindigkeits-Kommunikationsverbindungen zwischen jeder der PPUs 300 bereit. Obwohl in 23 eine bestimmte Anzahl von NVLINK™ 108- und Interconnect 302-Verbindungen dargestellt ist, kann die Anzahl der Verbindungen zu jeder GPU 102 und der CPU 150 variieren. Der Switch 555 bildet die Schnittstelle zwischen der Verbindung 302 und der CPU 150. Die PPUs 300, die Speicher 304 und die NVLinks 108 können auf einer einzigen Halbleiterplattform angeordnet sein, um ein Parallelverarbeitungsmodul 525 zu bilden. In einer Ausführungsform unterstützt der Switch 555 zwei oder mehr Protokolle, um zwischen verschiedenen Verbindungen und/oder Links zu vermitteln.
  • In einer anderen Ausführungsform (nicht gezeigt) stellt der NVLINK™ 108 eine oder mehrere Hochgeschwindigkeits-Kommunikationsverbindungen zwischen jeder der PPUs 300 und der CPU 150 bereit, und bildet der Switch 555 eine Schnittstelle zwischen der Verbindung 302 und jeder der PPUs 300. Die PPUs 300, die Speicher 304 und die Verbindung 302 können auf einer einzigen Halbleiterplattform angeordnet sein, um ein Parallelverarbeitungsmodul 525 zu bilden. In einer weiteren Ausführungsform (nicht gezeigt) stellt die Verbindung 302 eine oder mehrere Kommunikationsverbindungen zwischen jeder der PPUs 300 und der CPU 150 bereit, und bildet der Switch 555 eine Schnittstelle zwischen jeder der PPUs 300 unter Verwendung des NVLINK™ 108, um eine oder mehrere Hochgeschwindigkeits-Kommunikationsverbindungen zwischen den PPUs 300 bereitzustellen. In einer anderen Ausführungsform (nicht gezeigt) stellt der NVLINK™ 310 über den Switch 555 eine oder mehrere Hochgeschwindigkeits-Kommunikationsverbindungen zwischen den PPUs 300 und der CPU 150 bereit. In einer weiteren Ausführungsform (nicht gezeigt) stellt die Verbindung 302 eine oder mehrere Kommunikationsverbindungen zwischen den einzelnen PPUs 300 direkt bereit. Eine oder mehrere der NVLINK™ 108-Hochgeschwindigkeits-Kommunikationsverbindungen können als physische NVLINK™-Verbindung oder entweder als eine On-Chip- oder On-Die-Verbindung implementiert sein, die das gleiche Protokoll wie der NVLINK™ 108 verwendet.
  • Im Kontext der vorliegenden Beschreibung kann sich eine einzelne Halbleiterplattform auf eine einzige einheitliche halbleiterbasierte integrierte Schaltung beziehen, die auf einem Die oder Chip hergestellt ist. Es wird angemerkt, dass sich der Begriff einer einzelnen Halbleiterplattform auch auf Multi-Chip-Module mit erhöhter Konnektivität beziehen kann, die On-Chip-Betrieb simulieren und wesentliche Verbesserungen gegenüber der Verwendung einer herkömmlichen Busimplementierung mit sich bringen. Natürlich können die verschiedenen Schaltungen oder Vorrichtungen je nach den Wünschen des Anwenders auch separat oder in verschiedenen Kombinationen von Halbleiterplattformen angeordnet sein. Alternativ kann das Parallelverarbeitungsmodul 525 als ein Leiterplattensubstrat implementiert sein und können jede der PPUs 300 und/oder der Speicher 304 gehäuste Vorrichtungen sein. In einer Ausführungsform befinden sich die CPU 150, der Switch 555 und das Parallelverarbeitungsmodul 525 auf einer einzigen Halbleiterplattform.
  • In einer Ausführungsform beträgt die Signalisierungsrate jedes NVLINK™ 108 20 bis 25 Gigabit/Sekunde und beinhaltet jede GPU 102 sechs NVLINK™ 108-Schnittstellen (wie in 23 gezeigt, sind fünf oder zwölf NVLINK™ 108-Schnittstellen für jede GPU 102 enthalten). Jeder NVLINK™ 108 stellt eine Datenübertragungsrate von 25 Gigabyte/Sekunde in jeder Richtung bereit, wobei sechs Verbindungen 300 Gigabyte/Sekunde liefern. Die NVLinks 108 können ausschließlich für GPU-zu-GPU- und GPU-zu-FAM-Kommunikation verwendet werden, wie in 23 gezeigt ist, oder für eine Kombination aus GPU-zu-GPU- und GPU-zu-CPU-Kommunikation, wenn die CPU 150 auch eine oder mehrere NVLINK™ 108-Schnittstellen enthält.
  • In einer Ausführungsform erlaubt der NVLINK™ 108 einen direkten Lade/Speicher/atomischen Zugriff auf den Speicher 304 jeder PPU 300. In einer Ausführungsform unterstützt der NVLINK™ 108 Kohärenzoperationen, wodurch aus den Speichern 304 gelesene Daten in der Cache-Hierarchie der CPU 150 gespeichert werden können, was die Cache-Zugriffslatenz für die CPU 150 reduziert. In einer Ausführungsform beinhaltet der NVLINK™ 150 bzw. 108 Unterstützung für Adressübersetzungsdienste (adress translation services, ATS), wodurch die GPU 102 direkt auf Seitentabellen innerhalb der CPU 150 zugreifen kann. Einer oder mehrere der NVLinks 108 können auch dazu konfiguriert sein, in einem stromsparenden Modus zu arbeiten.
  • 24 veranschaulicht ein beispielhaftes System 565, in welchem die verschiedenen Architekturen und/oder Funktionalität der verschiedenen vorangehenden Ausführungsformen implementiert sein können. Das beispielhafte System 565 kann dazu konfiguriert sein, die in dieser Anwendung offenbarte Technologie zu implementieren.
  • Wie gezeigt ist, wird ein System 565 bereitgestellt, das zumindest eine zentrale Verarbeitungseinheit 150 beinhaltet, die mit einem Kommunikationsbus 575 verbunden ist. Der Kommunikationsbus 575 kann unter Verwendung jedes beliebigen geeigneten Protokolls implementiert sein, wie z.B. PCI (Peripheral Component Interconnect), PCI-Express, AGP (Accelerated Graphics Port), HyperTransport oder jedes beliebige andere Bus- oder Punkt-zu-Punkt-Kommunikationsprotokoll. Das System 565 beinhaltet auch einen Hauptspeicher 540. Steuerlogik (Software) und Daten werden im Hauptspeicher 540 gespeichert, welcher die Form eines Direktzugriffsspeichers (RAM) annehmen kann.
  • Das System 565 beinhaltet auch Eingabevorrichtungen 560, das Parallelverarbeitungssystem 525 und Anzeigevorrichtungen 545, z.B. eine herkömmliche CRT (Kathodenstrahlröhre), LCD (Flüssigkristallanzeige), LED (lichtemittierende Diode), Plasmaanzeige oder dergleichen. Benutzereingaben können von den Eingabevorrichtungen 560, z.B. einer Tastatur, einer Maus, einem Touchpad, einem Mikrofon und dergleichen, empfangen werden. Jedes der vorgenannten Module und/oder Geräte kann sogar auf einer einzigen Halbleiterplattform angeordnet sein, um das System 565 zu bilden. Alternativ können die verschiedenen Module je nach den Wünschen des Benutzers auch separat oder in verschiedenen Kombinationen von Halbleiterplattformen angeordnet sein.
  • Ferner kann das System 565 über eine Netzwerkschnittstelle 535 zu Kommunikationszwecken mit einem Netzwerk (z.B. einem Telekommunikationsnetzwerk, einem lokalen Netzwerk (LAN), einem drahtlosen Netzwerk, einem Weitverkehrsnetzwerk (WAN) wie dem Internet, einem Peer-to-Peer-Netzwerk, einem Kabelnetzwerk oder dergleichen) verbunden sein.
  • Das System 565 kann auch einen Sekundärspeicher (nicht gezeigt) beinhalten. Der Sekundärspeicher beinhaltet beispielsweise ein Festplattenlaufwerk und/oder ein Wechselspeicherlaufwerk, das ein Diskettenlaufwerk, ein Magnetbandlaufwerk, ein Compact-Disk-Laufwerk, ein DVD (Digital Versatile Disk)-Laufwerk, eine Aufzeichnungsvorrichtung oder einen USB (Universal Serial Bus)-Flash-Speicher repräsentiert. Das Wechselspeicherlaufwerk liest in bekannter Weise von einer und/oder schreibt auf eine Wechselspeichereinheit.
  • Computerprogramme oder Computersteuerungslogik-Algorithmen können in dem Hauptspeicher 540 und/oder in dem Sekundärspeicher gespeichert sein. Solche Computerprogramme ermöglichen, wenn sie ausgeführt werden, dem System 565, verschiedene Funktionen auszuführen. Der Speicher 540, der Speicher und/oder jeder beliebige andere Speicher sind mögliche Beispiele für computerlesbare Medien.
  • Die Architektur und/oder die Funktionalität der verschiedenen vorangehenden Figuren kann im Kontext eines allgemeinen Computersystems, eines Leiterplattensystems, eines Spielkonsolensystems, das für Unterhaltungszwecke bestimmt ist, eines anwendungsspezifischen Systems und/oder eines beliebigen anderen gewünschten Systems implementiert sein. Beispielsweise kann das System 565 die Form eines Desktop-Computers, eines Laptop-Computers, eines Tablet-Computers, eines Servers, eines Supercomputers, eines Smartphones (z.B. eines drahtlosen, handgehaltenen Geräts), eines persönlichen digitalen Assistenten (PDA), einer Digitalkamera, eines Fahrzeugs, eines Head-Mounted-Displays, eines handgehaltenen elektronischen Geräts, eines Mobiltelefongeräts, eines Fernsehers, einer Arbeitsstation, von Spielkonsolen, eines eingebetteten Systems und/oder jeder beliebigen anderen Art von Logik annehmen.
  • ******
  • Zusammenfassend ermöglicht Fabric-Attached-Speicher (FAM) eine viel höhere Kapazität bei hoher Bandbreite und niedriger Latenz. FAM erlaubt es, dass Speicherkapazität und Bandbreite unabhängig von GPUs und CPUs wachsen kann. FAM ermöglicht es Systemen auch, einen Speicher-„Disaggregations“-Pool mit mehreren TBs und mehreren TB/s Bandbreite zu erreichen. Es wird erwartet, dass solche Fähigkeiten im Wettbewerb von Rechenzentrum-Anwendungen besonders hilfreich sind, während sie bestehende Hardware- und Software-Technologien als Bausteine (z.B. NVLink/NVSwitch, CUDA, UVM usw.) nutzen. Beispielhafte Anwendungsfälle beinhalten:
    • • Big Data (z.B. In-Memory-Datenbanken, Graphen-Analytik, ETL (extraction, transform, load)-Analytik)
    • • HPC (Datenvisualisierung, Quantenchemie, Astrophysik (Square Kilometer Array von Radioteleskopen)
    • • KI (Recommender-Engines, Deep Learning-Datensätze, Parameter- & Temporaldatenspeicherung, Netzwerkaktivierungs-Auslagerung, berechnende Pathologie, medizinische Bildgebung
    • • Grafik-Rendering
    • • Überall dort, wo große Datenmengen vorhanden sind, auf die mit hoher Bandbreite zugegriffen werden muss.
  • Beispielhafte Merkmalskombinationen
  • Einige beispielhafte, nicht beschränkende Ausführungsformen stellen somit einen Fabric-Attached-Speicher bereit, der einen Grafikprozessor, der dazu konfiguriert ist, mit einem Interconnect-Fabric zu kommunizieren; und zumindest einen Speicher, der operativ mit dem Grafikprozessor gekoppelt ist, umfasst, wobei der Grafikprozessor dazu strukturiert ist, zumindest einen atomischen Lese-Änderungs-Schreib-Speicherzugriffsbefehl auf dem zumindest einen Speicher durchzuführen, wobei der Grafikprozessor ferner so konfiguriert ist, dass eine Rechenschaltungsfähigkeit defekt, deaktiviert oder nicht vorhanden ist.
  • Der Grafikprozessor-Rechenschaltkreis ist abgeschmolzen. Der Grafikprozessor umfasst zumindest einen Streaming-Multiprozessor. Das Interconnect-Fabric kann NVIDIA NVLINK™ umfassen.
  • Der Grafikprozessor kann eine Vielzahl von Fabric-Interconnect-Ports beinhalten, von denen nur eine Teilmenge dazu konfiguriert ist, mit dem Interconnect-Fabric verbunden zu sein. Der Speicher kann zumindest ein Dual-Inline-Speichermodul mit Halbleiter-Direktzugriffsspeicher umfassen.
  • Ein Fabric-Attached-Speichersystem kann ein Interconnect-Fabric; zumindest eine mit dem Interconnect-Fabric verbundene Source-GPU, wobei die Source-GPU eine Speicheradresse erzeugt; und mehrere Fabric-Attached-Speicher, die mit dem Interconnect-Fabric verbunden sind, wobei jeder der mehreren Fabric-Attached-Speicher einen Adressraum definiert, umfassen; wobei die Verbindung zwischen der Source-GPU und dem Interconnect-Fabric und die Verbindung zwischen jedem der Fabric-Attached-Speichervorrichtungen und dem Interconnect-Fabric asymmetrisch sind; und wobei zumindest eines der Source-GPU, des Interconnect-Fabric und der mehreren Fabric-Attached-Speicher einen Adresstransformator beinhaltet, der die Speicheradresse, die die zumindest eine Source-GPU erzeugt, in einen Adressraum des Fabric-Attached-Speichers transformiert.
  • Der Adresstransformator kann eine Teilungs- oder Verdichtungsschaltung umfassen. Der Adresstransformator kann einen Durchmischer und einen Adressverdichter beinhalten. Die zumindest eine GPU kann die generierte Adresse durchmischen, um eine Interconnect-Verbindung innerhalb des Interconnect-Fabric auszuwählen. Jeder Fabric-Attached-Speicher-Adressraum kann kleiner sein als ein Adressraum, der durch die von der GPU erzeugte Speicheradresse definiert ist.
  • Ein Interconnect-Fabric-Switch kann Eingangsports; Ausgangsports; und Routingtabellen umfassen, die es dem Switch ermöglichen, an Eingangsports empfangene Fabric-Attached-Speicher-Zugriffsanforderungen zu den Ausgangsports zu routen, wobei die Routingtabellen den Switch dazu steuern, Adressen innerhalb der Speicherzugriffsanforderungen selektiv zu verdichten, um Fabric-Attached-Speicher-Kapazität zu kompensieren.
  • Die Routingtabellen können ferner den Switch so steuern, dass er selektiv Adressen transformiert, um eine entropiebasierte Verteilung der Speicherzugriffsanforderungen an den Eingangsports zu kompensieren. Die Routingtabellen können ferner den Switch so steuern, dass er Adressen mischt, um Kollisionen von Speicherzugriffsanforderungen an verschiedenen Eingangsports zu verhindern, die auf denselben Fabric-Attached-Speicher konvergieren (in einigen Ausführungsformen ist das NVLINK™-Fabric bei FAM nicht vollständig konvergent, so dass eine bestimmte FAMM-Vorrichtung nur eine Teilmenge von Ebenen sehen muss). Die Routingtabellen können ferner eine Basis- und/oder Grenz-Adressprüfung für Adressen auswählen, die auf unregelmäßig große Bereiche des Fabric-Attached-Speichers abgebildet werden. Die Routingtabellen können ferner eine Adressversatzaddition ermöglichen, um eine andere Partition im Adressraum der Fabric-Attached-Speichervorrichtung auszuwählen.
  • Ein Verfahren zum Zugreifen auf einen Fabric-Attached-Speicher kann umfassen: Erzeugen einer Speicherzugriffsanforderung; Verwenden von Entropie, um eine Verbindung auszuwählen, über welche die Speicherzugriffsanforderung zu senden ist; Transformieren einer Adresse in der Speicherzugriffsanforderung, um die Entropieauswahl zu kompensieren; weiteres Transformieren der Adresse, um eine Abweichung zwischen der Größe der Adresse, die die transformierte Adresse definiert, und der Größe der Adresse eines Fabric-Attached-Speichers zu kompensieren; und Anwenden der weiter transformierten Adresse, um auf den Fabric-Attached-Speicher zuzugreifen.
  • Eine Fabric-Attached-Speicher-Basisplatte umfasst eine gedruckte Leiterplatte; eine Vielzahl von Fabric-Attached-Speichermodulen, die auf der gedruckten Leiterplatte angeordnet sind, wobei jedes der Vielzahl von Fabric-Attached-Speichermodulen mit einem Interconnect-Fabric verbunden ist, und einen Prozessor, der auf der gedruckten Leiterplatte angeordnet ist, wobei der Prozessor die Vielzahl von Fabric-Attached-Speichermodulen verwaltet; wobei die Vielzahl von Fabric-Attached-Speichermodulen jeweils in der Lage ist, atomische GPU-Speicheroperationen und Peer-to-Peer-GPU-Kommunikationen über das Interconnect-Fabric durchzuführen, während die Menge rechenfähiger GPUs von der Speicherkapazität, die von den Fabric-Attached-Speichermodulen bereitgestellt wird, disaggregiert ist.
  • Die Vielzahl der Fabric-Attached-Speichermodule kann jeweils eine Floor-Swept-GPU beinhalten, die zumindest teilweise defekt und/oder abgeschmolzen ist, um GPU-Rechenoperationen zu deaktivieren. Die Vielzahl der Fabric-Attached-Speichermodule kann jeweils einen Speichercontroller umfassen, der keine GPU-Rechenfähigkeit hat, aber umfasst: ein Boot-ROM; einen DDR-Speichercontroller, der in der Lage ist, die atomischen Funktionen ohne Emulation hardwarebeschleunigt durchzuführen; einen DRAM-Zeilen-Remapper; einen Daten-Cache; eine Crossbar-Verbindung; und eine Fabric-Interconnect-Schnittstelle, die zu Peer-zu-Peer-Kommunikation über das Interconnect-Fabric mit GPUs in der Lage ist.
  • * * * * * *
  • Alle Patente und gedruckten Veröffentlichungen, auf die vorstehend Bezug genommen wurde, werden durch Bezugnahme hierin wie als ausdrücklich dargelegt einbezogen.
  • Während die Erfindung in Verbindung mit derzeit als am praktischsten und bevorzugt erachteten Ausführungsformen beschrieben wurde, versteht sich, dass die Erfindung nicht auf die offenbarten Ausführungsformen zu beschränken ist, sondern im Gegenteil verschiedene Modifikationen und äquivalente Anordnungen abdecken soll, die innerhalb der Gedankens und des Schutzumfangs der beigefügten Ansprüche enthalten sind.
  • ZITATE ENTHALTEN IN DER BESCHREIBUNG
  • Diese Liste der vom Anmelder aufgeführten Dokumente wurde automatisiert erzeugt und ist ausschließlich zur besseren Information des Lesers aufgenommen. Die Liste ist nicht Bestandteil der deutschen Patent- bzw. Gebrauchsmusteranmeldung. Das DPMA übernimmt keinerlei Haftung für etwaige Fehler oder Auslassungen.
  • Zitierte Patentliteratur
    • US 7275123 [0008]
    • US 7627723 [0008]
    • US 7451259 [0008]
    • US 9767036 [0017]
    • US 9830210 [0017]
    • US 9424201 [0017]
    • US 9639474 [0017]
    • US 16/198649 [0053]
  • Zitierte Nicht-Patentliteratur
    • Achermann et al, „Separating Translation from Protection in Address Spaces with Dynamic Remapping“, Proceedings of the 16th Workshop on Hot Topics in Operating Systems Pages 118-124 (Whistler, BC, Canada, May 07 - 10, 2017) [0009]
    • Chen, Fei et al, „Billion node graph inference: iterative processing on The Machine“ (Hewlett Packard Labs HPE-2016-101, 2016) [0009]

Claims (35)

  1. Fabric-Attached-Speicher, umfassend: einen Prozessor, der dazu konfiguriert ist, mit einem Interconnect-Fabric zu kommunizieren; und zumindest einen Speicher, der operativ mit dem Prozessor gekoppelt ist, wobei der Prozessor dazu strukturiert ist, zumindest einen atomischen Lese-Änderung-Schreib-Speicherzugriffsbefehl auf dem zumindest einen Speicher auszuführen, wobei der Prozessor ferner derart konfiguriert ist, dass eine Rechenschaltungsfähigkeit desselben defekt, deaktiviert oder nicht vorhanden ist.
  2. Fabric-Attached-Speicher nach Anspruch 1, wobei die Rechenschaltung programmatisch deaktiviert, abgeschmolzen oder anderweitig absichtlich deaktiviert ist.
  3. Fabric-Attached-Speicher nach Anspruch 1 oder 2, wobei der Prozessor zumindest einen Streaming-Multiprozessor umfasst.
  4. Fabric-Attached-Speicher nach einem der vorangehenden Ansprüche, wobei das Interconnect-Fabric NVIDIA NVLINK™ umfasst.
  5. Fabric-Attached-Speicher nach einem der vorangehenden Ansprüche, wobei der Prozessor eine Vielzahl von Fabric-Interconnect-Ports aufweist, von welchen nur eine Teilmenge dazu konfiguriert ist, mit dem Interconnect Fabric verbunden zu sein.
  6. Fabric-Attached-Speicher nach einem der vorangehenden Ansprüche, wobei der Prozessor ein Boot-ROM, einen Speichercontroller, einen Zeilen-Remapper, einen Daten-Cache, eine Crossbar-Verbindung und eine Fabric-Verbindung beinhaltet.
  7. Fabric-Attached-Speicher nach einem der vorangehenden Ansprüche, wobei der Speicher ein Array von diskreten Halbleiter-Direktzugriffsspeichervorrichtungen umfasst.
  8. Fabric-Attached-Speichersystem, umfassend: ein Interconnect-Fabric zumindest eine Source-GPU, die mit dem Interconnect-Fabric verbunden ist, wobei die Source-GPU eine Speicheradresse erzeugt; und mehrere Fabric-Attached-Speicher, die mit dem Interconnect-Fabric verbunden sind, wobei die mehreren Fabric-Attached-Speicher jeweils einen Adressraum definieren; wobei die Verbindung zwischen der Source-GPU und dem Interconnect-Fabric und die Verbindung zwischen jedem der Fabric-Attached-Speicher und der Verbindungsstruktur asymmetrisch sind; und wobei zumindest eines der Source-GPU, des Interconnect-Fabric und der mehreren Fabric-Attached-Speicher einen Adresstransformator beinhaltet, der die Speicheradresse, die die zumindest eine Source-GPU erzeugt, in einen Adressraum des Fabric-Attached-Speichers transformiert.
  9. Fabric-Attached-Speichersystem nach Anspruch 8, wobei der Adresstransformator eine Teilungs- oder Verdichtungsschaltung umfasst.
  10. Fabric-Attached-Speichersystem nach Anspruch 8 oder 9, wobei der Adresstransformator einen Durchmischer, die mit dem von der Source-GPU durchgeführten Durchmischer übereinstimmt, und einen Adressverdichter beinhaltet.
  11. Fabric-Attached-Speichersystem nach Anspruch 10, wobei die zumindest eine GPU die erzeugte Adresse durchmischt, um eine Interconnect-Verbindung innerhalb des Interconnect-Fabric auszuwählen.
  12. Fabric-Attached-Speichersystem nach einem der Ansprüche 8 bis 11, wobei jeder Fabric-Attached-Speicher-Adressraum kleiner ist als ein Adressraum, der durch die von der GPU erzeugte Speicheradresse definiert ist.
  13. Interconnect-Fabric-Switch, umfassend: Eingangsports; Ausgangsports; und Routingtabellen, die den Switch in die Lage versetzen, an Eingangsports empfangene Fabric-Attached-Speicher-Zugriffsanforderungen zu den Ausgangsports zu routen, wobei die Routingtabellen den Switch dazu steuern, selektiv Adressen innerhalb der Speicherzugriffsanforderungen zu verdichten, um Fabric-Attached-Speicher-Kapazität zu kompensieren.
  14. Interconnect-Fabric-Switch nach Anspruch 13, wobei die Routingtabellen den Switch ferner dazu steuern, Adressen selektiv zu transformieren, um eine entropiebasierte Verteilung der Speicherzugriffsanforderungen an den Eingangsports zu kompensieren.
  15. Interconnect-Fabric-Switch nach Anspruch 13 oder 14, wobei die Routingtabellen den Switch ferner dazu steuern, die Adressen zu mischen, um Kollisionen von Speicherzugriffsanforderungen an verschiedenen Eingangsports, die an demselben Fabric-Attached-Speicher konvergieren, zu verhindern.
  16. Interconnect-Fabric-Switch nach einem der Ansprüche 13 bis 15, wobei die Routingtabellen ferner eine Basis- und/oder Grenzadressenprüfung für Adressen auswählen, die in unregelmäßig große Bereiche des Fabric-Attached-Speichers abgebildet werden.
  17. Interconnect-Fabric-Switch nach einem der Ansprüche 13 bis 16, wobei die Routingtabellen ferner eine Adress-Offset-Addition ermöglichen, um eine andere Partition in dem Fabric-Attached-Speicher auszuwählen und/oder um Adress-Kollisionen auf Fabric-Attached-Speichervorrichtungen zu vermeiden, die Ebenen konvergieren.
  18. Interconnect-Fabric-Switch nach einem der Ansprüche 13 bis 17, wobei der Switch eine automatische Zielidentifikationserzeugung/-inkrementierung auf verdichteten Adressen durchführt.
  19. Verfahren zum Zugriff auf einen Fabric-Attached-Speicher, umfassend: Erzeugen einer Speicherzugriffsanforderung; Verwenden von Entropie, um eine Verbindung auszuwählen, über welche die Speicherzugriffsanforderung zu senden ist; Transformieren einer Adresse innerhalb der Speicherzugriffsanforderung, um die Entropieauswahl zu kompensieren; weiteres Transformieren der Adresse, um eine Ungleichheit zwischen der Größe der Adresse, die die transformierte Adresse definiert, und der Größe der Adresse eines Fabric-Attached-Speichers zu kompensieren; und Anwenden der weiter transformierten Adresse, um auf den Fabric-Attached-Speicher zuzugreifen.
  20. Fabric-Attached-Speicher-Basisplatte, umfassend eine gedruckte Leiterplatte; eine Vielzahl von Fabric-Attached-Speichermodulen, die auf der gedruckten Leiterplatte angeordnet sind, wobei jedes der Vielzahl von Fabric-Attached-Speichermodulen mit einem Interconnect-Fabric verbunden ist, und einen Prozessor, der auf der gedruckten Leiterplatte angeordnet ist, wobei der Prozessor die Vielzahl von Fabric-Attached-Speichermodulen verwaltet; wobei jedes der Vielzahl von Fabric-Attached-Speichermodulen in der Lage ist, atomische GPU-Speicheroperationen und Peer-to-Peer-GPU-Kommunika-tionen über das Interconnect-Fabric durchzuführen, während die Menge rechenfähiger GPUs von Speicherkapazität, die von den Fabric-Attached-Speichermodulen bereitgestellt wird, disaggregiert wird.
  21. Fabric-Attached-Speicher-Basisplatte nach Anspruch 20, wobei jedes der Vielzahl von Fabric-Attached-Speichermodulen eine Floor-Swept-GPU enthält, die zumindest teilweise defekt und/oder abgeschmolzen ist, um GPU-Rechenoperationen zu deaktivieren.
  22. Fabric-Attached-Speicher-Basisplatte nach Anspruch 20 oder 21, wobei jedes der Vielzahl von Fabric-Attached-Speichermodulen einen Speichercontroller umfasst, der keine GPU-Rechenfähigkeit aufweist, aber zumindest umfasst: ein Boot-ROM; einen Speichercontroller, der in der Lage ist, atomische Speicherbefehle ohne Emulation in Hardware zu beschleunigen einen Zeilen-Remapper; einen Daten-Cache; eine Crossbar-Verbindung; und eine Fabric-Interconnect-Schnittstelle, die zur Peer-to-Peer-Kommunikation über das Interconnect-Fabric mit GPUs fähig ist.
  23. Verfahren zum Zugriff auf einen Fabric-Attached-Speicher, umfassend: Erzeugen einer Speicherzugriffsanforderung; Verwenden von Entropie, um eine Verbindung auszuwählen, über welche die Speicherzugriffsanforderung zu senden ist; und Verwenden einer Routingtabelle in einem Fabric-Switch, um eine Adresse innerhalb der Speicherzugriffsanforderung zu transformieren, um auswählbare Stripe-Schrittgrößen über einen Fabric-Attached-Speicher hinweg bereitzustellen.
  24. Verfahren zum Bereitstellen von Fabric-Attached-Speicher, umfassend: Anschließen zumindest einer Sammlung von Speichervorrichtungen an ein Interconnect-Fabric als einen Stripe, um mehr Gesamtkapazität oder Speicherbandbreite für zumindest eine Source-GPU bereitzustellen, auf der eine Anwendung läuft; und Ausführen von Speicherallozierungssoftware, um die Anzahl von Fabric-Attached-Speichervorrichtungen, die den Stripe umfassen, und die Anzahl von Verbindungen, über welche Zugriffe auf den Stripe erfolgen, in Abhängigkeit von den Kapazitäts- und Bandbreitenanforderungen der Anwendung zu konfigurieren.
  25. Verfahren nach Anspruch 24, ferner umfassend ein Dedizieren des Stripes an eine einzelne Source-GPU und/oder eine einzelne virtuelle Maschine durch Aufbauen des Stripes durch Entwurf eines Switching-Fabric und/oder durch Programmieren von Switch-Routingtabellen.
  26. Verfahren nach Anspruch 25, ferner umfassend ein Verhindern von Zugriffen von anderen GPUs als der einzelnen Source-GPU und/oder von anderen virtuellen Maschinen als der einzelnen virtuellen Maschine durch Sicherheitsüberprüfungen in dem Switching-Fabric.
  27. Verfahren nach einem der Ansprüche 24 bis 26, ferner eine gemeinsame Nutzung des Stripes zwischen mehreren Source-GPUs, die unter derselben virtuellen Maschine laufen, und/oder durch mehrere Source-GPUs, die unter verschiedenen virtuellen Maschinen laufen, abhängig von einem Data-Sharing-Modell für die Anwendung.
  28. Verfahren nach einem der Ansprüche 24 bis 27, ferner umfassend ein Steuern eines Engpasses, um sicherzustellen, dass Bandbreite für einen Satz von Fabric-Attached-Speicher-Stripes gleichmäßig unter mehreren Source-GPUs aufgeteilt wird.
  29. Verfahren nach Anspruch 28, wobei die Steuerung des Engpasses ein Verwenden einer Injektionsratenbegrenzung in der zumindest einen Source-GPU und/oder in Fabric-Switches umfasst.
  30. Verfahren zum Bereitstellen von Fabric-Attached-Speicher, umfassend: Anschließen zumindest einer Speichervorrichtung an ein Interconnect-Fabric, um mehr Gesamtkapazität oder Speicherbandbreite für zumindest eine Source-GPU bereitzustellen, auf der eine Anwendung läuft, wobei die zumindest eine Speichervorrichtung einer Spender-Rechenkomponente zugeordnet ist; und Signalisieren eines fatalen Fabric-Attached-Speicher-Fehlers in der Spender-Computerkomponente und/oder der Speichervorrichtung, wobei die Signalisierung genügend Informationen bereitstellt, um eine bestimmte Source-GPU und/oder virtuelle Maschine anzuzeigen, so dass Software zumindest eine chirurgische Korrekturmaßnahme durchführen kann.
  31. Verfahren nach Anspruch 30, wobei die Korrekturmaßnahme ein Herunterfahren nur einer von dem fatalen Fehler betroffenen GPU und/oder virtuellen Maschine umfasst, während andere GPUs oder virtuelle Maschinen von der Korrekturmaßnahme isoliert werden.
  32. Verfahren zum Bereitstellen Fabric-Attached-Speicher, umfassend Anschließen zumindest einer Speichervorrichtung und einer zugeordneten Spender-Rechenkomponente an ein Interconnect-Fabric, um mehr Gesamtkapazität oder Speicherbandbreite für zumindest eine Source-GPU, auf der eine Anwendung läuft, bereitzustellen; und Ermöglichen, dass die Spender-Rechenkomponente in Abhängigkeit von ihren Floor-Swept-Fähigkeiten bestimmte Housekeeping- und/oder Management-Aufgaben von einem zentralisierten Systemmanagement-Prozessor und/oder einer Host-CPU auslagert, einschließlich einer Durchführung einer Speicherdiagnose zur Systeminitialisierungszeit und/oder einer Sicherheitsmaßnahme.
  33. Verfahren nach Anspruch 32, wobei die Sicherheitsmaßnahme ein Löschen des Inhalts der zumindest einen Speichervorrichtung bei einer Verschiebung der Inhaberschaft an der zumindest einen Speichervorrichtung von einer virtuellen Maschine zu einer anderen umfasst.
  34. Fabric-Attached-Speichersystem, umfassend: ein Interconnect-Fabric; eine Grafikverarbeitungseinheit, die mit dem Interconnect-Fabric verbunden ist, wobei die Grafikverarbeitungseinheit dazu konfiguriert ist, eine Rechenfähigkeit bereitzustellen; einen ersten Speicher, der mit der Grafikverarbeitungseinheit verbunden ist; eine Verarbeitungsschaltung, die dazu konfiguriert ist, mit dem Interconnect-Fabric zu kommunizieren, wobei die Verarbeitungsschaltung ein Boot-ROM, einen Speichercontroller, einen Zeilen-Remapper, einen Daten-Cache, eine Crossbar-Verbindung und eine Interconnect-Fabric-Verbindung beinhaltet und so strukturiert ist, zumindest einen atomischen Lese-Änderung-Schreib-Speicherzugriffsbefehl ausführt, aber dazu konfiguriert ist, die Rechenfähigkeit nicht bereitzustellen; und einen zweiten Speicher, der mit der Verarbeitungsschaltung verbunden ist, wobei die Grafikverarbeitungseinheit in der Lage ist, über das Interconnect-Fabric und die Verarbeitungsschaltung atomisch auf den zweiten Speicher zuzugreifen.
  35. System nach Anspruch 34, wobei die Rechenfähigkeit eines oder mehreres des Folgenden umfasst: (a) atomische Addition, die auf Fließkommawerten in globalem und gemeinsam genutztem Speicher arbeitet; (b) Warp-Vote- und Ballot-Funktionen; (c) Memory-Fence-Funktionen; (d) Synchronisationsfunktionen; (e) Oberflächenfunktionen; (f) 3D-Gitter von Thread-Blöcken; (g) Trichterverschiebung; (h) dynamische Parallelität; (i) halbgenaue Gleitkommaoperationen: (j) Addition, Subtraktion, Multiplikation, Vergleich, Warp-Misch-Funktionen, Konvertierung; und (k) Tensorkern.
DE102020127705.9A 2019-11-04 2020-10-21 Techniken für einen effizienten fabric-attached-speicher Pending DE102020127705A1 (de)

Applications Claiming Priority (2)

Application Number Priority Date Filing Date Title
US16/673,537 2019-11-04
US16/673,537 US11182309B2 (en) 2019-11-04 2019-11-04 Techniques for an efficient fabric attached memory

Publications (1)

Publication Number Publication Date
DE102020127705A1 true DE102020127705A1 (de) 2021-05-06

Family

ID=75485592

Family Applications (1)

Application Number Title Priority Date Filing Date
DE102020127705.9A Pending DE102020127705A1 (de) 2019-11-04 2020-10-21 Techniken für einen effizienten fabric-attached-speicher

Country Status (3)

Country Link
US (2) US11182309B2 (de)
CN (1) CN112785485B (de)
DE (1) DE102020127705A1 (de)

Families Citing this family (14)

* Cited by examiner, † Cited by third party
Publication number Priority date Publication date Assignee Title
US10936492B2 (en) 2018-06-18 2021-03-02 FLC Technology Group, Inc. Method and apparatus for using a storage system as main memory
US11244222B2 (en) * 2018-06-27 2022-02-08 Sony Corporation Artificial intelligence-enabled device for network connectivity independent delivery of consumable information
US11182309B2 (en) * 2019-11-04 2021-11-23 Nvidia Corporation Techniques for an efficient fabric attached memory
US11914903B2 (en) * 2020-10-12 2024-02-27 Samsung Electronics Co., Ltd. Systems, methods, and devices for accelerators with virtualization and tiered memory
US20220156879A1 (en) * 2020-11-18 2022-05-19 Intel Corporation Multi-tile graphics processing unit
CN113674133B (zh) * 2021-07-27 2023-09-05 阿里巴巴新加坡控股有限公司 Gpu集群共享显存系统、方法、装置及设备
US20230112007A1 (en) * 2021-10-08 2023-04-13 Advanced Micro Devices, Inc. Global addressing for switch fabric
WO2023086574A1 (en) * 2021-11-11 2023-05-19 FLC Technology Group, Inc. Memory pooling bandwidth multiplier using final level cache system
US11989142B2 (en) 2021-12-10 2024-05-21 Samsung Electronics Co., Ltd. Efficient and concurrent model execution
CN114676011B (zh) * 2022-05-30 2022-08-12 芯耀辉科技有限公司 一种数据校验的验证方法、相关设备和存储介质
CN117632447A (zh) * 2022-08-09 2024-03-01 第四范式（北京）技术有限公司 Gpu资源使用方法、gpu虚拟化方法以及作业调度装置、集群
CN115981853A (zh) * 2022-12-23 2023-04-18 摩尔线程智能科技(北京)有限责任公司 Gpu互联架构、实现gpu互联架构的方法、计算设备
CN115933986B (zh) * 2022-12-29 2023-06-09 量子芯云(北京)微电子科技有限公司 一种具有高冗余量的计算存储阵列
CN117541896B (zh) * 2024-01-09 2024-04-09 深流微智能科技(深圳)有限公司 基于uvm的光栅化模块验证系统、方法及存储介质

Family Cites Families (22)

* Cited by examiner, † Cited by third party
Publication number Priority date Publication date Assignee Title
US7598958B1 (en) * 2004-11-17 2009-10-06 Nvidia Corporation Multi-chip graphics processing unit apparatus, system, and method
US7451259B2 (en) 2004-12-06 2008-11-11 Nvidia Corporation Method and apparatus for providing peer-to-peer data transfer within a computing environment
US7275123B2 (en) 2004-12-06 2007-09-25 Nvidia Corporation Method and apparatus for providing peer-to-peer data transfer within a computing environment
US7627723B1 (en) 2006-09-21 2009-12-01 Nvidia Corporation Atomic memory operators in a parallel processor
US9171264B2 (en) * 2010-12-15 2015-10-27 Microsoft Technology Licensing, Llc Parallel processing machine learning decision tree training
US9529712B2 (en) * 2011-07-26 2016-12-27 Nvidia Corporation Techniques for balancing accesses to memory having different memory types
US10133677B2 (en) 2013-03-14 2018-11-20 Nvidia Corporation Opportunistic migration of memory pages in a unified virtual memory system
US10445243B2 (en) 2013-03-14 2019-10-15 Nvidia Corporation Fault buffer for resolving page faults in unified virtual memory system
US9424201B2 (en) 2013-03-14 2016-08-23 Nvidia Corporation Migrating pages of different sizes between heterogeneous processors
US9830210B2 (en) 2013-03-15 2017-11-28 Nvidia Corporation CPU-to-GPU and GPU-to-GPU atomics
US9639474B2 (en) 2013-03-15 2017-05-02 Nvidia Corporation Migration of peer-mapped memory pages
WO2017086987A1 (en) * 2015-11-20 2017-05-26 Hewlett Packard Enterprise Development Lp In-memory data shuffling
US10180916B2 (en) * 2015-12-03 2019-01-15 Nvidia Corporation Managing copy operations in complex processor topologies
US10365843B2 (en) * 2017-04-10 2019-07-30 Intel Corporation Power management of memory chips based on working set size
US10241921B2 (en) * 2017-04-17 2019-03-26 Intel Corporation Avoid cache lookup for cold cache
US10346166B2 (en) * 2017-04-28 2019-07-09 Intel Corporation Intelligent thread dispatch and vectorization of atomic operations
US10719238B1 (en) * 2017-10-12 2020-07-21 EMC IP Holding Company LLC Memory fabric with reliability zone comprising two or more fabric attached memory endpoints
US10705951B2 (en) * 2018-01-31 2020-07-07 Hewlett Packard Enterprise Development Lp Shared fabric attached memory allocator
US11194722B2 (en) * 2018-03-15 2021-12-07 Intel Corporation Apparatus and method for improved cache utilization and efficiency on a many core processor
US10769076B2 (en) 2018-11-21 2020-09-08 Nvidia Corporation Distributed address translation in a multi-node interconnect fabric
US11579925B2 (en) * 2019-09-05 2023-02-14 Nvidia Corporation Techniques for reconfiguring partitions in a parallel processing system
US11182309B2 (en) * 2019-11-04 2021-11-23 Nvidia Corporation Techniques for an efficient fabric attached memory

Also Published As

Publication number Publication date
CN112785485A (zh) 2021-05-11
US20220043759A1 (en) 2022-02-10
US20210133123A1 (en) 2021-05-06
CN112785485B (zh) 2023-11-07
US11822491B2 (en) 2023-11-21
US11182309B2 (en) 2021-11-23

Similar Documents

Publication Publication Date Title
DE102020127705A1 (de) Techniken für einen effizienten fabric-attached-speicher
DE102019133028A1 (de) Für neuronale netzwerke geeignetes effizientes matrixformat
DE102018132468A1 (de) Multi-gpu-frame-rendern
DE102012213631B4 (de) Zwischenspeichern von Kontextdatenstrukturen in einer Vektorregisterdatei zum Beibehalten von Zustandsdaten in einer Multithread-Bildverarbeitungs-Pipeline
DE102020124932A1 (de) Vorrichtung und Verfahren zur Echtzeit-Grafikverarbeitung mittels lokaler und cloudbasierter Grafikverarbeitungsbetriebsmittel
DE112020003066T5 (de) Transponierungsoperationen mit verarbeitungselementarray
DE102009047200A1 (de) Ein Komprimierungs-Zustandsbit-Zwischenspeicher und Zusatzspeicher
DE102013017509A1 (de) Effiziente Speichervirtualisierung in mehrsträngigen Verarbeitungseinheiten
DE102013020968A1 (de) Technik zum Zugreifen auf einen inhaltsadressierbaren Speicher
DE102013205886A1 (de) Dynamische Bankmodus-Adressierung für Speicherzugriff
DE102013017511A1 (de) Effiziente speichervirtualisierung in mehrsträngigen verarbeitungseinheiten
DE112010003750T5 (de) Hardware für parallele Befehlslistenerzeugung
DE202015009260U1 (de) Effiziente Datenlesungen von verteilten Speichersystemen
DE102013202173A1 (de) Einheitliche Lade-Verarbeitung für Teilsätze von parallelen Threads
DE112020000865T5 (de) Speicherverwaltungssystem
DE102020107080A1 (de) Grafiksysteme und Verfahren zum Beschleunigen von Synchronisation mittels feinkörniger Abhängigkeitsprüfung und Planungsoptimierungen basierend auf verfügbarem gemeinsam genutztem Speicherplatz
DE102020118860A1 (de) Techniken zum vorladen von texturen beim rendering von graphik
DE102013208421A1 (de) Sharing einer Grafikverarbeitungseinheit unter vielen Anwendungen
DE102020132377A1 (de) Vorrichtung und Verfahren zur Drosselung einer Raytracing-Pipeline
DE102013020485A1 (de) Technik zur Ausführung von Speicherzugriffsoperationen über eine Textur-Hardware
DE102013020967B4 (de) Technik zur Ausführung von Speicherzugriffsoperationen über eine Textur-Hardware
DE102010044529A1 (de) Autonome Subsystem-Architektur
DE102023105565A1 (de) VERFAHREN UND VORRICHTUNG FÜR EFFIZIENTEN ZUGRIFF AUF MEHRDIMENSIONALE DATENSTRUKTUREN UND/ODER ANDERE GROßE DATENBLÖCKE
DE102020108526A1 (de) Adaptive pixelabtastreihenfolge für zeitlich dichtes rendern
DE102020103521A1 (de) Minimieren der Nutzung von Hardware-Zählern bei getriggerten Operationen für kollektive Kommunikation

Legal Events

Date Code Title Description
R012 Request for examination validly filed