Why NVIDIA’s First Data‑Center GPU Revolutionized Computing: Inside the Tesla G80 Architecture

Hello, I'm Fei! We all know NVIDIA started as a graphics card maker, but recent trends like Bitcoin mining and large AI models pushed it into general‑purpose GPU (GPGPU) computing, creating a company worth over $4.3 trillion.

We’ll explore NVIDIA’s first data‑center GPU, the Tesla series, and its original CUDA 1.0 toolkit to answer questions such as why NVIDIA shifted from gaming GPUs, what the unified shader architecture is, which hardware modules reside in a streaming multiprocessor (SM), how constant memory boosts performance, and how CUDA enables C/C++ GPU programming.

Why did NVIDIA transition from gaming GPUs to scientific computing?

What is the unified shader architecture in Tesla and why is it important?

What hardware modules are inside an SM?

How does proper use of constant memory dramatically improve GPU performance?

How does CUDA let developers program GPUs with C/C++?

Let’s begin the GPU learning journey!

1. NVIDIA’s Gaming Card Dilemma

In 2004 the graphics market was slowing down; AMD’s Radeon 9700 Pro even overtook NVIDIA in some areas. Scientists faced costly, low‑performance CPU clusters and began repurposing GPU graphics APIs and shader languages for tasks like image segmentation, CT reconstruction, FFT, and video codecs.

Seeing this, NVIDIA realized GPUs could serve a broader compute market. In 2006 it launched the Tesla GPU series and the CUDA programming suite, marking the shift from graphics to general‑purpose computing.

2. Tesla Architecture Overview

The first Tesla used a Unified Shader Architecture that merged vertex, pixel, and geometry shaders into programmable streaming processors (SPs). Although designed for graphics, these SPs could also execute non‑graphics compute instructions, laying the foundation for GPGPU.

Later Tesla‑named GPUs (e.g., Tesla P100, V100, V4) kept the “Tesla” branding for the architecture, but NVIDIA later renamed them “Data Center GPUs” to avoid confusion with the car brand.

Key GPU concepts:

Architecture name : denotes a generation (e.g., Tesla, Maxwell, Ampere, Hopper).

Core name : specific implementation within an architecture (e.g., GA102, AD104).

Product name : market‑facing model (e.g., GeForce 8800 Ultra, RTX 4090).

2.1 Tesla G80 Architecture

The G80 core introduced 128 CUDA cores. The architecture diagram shows 8 TPCs (Texture/Process Clusters) and 8 ROPs (Render Output Pipelines).

2.2 Inside a TPC

Each TPC contains a Geometry Controller, an SMC (Streaming Multiprocessor Controller), two SMs, and a Texture unit.

Geometry Controller : handles vertex processing, sharing instruction cache and registers with the SMs.

SMC : schedules threads, distributes thread blocks to the two SMs, and issues instructions at warp granularity.

SM (Streaming Multiprocessor) : core of parallel computation; see next section.

2.3 SM Cache Units

I cache : 16 KB instruction cache, reducing fetch latency.

C cache : 8 KB constant cache for fast access to __constant__ data.

Shared memory : 16 KB on‑chip memory for intra‑block communication.

__constant__ float coeff[1024]; // constant memory declaration

__global__ void kernel(float* data) {
    int idx = threadIdx.x + blockIdx.x * blockDim.x;
    data[idx] *= coeff[idx % 1024]; // use constant cache
}

2.4 SM Execution Units

Each SM contains 8 SPs (CUDA cores) and two SFUs (Special Function Units). An SP can perform two FP32 MAD operations per clock; SFUs can execute up to four transcendental operations per clock.

3. GeForce 8800 Ultra Performance

Peak FP32 FLOPS are calculated as:

GPU FLOPS = Shader Clock × SM Count × FP32 ops per SM per cycle

With a 1512 MHz shader clock, 128 SMs, and 20 FP32 ops per SM per cycle, the GeForce 8800 Ultra reaches ~387 GFLOPS, roughly 30× the theoretical FP32 performance of a 2007 Intel Core 2 Duo (≈10.6 GFLOPS).

4. CUDA Compute Unified Device Architecture

CUDA 1.0 (2007) introduced a C‑language programming model for GPUs, providing three main capabilities:

GPU memory management (cudaMalloc, cudaMemcpy, cudaFree).

Kernel function definition (using __global__).

GPU abstraction (threads, thread blocks, grids).

4.1 Simple CUDA Example

#include <stdio.h>
#include <cuda_runtime.h>

__global__ void vectorAdd(const float *a, const float *b, float *c, int n) {
    int idx = threadIdx.x + blockIdx.x * blockDim.x;
    if (idx < n) c[idx] = a[idx] + b[idx];
}

int main() {
    // allocate and initialize host memory ...
    // allocate device memory with cudaMalloc ...
    // copy data to device with cudaMemcpy ...
    int threadsPerBlock = 256;
    int blocksPerGrid = (n + threadsPerBlock - 1) / threadsPerBlock;
    vectorAdd<<<blocksPerGrid, threadsPerBlock>>>(d_a, d_b, d_c, n);
    cudaMemcpy(h_c, d_c, size, cudaMemcpyDeviceToHost);
}

Key steps:

GPU memory allocation and data transfer.

Define kernel and launch with grid‑block configuration.

Copy results back to host.

In CUDA, a thread maps to a streaming processor (SP), a block maps to an SM, and a grid maps to the whole GPU.

CUDA Software Abstraction

Corresponding Hardware

Demo Parameter

Thread

Streaming Processor (SP)

threadIdx.x

Thread Block

Streaming Multiprocessor (SM)

threadsPerBlock = 256

Grid

Entire GPU

blocksPerGrid = ceil(n/256)

Early CUDA versions only supported FP32 and required manual memory management, but they opened the door to GPU‑accelerated scientific computing and later AI workloads.

Conclusion

The Tesla architecture’s unified shader design and the launch of CUDA 1.0 gave NVIDIA a decisive lead in GPGPU, turning a graphics‑only company into a dominant compute platform that now powers AI, HPC, and many other fields.