nvidia cuda karten. CUDA We Roll: NVIDIA CUDA-Technologie. Videokonvertierung, Audiofilterung

Gehen wir zurück in die Geschichte – gehen Sie zurück ins Jahr 2003, als Intel und AMD in einem gemeinsamen Rennen um den leistungsstärksten Prozessor waren. In nur wenigen Jahren sind die Taktraten durch dieses Rennen deutlich gestiegen, insbesondere nach der Veröffentlichung des Intel Pentium 4.

Doch das Rennen näherte sich schnell dem Limit. Nach einer Welle gewaltiger Taktsteigerungen (zwischen 2001 und 2003 verdoppelte sich der Pentium-4-Takt von 1,5 auf 3 GHz) mussten sich Anwender mit Zehntel-Gigahertz begnügen, die die Hersteller herausquetschen konnten (von 2003 bis 2005 , Taktraten von nur 3 auf 3,8 GHz erhöht).

Auch auf hohe Taktraten optimierte Architekturen wie Prescott gerieten in Schwierigkeiten, diesmal nicht nur in der Produktion. Chiphersteller sind einfach auf die Gesetze der Physik gestoßen. Einige Analysten sagten sogar voraus, dass das Mooresche Gesetz nicht mehr gelten würde. Aber dazu kam es nicht. Die ursprüngliche Bedeutung des Gesetzes wird oft falsch dargestellt, bezieht sich aber auf die Anzahl der Transistoren auf der Oberfläche eines Siliziumkerns. Lange Zeit ging eine Erhöhung der Transistoranzahl in der CPU mit einer entsprechenden Leistungssteigerung einher – was zu einer Sinnverfälschung führte. Aber dann wurde die Situation komplizierter. Die Konstrukteure der CPU-Architektur näherten sich dem Gesetz der Verstärkungsreduzierung: Die Anzahl der Transistoren, die für die gewünschte Leistungssteigerung hinzugefügt werden mussten, wurde immer größer, was in eine Sackgasse führte.

Während CPU-Hersteller sich die Haare raufen, um eine Lösung für ihre Probleme zu finden, haben GPU-Hersteller weiterhin bemerkenswert von den Vorteilen des Mooreschen Gesetzes profitiert.

Warum landeten sie nicht in derselben Sackgasse wie die Designer der CPU-Architektur? Der Grund ist sehr einfach: CPUs sind so konzipiert, dass sie die beste Leistung bei einem Strom von Anweisungen erzielen, die unterschiedliche Daten (sowohl Ganzzahlen als auch Gleitkommazahlen) verarbeiten, wahlfreien Speicherzugriff ausführen und so weiter. Bisher haben Entwickler versucht, eine größere Befehlsparallelität bereitzustellen – also so viele Befehle wie möglich parallel auszuführen. So trat beispielsweise beim Pentium die superskalare Ausführung auf, als es unter bestimmten Bedingungen möglich war, zwei Befehle pro Takt auszuführen. Der Pentium Pro erhielt eine Out-of-Order-Ausführung von Anweisungen, die es ermöglichten, die Leistung von Recheneinheiten zu optimieren. Das Problem besteht darin, dass die parallele Ausführung eines sequentiellen Befehlsstroms offensichtliche Einschränkungen hat, sodass eine blinde Erhöhung der Anzahl von Recheneinheiten keinen Gewinn bringt, da sie die meiste Zeit noch im Leerlauf sind.

Im Gegenteil, die Arbeit der GPU ist relativ einfach. Es besteht darin, auf der einen Seite eine Gruppe von Polygonen zu nehmen und auf der anderen Seite eine Gruppe von Pixeln zu erzeugen. Polygone und Pixel sind voneinander unabhängig, sodass sie parallel verarbeitet werden können. So ist es in der GPU möglich, einen großen Teil des Kristalls für Recheneinheiten zu reservieren, die im Gegensatz zur CPU tatsächlich verwendet werden.

Zum Vergrößern auf das Bild klicken.

Die GPU unterscheidet sich nicht nur darin von der CPU. Der Speicherzugriff in der GPU ist sehr gekoppelt – wird ein Texel gelesen, dann wird nach einigen Zyklen das benachbarte Texel gelesen; Wenn ein Pixel geschrieben wird, wird das benachbarte nach einigen Zyklen geschrieben. Durch intelligentes Organisieren des Speichers erreichen Sie eine Leistung, die nahe an der theoretischen Bandbreite liegt. Dies bedeutet, dass die GPU im Gegensatz zur CPU keinen großen Cache benötigt, da ihre Aufgabe darin besteht, Texturierungsvorgänge zu beschleunigen. Alles, was man braucht, sind ein paar Kilobyte, die ein paar Texel enthalten, die in bilinearen und trilinearen Filtern verwendet werden.

Zum Vergrößern auf das Bild klicken.

Es lebe GeForce FX!

Die beiden Welten blieben lange getrennt. Wir haben die CPU (oder sogar mehrere CPUs) für Büroaufgaben und Internetanwendungen verwendet, und die GPU war nur zum Beschleunigen des Renderings gut geeignet. Aber ein Feature hat alles verändert: nämlich das Aufkommen programmierbarer GPUs. Zunächst hatten CPUs nichts zu befürchten. Die ersten sogenannten programmierbaren GPUs (NV20 und R200) ​​waren kaum eine Bedrohung. Die Anzahl der Befehle im Programm blieb auf etwa 10 beschränkt, sie bearbeiteten sehr exotische Datentypen, wie etwa 9- oder 12-Bit-Festkommazahlen.

Zum Vergrößern auf das Bild klicken.

Aber Moores Gesetz zeigte sich wieder von seiner besten Seite. Die Erhöhung der Anzahl von Transistoren erhöhte nicht nur die Anzahl der Recheneinheiten, sondern verbesserte auch deren Flexibilität. Das Erscheinen des NV30 kann aus mehreren Gründen als bedeutender Fortschritt angesehen werden. Natürlich mochten Gamer die NV30-Karten nicht wirklich, aber die neuen GPUs begannen, sich auf zwei Funktionen zu verlassen, die die Wahrnehmung von GPUs als mehr als nur Grafikbeschleuniger verändern sollten.

• Unterstützung für Gleitkommaberechnungen mit einfacher Genauigkeit (auch wenn sie nicht dem IEEE754-Standard entsprechen);
• Unterstützung für mehr als tausend Anweisungen.

Wir haben also alle Bedingungen, die wegweisende Forscher anziehen können, die immer auf der Suche nach zusätzlicher Rechenleistung sind.

Die Idee, Grafikbeschleuniger für mathematische Berechnungen zu verwenden, ist nicht neu. Die ersten Versuche wurden in den 90er Jahren des letzten Jahrhunderts unternommen. Natürlich waren sie sehr primitiv - größtenteils beschränkt auf die Verwendung einiger hardwarebasierter Funktionen wie Rasterung und Z-Puffer, um Aufgaben wie die Routensuche oder -ausgabe zu beschleunigen Voronoi-Diagramme .

Zum Vergrößern auf das Bild klicken.

Im Jahr 2003, mit dem Aufkommen von weiterentwickelten Shadern, wurde eine neue Leiste erreicht – diesmal bei der Durchführung von Matrixberechnungen. In diesem Jahr widmete sich eine ganze Rubrik der SIGGRAPH („Computations on GPUs“) dem neuen Bereich IT. Diese frühe Initiative hieß GPGPU (General-Purpose computing on GPU). Und die Entstehung von.

Um die Rolle von BrookGPU zu verstehen, müssen Sie verstehen, wie alles vor seinem Erscheinen passiert ist. Die einzige Möglichkeit, GPU-Ressourcen im Jahr 2003 zu erhalten, bestand darin, eine von zwei Grafik-APIs zu verwenden - Direct3D oder OpenGL. Folglich mussten sich Entwickler, die die Leistung der GPU für ihr Computing nutzen wollten, auf die beiden genannten APIs verlassen. Das Problem ist, dass sie nicht immer Experten in der Grafikkartenprogrammierung waren, was den Zugang zur Technologie sehr erschwerte. Arbeiten 3D-Programmierer mit Shadern, Texturen und Fragmenten, dann setzen Spezialisten im Bereich der parallelen Programmierung auf Threads, Cores, Scatter etc. Daher war es zunächst notwendig, Analogien zwischen den beiden Welten zu ziehen.

• Strom ist ein Strom von Elementen des gleichen Typs, in der GPU kann es durch eine Textur dargestellt werden. Im Prinzip gibt es in der klassischen Programmierung ein solches Analogon wie ein Array.
• Kernel- eine Funktion, die unabhängig auf jedes Element des Streams angewendet wird; ist das Äquivalent eines Pixel-Shaders. In der klassischen Programmierung können Sie eine Analogie für einen Zyklus angeben - er wird auf eine große Anzahl von Elementen angewendet.
• Um die Ergebnisse der Anwendung eines Kernels auf einen Stream zu lesen, muss eine Textur erstellt werden. Es gibt kein Äquivalent auf der CPU, da es einen vollen Speicherzugriff gibt.
• Die Stelle im Speicher, an die geschrieben werden soll (bei Scatter/Scatter-Operationen), wird durch den Vertex-Shader gesteuert, da der Pixel-Shader die Koordinaten des verarbeiteten Pixels nicht ändern kann.

Wie Sie sehen können, sieht die Aufgabe selbst unter Berücksichtigung der obigen Analogien nicht einfach aus. Und Brook kam zur Rettung. Dieser Name bezieht sich auf Erweiterungen der Sprache C ("C with streams", "C with streams"), wie die Entwickler in Stanford sie nannten. Im Kern bestand Brooks Aufgabe darin, alle Komponenten der 3D-API vor dem Programmierer zu verbergen, wodurch es möglich wurde, die GPU als Coprozessor für paralleles Rechnen darzustellen. Dazu verarbeitete der Brook-Compiler eine .br-Datei mit C++-Code und Erweiterungen und generierte dann C++-Code, der mit einer Bibliothek mit Unterstützung für verschiedene Ausgaben (DirectX, OpenGL ARB, OpenGL NV3x, x86) verknüpft wurde.

Zum Vergrößern auf das Bild klicken.

Brook hat mehrere Credits, von denen die erste darin besteht, GPGPU aus dem Schatten zu holen, damit die Technologie von der breiten Öffentlichkeit gesehen werden kann. Obwohl nach der Ankündigung des Projekts eine Reihe von IT-Sites zu optimistisch waren, dass die Veröffentlichung von Brook Zweifel an der Existenz von CPUs aufkommen lässt, die bald durch leistungsstärkere GPUs ersetzt werden. Aber wie wir sehen, ist dies auch nach fünf Jahren nicht geschehen. Um ehrlich zu sein, glauben wir nicht, dass es jemals passieren wird. Betrachtet man dagegen die erfolgreiche Entwicklung von CPUs, die sich zunehmend auf Parallelität ausrichten (mehr Kerne, SMT-Multithreading-Technologie, Erweiterung von SIMD-Blöcken), sowie GPUs, die im Gegenteil universeller werden (Unterstützung für Fließkomma-Berechnungen), Single Precision, Integer-Berechnungen, Unterstützung für Double-Precision-Berechnungen), sieht es so aus, als würden GPU und CPU bald einfach verschmelzen. Was wird dann passieren? Werden GPUs von CPUs geschluckt, wie es bei mathematischen Koprozessoren der Fall war? Gut möglich. Intel und AMD arbeiten heute an ähnlichen Projekten. Aber es kann sich noch viel ändern.

Aber zurück zu unserem Thema. Brooks Vorteil bestand darin, das Konzept von GPGPU bekannt zu machen, es vereinfachte den Zugriff auf GPU-Ressourcen erheblich, was es immer mehr Benutzern ermöglichte, das neue Programmiermodell zu beherrschen. Andererseits war es trotz aller Qualitäten von Brook noch ein langer Weg, bis GPU-Ressourcen für die Berechnung genutzt werden konnten.

Eines der Probleme hängt mit unterschiedlichen Abstraktionsebenen zusammen, insbesondere aber auch mit der zu starken Mehrbelastung durch die 3D-API, die sich durchaus bemerkbar machen kann. Als schwerwiegender kann jedoch ein Kompatibilitätsproblem angesehen werden, mit dem die Brook-Entwickler nichts anfangen konnten. Zwischen den GPU-Herstellern herrscht ein harter Wettbewerb, daher optimieren sie häufig ihre Treiber. Wenn solche Optimierungen größtenteils gut für Gamer sind, könnten sie die Brook-Kompatibilität im Handumdrehen beseitigen. Daher ist es schwierig, sich die Verwendung dieser API in industriellem Code vorzustellen, der irgendwo funktioniert. Und für lange Zeit blieb Brook der Haufen von Hobbyforschern und Programmierern.

Der Erfolg von Brook reichte jedoch aus, um die Aufmerksamkeit von ATI und Nvidia auf sich zu ziehen, und sie waren an einer solchen Initiative interessiert, da sie den Markt erweitern und einen neuen wichtigen Sektor für Unternehmen erschließen könnte.

Forscher, die ursprünglich am Brook-Projekt beteiligt waren, schlossen sich schnell den Entwicklungsteams in Santa Clara an, um eine globale Strategie zur Erschließung eines neuen Marktes vorzustellen. Die Idee war, eine Kombination aus Hard- und Software zu schaffen, die für GPGPU-Aufgaben geeignet ist. Da die Entwickler von nVidia alle Geheimnisse ihrer GPUs kennen, war es möglich, sich nicht auf die Grafik-API zu verlassen, sondern über den Treiber mit dem Grafikprozessor zu kommunizieren. Obwohl dies natürlich seine eigenen Probleme mit sich bringt. Daher hat das CUDA-Entwicklungsteam (Compute Unified Device Architecture) eine Reihe von Softwareschichten für die Arbeit mit der GPU erstellt.

Zum Vergrößern auf das Bild klicken.

Wie Sie im Diagramm sehen können, bietet CUDA zwei APIs.

• High-Level-API: CUDA Runtime API;
• Low-Level-API: CUDA-Treiber-API.

Da die High-Level-API über der Low-Level-API implementiert ist, wird jeder Laufzeitfunktionsaufruf in einfachere Anweisungen zerlegt, die von der Treiber-API verarbeitet werden. Beachten Sie, dass sich die beiden APIs gegenseitig ausschließen: Ein Programmierer kann die eine oder andere API verwenden, aber es ist nicht möglich, die Funktionsaufrufe der beiden APIs zu mischen. Im Allgemeinen ist der Begriff „High-Level-API“ relativ. Sogar die Laufzeit-API ist so, dass viele sie als Low-Level betrachten werden; Es bietet jedoch immer noch Funktionen, die für die Initialisierung oder das Kontextmanagement sehr praktisch sind. Erwarten Sie jedoch kein besonders hohes Abstraktionsniveau - Sie müssen immer noch über eine gute Menge Wissen über nVidia-GPUs und ihre Funktionsweise verfügen.

Die Treiber-API ist noch schwieriger zu handhaben; Sie benötigen mehr Aufwand, um die GPU-Verarbeitung auszuführen. Andererseits ist die Low-Level-API flexibler und gibt dem Programmierer bei Bedarf mehr Kontrolle. Zwei APIs können mit OpenGL- oder Direct3D-Ressourcen arbeiten (bisher nur die neunte Version). Der Vorteil dieser Funktion liegt auf der Hand – CUDA kann verwendet werden, um Ressourcen (Geometrie, prozedurale Texturen usw.) zu erstellen, die an die Grafik-API übergeben werden können, oder umgekehrt kann die 3D-API veranlasst werden, die Rendering-Ergebnisse an CUDA zu senden Programm, das wiederum die Nachbearbeitung durchführt. Es gibt viele Beispiele für solche Interaktionen, und der Vorteil besteht darin, dass Ressourcen weiterhin im GPU-Speicher gespeichert werden und nicht über den PCI-Express-Bus übertragen werden müssen, der immer noch ein Engpass ist.

Es sollte jedoch beachtet werden, dass die gemeinsame Nutzung von Ressourcen im Videospeicher nicht immer ideal ist und zu einigen "Kopfschmerzen" führen kann. Beim Ändern von Auflösung oder Farbtiefe haben beispielsweise Grafikdaten Vorrang. Wenn es daher erforderlich ist, die Ressourcen im Bildpuffer zu erhöhen, wird dies der Treiber leicht auf Kosten der Ressourcen der CUDA-Anwendungen tun, die einfach mit einem Fehler abstürzen. Natürlich nicht sehr elegant, aber diese Situation sollte nicht sehr oft vorkommen. Und da wir gerade von den Nachteilen gesprochen haben: Wenn Sie mehrere GPUs für CUDA-Anwendungen verwenden möchten, müssen Sie zuerst den SLI-Modus deaktivieren, da CUDA-Anwendungen sonst nur eine GPU "sehen" können.

Schließlich wird Bibliotheken die dritte Softwareebene gegeben - zwei, um genau zu sein.

• CUBLAS, das die notwendigen Blöcke für die Berechnung der linearen Algebra auf der GPU enthält;
• CUFFT, das die Berechnung von Fourier-Transformationen unterstützt, ist ein im Bereich der Signalverarbeitung weit verbreiteter Algorithmus.

Bevor wir in CUDA eintauchen, wollen wir einige Begriffe definieren, die in der nVidia-Dokumentation verstreut sind. Das Unternehmen hat eine sehr spezifische Terminologie gewählt, an die man sich nur schwer gewöhnen kann. Das stellen wir zunächst einmal fest Gewinde in CUDA hat bei weitem nicht die gleiche Bedeutung wie ein CPU-Thread, noch ist es das Äquivalent eines Threads in unseren GPU-Artikeln. Der GPU-Thread ist in diesem Fall der zugrunde liegende Datensatz, der verarbeitet werden muss. Im Gegensatz zu CPU-Threads sind CUDA-Threads sehr "leicht", was bedeutet, dass ein Kontextwechsel zwischen zwei Threads keineswegs eine ressourcenintensive Operation ist.

Der zweite Begriff, der häufig in der CUDA-Dokumentation vorkommt, ist Kette. Hier gibt es keine Verwirrung, da es kein Analogon auf Russisch gibt (es sei denn, Sie sind ein Fan von Start Trek- oder Warhammer-Spielen). Tatsächlich stammt der Begriff aus der Textilindustrie, wo Schussfaden durch den Kettfaden gezogen wird, der auf dem Webstuhl gespannt wird. Warp in CUDA ist eine Gruppe von 32 Threads und ist die minimale Datenmenge, die auf SIMD-Weise in CUDA-Multiprozessoren verarbeitet wird.

Aber eine solche "Körnigkeit" ist für den Programmierer nicht immer bequem. Daher können Sie in CUDA, anstatt direkt mit Warps zu arbeiten, mit arbeiten Block, mit 64 bis 512 Threads.

Schließlich werden diese Blöcke zusammengeführt Gitter. Der Vorteil dieser Gruppierung besteht darin, dass die Anzahl der gleichzeitig von der GPU verarbeiteten Blöcke eng mit den Hardwareressourcen zusammenhängt, wie wir weiter unten sehen werden. Durch das Gruppieren von Blöcken in Grids können Sie vollständig von dieser Einschränkung abstrahieren und den Kernel / Kernel in einem Aufruf auf mehrere Threads anwenden, ohne an feste Ressourcen denken zu müssen. Dafür sind die CUDA-Bibliotheken zuständig. Außerdem lässt sich dieses Modell gut skalieren. Wenn die GPU wenig Ressourcen hat, führt sie Blöcke sequentiell aus. Wenn die Anzahl der Rechenprozessoren groß ist, können Blöcke parallel ausgeführt werden. Das heißt, derselbe Code kann sowohl auf Einsteiger-GPUs als auch auf Top-End- und sogar zukünftigen Modellen ausgeführt werden.

Es gibt ein paar weitere Begriffe in der CUDA-API, die für CPU stehen ( Gastgeber/Gastgeber) und GPU ( Gerät/Gerät). Wenn Sie diese kleine Einführung nicht abgeschreckt hat, dann ist es an der Zeit, CUDA besser kennenzulernen.

Wenn Sie regelmäßig Tom's Hardware Guide lesen, dann ist Ihnen die Architektur der neusten GPUs von nVidia bekannt, falls nicht, empfehlen wir Ihnen die Lektüre des Artikels " nVidia GeForce GTX 260 und 280: eine neue Generation von Grafikkarten". In Bezug auf CUDA stellt nVidia die Architektur etwas anders dar und zeigt einige Details, die zuvor verborgen waren.

Wie Sie der obigen Abbildung entnehmen können, besteht der nVidia-Shader-Kern aus mehreren Clustern von Texturprozessoren. (Texturprozessor-Cluster, TPC). Die 8800 GTX verwendete beispielsweise acht Cluster, die 8800 GTS sechs und so weiter. Jeder Cluster besteht im Wesentlichen aus einer Textureinheit und zwei Streaming-Multiprozessoren. Letztere umfassen den Beginn der Pipeline (Frontend), der Anweisungen liest und dekodiert sowie sie zur Ausführung sendet, und das Ende der Pipeline (Backend), bestehend aus acht Computergeräten und zwei superfunktionalen Geräten. SFU (Super Function Unit), wobei Anweisungen nach dem SIMD-Prinzip ausgeführt werden, dh eine Anweisung wird auf alle Threads im Warp angewendet. nVidia nennt diese Vorgehensweise SIMT(einzelne Anweisung mehrere Threads, eine Anweisung, viele Threads). Es ist wichtig zu beachten, dass das Ende der Pipeline mit der doppelten Frequenz ihres Anfangs arbeitet. In der Praxis bedeutet dies, dass das Teil doppelt so "breiter" aussieht, als es tatsächlich ist (also wie ein 16-Kanal-SIMD-Block anstelle eines achtkanaligen). Streaming-Multiprozessoren funktionieren folgendermaßen: In jedem Zyklus wählt der Beginn der Pipeline einen zur Ausführung bereiten Warp aus und beginnt mit der Ausführung einer Anweisung. Damit eine Anweisung auf alle 32 Threads im Warp angewendet wird, würde das Ende der Pipeline vier Taktzyklen dauern, aber da sie mit der doppelten Frequenz des Starts läuft, würde es nur zwei Taktzyklen dauern (in Bezug auf den Anfang von die Rohrleitung). Damit der Anfang der Pipeline nicht für einen Zyklus im Leerlauf steht und die Hardware maximal ausgelastet ist, können Sie daher im Idealfall die Anweisungen in jedem Zyklus abwechseln – eine klassische Anweisung in einem Zyklus und eine Anweisung für SFU – in einem anderen .

Jeder Multiprozessor verfügt über einen bestimmten Satz von Ressourcen, die es wert sind, verstanden zu werden. Es gibt einen kleinen Speicherbereich namens "Geteilte Erinnerung", 16 KB pro Multiprozessor. Dies ist keineswegs ein Cache-Speicher: Der Programmierer kann ihn nach eigenem Ermessen verwenden. Das heißt, wir haben etwas in der Nähe des Local Store der SPU auf Cell-Prozessoren. Dieses Detail ist sehr interessant, da es betont, dass CUDA eine Kombination aus Software- und Hardwaretechnologien ist. Dieser Speicherbereich wird nicht für Pixel-Shader genutzt, wie Nvidia schlau darauf hinweist „wir mögen es nicht, wenn Pixel miteinander reden“.

Dieser Speicherbereich eröffnet die Möglichkeit, Informationen zwischen Threads auszutauschen. in einem Block. Es ist wichtig, diese Einschränkung hervorzuheben: Alle Threads in einem Block werden garantiert von einem einzigen Multiprozessor ausgeführt. Im Gegensatz dazu ist die Bindung von Blöcken an verschiedene Multiprozessoren überhaupt nicht spezifiziert, und zwei Threads aus verschiedenen Blöcken können während der Ausführung keine Informationen miteinander austauschen. Das heißt, die Verwendung von Shared Memory ist nicht so einfach. Shared Memory ist jedoch immer noch gerechtfertigt, außer wenn mehrere Threads versuchen, auf dieselbe Speicherbank zuzugreifen, was zu einem Konflikt führt. In anderen Situationen ist der Zugriff auf den gemeinsam genutzten Speicher so schnell wie der Registerzugriff.

Der gemeinsam genutzte Speicher ist nicht der einzige, auf den Multiprozessoren zugreifen können. Sie können Videospeicher verwenden, jedoch mit geringerer Bandbreite und höherer Latenz. Um die Zugriffshäufigkeit auf diesen Speicher zu reduzieren, hat nVidia daher Multiprozessoren mit einem Cache (ca. 8 KB pro Multiprozessor) ausgestattet, der Konstanten und Texturen speichert.

Der Multiprozessor hat 8.192 Register, die allen Threads aller auf dem Multiprozessor aktiven Blöcke gemeinsam sind. Die Anzahl aktiver Blöcke pro Multiprozessor darf acht nicht überschreiten, und die Anzahl aktiver Warps ist auf 24 (768 Threads) begrenzt. Daher kann die 8800 GTX bis zu 12.288 Threads gleichzeitig verarbeiten. Alle diese Einschränkungen sind erwähnenswert, da sie es ermöglichen, den Algorithmus basierend auf den verfügbaren Ressourcen zu optimieren.

Die Optimierung des CUDA-Programms besteht daher darin, ein optimales Gleichgewicht zwischen der Anzahl der Blöcke und ihrer Größe zu erhalten. Mehr Threads pro Block wären beim Reduzieren der Speicherlatenz hilfreich, aber die Anzahl der pro Thread verfügbaren Register würde ebenfalls reduziert werden. Außerdem wäre ein Block von 512 Threads ineffizient, da nur ein Block auf einem Multiprozessor aktiv sein könnte, was zu einem Verlust von 256 Threads führen würde. Daher empfiehlt nVidia die Verwendung von Blöcken mit 128 oder 256 Threads, was für die meisten Kerne / Kernel den besten Kompromiss zwischen geringerer Latenz und der Anzahl der Register darstellt.

Aus programmtechnischer Sicht besteht CUDA aus einer Reihe von Erweiterungen der C-Sprache, die BrookGPU ähnelt, sowie mehreren spezifischen API-Aufrufen. Zu den Erweiterungen gehören Typbezeichner, die sich auf Funktionen und Variablen beziehen. Es ist wichtig, sich das Schlüsselwort zu merken __global__, was, wenn es vor der Funktion angegeben wird, darauf hinweist, dass sich letzteres auf den Kernel / Kernel bezieht - diese Funktion wird von der CPU aufgerufen und auf der GPU ausgeführt. Präfix __Gerät__ gibt an, dass die Funktion auf der GPU ausgeführt wird (die CUDA übrigens "device/device" nennt), aber sie kann nur von der GPU aufgerufen werden (mit anderen Worten, von einer anderen __device__-Funktion oder von einer __global__-Funktion). Schließlich das Präfix __Gastgeber__ optional bezeichnet es eine Funktion, die von der CPU aufgerufen und von der CPU ausgeführt wird – mit anderen Worten, eine reguläre Funktion.

Mit den Funktionen __device__ und __global__ sind einige Einschränkungen verbunden: Sie können nicht rekursiv sein (d. h. sich selbst aufrufen) und sie können keine variable Anzahl von Argumenten haben. Da sich die __device__-Funktionen schließlich im GPU-Speicher befinden, ist es sinnvoll, dass ihre Adresse nicht abgerufen werden kann. Variablen haben auch eine Reihe von Kennzeichnern, die den Speicherort angeben, an dem sie gespeichert werden. Variable mit Präfix __geteilt__ bedeutet, dass es im gemeinsam genutzten Speicher des Streaming-Multiprozessors gespeichert wird. Der Aufruf der Funktion __global__ ist etwas anders. Die Sache ist, dass Sie beim Aufrufen die Ausführungskonfiguration festlegen müssen - genauer gesagt die Größe des Gitters / Gitters, auf das der Kernel / Kernel angewendet wird, sowie die Größe jedes Blocks. Nehmen Sie zum Beispiel einen Kernel mit der folgenden Signatur.

__global__ void Func(float* parameter);

Es wird als bezeichnet

Funkt<<< Dg, Db >>> (Parameter);

wobei Dg die Rastergröße und Db die Blockgröße ist. Diese beiden Variablen beziehen sich auf den neuen Vektortyp, der mit CUDA eingeführt wurde.

Die CUDA-API enthält Funktionen zum Arbeiten mit Speicher im VRAM: cudaMalloc zum Zuweisen von Speicher, cudaFree zum Freigeben und cudaMemcpy zum Kopieren von Speicher zwischen RAM und VRAM und umgekehrt.

Wir beenden diesen Bericht mit einer sehr interessanten Art und Weise, wie ein CUDA-Programm kompiliert wird: Die Kompilierung erfolgt in mehreren Schritten. Zunächst wird der CPU-spezifische Code extrahiert und an den Standard-Compiler übergeben. Code, der für die GPU bestimmt ist, wird zuerst in die PTX-Zwischensprache konvertiert. Es ähnelt der Assemblersprache und ermöglicht es Ihnen, den Code auf der Suche nach potenziell ineffizienten Abschnitten zu studieren. Schließlich besteht die letzte Phase darin, die Zwischensprache in GPU-spezifische Anweisungen zu übersetzen und die Binärdatei zu erstellen.

Beim Durchsehen der nVidia-Dokumentation bekomme ich Lust, diese Woche CUDA auszuprobieren. Was könnte in der Tat besser sein, als eine API zu evaluieren, indem Sie Ihr eigenes Programm erstellen? Dann sollten die meisten Probleme an die Oberfläche kommen, auch wenn auf dem Papier alles perfekt aussieht. Darüber hinaus zeigt die Praxis am besten, wie gut Sie alle in der CUDA-Dokumentation beschriebenen Prinzipien verstehen.

Es ist ganz einfach, in ein solches Projekt einzutauchen. Heute steht eine Vielzahl kostenloser, aber qualitativ hochwertiger Tools zum Download bereit. Für unseren Test haben wir Visual C++ Express 2005 verwendet, das alles hat, was Sie brauchen. Das Schwierigste war, ein Programm zu finden, das nicht Wochen brauchte, um es auf die GPU zu portieren, aber interessant genug war, dass unsere Bemühungen nicht umsonst waren. Am Ende haben wir uns für einen Codeabschnitt entschieden, der eine Höhenkarte nimmt und die entsprechende Normalkarte berechnet. Wir werden auf diese Funktion nicht im Detail eingehen, da sie in diesem Artikel kaum interessant ist. Kurz gesagt, das Programm befasst sich mit der Krümmung von Bereichen: Für jedes Pixel des Ausgangsbildes legen wir eine Matrix fest, die die Farbe des resultierenden Pixels im generierten Bild anhand einer mehr oder weniger komplexen Formel aus benachbarten Pixeln bestimmt. Der Vorteil dieser Funktion ist, dass sie sehr einfach zu parallelisieren ist, sodass dieser Test die Fähigkeiten von CUDA perfekt demonstriert.

Ein weiterer Vorteil ist, dass wir bereits eine Implementierung auf der CPU haben, sodass wir deren Ergebnis mit der CUDA-Version vergleichen können – und das Rad nicht neu erfinden.

Wir wiederholen noch einmal, dass der Zweck des Tests darin bestand, sich mit den CUDA SDK-Dienstprogrammen vertraut zu machen, und nicht, die Versionen für CPU und GPU zu vergleichen. Da dies unser erster Versuch war, ein CUDA-Programm zu erstellen, erwarteten wir nicht wirklich eine hohe Leistung. Da dieser Teil des Codes nicht kritisch ist, wurde die CPU-Version nicht optimiert, sodass ein direkter Vergleich der Ergebnisse kaum interessant ist.

Leistung

Wir haben jedoch die Ausführungszeit gemessen, um zu sehen, ob die Verwendung von CUDA selbst bei der rauesten Implementierung einen Vorteil hat oder ob wir lange und mühsame Übung benötigen, um bei der Verwendung der GPU einen Gewinn zu erzielen. Das Testgerät stammt aus unserem Entwicklungslabor - ein Laptop mit einem Core 2 Duo T5450 Prozessor und einer GeForce 8600M GT Grafikkarte mit Vista. Das ist noch lange kein Supercomputer, aber die Ergebnisse sind sehr interessant, da der Test nicht für die GPU "geschärft" wird. Es ist immer schön zu sehen, dass nVidia enorme Gewinne auf Systemen mit monströsen GPUs und viel Bandbreite zeigt, aber in der Praxis sind viele der 70 Millionen CUDA-fähigen GPUs auf dem heutigen PC-Markt bei weitem nicht so leistungsfähig, daher ist unser Test gerechtfertigt.

Für ein Bild mit 2048 x 2048 Pixeln haben wir die folgenden Ergebnisse erhalten.

• CPU 1-Thread: 1419 ms;
• CPU 2-Threads: 749 ms;
• CPU 4-Threads: 593 ms
• GPU (8600M GT) Blöcke mit 256 Threads: 109 ms;
• GPU (8600M GT) Blöcke mit 128 Threads: 94 ms;
• GPU (8800 GTX) Blöcke mit 128 Threads / 256 Threads: 31 ms.

Aus den Ergebnissen lassen sich mehrere Schlussfolgerungen ziehen. Beginnen wir mit der Tatsache, dass wir trotz aller Gerüchte über die offensichtliche Faulheit von Programmierern die ursprüngliche Version der CPU für mehrere Threads modifiziert haben. Wie bereits erwähnt, ist der Code für diese Situation ideal – es muss lediglich das ursprüngliche Bild in so viele Zonen aufgeteilt werden, wie es Streams gibt. Beachten Sie, dass die Beschleunigung beim Umschalten von einem Thread auf zwei Threads auf unserer Dual-Core-CPU nahezu linear ist, was auch auf die Parallelität des Testprogramms hinweist. Völlig unerwartet war auch die Version mit vier Threads schneller, was bei unserem Prozessor allerdings sehr merkwürdig ist – im Gegenteil, man hätte mit Effizienzeinbußen durch den Overhead der Verwaltung zusätzlicher Threads rechnen müssen. Wie ist ein solches Ergebnis zu erklären? Es ist schwer zu sagen, aber der Thread-Scheduler von Windows ist wahrscheinlich der Übeltäter; in jedem Fall ist das Ergebnis wiederholbar. Bei kleineren Texturen (512 x 512) war der Gewinn durch das Aufteilen nicht so ausgeprägt (ca. 35 % gegenüber 100 %), und das Verhalten der Vier-Thread-Version war logischer, ohne Gewinn im Vergleich zur Zwei-Thread-Version. Die GPU war immer noch schneller, aber nicht so ausgeprägt (die 8600M GT war dreimal schneller als die Dual-Thread-Version).

Zum Vergrößern auf das Bild klicken.

Die zweite wichtige Beobachtung ist, dass selbst die langsamste Implementierung der GPU fast sechsmal schneller war als die leistungsstärkste Version der CPU. Für das erste Programm und die nicht optimierte Version des Algorithmus ist das Ergebnis sehr ermutigend. Bitte beachten Sie, dass wir bei kleinen Blöcken ein merklich besseres Ergebnis erzielt haben, obwohl die Intuition etwas anderes vermuten lässt. Die Erklärung ist einfach – unser Programm verwendet 14 Register pro Thread, und bei 256-Thread-Blöcken sind 3.584 Register pro Block erforderlich, und 768 Threads sind für die volle Prozessorlast erforderlich, wie wir gezeigt haben. In unserem Fall sind dies drei Blöcke oder 10.572 Register. Aber der Multiprozessor hat nur 8.192 Register, kann also nur zwei Blöcke aktiv halten. Im Gegensatz dazu benötigen wir bei Blöcken von 128 Threads 1.792 Register pro Block; Wenn 8.192 durch 1.792 geteilt und auf die nächste ganze Zahl aufgerundet wird, erhalten wir vier Blöcke. In der Praxis bleibt die Anzahl der Threads gleich (512 pro Multiprozessor, obwohl 768 theoretisch für eine vollständige Auslastung benötigt werden), aber die Erhöhung der Anzahl der Blöcke gibt der GPU den Vorteil der Flexibilität beim Speicherzugriff – wenn eine Operation ausgeführt wird Bei großen Verzögerungen kann er mit der Ausführung der Anweisungen eines anderen Blocks beginnen und auf den Erhalt der Ergebnisse warten. Vier Blöcke reduzieren die Latenz deutlich, zumal unser Programm mehrere Speicherzugriffe verwendet.

Analyse

Trotz allem, was wir oben gesagt haben, konnten wir schließlich der Versuchung nicht widerstehen und ließen das Programm auf der 8800 GTX laufen, die dreimal schneller war als die 8600, unabhängig von der Blockgröße. Man könnte meinen, dass in der Praxis auf den entsprechenden Architekturen das Ergebnis viermal oder mehr höher ausfallen wird: 128 ALUs / Shader-Prozessoren gegenüber 32 und höheren Taktraten (1,35 GHz gegenüber 950 MHz), aber dies ist nicht geschehen. Der limitierende Faktor war höchstwahrscheinlich der Speicherzugriff. Genauer gesagt wird auf das ursprüngliche Bild als mehrdimensionales CUDA-Array zugegriffen – ein ziemlich komplizierter Begriff für etwas, das nichts anderes als eine Textur ist. Aber Essen ein paar Vorteile.

• Zugriffe profitieren vom Texturcache;
• Wir verwenden den Wrapping-Modus, der im Gegensatz zur CPU-Version keine Bildränder behandeln muss.

Außerdem können wir von einer "freien" Filterung mit normalisierter Adressierung zwischen statt und profitieren, aber das ist in unserem Fall kaum nützlich. Wie Sie wissen, hat der 8600 16 Textureinheiten, verglichen mit 32 beim 8800 GTX. Daher beträgt das Verhältnis zwischen den beiden Architekturen nur zwei zu eins. Fügen Sie dazu die Frequenzdifferenz hinzu und wir erhalten ein Verhältnis von (32 x 0,575) / (16 x 0,475) = 2,4 - nahe an den "drei zu eins", die wir tatsächlich erhalten haben. Diese Theorie erklärt auch, warum sich die Größe der Blöcke beim G80 nicht wesentlich ändert, da die ALU immer noch auf den Textureinheiten aufliegt.

Zum Vergrößern auf das Bild klicken.

Neben vielversprechenden Ergebnissen verlief unsere erste Begegnung mit CUDA angesichts der nicht so günstigen gewählten Bedingungen sehr gut. Die Entwicklung auf einem Vista-Laptop bedeutet die Verwendung von CUDA SDK 2.0, das sich noch in der Beta-Phase befindet, mit dem Treiber 174.55, der sich ebenfalls in der Beta-Phase befindet. Unangenehme Überraschungen können wir trotzdem nicht vermelden - lediglich anfängliche Fehler beim ersten Debuggen, als unser noch sehr "fehlerhaftes" Programm versuchte, Speicher außerhalb des zugewiesenen Speicherplatzes anzusprechen.

Der Monitor fing an, wild zu flackern, dann wurde der Bildschirm schwarz ... bis Vista den Treiber-Reparaturdienst ausführte und alles in Ordnung war. Aber es ist immer noch etwas überraschend zu sehen, ob Sie es gewohnt sind, den typischen Segmentierungsfehler bei Standardprogrammen wie unserem zu sehen. Abschließend noch eine kleine Kritik an nVidia: In allen Dokumentationen zu CUDA gibt es keine kleine Anleitung, die einem Schritt für Schritt erklärt, wie man eine Entwicklungsumgebung für Visual Studio aufsetzt. Eigentlich ist das Problem nicht groß, da das SDK einen vollständigen Satz von Beispielen enthält, die studiert werden können, um das Framework für CUDA-Anwendungen zu verstehen, aber eine Anleitung für Anfänger wäre nett.

Zum Vergrößern auf das Bild klicken.

Nvidia hat CUDA mit der Veröffentlichung der GeForce 8800 eingeführt. Und zu der Zeit schienen die Versprechungen sehr verlockend, aber wir haben unsere Begeisterung auf den echten Test beschränkt. Tatsächlich schien es damals eher so, als würde man Territorium markieren, um auf der GPGPU-Welle zu bleiben. Ohne ein verfügbares SDK ist es schwer zu sagen, dass wir nicht mit einem weiteren Marketing-Dummy konfrontiert sind, der nicht funktioniert. Dies ist nicht das erste Mal, dass eine gute Initiative zu früh angekündigt wurde und damals aufgrund mangelnder Unterstützung nicht zum Vorschein kam - insbesondere in einem so wettbewerbsintensiven Sektor. Jetzt, anderthalb Jahre nach der Ankündigung, können wir mit Sicherheit sagen, dass nVidia Wort gehalten hat.

Das SDK ging ziemlich schnell Anfang 2007 in die Beta-Phase und wurde seitdem schnell aktualisiert, was die Bedeutung dieses Projekts für nVidia beweist. Heute entwickelt sich CUDA sehr gut: Das SDK ist bereits in der Beta-Version 2.0 für die wichtigsten Betriebssysteme (Windows XP und Vista, Linux sowie 1.1 für Mac OS X) verfügbar, und nVidia hat ihm einen ganzen Bereich der Website gewidmet Entwickler.

Auf professioneller Ebene war der Eindruck der ersten Schritte mit CUDA sehr positiv. Selbst wenn Sie mit der GPU-Architektur vertraut sind, können Sie sie leicht herausfinden. Wenn eine API auf den ersten Blick klar aussieht, glauben Sie sofort, dass Sie überzeugende Ergebnisse erzielen werden. Aber wird nicht Rechenzeit durch mehrfache Übertragungen von der CPU zur GPU verschwendet? Und wie nutzt man diese Tausenden von Threads fast ohne Synchronisierungsprimitive? Wir begannen unsere Experimente mit all diesen Ängsten im Hinterkopf. Aber sie lösten sich schnell auf, als sich herausstellte, dass die erste Version unseres Algorithmus, wenn auch eine sehr triviale, deutlich schneller war als auf der CPU.

CUDA ist also kein Lebensretter für Forscher, die Universitätsbeamte davon überzeugen wollen, ihnen eine GeForce zu kaufen. CUDA ist bereits eine vollständig verfügbare Technologie, die jeder C-Programmierer verwenden kann, wenn er bereit ist, Zeit und Mühe zu investieren, um sich an das neue Programmierparadigma zu gewöhnen. Diese Bemühungen werden nicht verschwendet, wenn Ihre Algorithmen gut parallelisieren. Wir möchten uns auch bei nVidia für die Bereitstellung einer vollständigen und qualitativ hochwertigen Dokumentation bedanken, in der unerfahrene CUDA-Programmierer Antworten finden werden.

Was braucht es, damit CUDA eine erkennbare API wird? In einem Wort: Portabilität. Wir wissen, dass die Zukunft der IT im parallelen Rechnen liegt – schon heute bereiten sich alle auf solche Veränderungen vor, und alle Initiativen, sowohl Software als auch Hardware, zielen in diese Richtung. Wenn Sie sich jedoch die Entwicklung von Paradigmen ansehen, befinden wir uns im Moment noch in der Anfangsphase: Wir erstellen Threads manuell und versuchen, den Zugriff auf gemeinsam genutzte Ressourcen zu planen. All dies lässt sich irgendwie bewältigen, wenn die Anzahl der Kerne an den Fingern einer Hand gezählt werden kann. Aber in ein paar Jahren, wenn die Zahl der Prozessoren in die Hunderte geht, wird es diese Möglichkeit nicht mehr geben. Mit der Veröffentlichung von CUDA hat nVidia den ersten Schritt getan, um dieses Problem zu lösen – aber natürlich ist diese Lösung nur für GPUs dieser Firma geeignet, und selbst dann nicht für jedermann. Nur GF8 und 9 (und ihre Quadro/Tesla-Derivate) können heute CUDA-Programme ausführen. Und natürlich die neue 260/280-Linie.

Zum Vergrößern auf das Bild klicken.

Nvidia mag damit prahlen, weltweit 70 Millionen CUDA-kompatible GPUs verkauft zu haben, aber das reicht immer noch nicht, um zum De-facto-Standard zu werden. Unter Berücksichtigung der Tatsache, dass Wettbewerber nicht tatenlos zusehen. AMD bietet ein eigenes SDK (Stream Computing) an und Intel hat eine Lösung angekündigt (Ct), die allerdings noch nicht verfügbar ist. Ein Krieg um Standards steht bevor, und es wird auf dem Markt eindeutig keinen Platz für drei Konkurrenten geben, bis ein anderer Spieler wie Microsoft einen gemeinsamen API-Vorschlag vorlegt, der Entwicklern das Leben natürlich leichter machen wird.

Daher hat nVidia große Schwierigkeiten bei der CUDA-Zulassung. Obwohl wir technologisch zweifellos eine erfolgreiche Lösung vor uns haben, bleibt es, die Entwickler von ihren Perspektiven zu überzeugen - und das wird nicht einfach sein. Nach den vielen jüngsten API-Ankündigungen und Neuigkeiten zu urteilen, sieht die Zukunft jedoch überhaupt nicht düster aus.

Vorrichtungen zum Verwandeln von Personalcomputern in kleine Supercomputer sind seit langem bekannt. Bereits in den 80er Jahren des letzten Jahrhunderts wurden sogenannte Transputer auf dem Markt angeboten, die in die damals üblichen ISA-Erweiterungsslots gesteckt wurden. Ihre Leistung bei den entsprechenden Aufgaben war zunächst beeindruckend, aber dann beschleunigte sich das Leistungswachstum der Universalprozessoren, sie stärkten ihre Position im parallelen Rechnen, und Transputer machten keinen Sinn. Obwohl solche Geräte immer noch existieren, handelt es sich um eine Vielzahl spezialisierter Beschleuniger. Aber oft ist der Umfang ihrer Anwendung eng und solche Beschleuniger werden nicht weit verbreitet verwendet.

Aber in letzter Zeit hat sich der Staffelstab des parallelen Rechnens auf den Massenmarkt verlagert, auf die eine oder andere Weise verbunden mit dreidimensionalen Spielen. Allzweckgeräte mit Mehrkernprozessoren für paralleles Vektorcomputing, die in 3D-Grafik verwendet werden, erreichen eine hohe Spitzenleistung, die Allzweckprozessoren nicht erreichen können. Natürlich wird die maximale Geschwindigkeit nur bei einer Reihe bequemer Aufgaben erreicht und weist einige Einschränkungen auf, aber solche Geräte werden bereits in großem Umfang in Bereichen eingesetzt, für die sie ursprünglich nicht bestimmt waren. Ein hervorragendes Beispiel für einen solchen Parallelprozessor ist der von der Sony-Toshiba-IBM-Allianz entwickelte Cell-Prozessor, der in der Sony PlayStation 3-Spielekonsole sowie in allen modernen Grafikkarten der Marktführer Nvidia und AMD verwendet wird.

Wir werden Cell heute nicht anfassen, obwohl es früher erschien und ein universeller Prozessor mit zusätzlichen Vektorfähigkeiten ist, sprechen wir heute nicht darüber. Für 3D-Videobeschleuniger erschienen vor einigen Jahren die ersten nicht-grafischen Allzweck-Berechnungstechnologien GPGPU (General-Purpose computing on GPUs). Schließlich enthalten moderne Videochips Hunderte von mathematischen Ausführungseinheiten, und diese Leistung kann verwendet werden, um viele rechenintensive Anwendungen erheblich zu beschleunigen. Und die aktuellen GPU-Generationen verfügen über eine ausreichend flexible Architektur, die zusammen mit höheren Programmiersprachen und Hardware-Software-Architekturen wie der in diesem Artikel diskutierten diese Möglichkeiten eröffnet und sie viel zugänglicher macht.

Die Schaffung der GPCPU wurde durch das Aufkommen von ausreichend schnellen und flexiblen Shader-Programmen veranlasst, die in der Lage sind, moderne Videochips auszuführen. Die Entwickler entschieden sich dafür, dass die GPU nicht nur das Bild in 3D-Anwendungen berechnet, sondern auch für andere parallele Berechnungen verwendet wird. Die GPGPU nutzte dafür Grafik-APIs: OpenGL und Direct3D, wenn Daten in Form von Texturen auf den Videochip übertragen und Berechnungsprogramme in Form von Shadern geladen wurden. Die Nachteile dieser Methode sind die relativ hohe Komplexität der Programmierung, die geringe Geschwindigkeit des Datenaustauschs zwischen CPU und GPU und andere Einschränkungen, auf die wir später noch eingehen werden.

GPU-Computing hat sich weiterentwickelt und entwickelt sich sehr schnell. Und weiter haben zwei große Videochip-Hersteller, Nvidia und AMD, jeweilige Plattformen namens CUDA (Compute Unified Device Architecture) und CTM (Close To Metal oder AMD Stream Computing) entwickelt und angekündigt. Im Gegensatz zu früheren GPU-Programmiermodellen wurden diese mit direktem Zugriff auf die Hardwarefähigkeiten der Grafikkarten durchgeführt. Die Plattformen sind untereinander nicht kompatibel, CUDA ist eine Erweiterung der Programmiersprache C und CTM ist eine virtuelle Maschine, die Assemblercode ausführt. Aber beide Plattformen haben einige der wichtigen Einschränkungen früherer GPGPU-Modelle beseitigt, die die traditionelle Grafikpipeline und die entsprechenden Direct3D- oder OpenGL-Schnittstellen verwenden.

Offene Standards, die OpenGL verwenden, scheinen natürlich die portabelsten und universellsten zu sein, sie ermöglichen es Ihnen, denselben Code für Videochips verschiedener Hersteller zu verwenden. Aber solche Methoden haben viele Nachteile, sie sind viel weniger flexibel und nicht so bequem in der Anwendung. Darüber hinaus verhindern sie die Verwendung der spezifischen Funktionen bestimmter Grafikkarten, wie z. B. des schnellen gemeinsam genutzten (gemeinsamen) Speichers, der in modernen Rechenprozessoren vorhanden ist.

Aus diesem Grund hat Nvidia die CUDA-Plattform veröffentlicht, eine C-ähnliche Programmiersprache mit eigenem Compiler und Bibliotheken für GPU-Computing. Natürlich ist das Schreiben des optimalen Codes für Videochips gar nicht so einfach und diese Aufgabe erfordert lange manuelle Arbeit, aber CUDA offenbart einfach alle Möglichkeiten und gibt dem Programmierer mehr Kontrolle über die Hardwarefähigkeiten der GPU. Wichtig ist, dass Nvidia CUDA-Unterstützung für die G8x-, G9x- und GT2xx-Chips verfügbar ist, die in Grafikkarten der Serien Geforce 8, 9 und 200 verwendet werden, die sehr weit verbreitet sind. Jetzt wurde die finale Version von CUDA 2.0 veröffentlicht, die einige neue Funktionen enthält, wie z. B. die Unterstützung von Berechnungen mit doppelter Genauigkeit. CUDA ist auf 32-Bit- und 64-Bit-Linux-, Windows- und MacOS-X-Betriebssystemen verfügbar.

Unterschied zwischen CPU und GPU beim parallelen Rechnen

Das Wachstum der Frequenzen von Universalprozessoren stieß auf physikalische Grenzen und einen hohen Stromverbrauch, und ihre Leistung nimmt aufgrund der Platzierung mehrerer Kerne in einem einzigen Chip zunehmend zu. Jetzt verkaufte Prozessoren enthalten nur bis zu vier Kerne (weiteres Wachstum wird nicht schnell sein) und sie sind für allgemeine Anwendungen gedacht, verwenden MIMD - Multiple Instruction and Data Flow. Jeder Kern arbeitet getrennt von den anderen und führt unterschiedliche Anweisungen für unterschiedliche Prozesse aus.

Spezialisierte Vektorfähigkeiten (SSE2 und SSE3) für 4-Komponenten- (Gleitkommazahl mit einfacher Genauigkeit) und Zwei-Komponenten-Vektoren (mit doppelter Genauigkeit) sind in Allzweckprozessoren in erster Linie aufgrund der gestiegenen Anforderungen von Grafikanwendungen erschienen. Deshalb ist für bestimmte Aufgaben der Einsatz von GPUs rentabler, weil sie ursprünglich dafür gemacht wurden.

Beispielsweise ist die Haupteinheit in Nvidia-Videochips ein Multiprozessor mit insgesamt acht bis zehn Kernen und Hunderten von ALUs, mehreren tausend Registern und einer kleinen Menge an gemeinsam genutztem Speicher. Außerdem enthält die Videokarte einen schnellen globalen Speicher, auf den alle Multiprozessoren zugreifen können, einen lokalen Speicher in jedem Multiprozessor und einen speziellen Speicher für Konstanten.

Am wichtigsten ist, dass diese mehreren Multiprozessorkerne in der GPU SIMD-Kerne (Single Instruction Stream, Multiple Data Stream) sind. Und diese Kerne führen gleichzeitig dieselben Anweisungen aus. Dieser Programmierstil ist bei Grafikalgorithmen und vielen wissenschaftlichen Aufgaben üblich, erfordert jedoch eine spezifische Programmierung. Dieser Ansatz ermöglicht es Ihnen jedoch, die Anzahl der Ausführungseinheiten aufgrund ihrer Vereinfachung zu erhöhen.

Lassen Sie uns also die Hauptunterschiede zwischen CPU- und GPU-Architekturen auflisten. CPU-Kerne sind darauf ausgelegt, einen einzelnen Strom sequentieller Anweisungen mit maximaler Leistung auszuführen, während GPUs darauf ausgelegt sind, schnell eine große Anzahl paralleler Befehlsströme auszuführen. Allzweckprozessoren sind optimiert, um eine hohe Leistung mit einem einzigen Befehlsstrom zu erreichen, der sowohl Ganzzahlen als auch Gleitkommazahlen verarbeitet. Der Speicherzugriff ist zufällig.

CPU-Designer versuchen, so viele Befehle wie möglich parallel auszuführen, um die Leistung zu steigern. Zu diesem Zweck erschien ab den Intel Pentium-Prozessoren die superskalare Ausführung, die die Ausführung von zwei Anweisungen pro Takt ermöglichte, und der Pentium Pro zeichnete sich durch die Ausführung von Anweisungen außerhalb der Reihenfolge aus. Aber die parallele Ausführung eines sequentiellen Befehlsstroms hat gewisse grundlegende Beschränkungen, und durch Erhöhen der Anzahl von Ausführungseinheiten kann eine mehrfache Erhöhung der Geschwindigkeit nicht erreicht werden.

Videochips haben von Anfang an einen einfachen und parallelen Betrieb. Der Videochip nimmt eine Gruppe von Polygonen am Eingang, führt alle notwendigen Operationen durch und erzeugt Pixel am Ausgang. Die Verarbeitung von Polygonen und Pixeln ist unabhängig, sie können getrennt voneinander parallel verarbeitet werden. Daher werden aufgrund der inhärent parallelen Arbeitsorganisation in der GPU im Gegensatz zum sequentiellen Befehlsfluss für die CPU eine Vielzahl von Ausführungseinheiten verwendet, die einfach zu laden sind. Zudem können moderne GPUs auch mehr als einen Befehl pro Takt ausführen (Dual Issue). Somit startet die Tesla-Architektur unter bestimmten Bedingungen die MAD+MUL- oder MAD+SFU-Operationen gleichzeitig.

Die GPU unterscheidet sich von der CPU auch hinsichtlich der Prinzipien des Speicherzugriffs. In der GPU ist es verbunden und leicht vorhersehbar - wird ein Texturtexel aus dem Speicher gelesen, dann kommt nach einer Weile die Zeit für benachbarte Texel. Ja, und wenn Sie dasselbe aufnehmen, wird das Pixel in den Framebuffer geschrieben, und nach einigen Zyklen wird das daneben befindliche Pixel aufgezeichnet. Daher unterscheidet sich die Speicherorganisation von der in der CPU verwendeten. Und der Videochip benötigt im Gegensatz zu Universalprozessoren einfach keinen großen Cache-Speicher, und Texturen benötigen nur wenige (bis zu 128-256 in aktuellen GPUs) Kilobyte.

Und an sich ist die Arbeit mit Speicher für GPU und CPU etwas anders. So haben nicht alle CPUs eingebaute Speichercontroller, und alle GPUs haben normalerweise mehrere Controller, bis zu acht 64-Bit-Kanäle im Nvidia GT200-Chip. Außerdem verbrauchen Grafikkarten schnelleren Speicher, wodurch Videochips eine vielfach höhere Speicherbandbreite zur Verfügung haben, was auch für parallele Berechnungen, die mit riesigen Datenströmen arbeiten, sehr wichtig ist.

In Allzweckprozessoren gehen eine große Anzahl von Transistoren und Chipfläche in Befehlspuffer, Hardware-Verzweigungsvorhersage und riesige Mengen an On-Chip-Cache-Speicher. Alle diese Hardwareblöcke werden benötigt, um die Ausführung einiger Befehlsströme zu beschleunigen. Videochips geben Transistoren für Arrays von Ausführungseinheiten, Flusssteuereinheiten, kleinen gemeinsam genutzten Speichern und Mehrkanal-Speichercontrollern aus. Das Obige beschleunigt nicht die Ausführung einzelner Threads, es ermöglicht dem Chip, mehrere tausend Threads zu handhaben, die gleichzeitig auf dem Chip ausgeführt werden und eine hohe Speicherbandbreite erfordern.

Über Unterschiede beim Caching. Allzweck-CPUs verwenden Cache, um die Leistung zu steigern, indem sie die Latenz beim Speicherzugriff reduzieren, während GPUs Cache oder gemeinsam genutzten Speicher verwenden, um die Bandbreite zu erhöhen. CPUs reduzieren Speicherzugriffslatenzen mit großen Caches und Codeverzweigungsvorhersage. Diese Hardwareteile nehmen den größten Teil der Chipfläche ein und verbrauchen viel Strom. Videochips umgehen das Problem der Speicherzugriffsverzögerungen, indem sie Tausende von Threads gleichzeitig ausführen – während einer der Threads auf Daten aus dem Speicher wartet, kann der Videochip Berechnungen eines anderen Threads ohne Wartezeiten und Verzögerungen durchführen.

Auch bei der Multithreading-Unterstützung gibt es viele Unterschiede. Die CPU führt 1-2 Rechenthreads pro Prozessorkern aus, und Videochips können bis zu 1024 Threads pro Multiprozessor unterstützen, von denen es mehrere im Chip gibt. Und wenn das Wechseln von einem Thread zum anderen für die CPU Hunderte von Zyklen kostet, dann schaltet die GPU mehrere Threads in einem Zyklus um.

Darüber hinaus verwenden CPUs SIMD-Blöcke (Single Instruction, Multiple Data) für die Vektorberechnung und GPUs SIMT (Single Instruction, Multiple Threads) für die skalare Thread-Verarbeitung. SIMT verlangt vom Entwickler nicht, Daten in Vektoren umzuwandeln, und erlaubt willkürliche Verzweigungen in Streams.

Zusammenfassend kann man sagen, dass Videochips im Gegensatz zu modernen Universal-CPUs für paralleles Rechnen mit einer Vielzahl von Rechenoperationen ausgelegt sind. Und eine viel größere Anzahl von GPU-Transistoren arbeitet für ihren beabsichtigten Zweck – die Verarbeitung von Datenarrays und steuert nicht die Ausführung (Flusssteuerung) einiger weniger sequentieller Rechenthreads. Dies ist ein Diagramm, wie viel Platz in der CPU und GPU eine Vielzahl von Logiken benötigt:

Folglich ist die Grundlage für die effektive Nutzung der Leistung der GPU in wissenschaftlichen und anderen nicht-grafischen Berechnungen die Parallelisierung von Algorithmen in Hunderte von Ausführungseinheiten, die in Videochips verfügbar sind. Beispielsweise eignen sich viele Anwendungen der molekularen Modellierung gut für Berechnungen auf Videochips, sie erfordern eine große Rechenleistung und sind daher für paralleles Rechnen geeignet. Und die Verwendung mehrerer GPUs bietet noch mehr Rechenleistung zur Lösung solcher Probleme.

Das Durchführen von Berechnungen auf der GPU zeigt hervorragende Ergebnisse in Algorithmen, die eine parallele Datenverarbeitung verwenden. Das heißt, wenn dieselbe Folge mathematischer Operationen auf eine große Datenmenge angewendet wird. Dabei werden die besten Ergebnisse erzielt, wenn das Verhältnis der Anzahl der Rechenbefehle zur Anzahl der Speicherzugriffe groß genug ist. Dies stellt weniger Anforderungen an die Flusssteuerung, und die hohe Rechendichte und die große Datenmenge machen große Caches wie auf der CPU überflüssig.

Aufgrund all der oben beschriebenen Unterschiede übersteigt die theoretische Leistung von Videochips die Leistung der CPU erheblich. Nvidia stellt das folgende Diagramm des CPU- und GPU-Leistungswachstums in den letzten Jahren bereit:

Diese Daten sind natürlich nicht ohne Hinterlist. Auf der CPU ist es in der Tat viel einfacher, theoretische Zahlen in der Praxis zu erreichen, und die Zahlen werden im Fall der GPU für einfache Genauigkeit und im Fall der CPU für doppelte Genauigkeit angegeben. In jedem Fall reicht Single-Precision für einige parallele Aufgaben aus, und der Geschwindigkeitsunterschied zwischen Universal- und Grafikprozessoren ist sehr groß, und daher ist das Spiel die Kerze wert.

Die ersten Versuche, Berechnungen auf die GPU anzuwenden

Videochips werden seit langem in parallelen mathematischen Berechnungen verwendet. Die allerersten Versuche einer solchen Anwendung waren äußerst primitiv und beschränkten sich auf die Verwendung einiger Hardwarefunktionen wie Rasterung und Z-Pufferung. Aber im laufenden Jahrhundert, mit dem Aufkommen von Shadern, begannen sie, die Berechnung von Matrizen zu beschleunigen. Im Jahr 2003 wurde bei SIGGRAPH dem GPU-Computing ein separater Abschnitt zugewiesen, der als GPGPU (General-Purpose-Berechnung auf GPU) bezeichnet wurde – universelles GPU-Computing).

Die bekannteste BrookGPU ist der Brook-Stream-Programmiersprachen-Compiler, der entwickelt wurde, um nicht-grafische Berechnungen auf der GPU durchzuführen. Vor seinem Erscheinen entschieden sich Entwickler, die die Fähigkeiten von Videochips für Berechnungen nutzten, für eine von zwei gängigen APIs: Direct3D oder OpenGL. Dies schränkte die Verwendung der GPU stark ein, da 3D-Grafiken Shader und Texturen verwenden, die Parallelprogrammierer nicht kennen müssen, sie verwenden Threads und Kerne. Brook konnte helfen, ihre Aufgabe zu erleichtern. Diese an der Stanford University entwickelten Streaming-Erweiterungen der C-Sprache versteckten die 3D-API vor Programmierern und präsentierten den Videochip als parallelen Coprozessor. Der Compiler analysierte eine .br-Datei mit C++-Code und Erweiterungen und erzeugte Code, der mit einer DirectX-, OpenGL- oder x86-fähigen Bibliothek verknüpft war.

Natürlich hatte Brook viele Mängel, auf die wir näher eingehen und auf die wir später noch näher eingehen werden. Aber schon sein Erscheinen verursachte einen erheblichen Aufmerksamkeitsschub der gleichen Nvidia und ATI auf die GPU-Computing-Initiative, da die Entwicklung dieser Fähigkeiten den Markt in der Zukunft ernsthaft veränderte und einen ganz neuen Sektor davon erschloss – auf Parallel-Computing-Basis Videochips.

Darüber hinaus schlossen sich einige Forscher des Brook-Projekts dem Nvidia-Entwicklungsteam an, um eine Hardware-Software-Parallel-Computing-Strategie einzuführen und neue Marktanteile zu erschließen. Und der Hauptvorteil dieser Nvidia-Initiative war, dass die Entwickler alle Fähigkeiten ihrer GPUs bis ins kleinste Detail genau kennen und die Grafik-API nicht verwendet werden muss und Sie direkt mit der Hardware arbeiten können, indem Sie den Treiber verwenden. Das Ergebnis der Bemühungen dieses Teams ist Nvidia CUDA (Compute Unified Device Architecture), eine neue Hardware- und Softwarearchitektur für paralleles Rechnen auf der Nvidia-GPU, die Gegenstand dieses Artikels ist.

Anwendungsgebiete paralleler Berechnungen auf der GPU

Um zu verstehen, welche Vorteile die Übertragung von Berechnungen auf Videochips bringt, stellen wir die Durchschnittswerte vor, die von Forschern auf der ganzen Welt ermittelt wurden. Beim Übertragen von Berechnungen auf die GPU wird bei vielen Aufgaben im Durchschnitt eine 5- bis 30-fache Beschleunigung im Vergleich zu schnellen Universalprozessoren erreicht. Die größten Zahlen (in der Größenordnung von 100-facher Beschleunigung und noch mehr!) werden mit Code erzielt, der nicht sehr gut für Berechnungen mit SSE-Blöcken geeignet ist, aber für die GPU recht praktisch ist.

Dies sind nur einige Beispiele für Beschleunigungen von synthetischem Code auf der GPU im Vergleich zu SSE-vektorisiertem Code auf der CPU (laut Nvidia):

• Fluoreszenzmikroskopie: 12x;
• Molekulardynamik (nichtgebundene Kraftberechnung): 8-16x;
• Elektrostatik (direkte und mehrstufige Coulomb-Summation): 40-120x und 7x.

Und dies ist eine Platte, die Nvidia sehr liebt und die sie bei allen Präsentationen zeigt, auf die wir im zweiten Teil des Artikels näher eingehen werden, der konkreten Beispielen für praktische Anwendungen von CUDA-Computing gewidmet ist:

Wie Sie sehen können, sind die Zahlen sehr attraktiv, besonders die 100-150-fachen Gewinne sind beeindruckend. Im nächsten CUDA-Artikel werden wir uns einige dieser Zahlen genauer ansehen. Und jetzt listen wir die Hauptanwendungen auf, in denen GPU-Computing jetzt verwendet wird: Analyse und Verarbeitung von Bildern und Signalen, Physiksimulation, Computermathematik, Computerbiologie, Finanzberechnungen, Datenbanken, Dynamik von Gasen und Flüssigkeiten, Kryptographie, adaptive Strahlentherapie, Astronomie , Verarbeitung von Ton, Bioinformatik, biologische Simulationen, Computer Vision, Data Mining, digitales Kino und Fernsehen, elektromagnetische Simulationen, geografische Informationssysteme, militärische Anwendungen, Bergbauplanung, Molekulardynamik, Magnetresonanztomographie (MRI), neuronale Netze, ozeanographische Forschung, Partikel Physik, Proteinfaltungssimulation, Quantenchemie, Raytracing, Bildgebung, Radar, Reservoirsimulation, künstliche Intelligenz, Satellitendatenanalyse, seismische Erkundung, Chirurgie, Ultraschall, Videokonferenzen.

Details zu vielen Anwendungen finden Sie auf der Nvidia-Website im Abschnitt über . Wie Sie sehen können, ist die Liste ziemlich groß, aber das ist noch nicht alles! Es lässt sich fortsetzen, und wir können durchaus davon ausgehen, dass sich in Zukunft weitere Anwendungsgebiete paralleler Berechnungen auf Videochips finden werden, von denen wir noch keine Ahnung haben.

Nvidia CUDA-Fähigkeiten

Die CUDA-Technologie ist Nvidias Software- und Hardware-Computing-Architektur, die auf einer Erweiterung der C-Sprache basiert, die es ermöglicht, auf den Befehlssatz eines Grafikbeschleunigers zuzugreifen und seinen Speicher in Parallelverarbeitung zu verwalten. CUDA hilft bei der Implementierung von Algorithmen, die auf Grafikprozessoren von Geforce-Videobeschleunigern der achten Generation und älter (Geforce 8, Geforce 9, Geforce 200-Serie) sowie Quadro und Tesla implementiert werden können.

Die Komplexität der GPU-Programmierung mit CUDA ist zwar recht hoch, aber geringer als bei frühen GPGPU-Lösungen. Solche Programme erfordern eine Partitionierung der Anwendung über mehrere Multiprozessoren, ähnlich der MPI-Programmierung, aber ohne gemeinsame Nutzung der Daten, die im gemeinsam genutzten Videospeicher gespeichert sind. Und da die CUDA-Programmierung für jeden Multiprozessor der OpenMP-Programmierung ähnlich ist, erfordert sie ein gutes Verständnis der Speicherorganisation. Aber natürlich hängt die Komplexität der Entwicklung und Portierung nach CUDA stark von der Anwendung ab.

Das Developer Kit enthält viele Codebeispiele und ist gut dokumentiert. Der Lernprozess dauert etwa zwei bis vier Wochen für diejenigen, die bereits mit OpenMP und MPI vertraut sind. Die API basiert auf der erweiterten C-Sprache, und um Code aus dieser Sprache zu übersetzen, enthält das CUDA SDK den nvcc-Befehlszeilen-Compiler, der auf dem offenen Open64-Compiler basiert.

Wir listen die Hauptmerkmale von CUDA auf:

• einheitliche Software- und Hardwarelösung für paralleles Rechnen auf Nvidia-Videochips;
• eine große Auswahl an unterstützten Lösungen, von Mobilgeräten bis hin zu Multi-Chip
• die Standard-Programmiersprache C;
• Standardbibliotheken für die numerische Analyse FFT (Fast Fourier Transform) und BLAS (Linear Algebra);
• optimierter Datenaustausch zwischen CPU und GPU;
• Interaktion mit Grafik-API OpenGL und DirectX;
• Unterstützung für 32- und 64-Bit-Betriebssysteme: Windows XP, Windows Vista, Linux und MacOS X;
• die Fähigkeit, sich auf niedrigem Niveau zu entwickeln.

Zur Betriebssystemunterstützung sei noch hinzugefügt, dass alle wichtigen Linux-Distributionen offiziell unterstützt werden (Red Hat Enterprise Linux 3.x/4.x/5.x, SUSE Linux 10.x), CUDA aber laut Enthusiasten gut funktioniert auf anderen Builds: Fedora Core, Ubuntu, Gentoo usw.

Die CUDA-Entwicklungsumgebung (CUDA-Toolkit) umfasst:

• nvcc-Compiler;
• FFT- und BLAS-Bibliotheken;
• Profiler;
• gdb-Debugger für GPU;
• CUDA-Laufzeittreiber in Standard-Nvidia-Treibern enthalten
• Programmierhandbuch;
• CUDA Developer SDK (Quellcode, Dienstprogramme und Dokumentation).

In Quellcode-Beispielen: parallele bitonische Sortierung (bitonic sort), Matrix-Transposition, parallele Präfix-Summierung großer Arrays, Bildfaltung, diskrete Wavelet-Transformation, ein Beispiel für die Interaktion mit OpenGL und Direct3D, Verwendung von CUBLAS- und CUFFT-Bibliotheken, Optionspreisberechnung ( Black-Scholes, Binomialmodell, Monte-Carlo-Methode), paralleler Mersenne-Twister-Zufallszahlengenerator, Large-Array-Histogrammberechnung, Rauschunterdrückung, Sobel-Filter (Bound Detection).

Vorteile und Einschränkungen von CUDA

Aus der Sicht eines Programmierers ist die Grafikpipeline eine Reihe von Verarbeitungsstufen. Der Geometrieblock erzeugt Dreiecke, und der Rasterisierungsblock erzeugt Pixel, die auf dem Monitor angezeigt werden. Das traditionelle GPGPU-Programmiermodell sieht wie folgt aus:

Um im Rahmen eines solchen Modells Berechnungen auf die GPU zu übertragen, bedarf es eines speziellen Ansatzes. Sogar die elementweise Addition von zwei Vektoren erfordert das Zeichnen der Form auf dem Bildschirm oder in einem Puffer außerhalb des Bildschirms. Die Figur wird gerastert, die Farbe jedes Pixels wird nach einem vorgegebenen Programm (Pixel-Shader) berechnet. Das Programm liest die Eingabedaten aus den Texturen für jeden Pixel, addiert sie und schreibt sie in den Ausgabepuffer. Und all diese zahlreichen Operationen werden für das benötigt, was in einem einzigen Operator in einer herkömmlichen Programmiersprache geschrieben ist!

Daher hat die Verwendung von GPGPU für Allzweck-Computing eine Einschränkung in Form von zu viel Komplexität, die Entwickler lernen müssen. Und es gibt genug andere Einschränkungen, denn ein Pixel-Shader ist nur eine Formel für die Abhängigkeit der endgültigen Farbe eines Pixels von seinen Koordinaten, und die Pixel-Shader-Sprache ist eine Sprache zum Schreiben dieser Formeln mit einer C-ähnlichen Syntax. Die frühen GPGPU-Methoden sind ein cleverer Trick, um die Leistung der GPU zu nutzen, aber ohne jeglichen Komfort. Die Daten dort werden durch Bilder (Texturen) repräsentiert, und der Algorithmus wird durch einen Rasterisierungsprozess repräsentiert. Es sollte beachtet werden, und ein sehr spezifisches Modell des Gedächtnisses und der Ausführung.

Die Hardware- und Softwarearchitektur von Nvidia für das Rechnen auf GPUs von Nvidia unterscheidet sich von früheren GPGPU-Modellen darin, dass sie das Schreiben von Programmen für GPUs in echtem C mit Standardsyntax, Zeigern und der Notwendigkeit minimaler Erweiterungen für den Zugriff auf die Rechenressourcen von Videochips ermöglicht. CUDA ist nicht von Grafik-APIs abhängig und verfügt über einige Funktionen, die speziell für Allzweck-Computing entwickelt wurden.

Vorteile von CUDA gegenüber dem traditionellen Ansatz für GPGPU-Computing:

• die CUDA-Anbasiert auf der Standard-Programmiersprache C mit Erweiterungen, was das Erlernen und Implementieren der CUDA-Architektur vereinfacht;
• CUDA bietet Zugriff auf 16 KB gemeinsam genutzten Speicher pro Multiprozessor, der verwendet werden kann, um einen Cache mit einer höheren Bandbreite als Texturabrufe zu organisieren;
• effizientere Datenübertragung zwischen System und Videospeicher
• keine Notwendigkeit für Grafik-APIs mit Redundanz und Overhead;
• lineare Speicheradressierung und Sammeln und Streuen, die Fähigkeit, an beliebige Adressen zu schreiben;
• Hardwareunterstützung für Integer- und Bit-Operationen.

Haupteinschränkungen von CUDA:

• fehlende Rekursionsunterstützung für ausführbare Funktionen;
• die minimale Blockbreite beträgt 32 Threads;
• proprietäre CUDA-Architektur von Nvidia.

Die Schwächen der Programmierung mit früheren GPGPU-Methoden bestehen darin, dass diese Methoden keine Vertex-Shader-Ausführungseinheiten in früheren nicht einheitlichen Architekturen verwenden, Daten in Texturen gespeichert und an einen Off-Screen-Puffer ausgegeben werden und Multi-Pass-Algorithmen Pixel-Shader-Einheiten verwenden. Zu den GPGPU-Einschränkungen gehören: unzureichend effiziente Nutzung der Hardwarefunktionen, Einschränkungen der Speicherbandbreite, kein Scatter-Betrieb (nur Sammeln), obligatorische Verwendung der Grafik-API.

Die Hauptvorteile von CUDA gegenüber früheren GPGPU-Methoden ergeben sich aus der Tatsache, dass diese Architektur darauf ausgelegt ist, Nicht-Grafik-Computing auf der GPU effizient zu nutzen und die Programmiersprache C verwendet, ohne dass Algorithmen in eine Form portiert werden müssen, die für das Konzept der Grafik geeignet ist Pipeline. CUDA bietet einen neuen GPU-Rechenpfad, der keine Grafik-APIs verwendet und wahlfreien Speicherzugriff (Scatter oder Gather) bietet. Eine solche Architektur ist frei von den Nachteilen von GPGPU und verwendet alle Ausführungseinheiten und erweitert auch die Fähigkeiten durch ganzzahlige Mathematik und Bitverschiebungsoperationen.

Darüber hinaus eröffnet CUDA einige Hardwarefunktionen, die von den Grafik-APIs nicht verfügbar sind, wie z. B. gemeinsam genutzter Speicher. Dies ist eine kleine Speichermenge (16 Kilobyte pro Multiprozessor), auf die Blöcke von Threads zugreifen können. Es ermöglicht Ihnen, die am häufigsten aufgerufenen Daten zwischenzuspeichern und bietet eine schnellere Leistung als die Verwendung von Texturabrufen für diese Aufgabe. Dies reduziert wiederum die Durchsatzempfindlichkeit paralleler Algorithmen in vielen Anwendungen. Es ist beispielsweise nützlich für lineare Algebra, schnelle Fourier-Transformation und Bildverarbeitungsfilter.

Bequemer in CUDA und Speicherzugriff. Der Code in der Grafik-API gibt Daten als 32 Gleitkommawerte mit einfacher Genauigkeit (RGBA-Werte gleichzeitig zu acht Renderzielen) in vordefinierten Bereichen aus, und CUDA unterstützt Scatter-Recording – eine unbegrenzte Anzahl von Datensätzen an jeder Adresse. Solche Vorteile ermöglichen es, einige Algorithmen auf der GPU auszuführen, die mit GPGPU-Methoden basierend auf der Grafik-API nicht effizient implementiert werden können.

Außerdem speichern grafische APIs zwangsläufig Daten in Texturen, was ein vorheriges Packen großer Arrays in Texturen erfordert, was den Algorithmus verkompliziert und die Verwendung einer speziellen Adressierung erzwingt. Und mit CUDA können Sie Daten an jeder Adresse lesen. Ein weiterer Vorteil von CUDA ist die optimierte Kommunikation zwischen CPU und GPU. Und für Entwickler, die auf die Low-Level-Ebene zugreifen möchten (z. B. beim Schreiben einer anderen Programmiersprache), bietet CUDA die Möglichkeit der Low-Level-Assemblersprachenprogrammierung.

Geschichte der CUDA-Entwicklung

Die CUDA-Entwicklung wurde mit dem G80-Chip im November 2006 angekündigt, und eine öffentliche Beta-Version des CUDA SDK wurde im Februar 2007 veröffentlicht. Version 1.0 wurde im Juni 2007 veröffentlicht, um Tesla-Lösungen auf Basis des G80-Chips für den High Performance Computing-Markt einzuführen. Ende des Jahres wurde dann CUDA 1.1 Beta veröffentlicht, das trotz leicht steigender Versionsnummer einiges an Neuerungen brachte.

Aus dem, was in CUDA 1.1 erschien, können wir die Einbeziehung der CUDA-Funktionalität in reguläre Nvidia-Videotreiber feststellen. Dies bedeutete, dass es in den Anforderungen für jedes CUDA-Programm ausreichte, eine Grafikkarte der Geforce 8-Serie und höher sowie die minimale Treiberversion 169.xx anzugeben. Dies ist sehr wichtig für Entwickler, wenn diese Bedingungen erfüllt sind, funktionieren CUDA-Programme für jeden Benutzer. Außerdem wurde eine asynchrone Ausführung zusammen mit dem Kopieren von Daten (nur für G84-, G86-, G92- und höhere Chips), einer asynchronen Übertragung von Daten in den Videospeicher, atomaren Speicherzugriffsvorgängen, Unterstützung für 64-Bit-Versionen von Windows und der Möglichkeit von Multi hinzugefügt -Chip-CUDA-Betrieb im SLI-Modus.

Derzeit ist die aktuelle Version für Lösungen auf Basis des GT200 - CUDA 2.0 gedacht, das zusammen mit der Geforce GTX 200-Reihe veröffentlicht wurde.Die Beta-Version wurde bereits im Frühjahr 2008 veröffentlicht. Die zweite Version hat: Unterstützung für Berechnungen mit doppelter Genauigkeit (Hardwareunterstützung nur für GT200), Windows Vista (32- und 64-Bit-Versionen) und Mac OS X werden endlich unterstützt, Debugging- und Profiling-Tools wurden hinzugefügt, 3D-Texturen werden unterstützt, optimiert Datentransfer.

Bei Berechnungen mit doppelter Genauigkeit ist ihre Geschwindigkeit auf der aktuellen Hardwaregeneration um ein Vielfaches niedriger als bei einfacher Genauigkeit. Die Gründe werden in unserem diskutiert. Die Implementierung dieser Unterstützung im GT200 liegt in der Tatsache, dass FP32-Blöcke nicht verwendet werden, um Ergebnisse in einem viermal langsameren Tempo zu erhalten. Um FP64-Berechnungen zu unterstützen, hat Nvidia beschlossen, dedizierte Rechenblöcke zu erstellen. Und in GT200 gibt es zehnmal weniger davon als FP32-Blöcke (ein Block mit doppelter Genauigkeit für jeden Multiprozessor).

In der Realität kann die Leistung sogar noch geringer ausfallen, da die Architektur für das 32-Bit-Lesen von Speicher und Registern optimiert ist, zudem in Grafikanwendungen keine doppelte Genauigkeit benötigt wird und beim GT200 eher darauf ausgelegt ist. Ja, und moderne Quad-Core-Prozessoren zeigen nicht viel weniger reale Leistung. Da diese Unterstützung sogar 10-mal langsamer als Single Precision ist, ist sie für Mixed-Precision-Schaltungen nützlich. Eine gängige Technik besteht darin, anfänglich ungefähre Ergebnisse mit einfacher Genauigkeit zu erhalten und sie dann mit doppelter Genauigkeit zu verfeinern. Jetzt kann dies direkt auf der Grafikkarte erfolgen, ohne Zwischendaten an die CPU zu senden.

Eine weitere nützliche Funktion von CUDA 2.0 hat seltsamerweise nichts mit der GPU zu tun. Es ist gerade jetzt möglich, CUDA-Code in hocheffizienten Multithread-SSE-Code für eine schnelle Ausführung auf der CPU zu kompilieren. Das heißt, jetzt eignet sich diese Funktion nicht nur zum Debuggen, sondern auch für den realen Einsatz auf Systemen ohne Nvidia-Grafikkarte. Schließlich wird die Verwendung von CUDA in normalem Code durch die Tatsache eingeschränkt, dass Nvidia-Grafikkarten, obwohl sie die beliebtesten unter den dedizierten Videolösungen sind, nicht in allen Systemen verfügbar sind. Und vor Version 2.0 müssten in solchen Fällen zwei unterschiedliche Codes geschrieben werden: für CUDA und separat für die CPU. Und jetzt können Sie jedes CUDA-Programm mit hoher Effizienz auf der CPU ausführen, wenn auch mit einer geringeren Geschwindigkeit als auf Videochips.

Von Nvidia CUDA unterstützte Lösungen

Alle CUDA-fähigen Grafikkarten können dazu beitragen, die anspruchsvollsten Aufgaben zu beschleunigen, von der Audio- und Videoverarbeitung bis hin zur medizinischen und wissenschaftlichen Forschung. Die einzige wirkliche Einschränkung besteht darin, dass viele CUDA-Programme mindestens 256 Megabyte Videospeicher benötigen, und dies ist eine der wichtigsten Spezifikationen für CUDA-Anwendungen.

Eine aktuelle Liste der CUDA-fähigen Produkte finden Sie unter . Zum Zeitpunkt der Erstellung dieses Artikels unterstützten CUDA-Berechnungen alle Produkte der Serien Geforce 200, Geforce 9 und Geforce 8, einschließlich mobiler Produkte, beginnend mit Geforce 8400M, sowie Chipsätze Geforce 8100, 8200 und 8300. Moderne Quadro und alle Tesla: S1070, C1060, C870, D870 und S870.

Wir weisen besonders darauf hin, dass zusammen mit den neuen Grafikkarten Geforce GTX 260 und 280 die entsprechenden High-Performance-Computing-Lösungen angekündigt wurden: Tesla C1060 und S1070 (siehe Foto oben), die diesen Herbst zum Kauf angeboten werden. In ihnen wird dieselbe GPU verwendet - GT200, in C1060 ist es eine, in S1070 - vier. Aber anders als Gaming-Lösungen verbrauchen sie vier Gigabyte Speicher pro Chip. Von den Minuspunkten ist vielleicht die geringere Speicherfrequenz und Speicherbandbreite als bei Gaming-Karten, die 102 Gigabyte / s pro Chip bereitstellen.

Zusammensetzung von Nvidia CUDA

CUDA enthält zwei APIs: High-Level (CUDA Runtime API) und Low-Level (CUDA Driver API), obwohl es unmöglich ist, beide gleichzeitig in einem Programm zu verwenden, müssen Sie die eine oder andere verwenden. Die High-Level-Version arbeitet „auf“ der Low-Level-Version, alle Laufzeitaufrufe werden in einfache Anweisungen übersetzt, die von der Low-Level-Treiber-API verarbeitet werden. Aber selbst die „High-Level“-API setzt Wissen über das Design und die Funktionsweise von Nvidia-Videochips voraus, da gibt es kein zu hohes Abstraktionsniveau.

Es gibt eine andere Ebene, noch höher - zwei Bibliotheken:

CUBLAS- CUDA-Version von BLAS (Basic Linear Algebra Subprograms), entwickelt für die Berechnung von Problemen der linearen Algebra und den direkten Zugriff auf GPU-Ressourcen;

MANSCHETTE- CUDA-Version der Fast-Fourier-Transformationsbibliothek zur Berechnung der Fast-Fourier-Transformation, die in der Signalverarbeitung weit verbreitet ist. Die folgenden Transformationstypen werden unterstützt: Komplex-Komplex (C2C), Real-Komplex (R2C) und Komplex-Real (C2R).

Schauen wir uns diese Bibliotheken genauer an. CUBLAS sind Standardalgorithmen der linearen Algebra, die in die CUDA-Sprache übersetzt wurden. Momentan wird nur ein bestimmter Satz grundlegender CUBLAS-Funktionen unterstützt. Die Bibliothek ist sehr einfach zu bedienen: Sie müssen eine Matrix und Vektorobjekte im Videospeicher erstellen, mit Daten füllen, die erforderlichen CUBLAS-Funktionen aufrufen und die Ergebnisse aus dem Videospeicher zurück in den Systemspeicher laden. CUBLAS enthält spezielle Funktionen zum Erstellen und Zerstören von Objekten im GPU-Speicher sowie zum Lesen und Schreiben von Daten in diesen Speicher. Unterstützte BLAS-Funktionen: Level 1, 2 und 3 für reelle Zahlen, Level 1 CGEMM für Komplexe. Ebene 1 sind Vektor-Vektor-Operationen, Ebene 2 sind Vektor-Matrix-Operationen, Ebene 3 sind Matrix-Matrix-Operationen.

CUFFT - CUDA-Variante der schnellen Fourier-Transformation - weit verbreitet und sehr wichtig in der Signalanalyse, Filterung usw. CUFFT bietet eine einfache Schnittstelle für eine effiziente FFT-Berechnung auf Nvidia-GPUs, ohne dass eine benutzerdefinierte FFT für die GPU entwickelt werden muss. Die CUDA FFT-Variante unterstützt 1D-, 2D- und 3D-Transformationen von komplexen und realen Daten, Stapelausführung für mehrere 1D-Transformationen parallel, Größen von 2D- und 3D-Transformationen können innerhalb von liegen, für 1D wird eine Größe von bis zu 8 Millionen Elementen unterstützt.

Grundlagen zum Erstellen von Programmen auf CUDA

Um den folgenden Text zu verstehen, sollten Sie die grundlegenden Architekturmerkmale von Nvidia-Videochips verstehen. Die GPU besteht aus mehreren Clustern von Textureinheiten (Texture Processing Cluster). Jeder Cluster besteht aus einem vergrößerten Block von Texturabrufen und zwei oder drei Streaming-Multiprozessoren, von denen jeder aus acht Rechengeräten und zwei superfunktionalen Blöcken besteht. Alle Anweisungen werden nach dem SIMD-Prinzip ausgeführt, indem eine Anweisung auf alle Fäden einer Kette angewendet wird (ein Begriff aus der Textilindustrie, in CUDA ist dies eine Gruppe von 32 Fäden - die minimale Datenmenge, die von Multiprozessoren verarbeitet wird). Diese Ausführungsmethode wurde SIMT (Single Instruction Multiple Threads – One Instruction and Many Threads) genannt.

Jeder der Multiprozessoren hat bestimmte Ressourcen. So gibt es einen speziellen Shared Memory mit einer Kapazität von 16 Kilobyte pro Multiprozessor. Dies ist jedoch kein Cache, da der Programmierer ihn für jeden Bedarf verwenden kann, ähnlich wie der lokale Speicher in der SPU von Cell-Prozessoren. Dieser gemeinsam genutzte Speicher ermöglicht den Austausch von Informationen zwischen Threads desselben Blocks. Wichtig ist, dass alle Threads eines Blocks immer vom selben Multiprozessor ausgeführt werden. Und Threads aus verschiedenen Blöcken können keine Daten austauschen, und Sie müssen sich an diese Einschränkung erinnern. Gemeinsam genutzter Speicher ist oft nützlich, außer wenn mehrere Threads auf dieselbe Speicherbank zugreifen. Multiprozessoren können auch auf den Videospeicher zugreifen, jedoch mit höherer Latenz und geringerer Bandbreite. Um den Zugriff zu beschleunigen und die Häufigkeit des Zugriffs auf den Videospeicher zu reduzieren, haben Multiprozessoren 8 Kilobyte Cache für Konstanten und Texturdaten.

Der Multiprozessor verwendet 8192–16384 (für G8x/G9x bzw. GT2xx) Register, die allen Threads aller auf ihm ausgeführten Blöcke gemeinsam sind. Die maximale Anzahl von Blöcken pro Multiprozessor für G8x/G9x beträgt acht, und die Anzahl von Warps beträgt 24 (768 Threads pro Multiprozessor). Insgesamt können die Top-Grafikkarten der Geforce-8- und -9-Serie bis zu 12288 Threads gleichzeitig verarbeiten. Die auf dem GT200 basierende GeForce GTX 280 bietet bis zu 1024 Threads pro Multiprozessor, sie hat 10 Cluster aus drei Multiprozessoren, die bis zu 30720 Threads verarbeiten. Wenn Sie diese Einschränkungen kennen, können Sie Algorithmen für verfügbare Ressourcen optimieren.

Der erste Schritt bei der Portierung einer vorhandenen Anwendung auf CUDA besteht darin, sie zu profilieren und Codebereiche zu identifizieren, die Engpässe darstellen, die die Arbeit verlangsamen. Wenn sich unter solchen Abschnitten geeignete für eine schnelle parallele Ausführung befinden, werden diese Funktionen in C- und CUDA-Erweiterungen zur Ausführung auf der GPU übertragen. Das Programm wird mit dem von Nvidia bereitgestellten Compiler kompiliert, der Code sowohl für die CPU als auch für die GPU generiert. Wenn ein Programm ausgeführt wird, führt die CPU seine Teile des Codes aus, und die GPU führt den CUDA-Code mit den schwersten parallelen Berechnungen aus. Dieser für die GPU konzipierte Teil wird als Kernel (Kernel) bezeichnet. Der Kernel definiert die Operationen, die an den Daten durchgeführt werden sollen.

Der Videochip empfängt den Kern und erstellt Kopien für jedes Datenelement. Diese Kopien werden Threads genannt. Ein Stream enthält einen Zähler, Register und einen Zustand. Bei großen Datenmengen, etwa bei der Bildverarbeitung, werden Millionen von Threads gestartet. Threads werden in Gruppen von 32, sogenannten Warps, ausgeführt.Warps werden zugewiesen, um auf bestimmten Streaming-Multiprozessoren ausgeführt zu werden. Jeder Multiprozessor besteht aus acht Kernen – Stream-Prozessoren, die einen MAD-Befehl pro Taktzyklus ausführen. Um einen 32-Thread-Warp auszuführen, sind vier Multiprozessorzyklen erforderlich (wir sprechen von der Shader-Domain-Frequenz, die 1,5 GHz und mehr beträgt).

Der Multiprozessor ist kein herkömmlicher Mehrkernprozessor, er eignet sich gut für Multithreading und unterstützt bis zu 32 Warps gleichzeitig.Bei jedem Taktzyklus wählt die Hardware aus, welcher der Warps ausgeführt werden soll, und schaltet ohne von einem zum anderen um Zyklen verlieren. Wenn wir eine Analogie zum zentralen Prozessor ziehen, ist dies so, als würden 32 Programme gleichzeitig ausgeführt und bei jedem Taktzyklus zwischen ihnen gewechselt, ohne dass ein Kontextwechsel verloren geht. In Wirklichkeit unterstützen die CPU-Kerne die gleichzeitige Ausführung eines Programms und schalten mit einer Verzögerung von Hunderten von Zyklen auf andere um.

CUDA-Programmiermodell

Auch hier verwendet CUDA ein paralleles Berechnungsmodell, bei dem jeder der SIMD-Prozessoren dieselbe Anweisung an verschiedenen Datenelementen parallel ausführt. Die GPU ist ein Rechengerät, ein Coprozessor (Gerät) für den zentralen Prozessor (Host), der über einen eigenen Speicher verfügt und eine große Anzahl von Threads parallel verarbeitet. Der Kernel (Kernel) ist eine Funktion für die GPU, die von Threads ausgeführt wird (eine Analogie von 3D-Grafiken - ein Shader).

Wir haben oben gesagt, dass sich ein Videochip von einer CPU dadurch unterscheidet, dass er Zehntausende von Threads gleichzeitig verarbeiten kann, was normalerweise für gut parallelisierte Grafiken gilt. Jeder Stream ist skalar, es müssen keine Daten in 4-Komponenten-Vektoren gepackt werden, was für die meisten Aufgaben bequemer ist. Die Anzahl der logischen Threads und Thread-Blöcke übersteigt die Anzahl der physischen Ausführungseinheiten, was eine gute Skalierbarkeit für die gesamte Palette von Unternehmenslösungen bietet.

Das Programmiermodell in CUDA geht von einer Thread-Gruppierung aus. Threads werden zu Thread-Blöcken zusammengefasst – eindimensionale oder zweidimensionale Thread-Gitter, die über gemeinsam genutzten Speicher und Synchronisationspunkte miteinander interagieren. Das Programm (Kernel) wird über ein Raster von Thread-Blöcken ausgeführt, siehe Abbildung unten. Ein Grid wird gleichzeitig ausgeführt. Jeder Block kann eine ein-, zwei- oder dreidimensionale Form haben und auf aktueller Hardware aus 512 Threads bestehen.

Thread-Blöcke laufen in kleinen Gruppen, die Warps genannt werden und eine Größe von 32 Threads haben. Dies ist die minimale Datenmenge, die in Multiprozessoren verarbeitet werden kann. Und da dies nicht immer praktisch ist, ermöglicht Ihnen CUDA, mit Blöcken zu arbeiten, die 64 bis 512 Threads enthalten.

Das Gruppieren von Blöcken in Grids ermöglicht es Ihnen, sich von den Einschränkungen zu lösen und den Kernel in einem Aufruf auf eine größere Anzahl von Threads anzuwenden. Es hilft auch beim Skalieren. Wenn die GPU nicht über genügend Ressourcen verfügt, führt sie Blöcke sequentiell aus. Ansonsten können die Blöcke parallel ausgeführt werden, was für die optimale Verteilung der Arbeit auf Videochips unterschiedlicher Ebenen, von mobilen bis zu integrierten, wichtig ist.

CUDA-Speichermodell

Das Speichermodell in CUDA zeichnet sich durch die Möglichkeit der Byte-Adressierung, Unterstützung von Gather und Scatter aus. Für jeden Stromprozessor steht eine ziemlich große Anzahl von Registern zur Verfügung, bis zu 1024 Stück. Der Zugriff auf sie ist sehr schnell, Sie können 32-Bit-Ganzzahlen oder Gleitkommazahlen darin speichern.

Jeder Thread hat Zugriff auf die folgenden Arten von Speicher:

globales Gedächtnis- die größte verfügbare Speichermenge für alle Multiprozessoren auf einem Videochip, die Größe reicht von 256 Megabyte bis 1,5 Gigabyte für aktuelle Lösungen (und bis zu 4 GB für Tesla). Es hat einen hohen Durchsatz, mehr als 100 Gigabyte / s für Top-Nvidia-Lösungen, aber sehr große Verzögerungen von mehreren hundert Zyklen. Nicht zwischenspeicherbar, unterstützt generische Lade- und Speicheranweisungen und reguläre Speicherzeiger.

lokaler Speicher ist eine kleine Speichermenge, auf die nur ein Stream-Prozessor Zugriff hat. Es ist relativ langsam - genauso wie das globale.

Geteilte Erinnerung ist ein 16-Kilobyte-Speicherblock (in den Videochips der aktuellen Architektur) mit gemeinsamem Zugriff für alle Stream-Prozessoren im Multiprozessor. Dieser Speicher ist sehr schnell, genau wie Register. Es bietet Thread-Interaktion, wird direkt vom Entwickler verwaltet und hat eine geringe Latenz. Vorteile von Shared Memory: Nutzung in Form eines vom Programmierer verwalteten First-Level-Cache, Reduzierung von Verzögerungen beim Zugriff auf Daten durch Ausführungseinheiten (ALUs), Reduzierung der Anzahl globaler Speicherzugriffe.

Ständige Erinnerung- ein 64-Kilobyte-Speicherbereich (derselbe für aktuelle GPUs), der nur von allen Multiprozessoren gelesen werden kann. Es wird mit 8 Kilobyte pro Multiprozessor zwischengespeichert. Ziemlich langsam - eine Verzögerung von mehreren hundert Zyklen, wenn die erforderlichen Daten im Cache fehlen.

Texturspeicher- ein Speicherblock, der zum Lesen durch alle Multiprozessoren verfügbar ist. Die Datenabtastung erfolgt über die Textureinheiten des Videochips, sodass ohne zusätzliche Kosten die Möglichkeit der linearen Dateninterpolation gegeben ist. 8 Kilobyte zwischengespeichert pro Multiprozessor. Langsam wie global - Hunderte von Verzögerungszyklen, wenn keine Daten im Cache vorhanden sind.

Natürlich sind der globale, der lokale, der Textur- und der konstante Speicher physisch derselbe Speicher, der als lokaler Videospeicher der Videokarte bekannt ist. Ihre Unterschiede liegen in unterschiedlichen Caching-Algorithmen und Zugriffsmodellen. Die CPU kann nur externen Speicher aktualisieren und anfordern: global, konstant und texturiert.

Aus dem oben Geschriebenen geht hervor, dass CUDA einen besonderen Entwicklungsansatz impliziert, der nicht ganz derselbe ist wie bei Programmen für die CPU. Sie müssen sich an verschiedene Arten von Speicher erinnern, dass lokaler und globaler Speicher nicht zwischengespeichert werden und die Verzögerungen beim Zugriff darauf viel höher sind als beim registrierten Speicher, da sie sich physisch in separaten Mikroschaltkreisen befinden.

Ein typisches, aber nicht zwingendes Problemlösungsmuster:

• die Aufgabe ist in Teilaufgaben unterteilt;
• die Eingabedaten werden in Blöcke unterteilt, die in den gemeinsam genutzten Speicher passen;
• jeder Block wird von einem Thread-Block verarbeitet;
• der Unterblock wird aus dem globalen in den gemeinsam genutzten Speicher geladen;
• entsprechende Berechnungen werden an Daten im gemeinsam genutzten Speicher durchgeführt;
• die Ergebnisse werden aus dem gemeinsamen Speicher zurück in den globalen kopiert.

Programmierumgebung

CUDA enthält Laufzeitbibliotheken:

• ein gemeinsamer Teil, der integrierte Vektortypen und Teilmengen von RTL-Aufrufen bereitstellt, die auf der CPU und GPU unterstützt werden;
• CPU-Komponente zur Steuerung einer oder mehrerer GPUs;
• Eine GPU-Komponente, die GPU-spezifische Funktionen bereitstellt.

Der Hauptprozess der CUDA-Anwendung läuft auf einem generischen Prozessor (Host), er führt mehrere Kopien der Kernel-Prozesse auf der Grafikkarte aus. Der Code für die CPU tut Folgendes: Initialisiert die GPU, weist Speicher auf der Grafikkarte und dem System zu, kopiert die Konstanten in den Speicher der Grafikkarte, führt mehrere Kopien der Kernelprozesse auf der Grafikkarte aus, kopiert das Ergebnis von der Videospeicher, gibt den Speicher frei und beendet sich.

Als Beispiel zum Verständnis ist hier der CPU-Code für die Vektoraddition in CUDA dargestellt:

Die vom Videochip ausgeführten Funktionen haben die folgenden Einschränkungen: keine Rekursion, keine statischen Variablen innerhalb der Funktionen und keine variable Anzahl von Argumenten. Zwei Arten der Speicherverwaltung werden unterstützt: linearer Speicher, auf den über 32-Bit-Zeiger zugegriffen wird, und CUDA-Arrays, auf die nur über Texturabruffunktionen zugegriffen wird.

CUDA-Programme können mit Grafik-APIs interagieren: um im Programm generierte Daten zu rendern, Rendering-Ergebnisse zu lesen und sie mit CUDA-Tools zu verarbeiten (z. B. bei der Implementierung von Nachbearbeitungsfiltern). Um dies zu tun, können Grafik-API-Ressourcen in den globalen CUDA-Speicherraum abgebildet werden (Erhalten einer Ressourcenadresse). Die folgenden Arten von Grafik-API-Ressourcen werden unterstützt: Pufferobjekte (PBO / VBO) in OpenGL, Vertex-Puffer und Texturen (2D, 3D und Cubemaps) Direct3D9.

CUDA-Anwendungskompilierungsschritte:

CUDA-C-Quellcodedateien werden mit dem NVCC-Programm kompiliert, das andere Tools umschließt und aufruft: cudacc, g++, cl usw. NVCC generiert: CPU-Code, der zusammen mit dem Rest der in reinem C geschriebenen Anwendung kompiliert wird, und die PTX-Objektcode für den Videochip. Ausführbare Dateien mit CUDA-Code erfordern zwingend das Vorhandensein der CUDA-Laufzeitbibliothek (cudart) und der CUDA-Kernbibliothek (cuda).

Optimierung von Programmen auf CUDA

Im Rahmen eines Übersichtsartikels ist es natürlich unmöglich, ernsthafte Optimierungsfragen in der CUDA-Programmierung zu berücksichtigen. Deshalb werden wir nur kurz auf die grundlegenden Dinge eingehen. Um die Fähigkeiten von CUDA effektiv zu nutzen, müssen Sie die üblichen Methoden zum Schreiben von Programmen für die CPU vergessen und diese Algorithmen verwenden, die für Tausende von Threads gut parallelisiert sind. Es ist auch wichtig, den optimalen Ort zum Speichern von Daten (Register, gemeinsam genutzter Speicher usw.) zu finden, die Datenübertragung zwischen CPU und GPU zu minimieren und Pufferung zu verwenden.

Generell sollte man bei der Optimierung eines CUDA-Programms versuchen, ein optimales Gleichgewicht zwischen Größe und Anzahl der Blöcke zu erreichen. Mehr Threads in einem Block reduzieren die Auswirkungen der Speicherlatenz, reduzieren aber auch die verfügbare Anzahl von Registern. Außerdem ist ein Block von 512 Threads ineffizient, Nvidia selbst empfiehlt, Blöcke von 128 oder 256 Threads als Kompromisswert zu verwenden, um eine optimale Latenz und Anzahl der Register zu erreichen.

Zu den wichtigsten Optimierungspunkten von CUDA-Programmen gehören: Möglichst aktive Nutzung des gemeinsamen Speichers, da er viel schneller ist als der globale Videospeicher der Grafikkarte; Lese- und Schreibvorgänge aus dem globalen Speicher sollten nach Möglichkeit zusammengeführt werden. Dazu müssen Sie spezielle Datentypen zum gleichzeitigen Lesen und Schreiben von 32/64/128 Datenbits in einem Vorgang verwenden. Wenn Lesevorgänge schwierig zusammenzuführen sind, können Sie versuchen, Texturabrufe zu verwenden.

Ergebnisse

Die von Nvidia vorgestellte Hard- und Softwarearchitektur für Berechnungen auf CUDA-Videochips ist gut geeignet, um vielfältige Aufgaben mit hoher Parallelität zu lösen. CUDA läuft auf einer großen Anzahl von Nvidia-Videochips und verbessert das GPU-Programmiermodell, indem es es stark vereinfacht und viele Funktionen wie gemeinsam genutzten Speicher, die Fähigkeit zum Synchronisieren von Threads, Berechnungen mit doppelter Genauigkeit und Ganzzahloperationen hinzufügt.

CUDA ist eine Technologie, die jedem Softwareentwickler zur Verfügung steht und von jedem Programmierer verwendet werden kann, der die Sprache C beherrscht. Sie müssen sich nur an ein anderes Programmierparadigma gewöhnen, das dem parallelen Rechnen innewohnt. Aber wenn der Algorithmus im Prinzip gut parallelisiert ist, dann werden das Studium und die Zeit, die für die CUDA-Programmierung aufgewendet wird, in einem Vielfachen zurückkommen.

Es ist wahrscheinlich, dass aufgrund der weit verbreiteten Verwendung von Grafikkarten in der Welt die Entwicklung von parallelem Computing auf der GPU die High Performance Computing-Industrie stark beeinflussen wird. Diese Möglichkeiten haben bereits großes Interesse in wissenschaftlichen Kreisen geweckt, und nicht nur in ihnen. Schließlich ist das Potenzial, Algorithmen, die sich gut für die Parallelisierung eignen (auf erschwinglicher Hardware, was nicht weniger wichtig ist), auf einmal dutzendfach zu beschleunigen, nicht so verbreitet.

Allzweckprozessoren entwickeln sich recht langsam, sie haben solche Leistungsspitzen nicht. So laut es auch klingen mag, jeder, der schnelles Computing braucht, kann jetzt einen preiswerten persönlichen Supercomputer auf seinen Schreibtisch bekommen, manchmal ohne zusätzliches Geld zu investieren, da Nvidia-Grafikkarten weit verbreitet sind. Ganz zu schweigen von den Effizienzgewinnen in Bezug auf GFLOPS/$ und GFLOPS/W, die GPU-Hersteller so lieben.

Die Zukunft vieler Computer liegt eindeutig in parallelen Algorithmen, fast alle neuen Lösungen und Initiativen zielen in diese Richtung. Bisher befindet sich die Entwicklung neuer Paradigmen jedoch noch in einem frühen Stadium, man muss Threads manuell erstellen und Speicherzugriffe planen, was Aufgaben im Vergleich zur herkömmlichen Programmierung erschwert. Aber die CUDA-Technologie hat einen Schritt in die richtige Richtung gemacht und sieht eindeutig nach einer erfolgreichen Lösung aus, insbesondere wenn es Nvidia gelingt, die Entwickler so weit wie möglich von ihren Vorteilen und Perspektiven zu überzeugen.

Aber natürlich werden GPUs keine CPUs ersetzen. In ihrer jetzigen Form sind sie dafür nicht ausgelegt. Jetzt, da sich Videochips allmählich in Richtung CPU bewegen und immer universeller werden (Berechnungen mit Gleitkommazahlen mit einfacher und doppelter Genauigkeit, Ganzzahlberechnungen), werden die CPUs immer „paralleler“, erwerben eine große Anzahl von Kernen, Multithreading Technologien, ganz zu schweigen vom Auftreten von SIMD-Blöcken und heterogenen Prozessorprojekten. Höchstwahrscheinlich werden GPU und CPU in Zukunft einfach verschmelzen. Es ist bekannt, dass viele Unternehmen, darunter Intel und AMD, an ähnlichen Projekten arbeiten. Und es spielt keine Rolle, ob die GPU von der CPU verbraucht wird oder umgekehrt.

In dem Artikel haben wir hauptsächlich über die Vorteile von CUDA gesprochen. Aber es gibt auch ein Haar in der Suppe. Einer der wenigen Nachteile von CUDA ist seine schlechte Portabilität. Diese Architektur funktioniert nur auf den Videochips dieser Firma und nicht auf allen, sondern ab der Geforce 8- und 9-Serie und den entsprechenden Quadro und Tesla. Ja, es gibt viele solcher Lösungen auf der Welt, Nvidia gibt eine Zahl von 90 Millionen CUDA-kompatiblen Videochips an. Das ist einfach großartig, aber Konkurrenten bieten eigene Lösungen an, die sich von CUDA unterscheiden. AMD hat also Stream Computing, Intel wird in Zukunft Ct haben.

Welche der Technologien sich durchsetzen, sich durchsetzen und länger leben werden als die anderen – nur die Zeit wird es zeigen. Aber CUDA hat gute Chancen, weil es im Vergleich zu Stream Computing beispielsweise eine weiter entwickelte und einfacher zu bedienende Programmierumgebung in der regulären Sprache C bietet. Vielleicht hilft ein Dritter bei der Definition, indem er eine Art allgemeine Lösung veröffentlicht. So hat Microsoft im nächsten DirectX-Update unter Version 11 Compute-Shader versprochen, die zu einer Art Durchschnittslösung werden können, die jedem, oder fast jedem passt.

Nach den vorläufigen Daten zu urteilen, lehnt sich dieser neue Shader-Typ stark an das CUDA-Modell an. Und wenn Sie jetzt in dieser Umgebung programmieren, können Sie unmittelbar davon profitieren und die notwendigen Fähigkeiten für die Zukunft erwerben. Aus Sicht des High Performance Computing hat DirectX auch den entscheidenden Nachteil der schlechten Portabilität, da die API auf die Windows-Plattform beschränkt ist. Es wird jedoch ein weiterer Standard entwickelt – die offene Multiplattform-Initiative OpenCL, die von den meisten Unternehmen unterstützt wird, darunter Nvidia, AMD, Intel, IBM und viele andere.

Vergessen Sie nicht, dass Sie im nächsten CUDA-Artikel spezifische praktische Anwendungen wissenschaftlicher und anderer nicht-grafischer Berechnungen untersuchen werden, die von Entwicklern auf der ganzen Welt mit Nvidia CUDA durchgeführt werden.

– eine Reihe von Low-Level-Programmierschnittstellen ( API) zum Erstellen von Spielen und anderen leistungsstarken Multimedia-Anwendungen. Beinhaltet Unterstützung für hohe Leistung 2D- und 3D-Grafik, Sound und Eingabegeräte.

Direct3D (D3D) – 3D-Ausgabeschnittstelle Primitive(geometrische Körper). Eingeschlossen in .

OpenGL(aus dem Englischen. Öffnen Sie die Grafikbibliothek, wörtlich - eine offene Grafikbibliothek) ist eine Spezifikation, die eine programmiersprachenunabhängige plattformübergreifende Programmierschnittstelle zum Schreiben von Anwendungen definiert, die zweidimensionale und dreidimensionale Computergrafiken verwenden. Enthält über 250 Funktionen zum Zeichnen komplexer 3D-Szenen aus einfachen Primitiven. Es wird bei der Erstellung von Videospielen, virtueller Realität und Visualisierung in der wissenschaftlichen Forschung verwendet. Auf dem Bahnsteig Windows konkurriert mit .

OpenCL(aus dem Englischen. Offene Computersprache, wörtlich - eine offene Computersprache) - Rahmen(Framework eines Softwaresystems) zum Schreiben von Computerprogrammen im Zusammenhang mit parallelem Rechnen auf verschiedenen Grafiken ( Grafikkarte) und ( ). Zum Rahmen OpenCL enthält eine Programmiersprache und eine An( API). OpenCL bietet Parallelität auf der Befehlsebene und auf der Datenebene und ist die Implementierung der Technik GPGPU.

GPGPU(abgekürzt aus dem Englischen. Allzweck-Grafikverarbeitungseinheiten, buchstäblich - Grafikkarte Allzweck) - eine Technik zur Verwendung des Grafikprozessors einer Grafikkarte für allgemeine Berechnungen, die normalerweise ausgeführt wird.

Shader(Englisch) Shader) ist ein Programm zum Konstruieren von Schatten auf synthetisierten Bildern, das in dreidimensionalen Grafiken verwendet wird, um die endgültigen Parameter eines Objekts oder Bildes zu bestimmen. In der Regel umfasst es die Beschreibung von Lichtabsorption und -streuung, Texturabbildung, Reflexion und Brechung, Schattierung, Oberflächenverschiebung und Nachbearbeitungseffekten beliebiger Komplexität. Komplexe Oberflächen können mit einfachen geometrischen Formen gerendert werden.

Wiedergabe(Englisch) Wiedergabe) - Visualisierung, in der Computergrafik der Prozess, mit Hilfe von Software ein Bild von einem Modell zu erhalten.

SDK(abgekürzt aus dem Englischen. Softwareentwicklungskit) ist eine Reihe von Softwareentwicklungstools.

Zentralprozessor(abgekürzt aus dem Englischen. Zentraleinheit, wörtlich - central / main / main computing device) - central (micro); Gerät, das Maschinenbefehle ausführt; eine Hardware, die für die Durchführung von Rechenoperationen (vom Betriebssystem und der Anwendungssoftware vorgegeben) und die Koordination der Arbeit aller Geräte verantwortlich ist.

Grafikkarte(abgekürzt aus dem Englischen. Grafikverarbeitungseinheit, wörtlich - ein grafisches Rechengerät) - ein Grafikprozessor; ein separates Gerät oder eine Spielekonsole, die das Rendern von Grafiken (Visualisierung) durchführt. Moderne GPUs sind sehr effizient bei der realistischen Verarbeitung und Wiedergabe von Computergrafiken. Der Grafikprozessor in modernen Grafikkarten wird als 3D-Grafikbeschleuniger verwendet, kann aber in einigen Fällen auch für Berechnungen verwendet werden ( GPGPU).

Probleme Zentralprozessor

Lange Zeit war die Leistungssteigerung traditioneller hauptsächlich auf die sequentielle Erhöhung der Taktfrequenz (ca. 80% der Leistung wurde durch die Taktfrequenz bestimmt) bei gleichzeitiger Erhöhung der Anzahl der Transistoren auf einem Single zurückzuführen Chip. Eine weitere Erhöhung der Taktfrequenz (bei einer Taktfrequenz von mehr als 3,8 GHz überhitzen die Chips einfach!) widersteht jedoch einer Reihe grundlegender physikalischer Barrieren (da der technologische Prozess fast die Größe eines Atoms erreicht hat: . und die Größe eines Siliziumatoms beträgt ungefähr 0,543 nm):

Erstens steigt mit abnehmender Größe des Kristalls und mit zunehmender Taktfrequenz der Leckstrom von Transistoren. Dies führt zu einer Erhöhung des Stromverbrauchs und einer Erhöhung der Wärmeabgabe;

Zweitens werden die Vorteile höherer Taktgeschwindigkeiten teilweise durch Speicherzugriffslatenzen aufgehoben, da die Speicherzugriffszeiten nicht mit steigenden Taktgeschwindigkeiten übereinstimmen;

Drittens werden herkömmliche serielle Architekturen für einige Anwendungen mit steigenden Taktraten aufgrund des sogenannten „Von-Neumann-Engpasses“, eines Leistungsengpasses, der aus dem sequentiellen Rechenfluss resultiert, ineffizient. Gleichzeitig nehmen ohmsch-kapazitive Signalübertragungsverzögerungen zu, was ein zusätzlicher Flaschenhals ist, der mit einer Erhöhung der Taktfrequenz verbunden ist.

Entwicklung Grafikkarte

Parallel zur Entwicklung von Grafikkarte:

…

November 2008 - Intel eingeführt eine Linie von 4-Kern Intel-Core i7 basierend auf der Mikroarchitektur der nächsten Generation Nehalem. Prozessoren arbeiten mit einer Taktfrequenz von 2,6-3,2 GHz. Hergestellt in 45-nm-Prozesstechnologie.

Dezember 2008 – Quad-Core-Auslieferungen haben begonnen AMD Phenom II 940(Code Name - Deneb). Arbeitet mit einer Frequenz von 3 GHz, wird nach der 45-nm-Prozesstechnologie hergestellt.

…

Mai 2009 - Unternehmen AMD die GPU-Version eingeführt ATI Radeon HD4890 wobei der Kerntakt von 850 MHz auf 1 GHz erhöht wurde. Das ist das erste Grafik Prozessor läuft mit 1 GHz. Die Rechenleistung des Chips ist durch die Takterhöhung von 1,36 auf 1,6 Teraflops gewachsen. Der Prozessor enthält 800 (!) Kerne, unterstützt Videospeicher GDDR5, DirectX 10.1, ATI CrossFireX und alle anderen Technologien, die modernen Grafikkartenmodellen eigen sind. Der Chip wird auf Basis der 55-nm-Technologie hergestellt.

Hauptunterschiede Grafikkarte

Unterscheidungsmerkmale Grafikkarte(im Vergleich zu ) sind:

– eine Architektur, die maximal darauf abzielt, die Geschwindigkeit der Berechnung von Texturen und komplexen grafischen Objekten zu erhöhen;

ist die Spitzenleistung eines typischen Grafikkarte viel höher als ;

– dank einer dedizierten Pipeline-Architektur, Grafikkarte viel effizienter bei der Verarbeitung von grafischen Informationen als .

"Die Krise des Genres"

"Die Krise des Genres" für bis 2005 gereift - dann erschienen sie. Aber trotz der Entwicklung der Technologie, das Wachstum der konventionellen Produktivität stark zurückgegangen. Gleichzeitig Leistung Grafikkarte wächst weiter. Im Jahr 2003 kristallisierte sich diese revolutionäre Idee heraus - nutzen Sie die Rechenleistung der Grafik. GPUs werden aktiv für „nicht-grafisches“ Computing (Physiksimulation, Signalverarbeitung, Computermathematik/-geometrie, Datenbankoperationen, Computerbiologie, Computerökonomie, Computer Vision usw.) verwendet.

Das Hauptproblem war, dass es keine Standardschnittstelle für die Programmierung gab Grafikkarte. Die Entwickler verwendet OpenGL oder Direct3D aber es war sehr bequem. Konzern Nvidia(einer der größten Hersteller von Grafik-, Medien- und Kommunikationsprozessoren sowie drahtlosen Medienprozessoren; gegründet 1993) war an der Entwicklung eines einheitlichen und bequemen Standards beteiligt - und führte die Technologie ein KUDA.

Wie es begann

2006 - Nvidia demonstriert CUDA™; der Beginn einer Revolution in der Computertechnik Grafikkarte.

2007 - Nvidia veröffentlicht Architektur KUDA(Originalfassung CUDA-SDK wurde am 15. Februar 2007 präsentiert); Nominierung "Beste Neuheit" vom Magazin Populärwissenschaft und "Readers' Choice" aus der Veröffentlichung HPCWire.

2008 - Technologie Nvidia CUDA gewann in der Nominierung „Technical Excellence“ ab PC-Magazin.

Was KUDA

KUDA(abgekürzt aus dem Englischen. Compute Unified Device Architecture, wörtlich - eine einheitliche Computerarchitektur von Geräten) - eine Architektur (eine Reihe von Software und Hardware), mit der Sie produzieren können Grafikkarte Berechnungen für allgemeine Zwecke Grafikkarte fungiert tatsächlich als leistungsstarker Coprozessor.

Technologie NVIDIA CUDA™ ist die einzige Entwicklungsumgebung in einer Programmiersprache C, die es Entwicklern ermöglicht, dank der Rechenleistung von GPUs Software zur Lösung komplexer Rechenprobleme in kürzerer Zeit zu erstellen. Millionen arbeiten bereits auf der Welt Grafikkarte mit der Unterstützung KUDA, und Tausende von Programmierern verwenden bereits (kostenlose!) Tools KUDA um Anwendungen zu beschleunigen und die komplexesten ressourcenintensiven Aufgaben zu lösen - von der Video- und Audiokodierung bis zur Öl- und Gasexploration, Produktmodellierung, medizinischer Bildgebung und wissenschaftlicher Forschung.

KUDA gibt dem Entwickler die Möglichkeit, nach eigenem Ermessen den Zugriff auf den Befehlssatz des Grafikbeschleunigers zu organisieren und seinen Speicher zu verwalten, komplexe parallele Berechnungen darauf zu organisieren. Unterstützter Grafikbeschleuniger KUDA wird zu einer leistungsstarken programmierbaren offenen Architektur wie der heutigen . All dies bietet dem Entwickler einen Low-Level-, verteilten und High-Speed-Zugriff auf die Ausrüstung KUDA eine notwendige Grundlage für den Aufbau ernsthafter High-Level-Tools wie Compiler, Debugger, mathematische Bibliotheken, Softwareplattformen.

Uralsky, leitender Technologiespezialist Nvidia, vergleichen Grafikkarte und , sagt so: - Es ist ein Geländewagen. Er reist immer und überall, aber nicht sehr schnell. SONDERN Grafikkarte ist ein Sportwagen. Auf einer schlechten Straße wird er einfach nirgendwo hingehen, aber eine gute Abdeckung geben - und er wird seine ganze Geschwindigkeit zeigen, von der der SUV nie geträumt hat! ..».

Technologiefähigkeiten KUDA

Bei der Entwicklung moderner Prozessoren gibt es eine Tendenz zu einer schrittweisen Erhöhung der Anzahl von Kernen, was ihre Fähigkeiten im parallelen Rechnen erhöht. Allerdings gibt es längst GPUs, die den CPUs in dieser Hinsicht weit überlegen sind. Und diese Fähigkeiten von GPUs wurden bereits von einigen Unternehmen berücksichtigt. Seit Ende der 90er Jahre gibt es erste Versuche, Grafikbeschleuniger für nicht zielgerichtete Berechnungen einzusetzen. Aber erst das Erscheinen von Shadern wurde zum Anstoß für die Entwicklung einer völlig neuen Technologie, und 2003 erschien das Konzept der GPGPU (General-Purpose Graphics Processing Units). Eine wichtige Rolle bei der Entwicklung dieser Initiative spielte BrookGPU, eine spezielle Erweiterung für die Sprache C. Vor dem Aufkommen von BrookGPU konnten Programmierer nur über die Direct3D- oder OpenGL-APIs mit der GPU arbeiten. Brook ermöglichte es Entwicklern, mit der vertrauten Umgebung zu arbeiten, und der Compiler selbst implementierte mithilfe spezieller Bibliotheken die Interaktion mit der GPU auf niedrigem Niveau.

Solche Fortschritte konnten die Aufmerksamkeit der führenden Unternehmen dieser Branche auf sich ziehen - AMD und NVIDIA, die mit der Entwicklung ihrer eigenen Softwareplattformen für nicht-grafisches Computing auf ihren Grafikkarten begannen. Niemand kennt besser als GPU-Entwickler alle Nuancen und Funktionen ihrer Produkte, was es diesen Unternehmen ermöglicht, das Softwarepaket für bestimmte Hardwarelösungen so effizient wie möglich zu optimieren. Derzeit entwickelt NVIDIA die Plattform CUDA (Compute Unified Device Architecture), AMD nennt diese Technologie CTM (Close To Metal) oder AMD Stream Computing. Wir werden uns einige der CUDA-Features ansehen und die Rechenleistung des G92-Grafikchips der GeForce 8800 GT-Grafikkarte in der Praxis bewerten.

Aber lassen Sie uns zunächst einige der Nuancen der Durchführung von Berechnungen mit GPUs betrachten. Ihr Hauptvorteil liegt darin, dass der Grafikchip zunächst für die Ausführung vieler Threads ausgelegt ist und jeder Kern einer herkömmlichen CPU einen Strom sequentieller Anweisungen ausführt. Jede moderne GPU ist ein Multiprozessor, der aus mehreren Rechenclustern mit jeweils vielen ALUs besteht. Der leistungsstärkste moderne GT200-Chip besteht aus 10 solcher Cluster, von denen jeder 24 Stream-Prozessoren hat. Die getestete Grafikkarte GeForce 8800 GT auf Basis des G92-Chips verfügt über sieben große Recheneinheiten mit jeweils 16 Stream-Prozessoren. CPUs verwenden SIMD SSE-Blöcke für Vektorberechnungen (einzelne Anweisung, mehrere Daten – eine Anweisung wird für mehrere Daten ausgeführt), was eine Datentransformation in 4x-Vektoren erfordert. Die GPU verarbeitet Threads skalar, d.h. eine Anweisung wird über mehrere Threads hinweg angewendet (SIMT - Single Instruction Multiple Threads). Dies erspart Entwicklern das Konvertieren von Daten in Vektoren und ermöglicht willkürliche Verzweigungen in Streams. Jede GPU-Recheneinheit hat direkten Speicherzugriff. Und die Bandbreite des Videospeichers ist aufgrund der Verwendung mehrerer separater Speichercontroller (beim Top-G200 sind es 8 Kanäle mit jeweils 64 Bit) und hohen Arbeitsfrequenzen höher.

Im Allgemeinen sind GPUs bei bestimmten Aufgaben, wenn mit großen Datenmengen gearbeitet wird, viel schneller als CPUs. Unten sehen Sie eine Illustration dieser Aussage:

Das Diagramm zeigt die Entwicklung der CPU- und GPU-Leistung seit 2003. NVIDIA nennt diese Daten gerne als Werbung in seinen Dokumenten, aber sie sind nur eine theoretische Berechnung und tatsächlich kann die Lücke natürlich viel kleiner sein.

Aber wie dem auch sei, es gibt ein riesiges Potenzial an nutzbaren GPUs, das einen spezifischen Ansatz in der Softwareentwicklung erfordert. All dies ist in der CUDA-Hardware-Software-Umgebung implementiert, die aus mehreren Softwareschichten besteht – einer High-Level-CUDA-Laufzeit-API und einer Low-Level-CUDA-Treiber-API.

CUDA verwendet die Standardsprache C für die Programmierung, was einer der Hauptvorteile für Entwickler ist. CUDA enthält zunächst die Bibliotheken BLAS (Basic Linear Algebra) und FFT (Fourier Transform Calculation). CUDA hat auch die Fähigkeit, mit den OpenGL- oder DirectX-Grafik-APIs zu interagieren, die Fähigkeit, auf niedrigem Niveau zu entwickeln, und zeichnet sich durch eine optimierte Verteilung der Datenflüsse zwischen CPU und GPU aus. CUDA-Berechnungen werden gleichzeitig mit Grafikberechnungen durchgeführt, im Gegensatz zur ähnlichen AMD-Plattform, bei der im Allgemeinen eine spezielle virtuelle Maschine für GPU-Berechnungen gestartet wird. Aber auch ein solches „Zusammenleben“ ist fehlerbehaftet, wenn eine große Last durch die grafische API entsteht, während gleichzeitig CUDA läuft – schließlich haben grafische Operationen immer noch eine höhere Priorität. Die Plattform ist mit den 32-Bit- und 64-Bit-Betriebssystemen Windows XP, Windows Vista, MacOS X und verschiedenen Linux-Versionen kompatibel. Die Plattform ist offen und auf der Website können Sie neben speziellen Treibern für die Grafikkarte das CUDA Toolkit und die Softwarepakete CUDA Developer SDK herunterladen, die einen Compiler, Debugger, Standardbibliotheken und Dokumentation enthalten.

Was die praktische Umsetzung von CUDA betrifft, wurde diese Technologie lange Zeit nur für hochspezialisierte mathematische Berechnungen im Bereich der Elementarteilchenphysik, Astrophysik, Medizin oder Vorhersage von Veränderungen auf dem Finanzmarkt usw. verwendet. Aber diese Technologie nähert sich allmählich dem normalen Benutzer, insbesondere gibt es spezielle Plug-Ins für Photoshop, die die Rechenleistung der GPU nutzen können. Auf einer speziellen Seite können Sie die gesamte Liste der Programme durchsuchen, die die Fähigkeiten von NVIDIA CUDA nutzen.

Als Praxistest der neuen Technologie auf der Grafikkarte MSI NX8800GT-T2D256E-OC verwenden wir das Programm TMPGEnc. Dieses Produkt ist kommerziell (die Vollversion kostet 100 US-Dollar), aber für MSI-Grafikkarten gibt es es als Bonus in einer Testversion für 30 Tage. Sie können diese Version von der Entwicklerseite herunterladen, aber um TMPGEnc 4.0 XPress MSI Special Edition zu installieren, benötigen Sie die Originaldiskette mit Treibern von der MSI-Karte - das Programm kann ohne sie nicht installiert werden.

Um die umfassendsten Informationen über die Rechenleistung in CUDA anzuzeigen und mit anderen Grafikkarten zu vergleichen, können Sie das spezielle Dienstprogramm CUDA-Z verwenden. Hier sind die Informationen zu unserer GeForce 8800GT-Grafikkarte:

Im Vergleich zu den Referenzmodellen läuft unsere Instanz mit höheren Frequenzen: Die Rasterdomäne ist 63 MHz höher als der Nennwert, und die Shader-Einheiten sind 174 MHz schneller, der Speicher ist 100 MHz schneller.

Wir vergleichen die Konvertierungsgeschwindigkeit des gleichen HD-Videos bei Berechnung nur mit Hilfe der CPU und mit zusätzlicher Aktivierung von CUDA im Programm TMPGEnc bei folgender Konfiguration:

• Prozessor: Pentium Dual-Core E5200 2,5 GHz;
• Hauptplatine: Gigabyte P35-S3;
• Speicher: 2x1GB GoodRam PC6400 (5-5-5-18-2T)
• Grafikkarte: MSI NX8800GT-T2D256E-OC;
• Festplatte: 320 GB WD3200AAKS;
• Netzteil: CoolerMaster eXtreme Power 500-PCAP;
• Betriebssystem: Windows XP SP2;
• TMPGEnc 4.0 XPress 4.6.3.268;
• Grafikkartentreiber: ForceWare 180.60.

Für Tests wurde der Prozessor im ersten und zweiten Fall auf 3 GHz (in einer Konfiguration von 11,5 x 261 MHz) und bis zu 4 GHz (11,5 x 348 MHz) mit einer RAM-Frequenz von 835 MHz übertaktet. Video in Full HD-Auflösung von 1920 x 1080 für eine Minute und zwanzig Sekunden. Um eine zusätzliche Last zu erzeugen, wurde ein Rauschunterdrückungsfilter eingeschaltet, dessen Einstellungen standardmäßig belassen wurden.

Die Kodierung erfolgte mit dem Codec DivX 6.8.4. In den Qualitätseinstellungen dieses Codecs sind standardmäßig alle Werte belassen, Multithreading ist aktiviert.

Die Multithreading-Unterstützung in TMPGEnc ist zunächst auf der Registerkarte CPU/GPU-Einstellungen aktiviert. Im selben Abschnitt wird auch CUDA aktiviert.

Wie Sie im obigen Screenshot sehen können, ist die CUDA-Filterverarbeitung aktiviert und der Hardware-Videodecoder ist nicht aktiviert. Die Dokumentation des Programms warnt davor, dass die Aktivierung der letzten Option zu einer Verlängerung der Dateiverarbeitungszeit führt.

Basierend auf den Ergebnissen der Tests wurden die folgenden Daten erhalten:

Bei 4 GHz mit aktiviertem CUDA haben wir nur ein paar Sekunden (oder 2 %) gewonnen, was nicht besonders beeindruckend ist. Aber bei einer niedrigeren Frequenz können Sie durch die Aktivierung dieser Technologie etwa 13 % der Zeit einsparen, was sich bei der Verarbeitung großer Dateien deutlich bemerkbar macht. Dennoch sind die Ergebnisse nicht so beeindruckend wie erwartet.

Das Programm TMPGEnc verfügt über eine CPU- und CUDA-Lastanzeige, die in dieser Testkonfiguration die CPU-Last bei etwa 20 % und den Grafikkern bei den restlichen 80 % anzeigte. Als Ergebnis haben wir die gleichen 100% wie bei der Konvertierung ohne CUDA, und es gibt möglicherweise überhaupt keinen Zeitunterschied (der aber immer noch vorhanden ist). Auch der kleine Arbeitsspeicher von 256 MB ist kein limitierender Faktor. Den Messwerten von RivaTuner nach zu urteilen, wurden während des Betriebs nicht mehr als 154 MB Videospeicher verwendet.

Ergebnisse

Das TMPGEnc-Programm gehört zu denen, die die CUDA-Technologie der breiten Masse zugänglich machen. Die Verwendung der GPU in diesem Programm ermöglicht es Ihnen, den Videoverarbeitungsprozess zu beschleunigen und die CPU erheblich zu entlasten, wodurch der Benutzer bequem andere Aufgaben gleichzeitig erledigen kann. In unserem speziellen Beispiel hat die Grafikkarte GeForce 8800GT 256 MB die Zeitleistung beim Konvertieren von Videos basierend auf dem übertakteten Pentium Dual-Core E5200-Prozessor leicht verbessert. Aber es ist deutlich zu sehen, dass mit abnehmender Frequenz der Gewinn aus der CUDA-Aktivierung zunimmt; auf schwachen Prozessoren wird der Gewinn aus seiner Verwendung viel größer sein. Vor dem Hintergrund einer solchen Abhängigkeit ist es durchaus logisch anzunehmen, dass auch bei einer Erhöhung der Last (z. B. Verwendung sehr vieler zusätzlicher Videofilter) die Ergebnisse eines Systems mit CUDA durch a ausgezeichnet werden signifikanteres Delta des Unterschieds in der Zeit, die für den Codierungsprozess aufgewendet wird. Vergessen Sie auch nicht, dass der G92 derzeit nicht der leistungsstärkste Chip ist und modernere Grafikkarten in solchen Anwendungen eine deutlich höhere Leistung bieten. Während des Betriebs der Anwendung wird die GPU jedoch nicht vollständig ausgelastet, und wahrscheinlich hängt die Lastverteilung von jeder Konfiguration separat ab, nämlich von der Kombination aus Prozessor und Grafikkarte, was am Ende eine größere (oder kleinere) Steigerung ergeben kann in Prozent ab CUDA-Aktivierung. In jedem Fall wird diese Technologie für diejenigen, die mit großen Mengen an Videodaten arbeiten, immer noch erheblich Zeit sparen.

Zwar hat CUDA noch keine große Popularität erlangt, die Qualität der Software, die mit dieser Technologie arbeitet, muss verbessert werden. In dem von uns überprüften Programm TMPGEnc 4.0 XPress funktionierte diese Technologie nicht immer. Dasselbe Video konnte mehrmals neu kodiert werden, und dann war beim nächsten Start plötzlich die CUDA-Last schon 0%. Und dieses Phänomen war auf völlig unterschiedlichen Betriebssystemen völlig zufällig. Auch verweigerte das getestete Programm CUDA beim Enkodieren in das XviD-Format, mit dem beliebten DivX-Codec gab es aber keine Probleme.

Infolgedessen kann die CUDA-Technologie die Leistung von PCs nur bei bestimmten Aufgaben erheblich steigern. Aber der Umfang einer solchen Technologie wird sich erweitern, und der Prozess der Erhöhung der Anzahl von Kernen in konventionellen Prozessoren zeigt die wachsende Nachfrage nach parallelem Multithreaded-Computing in modernen Softwareanwendungen. Nicht umsonst haben alle Branchenführer in letzter Zeit die Idee in Brand gesetzt, CPU und GPU in einer einheitlichen Architektur zu vereinen (man erinnere sich zumindest an die beworbene AMD Fusion). Vielleicht ist CUDA eine der Phasen im Prozess dieser Vereinigung.

Wir danken folgenden Firmen für die Bereitstellung von Testequipment:

Und es wurde entwickelt, um den Hostcode (Haupt-, Steuercode) und den Gerätecode (Hardwarecode) (Dateien mit der Erweiterung .cu) in Objektdateien zu übersetzen, die zum Beispiel zum Erstellen des endgültigen Programms oder der Bibliothek in jeder Programmierumgebung geeignet sind , in NetBeans .

Die CUDA-Architektur verwendet das Grid-Speichermodell, Clustered-Thread-Modellierung und SIMD-Anweisungen. Anwendbar nicht nur für Hochleistungsgrafik-Computing, sondern auch für verschiedene wissenschaftliche Berechnungen mit nVidia-Grafikkarten. Wissenschaftler und Forscher nutzen CUDA ausgiebig in verschiedenen Bereichen, darunter Astrophysik, Computerbiologie und -chemie, Fluiddynamikmodellierung, elektromagnetische Wechselwirkungen, Computertomographie, seismische Analyse und mehr. CUDA kann eine Verbindung zu Anwendungen herstellen, die OpenGL und Direct3D verwenden. CUDA ist eine plattformübergreifende Software für Betriebssysteme wie Linux, Mac OS X und Windows.

Am 22. März 2010 veröffentlichte nVidia das CUDA Toolkit 3.0, das Unterstützung für OpenCL enthielt.

Ausrüstung

Die CUDA-Plattform kam erstmals mit der Veröffentlichung von NVIDIAs G80-Chip der achten Generation auf den Markt und war in allen nachfolgenden Serien von Grafikchips vorhanden, die in den Tesla-Beschleunigerfamilien GeForce, Quadro und NVidia verwendet wurden.

Die erste Hardwareserie, die das CUDA SDK unterstützte, das G8x, hatte einen 32-Bit-Vektorprozessor mit einfacher Genauigkeit, der das CUDA SDK als API verwendete (CUDA unterstützt den Double-Typ von C, aber jetzt wurde seine Genauigkeit auf 32-Bit-Floating herabgestuft Punkt). Spätere GT200-Prozessoren unterstützen 64-Bit-Präzision (nur für SFU), aber die Leistung ist erheblich schlechter als bei 32-Bit-Präzision (aufgrund von nur zwei SFUs pro Stream-Multiprozessor und acht Skalarprozessoren). Die GPU organisiert Hardware-Multithreading, wodurch Sie alle Ressourcen der GPU nutzen können. Damit eröffnet sich die Perspektive, die Funktionen eines physikalischen Beschleunigers auf einen Grafikbeschleuniger zu verlagern (Implementierungsbeispiel – nVidia PhysX). Es eröffnet auch breite Möglichkeiten für den Einsatz von Computergrafikgeräten zur Durchführung komplexer nichtgrafischer Berechnungen, beispielsweise in der Computerbiologie und anderen Wissenschaftszweigen.

Vorteile

Verglichen mit dem traditionellen Ansatz, Allzweck-Computing durch die Fähigkeiten von Grafik-APIs zu organisieren, hat die CUDA-Architektur in diesem Bereich die folgenden Vorteile:

Einschränkungen

• Alle vom Gerät ausführbaren Funktionen unterstützen keine Rekursion (in CUDA Toolkit 3.1 unterstützt es Zeiger und Rekursion) und haben einige andere Einschränkungen

Unterstützte GPUs und GPUs

Eine Liste von Geräten des Hardwareherstellers Nvidia mit erklärter voller Unterstützung der CUDA-Technologie findet sich auf der offiziellen Nvidia-Website: CUDA-Enabled GPU Products (englisch).

Tatsächlich unterstützen die folgenden Peripheriegeräte derzeit die CUDA-Technologie auf dem PC-Hardwaremarkt:

Spezifikationsversion	Grafikkarte	Grafikkarten
1.0	G80, G92, G92b, G94, G94b	GeForce 8800GTX/Ultra, 9400GT, 9600GT, 9800GT, Tesla C/D/S870, FX4/5600, 360M, GT 420
1.1	G86, G84, G98, G96, G96b, G94, G94b, G92, G92b	GeForce 8400GS/GT, 8600GT/GTS, 8800GT/GTS, 9600 GSO, 9800GTX/GX2, GTS250 /370M, 3/5/770M, 16/17/27/28/36/37/3800M, NVS420/50
1.2	GT218, GT216, GT215	GeForce 210, GT 220/40, FX380 LP, 1800M, 370/380M, NVS 2/3100M
1.3	GT200, GT200b	GeForce GTX 260, GTX 275, GTX 280, GTX 285, GTX 295, Tesla C/M1060, S1070, Quadro CX, FX 3/4/5800
2.0	GF100, GF110	GeForce (GF100) GTX 465, GTX 470, GTX 480, Tesla C2050, C2070, S/M2050/70, Quadro Plex 7000, Quadro 4000, 5000, 6000, GeForce (110) GTX GF560 TI 448, GTX570, GTX580, GTX590
2.1	GF104, GF114, GF116, GF108, GF106	GeForce 610M, GT 430, GT 440, GTS 450, GTX 460, GTX 550 Ti, GTX 560, GTX 560 Ti, 500M, Quadro 600, 2000
3.0	GK104, GK106, GK107	GeForce GTX 690, GTX 680, GTX 670, GTX 660 Ti, GTX 660, GTX 650 Ti, GTX 650, GT 640, GeForce GTX 680MX, GeForce GTX 680M, GeForce GTX 675MX, GeForce GTX 670MX, GTX 660M, GeForce GT 650M, GeForce GT 645M, GeForce GT 640M
3.5	GK110

Nvidia GeForce-Desktop Geforce GTX 590 Geforce GTX 580 Geforce GTX 570 GeForce GTX 560 Ti Geforce GTX 560 GeForce GTX 550 Ti Geforce GTX 520 Geforce GTX 480 Geforce GTX 470 Geforce GTX 465 Geforce GTX 460 GeForce GTS450 Geforce GTX 295 Geforce GTX 285 Geforce GTX 280 Geforce GTX 275 Geforce GTX 260 GeForce GTS250 Geforce GT 240 Geforce GT 220 Geforce 210 GeForce GTS150 Geforce GT 130 Geforce GT 120 GeForce G100 Geforce 9800 GX2 Geforce 9800 GTX+ Geforce 9800 GTX Geforce 9800GT GeForce 9600 GSO Geforce 9600GT Geforce 9500GT Geforce 9400GT GeForce 9400mGPU GeForce 9300mGPU GeForce 8800 GTS512 Geforce 8800GT Geforce 8600 GTS Geforce 8600GT Geforce 8500GT Geforce 8400GS

Nvidia GeForce für mobile Computer Geforce GTX 580M Geforce GTX 570M Geforce GTX 560M GeForce GT555M GeForce GT540M GeForce GT525M GeForce GT520M Geforce GTX 485M Geforce GTX 480M Geforce GTX 470M Geforce GTX 460M GeForce GT445M GeForce GT435M GeForce GT425M GeForce GT420M GeForce GT415M Geforce GTX 285M Geforce GTX 280M Geforce GTX 260M GeForce GTS 360M GeForce GTS350M GeForce GTS160M GeForce GTS150M GeForce GT335M GeForce GT330M GeForce GT325M GeForce GT240M Geforce GT 130M Geforce G210M Geforce G110M Geforce G105M Geforce 310M Geforce 305M GeForce 9800M GTX GeForce 9800MGT GeForce 9800M GTS GeForce 9700M GTS GeForce 9700MGT GeForce 9650M GS GeForce 9600MGT GeForce 9600M GS GeForce 9500M GS Geforce 9500MG GeForce 9300M GS Geforce 9300MG GeForce 9200M GS Geforce 9100MG GeForce 8800M GTS GeForce 8700MGT GeForce 8600MGT GeForce 8600M GS GeForce 8400MGT GeForce 8400M GS

Nvidia Tesla * Tesla C2050/C2070 Tesla M2050/M2070/M2090 Tesla S2050 Tesla S1070 Tesla M1060 Tesla C1060 Tesla C870 Tesla D870 Tesla S870

Nvidia Quadro-Desktop Quadro 6000 Quadro 5000 Quadro 4000 Quadro 2000 Quadro 600 QuadroFX5800 QuadroFX5600 QuadroFX4800 QuadroFX4700X2 QuadroFX4600 QuadroFX3700 QuadroFX 1700 QuadroFX570 QuadroFX470 Quadro FX 380 mit niedrigem Profil QuadroFX370 Quadro FX 370 mit niedrigem Profil Quadro CX Quadro NVS450 Quadro NVS420 Quadro NVS290 Quadro Plex 2100 D4 Quadro Plex 2200 D2 Quadro Plex 2100 S4 Quadro Plex 1000 Modell IV

Nvidia Quadro für mobile Computer Quadro 5010M Quadro5000M Quadro4000M Quadro3000M Quadro2000M Quadro 1000M QuadroFX3800M QuadroFX3700M QuadroFX3600M QuadroFX2800M QuadroFX2700M QuadroFX 1800M QuadroFX 1700M QuadroFX 1600M QuadroFX 880M QuadroFX770M QuadroFX570M QuadroFX380M QuadroFX370M QuadroFX 360M Quadro NVS5100M Quadro NVS4200M Quadro NVS3100M Quadro NVS2100M Quadro NVS320M Quadro NVS160M Quadro NVS150M Quadro NVS140M Quadro NVS135M Quadro NVS130M

• Die Modelle Tesla C1060, Tesla S1070, Tesla C2050/C2070, Tesla M2050/M2070, Tesla S2050 ermöglichen GPU-Berechnungen mit doppelter Genauigkeit.

Funktionen und Spezifikationen der verschiedenen Versionen

Funktionsunterstützung (nicht aufgeführte Funktionen sind für alle Rechenfunktionen unterstützt)	Rechenleistung (Version)
	1.0	1.1	1.2	1.3	2.x
32-Bit-Wörter im globalen Speicher	Nein	Ja
Gleitkommawerte im globalen Speicher
Ganzzahlige atomare Funktionen, die operieren 32-Bit-Wörter im gemeinsam genutzten Speicher	Nein		Ja
atomicExch() läuft auf 32-Bit Gleitkommawerte im gemeinsamen Speicher
Ganzzahlige atomare Funktionen, die operieren 64-Bit-Wörter im globalen Speicher
Warp-Abstimmungsfunktionen
Gleitkommaoperationen mit doppelter Genauigkeit	Nein			Ja
Atomare Funktionen, die auf 64-Bit ausgeführt werden ganzzahlige Werte im gemeinsamen Speicher	Nein				Ja
Fließkomma-Atomaddition, die auf arbeitet 32-Bit-Wörter im globalen und gemeinsam genutzten Speicher
_Abstimmung()
_threadfence_system()
_syncthreads_count(), _syncthreads_and(), _syncthreads_or()
Oberflächenfunktionen
3D-Gitter des Fadenblocks

technische Spezifikationen	Rechenleistung (Version)
	1.0	1.1	1.2	1.3	2.x
Maximale Abmessung des Rasters der Gewindeblöcke	2				3
Maximale x-, y- oder z-Dimension eines Gitters aus Fadenblöcken	65535
Maximale Abmessung des Gewindeblocks	3
Maximale x- oder y-Dimension eines Blocks	512				1024
Maximale z-Dimension eines Blocks	64
Maximale Anzahl von Threads pro Block	512				1024
Warp-Größe	32
Maximale Anzahl residenter Blöcke pro Multiprozessor	8
Maximale Anzahl residenter Warps pro Multiprozessor	24		32		48
Maximale Anzahl residenter Threads pro Multiprozessor	768		1024		1536
Anzahl der 32-Bit-Register pro Multiprozessor	8 TAUSEND		16K		32K
Maximale Menge an gemeinsam genutztem Speicher pro Multiprozessor	16 KB				48 KB
Anzahl der gemeinsam genutzten Speicherbänke	16				32
Menge des lokalen Arbeitsspeichers pro Thread	16 KB				512 KB
Konstante Speichergröße	64 KB
Cache-Arbeitssatz pro Multiprozessor für konstanten Speicher	8 KB
Cache-Arbeitssatz pro Multiprozessor für Texturspeicher	Geräteabhängig, zwischen 6 KB und 8 KB
Maximale Breite für 1D-Textur	8192				32768
Maximale Breite für 1D-Textur Referenz an linearen Speicher gebunden	2 27
Maximale Breite und Anzahl der Schichten für eine geschichtete 1D-Texturreferenz	8192x512				16384 x 2048
Maximale Breite und Höhe für 2D Texturreferenz gebunden an linearer Speicher oder ein CUDA-Array	65536 x 32768				65536 x 65535
Maximale Breite, Höhe und Anzahl von Ebenen für eine 2D-Ebenen-Texturreferenz	8192 x 8192 x 512				16384 x 16384 x 2048
Maximale Breite, Höhe und Tiefe für eine 3D-Texturreferenz, die an linear gebunden ist Arbeitsspeicher oder ein CUDA-Array	2048 x 2048 x 2048
Maximale Anzahl von Texturen, die kann an einen Kernel gebunden werden	128
Maximale Breite für eine 1D-Oberfläche Referenz an ein CUDA-Array gebunden	Nicht unterstützt				8192
Maximale Breite und Höhe für eine 2D Oberflächenreferenz, die an ein CUDA-Array gebunden ist					8192 x 8192
Maximale Anzahl von Oberflächen, die kann an einen Kernel gebunden werden					8
Maximale Anzahl von Anweisungen pro Kernel	2 Millionen

Beispiel

CudaArray* cu_array; Textur< float , 2 >Text; // Array zuweisen cudaMalloc( & cu_array, cudaCreateChannelDesc< float>() , Breite Höhe ) ; // Bilddaten in Array kopieren cudaMemcpy( cu_array, image, width* height, cudaMemcpyHostToDevice) ; // Das Array an die Textur binden cudaBindTexture( tex, cu_array) ; // Kernel ausführen dim3 blockDim(16 , 16 , 1 ) ; dim3 gridDim (Breite / BlockDim.x , Höhe / BlockDim.y , 1 ) ; Kernel<<< gridDim, blockDim, 0 >>> (d_odata, Breite, Höhe) ; cudaUnbindTexture(tex) ; __global__ void kernel(float * odata, int height, int width) ( unsigned int x = blockIdx.x * blockDim.x + threadIdx.x ; unsigned int y = blockIdx.y * blockDim.y + threadIdx.y ; float c = texfetch(tex, x, y) ; odata[ y* breite+ x] = c; )

pycuda.driver als drv importieren import numpy drv.init() dev = drv.Device(0) ctx = dev.make_context() mod = drv.SourceModule( """ __global__ void multiply_them(float *dest, float *a, float *b) ( const int i = threadIdx.x; dest[i] = a[i] * b[i]; ) """) multiply_them = mod.get_function ("multiply_them" ) a = numpy.random .randn (400 ) .astype (numpy.float32 ) b = numpy.random .randn (400 ) .astype (numpy.float32 ) dest = numpy.zeros_like (a) multiply_them( drv.Out (dest) , drv.In (a) , drv.In (b) , block= (400 , 1 , 1 ) ) print dest-a*b

CUDA als Fach an Universitäten

Seit Dezember 2009 wird das CUDA-Programmiermodell an 269 Universitäten weltweit gelehrt. In Russland werden Schulungen zu CUDA an der St. Petersburg Polytechnic University, Yaroslavl State University, durchgeführt. P. G. Demidov, Moskau, Nischni Nowgorod, St. Petersburg, Twer, Kasan, Nowosibirsk, Staatliche Technische Universität Nowosibirsk Staatliche Universitäten Omsk und Perm, Internationale Universität für die Natur der Gesellschaft und des Menschen "Dubna", Staatliche Universität für Energietechnik Iwanowo, Staatliche Universität Belgorod , MSTU ihnen. Bauman, RKhTU im. Mendeleev, Interregional Supercomputer Center of the Russian Academy of Sciences, . Darüber hinaus wurde im Dezember 2009 die Eröffnung des ersten russischen Wissenschafts- und Bildungszentrums „Parallel Computing“ in der Stadt Dubna angekündigt, dessen Aufgaben Schulungen und Beratung zur Lösung komplexer Rechenprobleme auf der GPU umfassen.

In der Ukraine werden Kurse zu CUDA am Kiev Institute for System Analysis unterrichtet.

Verknüpfungen

Offizielle Ressourcen

• CUDA-Zone (Russisch) - offizielle CUDA-Website
• CUDA GPU Computing (Englisch) – offizielle Webforen für CUDA-Computing

Inoffizielle Ressourcen

Toms Hardware

• Dmitri Tschkanow. nVidia CUDA: Grafikkarten-Computing oder CPU-Tod? . Toms Hardware (22. Juni 2008). Archiviert
• Dmitri Tschkanow. nVidia CUDA: GPU-Anwendungsbenchmarks für den Massenmarkt. Tom's Hardware (19. Mai 2009). Archiviert vom Original am 4. März 2012. Abgerufen am 19. Mai 2009.

iXBT.com

• Alexey Berillo. NVIDIA CUDA – Nicht-Grafik-Computing auf GPUs. Teil 1 . iXBT.com (23. September 2008). Archiviert vom Original am 4. März 2012. Abgerufen am 20. Januar 2009.
• Alexey Berillo. NVIDIA CUDA – Nicht-Grafik-Computing auf GPUs. Teil 2 . iXBT.com (22. Oktober 2008). - Beispiele für die Implementierung von NVIDIA CUDA. Archiviert vom Original am 4. März 2012. Abgerufen am 20. Januar 2009.

Andere Ressourcen

• Boreskov Alexey Viktorovich. Grundlagen von CUDA (20. Januar 2009). Archiviert vom Original am 4. März 2012. Abgerufen am 20. Januar 2009.
• Wladimir Frolow. Eine Einführung in die CUDA-Technologie. Online-Journal "Computergrafik und Multimedia" (19. Dezember 2008). Archiviert vom Original am 4. März 2012. Abgerufen am 28. Oktober 2009.
• Igor Oskolkow. NVIDIA CUDA ist eine erschwingliche Eintrittskarte für Big Computing. Computerra (30. April 2009). Abgerufen am 3. Mai 2009.
• Wladimir Frolow. Eine Einführung in die CUDA-Technologie (1. August 2009). Archiviert vom Original am 4. März 2012. Abgerufen am 3. April 2010.
• GPGPU.ru. Verwenden von Grafikkarten für Computer
• . Zentrum für paralleles Rechnen

Anmerkungen

siehe auch

NVIDIA
Grafik Prozessoren	Frühzeitig NV1 NV2 RIVA-Familie TNT TNT2 GeForce-Familie GeForce 256 GeForce 2 GeForce 3 GeForce 4 GeForce FX GeForce 6 GeForce 7 GeForce 8 GeForce 9 GeForce 100 GeForce 200 GeForce 400 GeForce 500 GeForce 600