Basic GEMM

Learn how to implement a basic General Matrix Multiply (GEMM) operation using CUTLASS

Basic GEMM Example

This example demonstrates how to call a CUTLASS GEMM kernel and provides a naive reference matrix multiply kernel to verify its correctness.

Overview

The CUTLASS GEMM template computes the general matrix product (GEMM) using single-precision floating-point arithmetic:

D = alpha * (A x B) + beta * C

This example assumes all matrices have column-major layout and uses a threadblock tile size of 128x128x8, which offers good performance for large matrices.

Key Concepts

• GEMM kernel instantiation: Defining and launching a CUTLASS GEMM kernel
• Template parameters: Configuring data types and matrix layouts
• Argument objects: Passing parameters to CUTLASS kernels
• Reference validation: Verifying correctness against a naive implementation

Implementation

Building and Running

Build the example

cd /path/to/cutlass
mkdir build && cd build
cmake .. -DCUTLASS_NVCC_ARCHS='75;80;86'
make 00_basic_gemm

Run the example

# Run with default dimensions (128x128x128)
./examples/00_basic_gemm/00_basic_gemm

# Run with custom dimensions M=512, N=256, K=128
./examples/00_basic_gemm/00_basic_gemm 512 256 128

# Run with custom dimensions and scalars (alpha=2.0, beta=1.0)
./examples/00_basic_gemm/00_basic_gemm 512 256 128 2.0 1.0

Source Code Location

The complete source code for this example is available at:

• examples/00_basic_gemm/basic_gemm.cu

What This Example Demonstrates

1. Minimal CUTLASS usage: This example deliberately uses minimal CUTLASS components to show the simplest path to getting started
2. Template-based kernel instantiation: How to define a GEMM kernel using CUTLASS templates
3. Argument construction: The pattern for passing arguments to CUTLASS kernels
4. Correctness verification: How to validate CUTLASS output against a reference implementation

Key Takeaways

• CUTLASS provides simplified abstractions for high-performance GEMM operations
• The cutlass::gemm::device::Gemm template handles kernel instantiation
• Argument objects provide a structured way to pass parameters to kernels
• CUTLASS defaults (e.g., 128x128x8 tile size) are chosen for good general performance

Next Steps

• Explore Batched GEMM for processing multiple independent matrix multiplications
• Learn about Fused Operations to combine GEMM with element-wise operations
• Check out Convolution for convolutional neural network operations