What is FLOP (Arithmetic operation on floating-point numbers)

FLOP (Floating-Point Operation) is a single computational operation of addition, subtraction, multiplication or division on fractional numbers. This unit measures not the speed of programs, but the volume of mathematical work that a processor or video card is capable of performing.

The FLOP metric underpins high-performance computing, machine learning and scientific modeling. It is used to evaluate the performance of supercomputers (TOP500 ranking), the required time for training large neural networks such as LLMs, as well as GPU capabilities when rendering graphics or calculating physics in real time. Engineers operate with gigaflops and teraflops to determine the suitability of hardware for resource-intensive tasks.

A straightforward pursuit of peak FLOP values is often misleading. Real performance hits a bottleneck in memory bandwidth and data transfer latencies, causing the processor to idle. Rounding errors and loss of precision occur when adding numbers of different orders. Different architectures (CPU vs GPU) achieve the same theoretical number of operations but differ dramatically in efficiency on specific algorithms.

How FLOP works

Unlike integer operations, a FLOP is a multi-stage pipeline working with the components of a number: sign, exponent (order) and mantissa. For example, when adding two floating-point numbers, the device first aligns their exponents, shifting the mantissa of the smaller number to the right to bring it to a common scale. Then the integer adder sums the mantissas, after which the result is normalized — shifted so that the leading bit of the mantissa becomes one, with the corresponding adjustment of the exponent. Multiplication, on the contrary, does not require order alignment: mantissas are multiplied as integers, and exponents are added, after which the result is rounded according to the mode set by the IEEE 754 standard. This is fundamentally more complex than the bitwise shifts and logical operations used in integer calculations. Modern processors execute this in hardware within FPU modules, and video cards on thousands of cores simultaneously. The terms MIPS (millions of instructions per second) and FLOPS are often confused, but MIPS reflects the total number of any machine commands, including data loads and branches, whereas FLOPS accounts only for useful mathematics, abstracting away from processor architecture. The distinction between FLOP (unit of work) and FLOPS (speed, operations per second) is analogous to the difference between a joule and a watt: the first is the volume of computation, the second is the rate of execution.

MIPS (Simplified pipelined RISC architecture without interlocks)FPU (Hardware acceleration of floating point computations)

FLOP functionality

Nature of a FLOP operation. A FLOP represents a single arithmetic operation on floating-point numbers in IEEE 754 format. Basic operations include addition, subtraction, multiplication, and division. Transcendental functions such as sine or exponent typically require several FLOPs, as they are implemented through sequences of polynomial approximations and table lookups.
Addition and subtraction as a FLOP. The addition operation includes alignment of mantissa orders, the actual addition, and normalization of the result. The FPU hardware handles denormalized numbers and overflow. Subtraction of numbers close in magnitude leads to catastrophic loss of significance, requiring additional cycles from the processor to restore precision.
Multiplication as a FLOP. Multiplication is performed by multiplying mantissas as integers, followed by adding orders and normalization. Hardware multipliers in modern cores are deeply pipelined. The FMA (Fused Multiply-Add) operation combines multiplication and addition into a single indivisible FLOP with a single rounding, which critically reduces error accumulation in calculations.
Division and square root extraction. Division is historically implemented by iterative SRT algorithms, requiring many times more cycles than multiplication. Square root extraction uses similar mechanisms of initial approximation and refinement by the Newton-Raphson method. In high-performance rankings, these operations are counted with a weight exceeding a single FLOP.
Scalar integer analogues. In the context of performance, the concept of FLOP cannot be substituted with integer operations (OPS). Integer multiplication and division have deterministic precision and different latency. When analyzing scientific calculations based on real arithmetic, integer array addressing logic is not accounted for in the total FLOP pool.
Scalar (Converting a multidimensional tensor into a single number)
Precision as a FLOP classifier. Distinctions are made between FP64 (double precision), FP32 (single precision), FP16 and BF16. Vector extensions like AVX-512 process different volumes of operations depending on the bit width of the data type. Performance in TFLOPS for neural network accelerators is often specified for mixed-precision formats, where FP16 multiplication is combined with FP32 accumulation.
AVX-512 (Processing 16 numbers per instruction)TFLOPS (Computational performance meter)Vector (Ordered storage of numbers in continuous memory)FP16 (Compact representation of Floating-Point numbers)BF16 (Floating-point number format for machine learning)
Theoretical peak performance. Processor clock frequency multiplied by the number of cores and the width of the SIMD block gives the theoretical maximum FLOP per second. Achieving this peak on real code is impossible without an ideal ratio of operations to bytes of data transfer due to the Memory Wall effect.
FLOPS metric for systems. Aggregated performance is measured in FLOPS (operations per second). Exaflop supercomputers operate with quintillions of operations. When building the Top500 ranking, the execution of the HPL (High-Performance Linpack) test is measured, solving a dense matrix system of linear equations, where the FLOP share is strictly determined by the task dimension.
Arithmetic intensity of code. A fundamental metric is the ratio of the total number of executed FLOPs to the volume of data moved in bytes. If the intensity is below the machine balance value (the ratio of peak performance to memory bandwidth), the application is memory-bound, and the processor idles waiting for data.
FLOP in algorithm complexity evaluation. For matrix multiplication of geometric dimension N, the complexity is O(N³) FLOP. Fast Fourier Transform reduces complexity from O(N²) to O(N log N). Estimating workload in FLOPs enables platform-independent comparison of algorithms without running benchmarks on specific hardware.
Software code instrumentation. For precise operation counting, binary instrumentation frameworks intercepting FPU instructions are used. Analyzers such as Intel SDE or Likwid allow simulation of execution on various microarchitectures, outputting the exact number of performed FLOPs and verifying the vectorization of critical loops.
Code vectorization by the compiler. Automatic vectorization of scalar loops turns sequential FLOPs into parallel SIMD instructions. The compiler analyzes data dependencies and memory alignment. Pragmas such as #pragma omp simd forcibly inform the compiler about the absence of pointer overlaps to multiply the density of FLOPs per cycle.
The role of FMA in counting. A single FMA instruction formally executes two FLOPs per cycle (multiplication and addition). The marketing performance of accelerators is often doubled by counting FMA as two separate operations. The Linpack test allows achieving near-theoretical peak precisely due to the dominance of FMA instructions in the DGEMM assembly kernel.
Energy efficiency of operations. Measuring FLOP per watt has become a key metric in the race for exaflops. Mobile GPUs and NPUs achieve extreme efficiency on low-precision calculations. Each joule of energy is spent not only on the operation itself but also on moving data across the cache hierarchy, so bringing computation closer to memory (Processing-in-Memory) increases the useful FLOP output.
Specialized number formats. Modern AI accelerators operate with non-standard floating-point blocks. The FP8 format (E4M3 and E5M2) doubles the peak FLOP rate relative to FP16. BF16 numbers preserve the dynamic range of FP32 by truncating the mantissa, which ensures identical exponent dimensions and simplifies conversion in mixed calculations.
Sparse operations (Sparse FLOP). Hardware support for structured sparsity (2:4) in tensor cores allows ignoring zero weights. FLOP performance for sparse matrices doubles by skipping computations with guaranteed zeros, but the real gain heavily depends on the distribution pattern of non-zero elements in the tensor.
Asynchronous rounding errors. The non-determinism of parallel FLOP reduction arises from the non-associativity of floating-point operations. Changing the summation order of vectors in multi-threaded implementations yields a different bit picture of the result. The FLOP metric does not fix the fact of precision loss but merely confirms the fact of instruction execution.
Software performance counters. The PMU subsystem of modern processors contains the retired fp_arith_inst and simd_fp_retired counters. Reading these MSR registers via the perf interface allows obtaining the exact number of executed scalar and vector FLOPs without injecting into the application source code for profiling purposes.
Emulation of high-precision FLOPs. Software implementation of quadruple (FP128) or arbitrary precision emulates a single high-precision FLOP through dozens of hardware ones. Libraries like MPFR split the mantissa into several machine words. Peak useful performance falls exponentially, so this approach is used only to verify reference results.
Quantization and integer approximation. To reduce the number of energy-consuming FLOPs, quantization of neural networks into INT8 and INT4 is applied. The multiply-accumulate operation is replaced by integer MAC. Hardware blocks lacking an FPU use FLOP emulation via soft-float libraries, which is orders of magnitude slower than native integer arithmetic.
The concept of FLOP/s of bytecode. Virtual machines executing bytecode add an interpretation layer that does not create useful FLOPs from the point of view of applied mathematics. Evaluating virtual machine efficiency requires distinguishing system operations supporting execution from target instructions working with real data. Computational physics operates exclusively with useful operations on problem elements.
The impact of speculation and branching. Predication in GPU cores and branch prediction in CPUs affect the final rate of FLOP execution. A pipeline flush on misprediction nullifies the results of speculatively executed but unconfirmed floating-point operations. Hardware counters record only successfully completed (retired) instructions, excluding junk FLOPs executed on the wrong path.

Comparisons

FLOP vs FLOPS. FLOP is a unit of measurement for the volume of computation, equivalent to a single floating-point operation, whereas FLOPS is a derived performance metric denoting the number of such operations per second. The fundamental difference lies in the fact that FLOP records completed work, while FLOPS characterizes the data processing speed of a computing system in real time.
FLOP vs MAC. In processor microarchitecture, a multiply-accumulate (MAC) operation combines multiplication and addition in a single instruction, while an elementary FLOP counts these actions separately. Modern tensor cores and neuroprocessors traditionally measure performance in MAC, however, to unify benchmark metrics, one MAC cycle is converted into two FLOPs.
FLOP vs OPS. Floating-point operations are a subset of the broader OPS class, encompassing any computational instructions, including integer arithmetic and logical shifts. In heterogeneous computing, the OPS metric is used for INT8 or INT4 quantization, demonstrating significantly larger absolute values, while FLOP is critically important for high-precision FP64 scientific simulations.
FLOP vs FLOP/s per watt. If the absolute number of FLOPs measures the computational complexity of an algorithm without resource binding, then the FLOP/s per watt metric characterizes the energy efficiency of the architecture. This ratio is fundamental for data centers, where direct operation counting gives way to analyzing the total cost of equipment ownership converted into useful work.
FLOP vs Synaptic Operations (SynOps). In neuromorphic chips, standard floating-point arithmetic is replaced by spike events measured in SynOps. Unlike deterministic FLOP, SynOps reflects the event-driven accumulation of charge by the neuron membrane, making direct comparison incorrect due to a fundamentally different, impulse-based paradigm of sparse information transmission.

OS and driver support

The operating system interacts with floating-point operations by saving and restoring the FPU (Floating-Point Unit) context during thread switching using FXSAVE/FXRSTOR or XSAVE/XRSTOR instructions; kernel-mode drivers are required to wrap the use of XMM/YMM registers in KeSaveFloatingPointState/KeRestoreFloatingPointState; and user applications receive an isolated context automatically, while the exposure of hardware capabilities to virtualized environments is realized through the disclosure of CPUID flags for AVX, SSE, FMA by the hypervisor and the guest additions driver.

AVX (SIMD data processing with 256/512-bit registers)SSE (Simultaneous processing of multiple data with a single instruction)

Security

Vulnerabilities at the FLOP level arise through side channels in speculative execution of the FPU block, leakage of register state during lazy context switching, and attacks on floating-point algorithms; countermeasures include clearing AVX registers with a hardware scrubber with a microarchitectural barrier, forced eager switching of the FPU state by turning off the TS bit in CR0, and implementing constant-time algorithms that exclude branching based on mantissa values.

Logging

Registration of floating-point operations is performed by intercepting the #NM (Device Not Available) exception upon the first touch of the FPU by a thread; logging is carried out by the driver via WMI tracing with fixation of precision control word and rounding values; and detailed tracing of computation results is achieved by setting a processor breakpoint on writing to the .data area using DR0–DR3 debug registers and a debug trap handler.

Limitations

FLOP implementation is limited by the deterministic error of the IEEE 754 standard; subnormal numbers cause microcode assistance with a tenfold performance drop; mixing AVX-512 instructions with legacy SSE leads to a frequency penalty under the AVX-Software License; and execution in kernel mode prohibits using the upper 256 bits of ZMM registers due to the risk of data loss during interrupts without explicit opmask register saving.

History and development

Evolution began with the x87 coprocessor and its register stack model, then Intel SSE introduced flat XMM registers and scalar/packed operations, the AVX revolution added three-operand non-destructive syntax and 256-bit vectorization, and modern AVX-512 and AVX10 instruction sets implement eightfold precision through fused multiply-add, operation masking, and a scalable architecture that unifies server and client extensions into a single convergent standard with automatic power management.