Lesson 52 of 73
Quantized Kernels, the Vector API, and Threads
A language model produces text one token at a time. For each token, the program takes a vector of numbers that describes the current state (the activations) and multiplies it by every weight matrix in the model. This operation is called a matrix-vector product. It is built from dot products: a dot product multiplies two lists of numbers pair by pair and adds the results. For example, the dot product of [1, 2, 3] and [4, 5, 6] is 1 times 4 plus 2 times 5 plus 3 times 6, which is 32. A matrix-vector product is one dot product per matrix row. These products take almost all of the time of one forward pass from The Forward Pass in Java, so the speed of the whole engine is the speed of this one operation.
The small function that runs this operation in the innermost loop is called a kernel. This lesson writes kernels for three weight formats: F32 (plain 4-byte floats), Q8_0 (one byte per weight plus a scale for every 32 weights), and Q4_0 (half a byte per weight plus a scale for every 32 weights). It starts with simple scalar kernels that read quantized blocks from a MemorySegment. It then rewrites them with the Java Vector API, which lets one CPU instruction process several numbers at once. After that it splits the rows of a matrix across threads and measures how close the result comes to the memory bandwidth limit.
Every fast kernel in the lesson is compared with a simple scalar reference, because a fast kernel that returns wrong numbers is worse than a slow one. Afterwards, a reader can write a SIMD kernel in Java, can say when its result is bit-identical to a plain loop and when only a tolerance applies, and can tell whether a kernel is limited by memory or by the instructions it executes.
| Idea | What it means | Examples |
|---|---|---|
| Quantized dot products | add up integer products inside each block, then multiply once by both FP16 scales | 1, 2 |
| Species and lanes | a species fixes the vector size; FloatVector.SPECIES_256, fromArray, mul, add, reduceLanes |
3 |
| Tails | what to do with the last few elements: loopBound, masks from indexInRange |
4 |
| FMA and accumulators | fma, and several independent partial sums to keep the CPU busy |
5, 6 |
| Widening | turn bytes into ints or floats before multiplying: convertShape(B2I, ...), castShape |
7, 8 |
| Nibbles | split a byte into two 4-bit values: and, lanewise(LSHR, 4) |
9 |
| Off-heap loads | read vectors straight from a memory-mapped file: ByteVector.fromMemorySegment |
10 |
| Threads | 64-row chunks handed out by an atomic counter | 11 |
| Project 19 | a Kernels class tested against scalar references |
12 |
| Speed and profiling | effective bandwidth, jfr view hot-methods, allocation per call |
13, 14, 15 |
What the lesson builds on
The block layouts come from Building Q8_0 and Q4_0 Like ggml. In both formats, 32 weights share one scale d that is stored as a 16-bit float (FP16). The diagrams show where each byte sits.
Q8_0 block: 34 bytes, 32 values
bytes 0-1 : scale d (FP16)
bytes 2-33 : 32 signed bytes q[0] ... q[31]
value i = q[i] * d
Q4_0 block: 18 bytes, 32 values
bytes 0-1 : scale d (FP16)
bytes 2-17 : 16 bytes; byte j holds two 4-bit codes
low nibble = code for value j
high nibble = code for value j + 16
value i = (code - 8) * dA nibble is half a byte, 4 bits. Arena, MemorySegment, and little-endian ValueLayout constants come from Loading GGUF with the FFM API. Briefly, a MemorySegment is a Java handle to a range of memory outside the Java heap (for example a memory-mapped file), and an Arena decides when that memory is released. SIMD (single instruction, multiple data) means one CPU instruction works on several numbers at once. The bandwidth measurements of the reference machine, and the reason decoding is limited by memory bandwidth, come from CPUs, GPUs, and Memory for Inference and FLOPs, Bandwidth, and the Roofline Model.
Examples 3 to 15 use the Vector API, which is an incubator module in JDK 25. An incubator module is a feature that ships with the JDK but is not final yet, so it must be switched on. Programs compile and run with --add-modules jdk.incubator.vector, and the JVM prints WARNING: Using incubator modules: jdk.incubator.vector to standard error at startup. That warning is expected.
Example 1: scalar Q8_0 dot product on a segment
The first question is what a quantized dot product reads and in what order. The key idea is that the scale can be pulled out of the sum. Suppose a block of three weights has codes [2, -1, 3] and scale 0.5, and the activations have codes [4, 5, -2] and scale 0.1. The real values are [1, -0.5, 1.5] and [0.4, 0.5, -0.2], and their dot product is 0.4 - 0.25 - 0.3 = -0.15. The same result comes from the integer codes alone:
sumi = 2*4 + (-1)*5 + 3*(-2) = 8 - 5 - 6 = -3
result = sumi * (0.5 * 0.1) = -3 * 0.05 = -0.15So a block costs many cheap integer multiplications and one float multiplication. The program quantizes 64 seeded weights and 64 activations into two Q8_0 segments of two blocks each. It then computes the dot product the way ggml_vec_dot_q8_0_q8_0_generic in ggml/src/ggml-cpu/quants.c does (ggml is the C library behind llama.cpp): per block, an integer sum of 32 byte products, then sumf += sumi * (dw * dx). The FP16 scale is read with a little-endian unaligned short layout, which works at any byte offset, so the kernel does not depend on where a block starts. Look at the printed sumi, dw, and dx of each block, and at how the three final numbers compare.
import java.lang.foreign.Arena;
import java.lang.foreign.MemorySegment;
import java.lang.foreign.ValueLayout;
import java.nio.ByteOrder;
import java.util.Locale;
import java.util.Random;
public class ScalarQ8Dot {
static final ValueLayout.OfShort F16 = ValueLayout.JAVA_SHORT_UNALIGNED.withOrder(ByteOrder.LITTLE_ENDIAN);
static final int QK = 32, Q8_BYTES = 34;
static int roundf(float v) { return (int) Math.copySign(Math.floor(Math.abs((double) v) + 0.5), v); }
static void quantizeQ8_0(float[] x, MemorySegment out) { // Lesson 18, written to a segment
for (int b = 0; b < x.length / QK; b++) {
long o = (long) b * Q8_BYTES;
float amax = 0;
for (int j = 0; j < QK; j++) amax = Math.max(amax, Math.abs(x[b * QK + j]));
float d = amax / 127f, id = d != 0 ? 1f / d : 0f;
out.set(F16, o, Float.floatToFloat16(d));
for (int j = 0; j < QK; j++) out.set(ValueLayout.JAVA_BYTE, o + 2 + j, (byte) roundf(x[b * QK + j] * id));
}
}
static float dotQ8_0(MemorySegment w, MemorySegment x, int n) { // ggml_vec_dot_q8_0_q8_0_generic
float sumf = 0;
for (int b = 0; b < n / QK; b++) {
long o = (long) b * Q8_BYTES;
int sumi = 0;
for (int j = 0; j < QK; j++)
sumi += w.get(ValueLayout.JAVA_BYTE, o + 2 + j) * x.get(ValueLayout.JAVA_BYTE, o + 2 + j);
float dw = Float.float16ToFloat(w.get(F16, o)), dx = Float.float16ToFloat(x.get(F16, o));
System.out.printf(Locale.ROOT, "block %d: sumi = %6d, dw = %.6f, dx = %.6f, adds %.6f%n", b, sumi, dw, dx, sumi * (dw * dx));
sumf += sumi * (dw * dx);
}
return sumf;
}
public static void main(String[] args) {
int n = 64;
Random rnd = new Random(49);
float[] w = new float[n], x = new float[n];
for (int i = 0; i < n; i++) { w[i] = (float) rnd.nextGaussian() * 0.1f; x[i] = (float) rnd.nextGaussian(); }
try (Arena arena = Arena.ofConfined()) {
MemorySegment ws = arena.allocate((long) n / QK * Q8_BYTES), xs = arena.allocate((long) n / QK * Q8_BYTES);
quantizeQ8_0(w, ws);
quantizeQ8_0(x, xs);
System.out.println("segment bytes = " + ws.byteSize() + " for " + n + " values");
float q = dotQ8_0(ws, xs, n);
double deq = 0, exact = 0;
for (int i = 0; i < n; i++) {
long o = (long) (i / QK) * Q8_BYTES + 2 + i % QK, s = (long) (i / QK) * Q8_BYTES;
deq += (double) ws.get(ValueLayout.JAVA_BYTE, o) * Float.float16ToFloat(ws.get(F16, s))
* xs.get(ValueLayout.JAVA_BYTE, o) * Float.float16ToFloat(xs.get(F16, s));
exact += (double) w[i] * x[i];
}
System.out.printf(Locale.ROOT, "Q8_0 kernel = %.6f, dequantized in double = %.6f, exact = %.6f%n", q, deq, exact);
}
}
}segment bytes = 68 for 64 values
block 0: sumi = -19010, dw = 0.001895, dx = 0.024811, adds -0.893760
block 1: sumi = -14543, dw = 0.001422, dx = 0.022369, adds -0.462579
Q8_0 kernel = -1.356338, dequantized in double = -1.356338, exact = -1.354703The first line shows the size: 64 values in Q8_0 take two blocks of 34 bytes, 68 bytes in total, against 256 bytes in F32.
Each block line shows one block's work. sumi is the sum of 32 products of signed bytes. It is an int, and it cannot overflow: the largest possible magnitude is 32 times 127 times 127, which is 516,128, far below the int limit of about 2.1 billion. dw and dx are the two scales read back from FP16. The last number is that block's contribution, sumi * (dw * dx). For block 0 that is -19,010 times 0.001895 times 0.024811, which gives -0.893760. The weights are about 10 times smaller than the activations (they were drawn with a standard deviation of 0.1, the activations with 1), so dw is about 10 times smaller than dx. A scale is the largest magnitude in the block divided by 127, so smaller values mean a smaller scale.
The last line has three numbers:
- The kernel result, -1.356338, is the sum of the two block contributions (-0.893760 and -0.462579).
- The second number dequantizes every value (code times scale) and multiplies in
double. It agrees with the kernel to all six printed decimals, which shows that the kernel computes exactly the dot product of the dequantized vectors. - The third number is the exact dot product of the original floats. The gap of about 0.0016 between the kernel and this value is the quantization error of both vectors, not an error of the kernel.
Most of the work is integer arithmetic: 64 byte multiplications, and only two float multiply-adds, one per block.
Example 2: Q4_0 by Q8_0 scalar dot product
Weights are stored in the file, so they are quantized once, in advance, to save space. Activations are computed fresh for every token, so the engine quantizes them at run time. Q4_0 weights are multiplied with Q8_0 activations (not Q4_0 activations), because the CPU backend's type table in ggml-cpu.c sets vec_dot_type = GGML_TYPE_Q8_0 for Q4_0. The vec_dot_type is the format that the activation vector is converted to before the dot product. Eight bits keep more precision for the activations at a small cost.
The program writes the same seeded weights as Q4_0 (18-byte blocks) and the activations as Q8_0. It runs the loop of ggml_vec_dot_q4_0_q8_0_generic, which keeps two integer sums per block: sumi0 for the low nibbles against activations 0 to 15, and sumi1 for the high nibbles against activations 16 to 31. Compare the output with Example 1: the same data, a smaller weight format, and a different error.
import java.lang.foreign.Arena;
import java.lang.foreign.MemorySegment;
import java.lang.foreign.ValueLayout;
import java.nio.ByteOrder;
import java.util.Locale;
import java.util.Random;
public class ScalarQ4Dot {
static final ValueLayout.OfShort F16 = ValueLayout.JAVA_SHORT_UNALIGNED.withOrder(ByteOrder.LITTLE_ENDIAN);
static final int QK = 32, Q8_BYTES = 34, Q4_BYTES = 18;
static int roundf(float v) { return (int) Math.copySign(Math.floor(Math.abs((double) v) + 0.5), v); }
static void quantizeQ8_0(float[] x, MemorySegment out) {
for (int b = 0; b < x.length / QK; b++) {
long o = (long) b * Q8_BYTES;
float amax = 0;
for (int j = 0; j < QK; j++) amax = Math.max(amax, Math.abs(x[b * QK + j]));
float d = amax / 127f, id = d != 0 ? 1f / d : 0f;
out.set(F16, o, Float.floatToFloat16(d));
for (int j = 0; j < QK; j++) out.set(ValueLayout.JAVA_BYTE, o + 2 + j, (byte) roundf(x[b * QK + j] * id));
}
}
static void quantizeQ4_0(float[] x, MemorySegment out) { // quantize_row_q4_0_ref
for (int b = 0; b < x.length / QK; b++) {
long o = (long) b * Q4_BYTES;
float amax = 0, max = 0;
for (int j = 0; j < QK; j++) { float v = x[b * QK + j]; if (amax < Math.abs(v)) { amax = Math.abs(v); max = v; } }
float d = max / -8f, id = d != 0 ? 1f / d : 0f;
out.set(F16, o, Float.floatToFloat16(d));
for (int j = 0; j < 16; j++) {
int lo = Math.min(15, (int) (x[b * QK + j] * id + 8.5f)), hi = Math.min(15, (int) (x[b * QK + j + 16] * id + 8.5f));
out.set(ValueLayout.JAVA_BYTE, o + 2 + j, (byte) (lo | (hi << 4)));
}
}
}
static float dotQ4_0Q8_0(MemorySegment w, MemorySegment x, int n) { // ggml_vec_dot_q4_0_q8_0_generic
float sumf = 0;
for (int b = 0; b < n / QK; b++) {
long ow = (long) b * Q4_BYTES, ox = (long) b * Q8_BYTES;
int sumi0 = 0, sumi1 = 0;
for (int j = 0; j < 16; j++) {
int q = w.get(ValueLayout.JAVA_BYTE, ow + 2 + j) & 0xFF;
sumi0 += ((q & 0x0F) - 8) * x.get(ValueLayout.JAVA_BYTE, ox + 2 + j);
sumi1 += ((q >> 4) - 8) * x.get(ValueLayout.JAVA_BYTE, ox + 2 + j + 16);
}
float dw = Float.float16ToFloat(w.get(F16, ow)), dx = Float.float16ToFloat(x.get(F16, ox));
System.out.printf(Locale.ROOT, "block %d: sumi0 = %5d, sumi1 = %5d, dw = %+.6f, dx = %.6f%n", b, sumi0, sumi1, dw, dx);
sumf += (sumi0 + sumi1) * dw * dx;
}
return sumf;
}
public static void main(String[] args) {
int n = 64;
Random rnd = new Random(49);
float[] w = new float[n], x = new float[n];
for (int i = 0; i < n; i++) { w[i] = (float) rnd.nextGaussian() * 0.1f; x[i] = (float) rnd.nextGaussian(); }
try (Arena arena = Arena.ofConfined()) {
MemorySegment ws = arena.allocate((long) n / QK * Q4_BYTES), xs = arena.allocate((long) n / QK * Q8_BYTES);
quantizeQ4_0(w, ws);
quantizeQ8_0(x, xs);
System.out.println("weight bytes = " + ws.byteSize() + ", activation bytes = " + xs.byteSize());
float q = dotQ4_0Q8_0(ws, xs, n);
double deq = 0, exact = 0;
for (int i = 0; i < n; i++) {
long bw = (long) (i / QK) * Q4_BYTES, bx = (long) (i / QK) * Q8_BYTES;
int packed = ws.get(ValueLayout.JAVA_BYTE, bw + 2 + i % 16) & 0xFF;
int code = (i % QK < 16 ? packed & 0x0F : packed >> 4) - 8;
deq += (double) code * Float.float16ToFloat(ws.get(F16, bw))
* xs.get(ValueLayout.JAVA_BYTE, bx + 2 + i % QK) * Float.float16ToFloat(xs.get(F16, bx));
exact += (double) w[i] * x[i];
}
System.out.printf(Locale.ROOT, "kernel = %.6f, dequantized in double = %.6f, exact = %.6f%n", q, deq, exact);
}
}
}weight bytes = 36, activation bytes = 68
block 0: sumi0 = 537, sumi1 = 475, dw = -0.030075, dx = 0.024811
block 1: sumi0 = -9, sumi1 = 732, dw = -0.022568, dx = 0.022369
kernel = -1.120130, dequantized in double = -1.120130, exact = -1.354703The weight segment is 36 bytes for 64 values, about half of the 68 bytes of the Q8_0 activations. Each packed byte is read with & 0xFF, so that q >> 4 sees an unsigned value from 0 to 255. Java bytes are signed, and Java widens a byte to an int with its sign. Without the mask, a byte such as 0xF3 (243 as an unsigned number) would become the int -13, and the shift -13 >> 4 would give -1 instead of 15, the real high nibble.
The block lines print the two integer sums separately. In block 0, the low nibbles contribute 537 and the high nibbles 475. The scale dw is negative in both blocks. That is the Q4_0 rule from the earlier lesson: d = max / -8, where max is the signed value with the largest magnitude. A block whose largest value is positive therefore gets a negative scale. The code 15 then stands for (15 - 8) times a negative scale, which is a negative number, and the sign of the largest value is restored through the scale. The multiplication (sumi0 + sumi1) * dw * dx handles the sign without any special case.
The kernel and the double-precision dequantized dot product agree again, -1.120130. The exact float result is -1.354703, so the error is much larger than in Example 1. Four-bit weights have only 16 levels per block instead of 255. With only 64 values there is also no averaging over thousands of blocks, which would make the error small relative to the total. The kernel itself is exact; the error comes from the format.
The evaluation order also follows ggml. Q4_0 multiplies sumi * dw * dx from left to right, while Q8_0 multiplies sumi * (dw * dx). The two orders can differ in the last bit of a float, because float multiplication is not associative (changing the grouping can change the rounding). A Java port that is meant to match ggml bit for bit keeps each order.
Example 3: Vector API basics
The scalar loops above handle one number per step. A modern CPU can load, multiply, and add 8 floats (or more) in one instruction. The Vector API (package jdk.incubator.vector) lets Java code ask for these SIMD operations directly, and the JIT compiler turns them into the matching CPU instructions.
Four terms are needed:
- A species is a fixed combination of element type and vector size.
FloatVector.SPECIES_256is a vector offloatvalues that is 256 bits wide. - A lane is one slot of a vector. A float takes 32 bits, so 256 bits hold 8 float lanes.
fromArrayloads 8 consecutive floats from an array into one vector.mulandaddwork lane by lane: lane 0 with lane 0, lane 1 with lane 1, and so on.reduceLanes(VectorOperators.ADD)adds all lanes of one vector into a single number.
The program computes a 16-element dot product in two vector steps and prints every intermediate vector. Read the output as a table of 8 columns, one per lane.
import jdk.incubator.vector.FloatVector;
import jdk.incubator.vector.VectorOperators;
import jdk.incubator.vector.VectorSpecies;
public class VectorBasics {
static final VectorSpecies<Float> S = FloatVector.SPECIES_256;
public static void main(String[] args) {
System.out.println("species " + S + ": " + S.length() + " lanes of " + S.elementSize() + " bits");
float[] a = new float[16], b = new float[16];
for (int i = 0; i < 16; i++) { a[i] = i + 1; b[i] = (i % 4) - 1; }
FloatVector acc = FloatVector.zero(S);
for (int i = 0; i < a.length; i += S.length()) {
FloatVector va = FloatVector.fromArray(S, a, i);
FloatVector vb = FloatVector.fromArray(S, b, i);
FloatVector p = va.mul(vb);
acc = acc.add(p);
System.out.println("i = " + i);
System.out.println(" a " + va);
System.out.println(" b " + vb);
System.out.println(" a * b " + p);
System.out.println(" acc " + acc);
}
float vector = acc.reduceLanes(VectorOperators.ADD);
float scalar = 0;
for (int i = 0; i < a.length; i++) scalar += a[i] * b[i];
System.out.println("reduceLanes(ADD) = " + vector + ", scalar loop = " + scalar);
}
}species Species[float, 8, S_256_BIT]: 8 lanes of 32 bits
i = 0
a [1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0]
b [-1.0, 0.0, 1.0, 2.0, -1.0, 0.0, 1.0, 2.0]
a * b [-1.0, 0.0, 3.0, 8.0, -5.0, 0.0, 7.0, 16.0]
acc [-1.0, 0.0, 3.0, 8.0, -5.0, 0.0, 7.0, 16.0]
i = 8
a [9.0, 10.0, 11.0, 12.0, 13.0, 14.0, 15.0, 16.0]
b [-1.0, 0.0, 1.0, 2.0, -1.0, 0.0, 1.0, 2.0]
a * b [-9.0, 0.0, 11.0, 24.0, -13.0, 0.0, 15.0, 32.0]
acc [-10.0, 0.0, 14.0, 32.0, -18.0, 0.0, 22.0, 48.0]
reduceLanes(ADD) = 88.0, scalar loop = 88.0The first line confirms the species: 8 lanes of 32 bits. On this machine FloatVector.SPECIES_PREFERRED, the largest species the CPU supports well, is 512 bits with 16 float lanes (the CPU has AVX-512, an x86 instruction set with 512-bit vectors). On a CPU with AVX2 only, it is 256 bits. The examples in this lesson use SPECIES_256 explicitly, so that every lane count, every intermediate vector, and every printed number is the same on any machine. A kernel for production picks SPECIES_PREFERRED and accepts that the order of float additions then depends on the hardware.
At i = 0 the program loads a[0..7] and b[0..7] and multiplies them lane by lane. It adds the products to an accumulator that starts at zero, so the accumulator equals the products. At i = 8 it loads the next 8 values of each array. The lane-wise products are added to the accumulator. Lane 0 now holds 1 * -1 + 9 * -1 = -10, and lane 3 holds 4 * 2 + 12 * 2 = 32. Each lane is an independent partial sum over the elements whose index has the same remainder modulo 8.
lane: 0 1 2 3 4 5 6 7
step i = 0: -1 0 3 8 -5 0 7 16
step i = 8: -9 0 11 24 -13 0 15 32
acc after: -10 0 14 32 -18 0 22 48 sum = 88reduceLanes(ADD) adds the 8 lanes: -10 + 0 + 14 + 32 - 18 + 0 + 22 + 48 = 88, the same as the scalar loop. The two results agree exactly here because every value is a small integer, and float arithmetic on small integers has no rounding. With arbitrary floats, the vector version adds in a different order (eight partial sums, then a final sum), so the results can differ in the last bits. Example 6 measures that difference.
Example 4: the tail, loopBound and masks
A vector step always takes 8 elements, but most array lengths are not a multiple of 8. A length of 19 gives two full steps (16 elements) and 3 elements left over. The part left over after the last full vector is called the tail. There are two ways to handle it.
- Scalar tail.
S.loopBound(n)returns the largest multiple of the lane count that is not greater thann. The vector loop runs up to that bound, and an ordinary scalar loop finishes the remaining elements. - Mask. A mask is a vector of booleans that turns lanes on or off.
S.indexInRange(i, n)sets lanekwheni + k < n. A maskedfromArrayloads only the lanes that are on and fills the others with zero, so a final vector step can run on a partial group.
The program computes a 19-element dot product both ways and prints the bound and the mask.
import jdk.incubator.vector.FloatVector;
import jdk.incubator.vector.VectorMask;
import jdk.incubator.vector.VectorOperators;
import jdk.incubator.vector.VectorSpecies;
public class VectorTail {
static final VectorSpecies<Float> S = FloatVector.SPECIES_256;
static float tailLoop(float[] a, float[] b) {
int upper = S.loopBound(a.length);
FloatVector acc = FloatVector.zero(S);
int i = 0;
for (; i < upper; i += S.length()) acc = acc.add(FloatVector.fromArray(S, a, i).mul(FloatVector.fromArray(S, b, i)));
float sum = acc.reduceLanes(VectorOperators.ADD);
System.out.println(" loopBound(" + a.length + ") = " + upper + ", scalar tail covers " + (a.length - upper) + " values");
for (; i < a.length; i++) sum += a[i] * b[i];
return sum;
}
static float masked(float[] a, float[] b) {
FloatVector acc = FloatVector.zero(S);
for (int i = 0; i < a.length; i += S.length()) {
VectorMask<Float> m = S.indexInRange(i, a.length);
if (i + S.length() > a.length) System.out.println(" mask at i = " + i + ": " + m);
acc = acc.add(FloatVector.fromArray(S, a, i, m).mul(FloatVector.fromArray(S, b, i, m)));
}
return acc.reduceLanes(VectorOperators.ADD);
}
public static void main(String[] args) {
int n = 19;
float[] a = new float[n], b = new float[n];
for (int i = 0; i < n; i++) { a[i] = i + 1; b[i] = (i % 3) - 1; }
float scalar = 0;
for (int i = 0; i < n; i++) scalar += a[i] * b[i];
System.out.println("scalar loop: " + scalar);
System.out.println("vector + scalar tail:");
System.out.println(" result " + tailLoop(a, b));
System.out.println("masked last step:");
System.out.println(" result " + masked(a, b));
}
}scalar loop: -7.0
vector + scalar tail:
loopBound(19) = 16, scalar tail covers 3 values
result -7.0
masked last step:
mask at i = 16: Mask[TTT.....]
result -7.0The scalar loop gives -7.0, the reference.
In the first method, loopBound(19) is 16: two full vectors of 8 lanes cover elements 0 to 15, and the scalar loop adds elements 16, 17, and 18. The program prints that the tail covers 3 values, and the result is -7.0.
In the second method, every step is a vector step. At i = 16 only three elements remain, and the mask prints as Mask[TTT.....]: lanes 0 to 2 are on (T), lanes 3 to 7 are off (.). The masked loads read a[16..18] and b[16..18] and put zeros in the other lanes, so the extra lanes add nothing to the sum. A masked load never reads past the end of the array. An unmasked fromArray at i = 16 would throw IndexOutOfBoundsException, because it needs 8 elements starting at index 16 and the array has only 3.
Both methods agree with the scalar loop. The scalar tail keeps every step of the main loop identical, and the masked form avoids a second loop. Project 19 uses the scalar tail. Quantized kernels have no tail at all: GGUF requires the row length of Q8_0 and Q4_0 tensors to be a multiple of 32, and the vector code works on whole blocks.
Example 5: FMA and several accumulators
Two ideas make a vector dot-product loop fast. The first is the fused multiply-add (FMA). It computes a * b + c as one operation with a single rounding at the end, while mul followed by add rounds twice (once after the multiplication and once after the addition). In the Vector API it is a.fma(b, c), lane by lane. On CPUs with an FMA instruction it is also one instruction instead of two.
The second idea is to use several accumulators, which are the vectors that hold the running partial sums. The program has two parts. The first part picks values where the difference between FMA and mul plus add is visible. The second part computes the same dot product with one accumulator and with four, and shows that both give the same value on exact data. The reason for four accumulators is explained after the output.
import jdk.incubator.vector.FloatVector;
import jdk.incubator.vector.VectorOperators;
import jdk.incubator.vector.VectorSpecies;
public class FmaAccumulators {
static final VectorSpecies<Float> S = FloatVector.SPECIES_256;
static float oneAccumulator(float[] a, float[] b) {
FloatVector acc = FloatVector.zero(S);
for (int i = 0; i < a.length; i += S.length())
acc = FloatVector.fromArray(S, a, i).fma(FloatVector.fromArray(S, b, i), acc);
return acc.reduceLanes(VectorOperators.ADD);
}
static float fourAccumulators(float[] a, float[] b) { // a.length must be a multiple of 4 * lanes
int L = S.length();
FloatVector s0 = FloatVector.zero(S), s1 = s0, s2 = s0, s3 = s0;
for (int i = 0; i < a.length; i += 4 * L) {
s0 = FloatVector.fromArray(S, a, i).fma(FloatVector.fromArray(S, b, i), s0);
s1 = FloatVector.fromArray(S, a, i + L).fma(FloatVector.fromArray(S, b, i + L), s1);
s2 = FloatVector.fromArray(S, a, i + 2 * L).fma(FloatVector.fromArray(S, b, i + 2 * L), s2);
s3 = FloatVector.fromArray(S, a, i + 3 * L).fma(FloatVector.fromArray(S, b, i + 3 * L), s3);
}
return s0.add(s1).add(s2.add(s3)).reduceLanes(VectorOperators.ADD);
}
public static void main(String[] args) {
float e = 1f + 0x1p-23f; // 1.0000001, the float after 1
FloatVector va = FloatVector.broadcast(S, e), vc = FloatVector.broadcast(S, -(1f + 0x1p-22f));
System.out.println("a = " + e + ", c = " + -(1f + 0x1p-22f));
System.out.println("mul then add, lane 0: " + va.mul(va).add(vc).lane(0));
System.out.println("fma, lane 0: " + va.fma(va, vc).lane(0));
System.out.println("exact a*a + c: " + 0x1p-46);
int n = 128;
float[] a = new float[n], b = new float[n];
for (int i = 0; i < n; i++) { a[i] = (i % 7) - 3; b[i] = (i % 5) - 1; }
float scalar = 0;
for (int i = 0; i < n; i++) scalar += a[i] * b[i];
System.out.println("dot of " + n + " small integers: scalar " + scalar + ", one accumulator " + oneAccumulator(a, b)
+ ", four accumulators " + fourAccumulators(a, b));
}
}a = 1.0000001, c = -1.0000002
mul then add, lane 0: 0.0
fma, lane 0: 1.4210855E-14
exact a*a + c: 1.4210854715202004E-14
dot of 128 small integers: scalar -3.0, one accumulator -3.0, four accumulators -3.0The first four lines show what one rounding means. a is 1.0000001, the float right after 1 (that is, 1 + 2^-23). Its exact square is 1 + 2^-22 + 2^-46. The float closest to that square is 1 + 2^-22, because 2^-46 is far below the float precision at 1 (floats near 1 can only represent steps of 2^-23). c is minus that rounded value, -1.0000002.
With mul then add, the product rounds to 1 + 2^-22 first, and adding c gives exactly 0.0. The tiny 2^-46 was thrown away by the first rounding. With fma, the product is kept exactly inside the operation, c is added, and only the final result, 2^-46, is rounded. The program prints 1.4210855E-14, and the exact value 2^-46 printed in double is 1.4210854715202004E-14. FMA is more accurate, and on x86 CPUs with the FMA extension (this machine lists fma among its CPU flags) it is one instruction instead of two. Java's FMA is defined to round once on every platform, so it gives the same bits everywhere.
The last line compares one accumulator with four. All three methods give -3.0, because the data is small integers. The reason to use four accumulators is speed, not accuracy. With one accumulator, every FMA needs the result of the previous FMA, so the loop is a chain of dependent instructions. An FMA takes several clock cycles to finish, and the CPU waits for each one before it starts the next. A core can start new FMAs much more often than one FMA takes to finish, so most of its capacity sits idle. With four accumulators, four chains are independent, and the CPU can have several FMAs in progress at the same time.
one accumulator: acc -> fma -> fma -> fma -> fma ... (each waits for the one before)
four accumulators: s0 -> fma -> fma ...
s1 -> fma -> fma ... (four chains overlap)
s2 -> fma -> fma ...
s3 -> fma -> fma ...The four partial vectors are added at the end with s0.add(s1).add(s2.add(s3)). The lesson on CPUs earlier in the course showed the same idea with scalar code; Example 13 shows how much it matters in the vector kernels.
Example 6: vectorized F32 dot product against scalar
With real data, the vector dot product and the scalar loop add in different orders, so their results differ slightly. The question is by how much, and how to test a vector kernel when exact equality is not expected.
The program computes dot products of 4,096 and 4,100 seeded Gaussian values (random numbers shaped like a bell curve) in three ways: with a scalar float loop, with a vector kernel that uses FMA and four accumulators, and in double as the reference. It reports each float error relative to the sum of absolute products. Watch two things in the output: neither float result equals the reference, and neither method is always the closer one.
import java.util.Locale;
import java.util.Random;
import jdk.incubator.vector.FloatVector;
import jdk.incubator.vector.VectorSpecies;
public class DotF32 {
static final VectorSpecies<Float> S = FloatVector.SPECIES_256;
static float scalar(float[] a, float[] b) {
float s = 0;
for (int i = 0; i < a.length; i++) s += a[i] * b[i];
return s;
}
static float vector(float[] a, float[] b) {
int L = S.length(), i = 0, upper = a.length - a.length % (4 * L);
FloatVector s0 = FloatVector.zero(S), s1 = s0, s2 = s0, s3 = s0;
for (; i < upper; i += 4 * L) {
s0 = FloatVector.fromArray(S, a, i).fma(FloatVector.fromArray(S, b, i), s0);
s1 = FloatVector.fromArray(S, a, i + L).fma(FloatVector.fromArray(S, b, i + L), s1);
s2 = FloatVector.fromArray(S, a, i + 2 * L).fma(FloatVector.fromArray(S, b, i + 2 * L), s2);
s3 = FloatVector.fromArray(S, a, i + 3 * L).fma(FloatVector.fromArray(S, b, i + 3 * L), s3);
}
float[] lanes = s0.add(s1).add(s2.add(s3)).toArray();
float sum = 0;
for (float v : lanes) sum += v; // fixed order, same bits on every machine
for (; i < a.length; i++) sum += a[i] * b[i];
return sum;
}
public static void main(String[] args) {
Random rnd = new Random(6);
for (int n : new int[] {4096, 4100}) {
float[] a = new float[n], b = new float[n];
for (int i = 0; i < n; i++) { a[i] = (float) rnd.nextGaussian(); b[i] = (float) rnd.nextGaussian(); }
double exact = 0, scale = 0;
for (int i = 0; i < n; i++) { exact += (double) a[i] * b[i]; scale += Math.abs((double) a[i] * b[i]); }
float s = scalar(a, b), v = vector(a, b);
System.out.printf(Locale.ROOT, "n = %d: double %.6f, scalar float %.6f, vector float %.6f%n", n, exact, s, v);
System.out.printf(Locale.ROOT, " error / sum|a*b|: scalar %.2e, vector %.2e, within 1e-6: %b%n",
Math.abs(s - exact) / scale, Math.abs(v - exact) / scale, Math.abs(v - exact) / scale < 1e-6);
}
}
}n = 4096: double 20.848300, scalar float 20.848284, vector float 20.848335
error / sum|a*b|: scalar 6.32e-09, vector 1.36e-08, within 1e-6: true
n = 4100: double -13.482397, scalar float -13.482470, vector float -13.482407
error / sum|a*b|: scalar 2.78e-08, vector 3.78e-09, within 1e-6: trueFor n = 4096, the double reference is 20.848300, the scalar float loop gives 20.848284, and the vector kernel gives 20.848335. Neither float result equals the reference, and they do not equal each other. Both errors are tiny: about 6 and 14 parts per billion of the sum of the absolute products. The vector kernel is not more wrong than the scalar loop as a rule. For n = 4100 the order is reversed: the scalar loop's error is 2.78e-08 and the vector kernel's is 3.78e-09. Which one lands closer to the exact value depends on the data.
The error is divided by the sum of |a[i] * b[i]|, not by the result. A dot product of random vectors can be close to zero even when its terms are large (positive and negative terms cancel). Dividing by a near-zero result would make any rounding look like a huge relative error. The sum of absolute products is the natural scale of the rounding errors, and a test with a tolerance of 1e-6 against it passes for any correct float kernel at these lengths.
The n = 4100 case exercises the tail: the main loop covers 4,096 elements in steps of 32 (four vectors of 8), and the scalar loop adds the last 4.
The kernel adds the final 8 lanes with toArray() and a plain loop instead of reduceLanes(ADD). The Vector API documentation states that a floating-point ADD reduction may use any order of additions, and the order may vary between runs and machines. A fixed loop over the lanes gives the same bits on every machine, which keeps the printed output of this example stable. Integer reductions have no such issue, because integer addition gives the same result in any order. The next example uses this.
Example 7: vectorized Q8_0 dot product with integer lanes
The Q8_0 dot product multiplies bytes as integers. A product of two signed bytes can be as large as 127 times 127 = 16,129, which does not fit in a byte, and the Vector API has no operation that takes two byte vectors and returns 32-bit sums. So the kernel first widens the bytes, which means it converts each byte lane to a wider type before multiplying.
convertShape(VectorOperators.B2I, IntVector.SPECIES_256, part) converts 8 of the 32 bytes into 8 int lanes, with the sign kept. A block has 32 codes, which is 4 groups of 8, so the part argument selects the group: part 0 is bytes 0 to 7, part 3 is bytes 24 to 31. The kernel multiplies the widened vectors, adds the four results into an integer accumulator, and reduces it once per block.
The program computes a 4,096-value Q8_0 by Q8_0 dot product with the scalar kernel and with the vector kernel, and compares the bits of the two results. It also prints the widened lanes of the first block so the conversion is visible.
import java.util.Locale;
import java.util.Random;
import jdk.incubator.vector.ByteVector;
import jdk.incubator.vector.IntVector;
import jdk.incubator.vector.VectorOperators;
public class DotQ8Vector {
static int roundf(float v) { return (int) Math.copySign(Math.floor(Math.abs((double) v) + 0.5), v); }
static float[] quantize(float[] x, byte[] qs) { // Q8_0 codes and FP16-rounded scales
float[] d = new float[x.length / 32];
for (int b = 0; b < d.length; b++) {
float amax = 0;
for (int j = 0; j < 32; j++) amax = Math.max(amax, Math.abs(x[b * 32 + j]));
float s = amax / 127f, id = s != 0 ? 1f / s : 0f;
d[b] = Float.float16ToFloat(Float.floatToFloat16(s));
for (int j = 0; j < 32; j++) qs[b * 32 + j] = (byte) roundf(x[b * 32 + j] * id);
}
return d;
}
static float scalar(byte[] qw, float[] dw, byte[] qx, float[] dx) {
float sumf = 0;
for (int b = 0; b < dw.length; b++) {
int sumi = 0;
for (int j = 0; j < 32; j++) sumi += qw[b * 32 + j] * qx[b * 32 + j];
sumf += sumi * (dw[b] * dx[b]);
}
return sumf;
}
static float vector(byte[] qw, float[] dw, byte[] qx, float[] dx) {
float sumf = 0;
for (int b = 0; b < dw.length; b++) {
ByteVector w = ByteVector.fromArray(ByteVector.SPECIES_256, qw, b * 32); // 32 codes = one block
ByteVector x = ByteVector.fromArray(ByteVector.SPECIES_256, qx, b * 32);
IntVector acc = IntVector.zero(IntVector.SPECIES_256);
for (int part = 0; part < 4; part++) { // 4 x 8 int lanes
IntVector wi = (IntVector) w.convertShape(VectorOperators.B2I, IntVector.SPECIES_256, part);
IntVector xi = (IntVector) x.convertShape(VectorOperators.B2I, IntVector.SPECIES_256, part);
acc = acc.add(wi.mul(xi));
}
int sumi = acc.reduceLanes(VectorOperators.ADD); // integer sum: exact in any order
sumf += sumi * (dw[b] * dx[b]);
}
return sumf;
}
public static void main(String[] args) {
int n = 4096;
Random rnd = new Random(7);
float[] w = new float[n], x = new float[n];
for (int i = 0; i < n; i++) { w[i] = (float) rnd.nextGaussian() * 0.02f; x[i] = (float) rnd.nextGaussian(); }
byte[] qw = new byte[n], qx = new byte[n];
float[] dw = quantize(w, qw), dx = quantize(x, qx);
ByteVector first = ByteVector.fromArray(ByteVector.SPECIES_256, qw, 0);
System.out.println("block 0 codes 0..7 as bytes: " + first.convertShape(VectorOperators.B2I, IntVector.SPECIES_256, 0));
System.out.println("block 0 codes 8..15 as ints: " + first.convertShape(VectorOperators.B2I, IntVector.SPECIES_256, 1));
float s = scalar(qw, dw, qx, dx), v = vector(qw, dw, qx, dx);
System.out.printf(Locale.ROOT, "scalar %.6f, vector %.6f, same bits: %b%n", s, v, Float.floatToIntBits(s) == Float.floatToIntBits(v));
}
}block 0 codes 0..7 as bytes: [43, -15, 68, 31, 127, -72, 36, -21]
block 0 codes 8..15 as ints: [-48, 3, 30, -19, -85, 22, 77, 1]
scalar -0.638400, vector -0.638400, same bits: trueThe first two lines show widening. Part 0 of block 0's weight codes is the first 8 bytes as int lanes, and part 1 is the next 8. Negative codes stay negative (-15, -72, -21), because B2I sign-extends: the byte 0xF1 becomes the int -15, not 241. The value 127 at lane 4 of part 0 is the block's largest magnitude, which Q8_0 always maps to 127 or -127.
The last line shows the scalar and vector results: both -0.638400, with identical bits. That equality is not a coincidence of rounding. Inside each block, the vector kernel adds 32 integer products in a different order from the scalar loop, but integer addition is exact, so sumi is the same number in both kernels. After that, both kernels run the same float operations in the same order: sumf += sumi * (dw[b] * dx[b]), block 0 first, block 127 last. Same inputs and same float operations in the same order give the same bits.
This property makes quantized kernels straightforward to test: the vector kernel can be required to match the scalar reference exactly, not within a tolerance. The cost of this design is one reduceLanes per block, which Example 13 shows to be one reason the kernel does not reach the memory bandwidth limit on one thread.
ggml's x86 kernels take a different route. They convert each block's integer sums to floats and accumulate them in float vector lanes across blocks (_mm256_fmadd_ps with the block scale). The integer products are computed by _mm256_maddubs_epi16, a CPU instruction that multiplies pairs of bytes and adds neighbors, or by a VNNI dot-product instruction (_mm256_dpbusd_epi32 or _mm256_dpbusd_avx_epi32) when the build targets a CPU with VNNI (mul_sum_us8_pairs_float in ggml-cpu/arch/x86/quants.c). VNNI is an extension that multiplies and sums groups of bytes in one step. Their results therefore differ from ggml's generic C kernel in the last bits.
Example 8: Q8_0 weights by F32 activations, widening to float
Example 7 quantized the activations too. Some engines keep the activations in F32 and multiply them directly with Q8_0 weights, which avoids quantizing the activation vector. Then the weight bytes are widened to floats instead of ints. castShape(FloatVector.SPECIES_256, part) converts 8 bytes into 8 float lanes.
Per block, the kernel accumulates code * x in a float vector. It then multiplies the block's vector by the block scale with one more FMA into an accumulator that runs across all blocks. The program compares this vector kernel with a scalar loop that uses the same per-block formula, and with a double-precision reference. Look at the last line: the answer is accurate, but the bits differ from the scalar loop, unlike Example 7.
import java.util.Locale;
import java.util.Random;
import jdk.incubator.vector.ByteVector;
import jdk.incubator.vector.FloatVector;
import jdk.incubator.vector.VectorSpecies;
public class DotQ8F32 {
static final VectorSpecies<Float> F = FloatVector.SPECIES_256;
static int roundf(float v) { return (int) Math.copySign(Math.floor(Math.abs((double) v) + 0.5), v); }
static float scalar(byte[] qw, float[] dw, float[] x) {
float sumf = 0;
for (int b = 0; b < dw.length; b++) {
float block = 0;
for (int j = 0; j < 32; j++) block += qw[b * 32 + j] * x[b * 32 + j];
sumf += block * dw[b];
}
return sumf;
}
static float vector(byte[] qw, float[] dw, float[] x) {
FloatVector acc = FloatVector.zero(F);
for (int b = 0; b < dw.length; b++) {
ByteVector w = ByteVector.fromArray(ByteVector.SPECIES_256, qw, b * 32);
FloatVector block = FloatVector.zero(F);
for (int part = 0; part < 4; part++) {
FloatVector wf = (FloatVector) w.castShape(F, part); // 8 bytes -> 8 floats
block = wf.fma(FloatVector.fromArray(F, x, b * 32 + part * 8), block);
}
acc = block.fma(FloatVector.broadcast(F, dw[b]), acc); // times the block scale
}
float sum = 0;
for (float v : acc.toArray()) sum += v;
return sum;
}
public static void main(String[] args) {
int n = 4096;
Random rnd = new Random(8);
float[] w = new float[n], x = new float[n];
for (int i = 0; i < n; i++) { w[i] = (float) rnd.nextGaussian() * 0.02f; x[i] = (float) rnd.nextGaussian(); }
byte[] qw = new byte[n];
float[] dw = new float[n / 32];
for (int b = 0; b < dw.length; b++) {
float amax = 0;
for (int j = 0; j < 32; j++) amax = Math.max(amax, Math.abs(w[b * 32 + j]));
float d = amax / 127f, id = d != 0 ? 1f / d : 0f;
dw[b] = Float.float16ToFloat(Float.floatToFloat16(d));
for (int j = 0; j < 32; j++) qw[b * 32 + j] = (byte) roundf(w[b * 32 + j] * id);
}
double deq = 0, scale = 0;
for (int i = 0; i < n; i++) { deq += (double) qw[i] * dw[i / 32] * x[i]; scale += Math.abs((double) qw[i] * dw[i / 32] * x[i]); }
float s = scalar(qw, dw, x), v = vector(qw, dw, x);
System.out.printf(Locale.ROOT, "double reference %.6f, scalar %.6f, vector %.6f%n", deq, s, v);
System.out.printf(Locale.ROOT, "vector error / sum|w*x| = %.2e, same bits as scalar: %b%n",
Math.abs(v - deq) / scale, Float.floatToIntBits(s) == Float.floatToIntBits(v));
}
}double reference -1.527944, scalar -1.527944, vector -1.527944
vector error / sum|w*x| = 2.92e-09, same bits as scalar: falseAll three results print as -1.527944 to six decimals. The vector result has an error of about 3 parts per billion against the double reference, but its bits differ from the scalar loop (same bits as scalar: false). Here every product is a float operation. The vector kernel computes partial sums in 8 lanes and multiplies each lane by the block scale before adding across blocks, while the scalar loop sums the whole block first. The orders differ, so the last bits differ, and the test needs a tolerance, as in Example 6.
The comparison with Example 7 shows the trade-off between the two designs:
| Activations | Inner arithmetic | Result vs scalar reference | Extra work per matrix-vector product |
|---|---|---|---|
| Q8_0 (Example 7) | integer products, one integer sum per block | identical bits | quantize the activation vector once |
| F32 (this example) | float FMAs | within a tolerance | none |
The extra quantization in the first row is cheap. The activation vector has a few thousand values and is quantized once, then reused by every row of the matrix. A 4,096 by 4,096 matrix has 4,096 rows, so the cost is spread over 4,096 dot products. llama.cpp uses the first design on the CPU, and so do the rest of this lesson and the capstone engine.
Example 9: unpacking Q4_0 nibbles with vectors
Q4_0 needs one more step before any multiplication: each byte holds two 4-bit codes, and the multiplication needs them as separate numbers. The block stores the codes of values 0 to 15 in the low nibbles and the codes of values 16 to 31 in the high nibbles. With vectors, 16 packed bytes are unpacked with two lane-wise operations:
and((byte) 0x0F)keeps the low 4 bits of every byte (the low nibbles).lanewise(VectorOperators.LSHR, 4)shifts each byte right by 4 bits, filling with zeros, which leaves the high nibble (a logical shift right).
Subtracting 8 turns the codes 0 to 15 into the values -8 to 7. For example, the byte 0xA3 has low nibble 3 and high nibble 10 (0xA), so it stands for the values 3 - 8 = -5 and 10 - 8 = 2.
The program unpacks a block whose nibbles form a visible pattern and shows why the shift must be logical. It then checks the vector integer sum against the scalar one on 1,000 random blocks.
import java.util.Locale;
import java.util.Random;
import jdk.incubator.vector.ByteVector;
import jdk.incubator.vector.IntVector;
import jdk.incubator.vector.VectorOperators;
import jdk.incubator.vector.VectorSpecies;
public class UnpackQ4 {
static final VectorSpecies<Integer> I = IntVector.SPECIES_256;
static int sumiScalar(byte[] q4, byte[] q8) { // one block: 16 packed bytes, 32 codes
int s0 = 0, s1 = 0;
for (int j = 0; j < 16; j++) { s0 += ((q4[j] & 0x0F) - 8) * q8[j]; s1 += (((q4[j] & 0xFF) >> 4) - 8) * q8[j + 16]; }
return s0 + s1;
}
static int sumiVector(byte[] q4, byte[] q8) {
ByteVector packed = ByteVector.fromArray(ByteVector.SPECIES_128, q4, 0);
ByteVector lo = packed.and((byte) 0x0F).sub((byte) 8); // values 0..15
ByteVector hi = packed.lanewise(VectorOperators.LSHR, 4).sub((byte) 8); // values 16..31
ByteVector xlo = ByteVector.fromArray(ByteVector.SPECIES_128, q8, 0);
ByteVector xhi = ByteVector.fromArray(ByteVector.SPECIES_128, q8, 16);
IntVector acc = IntVector.zero(I);
for (int part = 0; part < 2; part++) {
acc = acc.add(((IntVector) lo.convertShape(VectorOperators.B2I, I, part)).mul((IntVector) xlo.convertShape(VectorOperators.B2I, I, part)));
acc = acc.add(((IntVector) hi.convertShape(VectorOperators.B2I, I, part)).mul((IntVector) xhi.convertShape(VectorOperators.B2I, I, part)));
}
return acc.reduceLanes(VectorOperators.ADD);
}
public static void main(String[] args) {
byte[] q4 = new byte[16];
for (int j = 0; j < 16; j++) q4[j] = (byte) ((j << 4) | (15 - j)); // low nibble 15-j, high nibble j
ByteVector packed = ByteVector.fromArray(ByteVector.SPECIES_128, q4, 0);
System.out.println("packed (signed bytes) " + packed);
System.out.println("low - 8 " + packed.and((byte) 0x0F).sub((byte) 8));
System.out.println("high - 8 (LSHR) " + packed.lanewise(VectorOperators.LSHR, 4).sub((byte) 8));
System.out.println("high - 8 (ASHR) " + packed.lanewise(VectorOperators.ASHR, 4).sub((byte) 8));
Random rnd = new Random(9);
int blocks = 1000, equal = 0;
for (int b = 0; b < blocks; b++) {
byte[] w = new byte[16], x = new byte[32];
rnd.nextBytes(w);
for (int j = 0; j < 32; j++) x[j] = (byte) (rnd.nextInt(255) - 127);
if (sumiScalar(w, x) == sumiVector(w, x)) equal++;
}
System.out.printf(Locale.ROOT, "random blocks with equal integer sums: %d of %d%n", equal, blocks);
}
}packed (signed bytes) [15, 30, 45, 60, 75, 90, 105, 120, -121, -106, -91, -76, -61, -46, -31, -16]
low - 8 [7, 6, 5, 4, 3, 2, 1, 0, -1, -2, -3, -4, -5, -6, -7, -8]
high - 8 (LSHR) [-8, -7, -6, -5, -4, -3, -2, -1, 0, 1, 2, 3, 4, 5, 6, 7]
high - 8 (ASHR) [-8, -7, -6, -5, -4, -3, -2, -1, -16, -15, -14, -13, -12, -11, -10, -9]
random blocks with equal integer sums: 1000 of 1000The test block stores 15 - j in the low nibble of byte j and j in the high nibble. The first line prints the packed bytes as Java's signed byte values: byte 0 is 0x0F = 15, byte 1 is 0x1E = 30, and from byte 8 on the top bit is set, so 0x87 prints as -121.
The second line is the low nibbles minus 8: byte 0 has low nibble 15, so value 0 of the block is 7, and byte 15 has low nibble 0, so value 15 is -8. The third line is the high nibbles minus 8: byte j has high nibble j, so values 16 to 31 run from -8 up to 7. That is the Q4_0 order from the earlier lesson: low nibbles are the first 16 values of the block, high nibbles the last 16.
The fourth line shows the bug that the wrong shift causes. ASHR is an arithmetic shift, which copies the sign bit into the vacated bits. For bytes 8 to 15 the sign bit is set, so 0x87 >> 4 gives 0xF8 = -8 instead of 8, and after subtracting 8 the value is -16, outside the valid range of -8 to 7. In the Vector API, LSHR on a byte vector treats each lane as an unsigned 8-bit value, the same as (b & 0xFF) >>> 4 in scalar code. It is the vector form of the & 0xFF mask in Example 2.
The last line runs sumiVector and sumiScalar on 1,000 blocks of random packed bytes and random activation codes. All 1,000 integer sums are equal. The vector version widens each 16-byte group in two parts of 8 lanes, multiplies low values with activations 0 to 15 and high values with activations 16 to 31, and reduces once per block. Since the integer sums are equal, a full Q4_0 kernel built on this function gives the same bits as the scalar kernel, as Example 10 shows.
Example 10: loading straight from a MemorySegment
In the engine, weights live in a memory-mapped GGUF file, not in a Java array. Memory mapping means the operating system makes the file's bytes appear as memory, loading pages as they are touched. Copying the weights into Java arrays would double the memory use and take time. ByteVector.fromMemorySegment(species, segment, offset, ByteOrder.LITTLE_ENDIAN) loads a vector straight from a segment, at any byte offset. The byte order matters only for multi-byte element types.
The program writes 8 rows of random Q4_0 blocks to a temporary file and maps it with FileChannel.map(..., arena). It keeps Q8_0 activations in an off-heap segment, and it runs a scalar and a vector Q4_0 kernel that both read from the segments. No byte is copied into a Java array. Check that the scalar and vector results are bit-identical for every row.
import java.io.IOException;
import java.lang.foreign.Arena;
import java.lang.foreign.MemorySegment;
import java.lang.foreign.ValueLayout;
import java.nio.ByteBuffer;
import java.nio.ByteOrder;
import java.nio.channels.FileChannel;
import java.nio.file.Files;
import java.nio.file.Path;
import java.nio.file.StandardOpenOption;
import java.util.Locale;
import java.util.Random;
import jdk.incubator.vector.ByteVector;
import jdk.incubator.vector.IntVector;
import jdk.incubator.vector.VectorOperators;
public class SegmentKernel {
static final ValueLayout.OfShort F16 = ValueLayout.JAVA_SHORT_UNALIGNED.withOrder(ByteOrder.LITTLE_ENDIAN);
static final ByteOrder LE = ByteOrder.LITTLE_ENDIAN;
static float dotScalar(MemorySegment w, long row, MemorySegment x, int nb) {
float sumf = 0;
for (int b = 0; b < nb; b++) {
long ow = row + b * 18L, ox = b * 34L;
int sumi = 0;
for (int j = 0; j < 16; j++) {
int q = w.get(ValueLayout.JAVA_BYTE, ow + 2 + j) & 0xFF;
sumi += ((q & 0x0F) - 8) * x.get(ValueLayout.JAVA_BYTE, ox + 2 + j) + ((q >> 4) - 8) * x.get(ValueLayout.JAVA_BYTE, ox + 18 + j);
}
sumf += sumi * Float.float16ToFloat(w.get(F16, ow)) * Float.float16ToFloat(x.get(F16, ox));
}
return sumf;
}
static float dotVector(MemorySegment w, long row, MemorySegment x, int nb) {
float sumf = 0;
for (int b = 0; b < nb; b++) {
long ow = row + b * 18L, ox = b * 34L;
ByteVector p = ByteVector.fromMemorySegment(ByteVector.SPECIES_128, w, ow + 2, LE); // 16 packed bytes
ByteVector lo = p.and((byte) 0x0F).sub((byte) 8), hi = p.lanewise(VectorOperators.LSHR, 4).sub((byte) 8);
ByteVector xlo = ByteVector.fromMemorySegment(ByteVector.SPECIES_128, x, ox + 2, LE);
ByteVector xhi = ByteVector.fromMemorySegment(ByteVector.SPECIES_128, x, ox + 18, LE);
IntVector acc = IntVector.zero(IntVector.SPECIES_256);
for (int part = 0; part < 2; part++)
acc = acc.add(((IntVector) lo.convertShape(VectorOperators.B2I, IntVector.SPECIES_256, part)).mul((IntVector) xlo.convertShape(VectorOperators.B2I, IntVector.SPECIES_256, part)))
.add(((IntVector) hi.convertShape(VectorOperators.B2I, IntVector.SPECIES_256, part)).mul((IntVector) xhi.convertShape(VectorOperators.B2I, IntVector.SPECIES_256, part)));
sumf += acc.reduceLanes(VectorOperators.ADD) * Float.float16ToFloat(w.get(F16, ow)) * Float.float16ToFloat(x.get(F16, ox));
}
return sumf;
}
public static void main(String[] args) throws IOException {
int rows = 8, cols = 256, nb = cols / 32;
long rowBytes = nb * 18L;
Random rnd = new Random(10);
ByteBuffer file = ByteBuffer.allocate((int) (rows * rowBytes)).order(LE);
for (int b = 0; b < rows * nb; b++) { // random but valid Q4_0 blocks
file.putShort(Float.floatToFloat16(0.001f + rnd.nextFloat() * 0.002f));
for (int j = 0; j < 16; j++) file.put((byte) rnd.nextInt(256));
}
Path path = Files.createTempFile("q4_0-", ".bin");
Files.write(path, file.array());
try (Arena arena = Arena.ofConfined(); FileChannel ch = FileChannel.open(path, StandardOpenOption.READ)) {
MemorySegment w = ch.map(FileChannel.MapMode.READ_ONLY, 0, ch.size(), arena);
MemorySegment x = arena.allocate(nb * 34L); // Q8_0 activations, off heap
for (int b = 0; b < nb; b++) {
x.set(F16, b * 34L, Float.floatToFloat16(0.01f));
for (int j = 0; j < 32; j++) x.set(ValueLayout.JAVA_BYTE, b * 34L + 2 + j, (byte) (rnd.nextInt(255) - 127));
}
System.out.println("mapped " + w.byteSize() + " bytes: " + rows + " rows x " + rowBytes + " bytes, mapped = " + w.isMapped());
boolean same = true;
for (int r = 0; r < rows; r++) {
float s = dotScalar(w, r * rowBytes, x, nb), v = dotVector(w, r * rowBytes, x, nb);
same &= Float.floatToIntBits(s) == Float.floatToIntBits(v);
if (r < 3) System.out.printf(Locale.ROOT, "row %d: scalar %.6f, vector %.6f%n", r, s, v);
}
System.out.println("all rows identical: " + same);
} finally {
Files.delete(path);
}
}
}mapped 1152 bytes: 8 rows x 144 bytes, mapped = true
row 0: scalar -0.014319, vector -0.014319
row 1: scalar -0.218860, vector -0.218860
row 2: scalar 0.149180, vector 0.149180
all rows identical: trueThe first line confirms the mapping: 8 rows of 256 values are 8 blocks per row, 8 times 18 = 144 bytes per row, and 1,152 bytes in the file. isMapped() returns true for a segment that comes from FileChannel.map.
The three row lines show scalar and vector results for the first rows, and all rows identical: true confirms that all 8 rows agree bit for bit. The vector kernel is Example 9's unpacking with a different source. fromMemorySegment at offset ow + 2 skips the 2-byte scale of the block and loads the 16 packed bytes. The activation codes are loaded from offsets ox + 2 (values 0 to 15) and ox + 18 (values 16 to 31) of each 34-byte Q8_0 block. None of these offsets is a multiple of 16, and that is allowed: vector loads from a segment do not need aligned addresses.
Each access through a segment is bounds-checked against the segment's size, and checked against the arena's state. Reading after the arena is closed therefore throws an exception instead of reading freed memory. Example 14 shows how much time these checks take in a profile.
Any 18 bytes are a valid Q4_0 block, because every nibble value from 0 to 15 is a legal code. Random bytes therefore test the kernel on more patterns than quantized smooth data would.
Example 11: parallel rows across threads
The rows of a matrix-vector product are independent: row 5 does not need the result of row 4. Threads can therefore share the rows. The open question is how to divide them. Giving each thread one equal block of rows has a weakness: if one thread is slowed down (another program uses its core), everyone waits for it.
ggml's CPU backend (ggml_compute_forward_mul_mat in ggml-cpu.c) cuts the rows into small chunks instead. It uses 64 rows per chunk when the other dimension is 1 (decoding one token). Threads take the next chunk from a shared atomic counter, a number that many threads can increment safely, each getting a different value. A thread that finishes early takes more chunks, so the load balances itself.
The program does the same with an AtomicInteger and a fixed thread pool, over a 1,000 by 1,024 Q8_0 matrix stored in arrays. It compares the outputs of 1, 2, 4, and 8 threads. Each multi-threaded output should equal the single-threaded one exactly.
import java.util.Arrays;
import java.util.Locale;
import java.util.Random;
import java.util.concurrent.ExecutorService;
import java.util.concurrent.Executors;
import java.util.concurrent.Future;
import java.util.concurrent.atomic.AtomicInteger;
import java.util.ArrayList;
import java.util.List;
import jdk.incubator.vector.ByteVector;
import jdk.incubator.vector.IntVector;
import jdk.incubator.vector.VectorOperators;
public class ParallelRows {
static final int CHUNK = 64; // rows per chunk, as ggml uses for one token
static float dotRow(byte[] qw, float[] dw, int row, int nb, byte[] qx, float[] dx) { // Q8_0 x Q8_0, Example 7
float sumf = 0;
for (int b = 0; b < nb; b++) {
int o = (row * nb + b) * 32;
ByteVector w = ByteVector.fromArray(ByteVector.SPECIES_256, qw, o), x = ByteVector.fromArray(ByteVector.SPECIES_256, qx, b * 32);
IntVector acc = IntVector.zero(IntVector.SPECIES_256);
for (int p = 0; p < 4; p++)
acc = acc.add(((IntVector) w.convertShape(VectorOperators.B2I, IntVector.SPECIES_256, p)).mul((IntVector) x.convertShape(VectorOperators.B2I, IntVector.SPECIES_256, p)));
sumf += acc.reduceLanes(VectorOperators.ADD) * (dw[row * nb + b] * dx[b]);
}
return sumf;
}
static float[] matvec(byte[] qw, float[] dw, int rows, int nb, byte[] qx, float[] dx, int threads) throws Exception {
float[] out = new float[rows];
int chunks = (rows + CHUNK - 1) / CHUNK;
AtomicInteger next = new AtomicInteger();
try (ExecutorService pool = Executors.newFixedThreadPool(threads)) {
List<Future<?>> done = new ArrayList<>();
for (int t = 0; t < threads; t++)
done.add(pool.submit(() -> {
for (int c; (c = next.getAndIncrement()) < chunks; )
for (int r = c * CHUNK; r < Math.min(rows, (c + 1) * CHUNK); r++) out[r] = dotRow(qw, dw, r, nb, qx, dx);
}));
for (Future<?> f : done) f.get();
}
return out;
}
public static void main(String[] args) throws Exception {
int rows = 1000, cols = 1024, nb = cols / 32;
Random rnd = new Random(11);
byte[] qw = new byte[rows * cols], qx = new byte[cols];
float[] dw = new float[rows * nb], dx = new float[nb];
for (int i = 0; i < qw.length; i++) qw[i] = (byte) (rnd.nextInt(255) - 127);
for (int i = 0; i < dw.length; i++) dw[i] = Float.float16ToFloat(Float.floatToFloat16(rnd.nextFloat() * 0.001f));
for (int i = 0; i < cols; i++) qx[i] = (byte) (rnd.nextInt(255) - 127);
for (int i = 0; i < nb; i++) dx[i] = Float.float16ToFloat(Float.floatToFloat16(0.01f + rnd.nextFloat() * 0.01f));
float[] one = matvec(qw, dw, rows, nb, qx, dx, 1);
System.out.println("rows " + rows + ", chunks of " + CHUNK + " rows: " + (rows + CHUNK - 1) / CHUNK);
for (int threads : new int[] {2, 4, 8}) {
float[] many = matvec(qw, dw, rows, nb, qx, dx, threads);
System.out.println(threads + " threads: identical to 1 thread = " + Arrays.equals(one, many));
}
System.out.printf(Locale.ROOT, "out[0] = %.6f, out[999] = %.6f%n", one[0], one[999]);
}
}rows 1000, chunks of 64 rows: 16
2 threads: identical to 1 thread = true
4 threads: identical to 1 thread = true
8 threads: identical to 1 thread = true
out[0] = 0.874649, out[999] = 0.103208The first line shows the chunking: 1,000 rows in chunks of 64 is 16 chunks (15 full chunks and one of 40 rows), computed as (rows + 63) / 64. Each thread runs a loop that calls next.getAndIncrement(), gets a chunk number, and computes that chunk's rows, until the counter passes 15.
All three multi-threaded runs give arrays equal to the single-threaded run. Which thread computes which chunk changes from run to run, but every row is computed by one thread with the same kernel and the same order of operations, so each output value has the same bits no matter who computed it. A different split, where several threads each compute part of one dot product and add their partial sums, would change the order of additions and could change the last bits of the result.
Eight threads on this 4-core machine give the same results; the extra threads take turns on the cores. ggml has one more rule that this program does not need: when there would be fewer than 4 chunks per thread, it gives up on 64-row chunks and splits the rows into one equal part per thread. With 4 threads and 16 chunks, that rule does not apply.
The output array is shared by all threads without locks. That is safe because each element is written by exactly one thread. Future.get() (here inside the pool's close() and the f.get() loop) makes the writes visible to the main thread before it reads the array.
Example 12: Project 19, a kernel library with tests
Project 19 packages the lesson into a kernel library for the capstone engine, together with a test that proves the fast code is correct. It has three parts:
Type, an enum with the block size and bytes per block of F32, Q8_0, and Q4_0, androwBytes(cols), which rejects a row length that is not a multiple of the block size.Kernels, which holds a thread pool and a shared arena. It has the vector kernels from Examples 6, 7, and 10, all reading weights from aMemorySegment, the Q8_0 activation quantizer, andmatvec, which quantizes the activation vector once and hands out 64-row chunks as in Example 11.Reference, the scalar kernels from Examples 1, 2, and 6, and a test that runs every type at several shapes. The test requires Q8_0 and Q4_0 rows to match the reference bit for bit, F32 rows to match within1e-6of the sum of absolute products, and 1 thread and 4 threads to give equal arrays.
The test tensors are random bytes with random FP16 scales, and the arena is shared, because the pool threads read the segments. A confined arena, as in Examples 1 and 2, belongs to the thread that created it. Using one from pool threads would make every pool thread fail with WrongThreadException, as Loading GGUF with the FFM API showed. When reading the output table, look at the "same bits" column for the exact comparisons and at the "max rel" column for the F32 tolerance.
import java.lang.foreign.Arena;
import java.lang.foreign.MemorySegment;
import java.lang.foreign.ValueLayout;
import java.nio.ByteOrder;
import java.util.ArrayList;
import java.util.Arrays;
import java.util.List;
import java.util.Locale;
import java.util.Random;
import java.util.concurrent.Callable;
import java.util.concurrent.ExecutorService;
import java.util.concurrent.Executors;
import java.util.concurrent.atomic.AtomicInteger;
import jdk.incubator.vector.ByteVector;
import jdk.incubator.vector.FloatVector;
import jdk.incubator.vector.IntVector;
import jdk.incubator.vector.VectorOperators;
import jdk.incubator.vector.VectorSpecies;
public class KernelLibrary {
enum Type {
F32(1, 4), Q8_0(32, 34), Q4_0(32, 18);
final int block, bytes;
Type(int block, int bytes) { this.block = block; this.bytes = bytes; }
long rowBytes(int cols) {
if (cols % block != 0) throw new IllegalArgumentException(this + " row length " + cols + " is not a multiple of " + block);
return (long) cols / block * bytes;
}
}
static final class Kernels implements AutoCloseable {
static final ByteOrder LE = ByteOrder.LITTLE_ENDIAN;
static final ValueLayout.OfShort F16 = ValueLayout.JAVA_SHORT_UNALIGNED.withOrder(LE);
static final ValueLayout.OfFloat F32 = ValueLayout.JAVA_FLOAT_UNALIGNED.withOrder(LE);
static final VectorSpecies<Float> FS = FloatVector.SPECIES_256;
static final VectorSpecies<Integer> IS = IntVector.SPECIES_256;
static final int CHUNK = 64;
final int threads;
final ExecutorService pool;
final Arena arena = Arena.ofShared();
MemorySegment xq = MemorySegment.NULL;
Kernels(int threads) { this.threads = threads; this.pool = Executors.newFixedThreadPool(threads); }
static int roundf(float v) { return (int) Math.copySign(Math.floor(Math.abs((double) v) + 0.5), v); }
static void quantizeQ8_0(float[] x, MemorySegment out) { // quantize_row_q8_0_ref
for (int b = 0; b < x.length / 32; b++) {
float amax = 0;
for (int j = 0; j < 32; j++) amax = Math.max(amax, Math.abs(x[b * 32 + j]));
float d = amax / 127f, id = d != 0 ? 1f / d : 0f;
out.set(F16, b * 34L, Float.floatToFloat16(d));
for (int j = 0; j < 32; j++) out.set(ValueLayout.JAVA_BYTE, b * 34L + 2 + j, (byte) roundf(x[b * 32 + j] * id));
}
}
static float dotF32(MemorySegment w, long off, float[] x) {
int n = x.length, L = FS.length(), i = 0, upper = FS.loopBound(n);
FloatVector acc = FloatVector.zero(FS);
for (; i < upper; i += L) acc = FloatVector.fromMemorySegment(FS, w, off + 4L * i, LE).fma(FloatVector.fromArray(FS, x, i), acc);
float sum = 0;
for (float v : acc.toArray()) sum += v;
for (; i < n; i++) sum += w.get(F32, off + 4L * i) * x[i];
return sum;
}
static float dotQ8_0(MemorySegment w, long off, MemorySegment x, int n) {
float sumf = 0;
for (int b = 0; b < n / 32; b++) {
long ow = off + b * 34L, ox = b * 34L;
ByteVector wv = ByteVector.fromMemorySegment(ByteVector.SPECIES_256, w, ow + 2, LE);
ByteVector xv = ByteVector.fromMemorySegment(ByteVector.SPECIES_256, x, ox + 2, LE);
IntVector acc = IntVector.zero(IS);
for (int p = 0; p < 4; p++)
acc = acc.add(((IntVector) wv.convertShape(VectorOperators.B2I, IS, p)).mul((IntVector) xv.convertShape(VectorOperators.B2I, IS, p)));
sumf += acc.reduceLanes(VectorOperators.ADD) * (Float.float16ToFloat(w.get(F16, ow)) * Float.float16ToFloat(x.get(F16, ox)));
}
return sumf;
}
static float dotQ4_0(MemorySegment w, long off, MemorySegment x, int n) {
float sumf = 0;
for (int b = 0; b < n / 32; b++) {
long ow = off + b * 18L, ox = b * 34L;
ByteVector p = ByteVector.fromMemorySegment(ByteVector.SPECIES_128, w, ow + 2, LE);
ByteVector lo = p.and((byte) 0x0F).sub((byte) 8), hi = p.lanewise(VectorOperators.LSHR, 4).sub((byte) 8);
ByteVector xlo = ByteVector.fromMemorySegment(ByteVector.SPECIES_128, x, ox + 2, LE);
ByteVector xhi = ByteVector.fromMemorySegment(ByteVector.SPECIES_128, x, ox + 18, LE);
IntVector acc = IntVector.zero(IS);
for (int q = 0; q < 2; q++)
acc = acc.add(((IntVector) lo.convertShape(VectorOperators.B2I, IS, q)).mul((IntVector) xlo.convertShape(VectorOperators.B2I, IS, q)))
.add(((IntVector) hi.convertShape(VectorOperators.B2I, IS, q)).mul((IntVector) xhi.convertShape(VectorOperators.B2I, IS, q)));
sumf += acc.reduceLanes(VectorOperators.ADD) * Float.float16ToFloat(w.get(F16, ow)) * Float.float16ToFloat(x.get(F16, ox));
}
return sumf;
}
/** out[r] = row r of w (rows x cols, GGUF type t) times x. Q types quantize x to Q8_0 once. */
void matvec(Type t, MemorySegment w, float[] x, float[] out, int rows) throws Exception {
int cols = x.length;
long rb = t.rowBytes(cols);
if (t != Type.F32) {
if (xq.byteSize() < Type.Q8_0.rowBytes(cols)) xq = arena.allocate(Type.Q8_0.rowBytes(cols));
quantizeQ8_0(x, xq);
}
MemorySegment xs = xq;
int chunks = (rows + CHUNK - 1) / CHUNK;
AtomicInteger next = new AtomicInteger();
List<Callable<Void>> tasks = new ArrayList<>();
for (int i = 0; i < threads; i++) tasks.add(() -> {
for (int c; (c = next.getAndIncrement()) < chunks; )
for (int r = c * CHUNK; r < Math.min(rows, (c + 1) * CHUNK); r++)
out[r] = switch (t) {
case F32 -> dotF32(w, r * rb, x);
case Q8_0 -> dotQ8_0(w, r * rb, xs, cols);
case Q4_0 -> dotQ4_0(w, r * rb, xs, cols);
};
return null;
});
for (var f : pool.invokeAll(tasks)) f.get();
}
@Override public void close() { pool.close(); arena.close(); }
}
static final class Reference { // scalar kernels of Examples 1, 2, and 6
static float dotF32(MemorySegment w, long off, float[] x) {
float s = 0;
for (int i = 0; i < x.length; i++) s += w.get(Kernels.F32, off + 4L * i) * x[i];
return s;
}
static float dotQ8_0(MemorySegment w, long off, MemorySegment x, int n) {
float sumf = 0;
for (int b = 0; b < n / 32; b++) {
long ow = off + b * 34L, ox = b * 34L;
int sumi = 0;
for (int j = 0; j < 32; j++) sumi += w.get(ValueLayout.JAVA_BYTE, ow + 2 + j) * x.get(ValueLayout.JAVA_BYTE, ox + 2 + j);
sumf += sumi * (Float.float16ToFloat(w.get(Kernels.F16, ow)) * Float.float16ToFloat(x.get(Kernels.F16, ox)));
}
return sumf;
}
static float dotQ4_0(MemorySegment w, long off, MemorySegment x, int n) {
float sumf = 0;
for (int b = 0; b < n / 32; b++) {
long ow = off + b * 18L, ox = b * 34L;
int sumi0 = 0, sumi1 = 0;
for (int j = 0; j < 16; j++) {
int q = w.get(ValueLayout.JAVA_BYTE, ow + 2 + j) & 0xFF;
sumi0 += ((q & 0x0F) - 8) * x.get(ValueLayout.JAVA_BYTE, ox + 2 + j);
sumi1 += ((q >> 4) - 8) * x.get(ValueLayout.JAVA_BYTE, ox + 18 + j);
}
sumf += (sumi0 + sumi1) * Float.float16ToFloat(w.get(Kernels.F16, ow)) * Float.float16ToFloat(x.get(Kernels.F16, ox));
}
return sumf;
}
static float dot(Type t, MemorySegment w, long off, float[] x, MemorySegment xq) {
return switch (t) {
case F32 -> dotF32(w, off, x);
case Q8_0 -> dotQ8_0(w, off, xq, x.length);
case Q4_0 -> dotQ4_0(w, off, xq, x.length);
};
}
}
static MemorySegment randomTensor(Arena arena, Type t, int rows, int cols, Random rnd) {
MemorySegment s = arena.allocate(rows * t.rowBytes(cols));
for (long o = 0; o < s.byteSize(); o += t.bytes) {
if (t == Type.F32) { s.set(Kernels.F32, o, (rnd.nextFloat() - 0.5f) * 0.04f); continue; }
s.set(Kernels.F16, o, Float.floatToFloat16((rnd.nextBoolean() ? 1 : -1) * (0.0005f + rnd.nextFloat() * 0.002f)));
for (int j = 2; j < t.bytes; j++) s.set(ValueLayout.JAVA_BYTE, o + j, (byte) (t == Type.Q8_0 ? rnd.nextInt(255) - 127 : rnd.nextInt(256)));
}
return s;
}
static boolean test(Kernels one, Kernels four, Arena arena, Type t, int rows, int cols, Random rnd) throws Exception {
MemorySegment w = randomTensor(arena, t, rows, cols, rnd);
float[] x = new float[cols], a = new float[rows], b = new float[rows];
for (int i = 0; i < cols; i++) x[i] = (float) rnd.nextGaussian();
MemorySegment xq = t == Type.F32 ? MemorySegment.NULL : arena.allocate(Type.Q8_0.rowBytes(cols));
if (t != Type.F32) Kernels.quantizeQ8_0(x, xq);
one.matvec(t, w, x, a, rows);
four.matvec(t, w, x, b, rows);
int sameBits = 0;
double maxRel = 0;
for (int r = 0; r < rows; r++) {
long off = r * t.rowBytes(cols);
float ref = Reference.dot(t, w, off, x, xq);
if (Float.floatToIntBits(a[r]) == Float.floatToIntBits(ref)) sameBits++;
if (t == Type.F32) {
double scale = 0;
for (int i = 0; i < cols; i++) scale += Math.abs(w.get(Kernels.F32, off + 4L * i) * x[i]);
maxRel = Math.max(maxRel, Math.abs(a[r] - ref) / scale);
}
}
boolean threadsEqual = Arrays.equals(a, b);
boolean pass = threadsEqual && (t == Type.F32 ? maxRel < 1e-6 : sameBits == rows);
System.out.printf(Locale.ROOT, "%-5s %3d x %-5d %4d of %-4d %9.1e %-6b %s%n",
t, rows, cols, sameBits, rows, maxRel, threadsEqual, pass ? "PASS" : "FAIL");
return pass;
}
public static void main(String[] args) throws Exception {
Random rnd = new Random(19);
int passed = 0, total = 0;
try (Arena arena = Arena.ofShared(); Kernels one = new Kernels(1); Kernels four = new Kernels(4)) {
System.out.println("type shape same bits max rel 1 = 4");
for (Type t : Type.values())
for (int cols : t == Type.F32 ? new int[] {7, 100, 4100} : new int[] {32, 96, 4096}) {
total++;
if (test(one, four, arena, t, 130, cols, rnd)) passed++;
}
total++;
try { Type.Q4_0.rowBytes(100); } catch (IllegalArgumentException e) { passed++; System.out.println("rejected: " + e.getMessage()); }
}
System.out.println("tests passed: " + passed + " of " + total);
}
}type shape same bits max rel 1 = 4
F32 130 x 7 130 of 130 0.0e+00 true PASS
F32 130 x 100 25 of 130 7.9e-08 true PASS
F32 130 x 4100 4 of 130 1.2e-07 true PASS
Q8_0 130 x 32 130 of 130 0.0e+00 true PASS
Q8_0 130 x 96 130 of 130 0.0e+00 true PASS
Q8_0 130 x 4096 130 of 130 0.0e+00 true PASS
Q4_0 130 x 32 130 of 130 0.0e+00 true PASS
Q4_0 130 x 96 130 of 130 0.0e+00 true PASS
Q4_0 130 x 4096 130 of 130 0.0e+00 true PASS
rejected: Q4_0 row length 100 is not a multiple of 32
tests passed: 10 of 10Each row of the table is one test. The shape column is rows by columns; 130 rows means two full 64-row chunks and one chunk of 2 rows, so the last chunk is short in every test.
For F32, the row lengths 7, 100, and 4,100 are not multiples of 8, so each one runs the scalar tail. With 7 columns, the whole row is tail: loopBound(7) is 0, the kernel adds 7 products in the same order as the reference, and all 130 rows have the same bits. With 100 and 4,100 columns, the vector kernel adds in a different order, so only some rows match bit for bit (25 and 4 of 130). The largest difference is about 1e-7 of the sum of absolute products, far inside the tolerance of 1e-6.
For Q8_0 and Q4_0, every row of every shape matches the scalar reference exactly, as Examples 7 and 10 predicted. The 32-column case is a single block per row, and the 96-column case is three blocks.
The 1 = 4 column shows that 1 and 4 threads give equal arrays for every type, as in Example 11. The last test checks that a Q4_0 row of 100 values is rejected with a message that names the type and the length. All 10 tests pass.
The library leaves out SPECIES_PREFERRED, float accumulation across blocks, kernels that process several rows per call, and K-quant kernels (the newer, more complex quantization formats in ggml). The capstone uses the class as it is.
Example 13: speed and the bandwidth limit
Correct kernels are only useful if they are fast, so this example measures speed. A decode step reads every weight once, so the useful measure for a matrix-vector kernel is effective bandwidth: the weight bytes it processes per second, here in gigabytes per second (GB/s). If a kernel reaches the memory bandwidth of the machine, memory is the limit, and only reading fewer bytes would make it faster. If it stays far below, the kernel is limited by the instructions it runs.
The benchmark runs matvec from Project 19 over a stack of 24 matrices of 4,096 by 4,096 values, like the layers of a model, with 1, 2, and 4 threads. One pass is one matrix-vector product per matrix, the work of one decode step. Each configuration runs 3 warm-up passes (so the JIT compiler has finished its work) and keeps the fastest of the next 7. It also times the scalar reference kernels on one matrix. The program prints timings and needs the KernelLibrary class from Example 12 compiled next to it, so it is not run by the checker.
import java.lang.foreign.Arena;
import java.lang.foreign.MemorySegment;
import java.util.Locale;
import java.util.Random;
public class KernelBench {
public static void main(String[] args) throws Exception {
int rows = 4096, cols = 4096, mats = Integer.parseInt(args[0]);
Random rnd = new Random(13);
float[] x = new float[cols], out = new float[rows];
for (int i = 0; i < cols; i++) x[i] = (float) rnd.nextGaussian();
System.out.println("type MB threads ms/pass GB/s");
for (KernelLibrary.Type t : KernelLibrary.Type.values()) {
try (Arena arena = Arena.ofShared()) {
MemorySegment[] w = new MemorySegment[mats];
for (int m = 0; m < mats; m++) w[m] = KernelLibrary.randomTensor(arena, t, rows, cols, rnd);
long bytes = mats * rows * t.rowBytes(cols);
for (int threads : new int[] {1, 2, 4}) {
try (KernelLibrary.Kernels k = new KernelLibrary.Kernels(threads)) {
double best = Double.MAX_VALUE;
for (int rep = 0; rep < 10; rep++) { // 3 warm-up passes, keep the fastest of 7
long t0 = System.nanoTime();
for (MemorySegment s : w) k.matvec(t, s, x, out, rows);
if (rep >= 3) best = Math.min(best, (System.nanoTime() - t0) / 1e6);
}
System.out.printf(Locale.ROOT, "%-5s %-7.1f %-8d %-8.1f %.1f%n", t, bytes / 1e6, threads, best, bytes / best / 1e6);
}
}
}
}
for (KernelLibrary.Type t : new KernelLibrary.Type[] {KernelLibrary.Type.Q8_0, KernelLibrary.Type.Q4_0}) {
try (Arena arena = Arena.ofShared()) { // scalar reference on one matrix
MemorySegment w = KernelLibrary.randomTensor(arena, t, rows, cols, rnd);
MemorySegment xq = arena.allocate(KernelLibrary.Type.Q8_0.rowBytes(cols));
KernelLibrary.Kernels.quantizeQ8_0(x, xq);
double best = Double.MAX_VALUE;
for (int rep = 0; rep < 8; rep++) {
long t0 = System.nanoTime();
for (int r = 0; r < rows; r++) out[r] = KernelLibrary.Reference.dot(t, w, r * t.rowBytes(cols), x, xq);
if (rep >= 3) best = Math.min(best, (System.nanoTime() - t0) / 1e6);
}
long bytes = rows * t.rowBytes(cols);
System.out.printf(Locale.ROOT, "scalar %-5s %.1f MB, 1 thread: %.1f ms, %.2f GB/s%n", t, bytes / 1e6, best, bytes / best / 1e6);
}
}
}
}javac --add-modules jdk.incubator.vector -d out KernelLibrary.java KernelBench.java
java --add-modules jdk.incubator.vector -cp out KernelBench 24Representative output from the 4-core reference machine (timings differ by machine and by run). Other processes were using the cores during the measurement; the load average was about 10. The columns are the matrix type, the total weight size in megabytes, the thread count, the milliseconds per pass, and the effective bandwidth.
type MB threads ms/pass GB/s
F32 1610.6 1 148.5 10.8
F32 1610.6 2 80.3 20.1
F32 1610.6 4 40.8 39.4
Q8_0 427.8 1 135.8 3.2
Q8_0 427.8 2 72.3 5.9
Q8_0 427.8 4 55.6 7.7
Q4_0 226.5 1 176.5 1.3
Q4_0 226.5 2 86.9 2.6
Q4_0 226.5 4 43.3 5.2
scalar Q8_0 17.8 MB, 1 thread: 12.7 ms, 1.40 GB/s
scalar Q4_0 9.4 MB, 1 thread: 16.8 ms, 0.56 GB/sEvery row reads more data than the 260 MB L3 cache of this machine (the L3 cache is fast memory on the CPU chip, shared by all cores), so the weights come from main memory: 1.6 GB for F32, 428 MB for Q8_0, and 227 MB for Q4_0, plus the cache traffic of the other processes.
The single-thread rows are the clearest. F32 processes 10.8 GB/s, Q8_0 3.2 GB/s, and Q4_0 1.3 GB/s. In time per pass, all three are close: 148.5 ms, 135.8 ms, and 176.5 ms for the same 402,653,184 weights (24 times 4,096 times 4,096), which is about 0.37, 0.34, and 0.44 nanoseconds per weight. On one thread, the time depends on the number of weights, not on their size in bytes. Each kernel is limited by the instructions it runs per weight:
- F32 needs one load and one FMA per weight.
- Q8_0 needs a widening conversion, an integer multiply and add, and a share of the per-block reduction.
- Q4_0 adds the nibble unpacking on top of that.
Quantization makes these weights 3.8 and 7.1 times smaller (4 bytes against 34/32 and 18/32 bytes per weight), but it does not make the Java kernels faster on one thread. The smaller size would only pay off if memory were the limit.
The multi-threaded rows show that memory was not the limit at 4 threads either. F32 goes from 10.8 to 20.1 to 39.4 GB/s, nearly doubling with each doubling of threads. If the memory system could deliver only, for example, 20 GB/s, the 4-thread F32 row would stay near that value. It reached 39.4 GB/s, so this machine delivers at least that much. The scalar float loop in CPUs, GPUs, and Memory for Inference reached 12.8 GB/s with four threads on the same machine. That loop was limited by its own arithmetic, and the vector kernel moves three times as many bytes per second. The quantized kernels also gain with threads (Q4_0 from 1.3 to 5.2 GB/s), but each thread still runs out of instructions long before memory runs out of bytes.
The roofline model from FLOPs, Bandwidth, and the Roofline Model predicts decode speed from bytes and bandwidth, and it assumes a kernel that keeps up with memory. These Java kernels do not, so their speed is better described in weights per second per thread. On a machine with fewer memory channels per core (a typical desktop has two channels shared by all cores), the F32 kernel reaches the bandwidth limit with fewer threads, and there the quantized kernels win by reading fewer bytes.
The scalar lines show what the Vector API gained. The scalar Q8_0 kernel takes 12.7 ms per matrix, and the vector kernel about 5.7 ms (135.8 ms divided by 24 matrices), which is 2.2 times faster. For Q4_0, 16.8 ms against about 7.4 ms is 2.3 times faster.
In three other runs on this shared machine, the single-thread rows stayed within about 20% of these values, while some 4-thread rows were up to 3.4 times slower when other processes took cores. A measurement for a decision repeats the run on an idle machine with a harness such as JMH (Benchmarking with JMH). llama.cpp's kernels for the same formats use instructions that the Vector API does not expose, such as the byte multiply-add pairs of Example 7; the capstone compares the two engines with llama-bench.
Example 14: JFR on the kernels
When a program is slower than expected, a profiler shows where the time goes. JDK Flight Recorder (JFR) is the profiler built into the JVM. It records events into a file with low overhead. The events include stack samples of running threads taken at a fixed interval (a snapshot of what each thread is executing right now), allocation samples, and garbage collections. A method that appears in many samples is a method where the program spends a lot of time.
-XX:StartFlightRecording starts a recording when the JVM starts. settings=profile takes a Java stack sample every 10 ms instead of every 20 ms (the values in lib/jfr/profile.jfc and lib/jfr/default.jfc of JDK 25). The jfr tool in the JDK prints predefined views of a recording. The commands below record the benchmark from Example 13 with 8 matrices per type and print the methods with the most samples.
java --add-modules jdk.incubator.vector -XX:StartFlightRecording=filename=kernels.jfr,settings=profile -cp out KernelBench 8
jfr view hot-methods kernels.jfrOutput of jfr view hot-methods on the reference machine (the first 12 of the listed methods; sample counts differ by run):
Java Methods that Execute the Most
Method Samples Percent
------------------------------------------------------------------------------------------------------- ------- -------
KernelLibrary.randomTensor(Arena, KernelLibrary$Type, int, int, Random) 320 29.57%
KernelLibrary$Kernels.dotQ8_0(MemorySegment, long, MemorySegment, int) 132 12.20%
KernelLibrary$Kernels.dotQ4_0(MemorySegment, long, MemorySegment, int) 119 11.00%
KernelLibrary$Kernels.dotF32(MemorySegment, long, float[]) 111 10.26%
java.util.Random.nextInt(int) 102 9.43%
jdk.internal.foreign.MemorySessionImpl.checkValidStateRaw() 90 8.32%
jdk.internal.foreign.SegmentFactories.initNativeMemory(long, long) 34 3.14%
java.util.Random.next(int) 23 2.13%
KernelLibrary$Reference.dotQ4_0(MemorySegment, long, MemorySegment, int) 15 1.39%
KernelLibrary$Kernels.lambda$matvec$0(...) 15 1.39%
KernelLibrary$Reference.dotQ8_0(MemorySegment, long, MemorySegment, int) 14 1.29%
java.util.Objects.checkIndex(long, long) 9 0.83%Each line is a method and the number of stack samples in which it was the top frame, the method actually running when the sample was taken.
The three vector kernels together take 33.5% of the samples: dotQ8_0 12.2%, dotQ4_0 11.0%, and dotF32 10.3%. The scalar reference kernels take under 3%, because they ran on only one matrix each.
The largest single entry is randomTensor, 29.6%, and several entries below it belong to it as well. This is the code that fills the test matrices, not the kernels:
Random.nextIntandRandom.nextproduce the random bytes.SegmentFactories.initNativeMemoryzeroes each newly allocated off-heap segment, becauseArena.allocatereturns zeroed memory.MemorySessionImpl.checkValidStateRawis the check that a segment's arena is still open. Everygetandseton a segment runs it.
To see where checkValidStateRaw was called from, jfr print --events jdk.ExecutionSample --stack-depth 12 kernels.jfr prints each sample with its stack. Grouping the samples of checkValidStateRaw by the first KernelLibrary frame shows that all 90 came from randomTensor, which writes the test tensors one value at a time. In the kernels, checks take only a few samples: Objects.checkIndex, the index check of vector loads, has 9 samples, called from dotQ8_0 and dotQ4_0. Filling hundreds of megabytes with test data is a cost of the benchmark; an engine maps the model file and pays no such cost.
The bottom lines of the full listing (not shown) include ByteVector.toArray, ByteVector.toLongArray, and ByteVector.bOpTemplate with a few samples each, called from dotQ8_0 and dotQ4_0. These are the Java fallback implementations of vector operations. The JIT compiler replaces Vector API calls with SIMD instructions only when it compiles a method with its optimizing compiler (C2). Until then, each vector operation runs as a Java loop over a temporary array. A few fallback samples are expected during warm-up. A profile in which such methods stay near the top after warm-up means the kernel is not being compiled to SIMD instructions. That can happen when a vector is stored in a field or collection, or when the species is not a constant the compiler can see. The next example measures the warm-up effect directly.
Example 15: allocation before and after warm-up
Example 14 hinted that before the JIT compiles a kernel, each vector operation creates temporary objects. A compiled vector kernel keeps its vectors in CPU registers and allocates nothing. The fallback code before compilation allocates on every operation. This matters because allocation creates work for the garbage collector and slows the loop.
com.sun.management.ThreadMXBean.getCurrentThreadAllocatedBytes() returns the number of bytes the current thread has allocated so far, so the difference around a call is that call's allocation. The program runs each kernel of Project 19 over 1,024 rows for 30 passes on one thread, and prints the bytes allocated per row for selected passes. It prints JIT-dependent numbers, so it is not run by the checker.
import java.lang.foreign.Arena;
import java.lang.foreign.MemorySegment;
import java.lang.management.ManagementFactory;
import java.util.Random;
import com.sun.management.ThreadMXBean;
public class KernelAllocation {
public static void main(String[] args) {
int rows = 1024, cols = 4096;
Random rnd = new Random(15);
ThreadMXBean mx = (ThreadMXBean) ManagementFactory.getThreadMXBean();
try (Arena arena = Arena.ofShared()) {
float[] x = new float[cols];
for (int i = 0; i < cols; i++) x[i] = (float) rnd.nextGaussian();
MemorySegment xq = arena.allocate(KernelLibrary.Type.Q8_0.rowBytes(cols));
KernelLibrary.Kernels.quantizeQ8_0(x, xq);
for (KernelLibrary.Type t : KernelLibrary.Type.values()) {
MemorySegment w = KernelLibrary.randomTensor(arena, t, rows, cols, rnd);
long rb = t.rowBytes(cols);
StringBuilder line = new StringBuilder(t + " bytes allocated per row, passes 1 to 30:");
for (int pass = 1; pass <= 30; pass++) {
long before = mx.getCurrentThreadAllocatedBytes();
for (int r = 0; r < rows; r++) {
switch (t) {
case F32 -> KernelLibrary.Kernels.dotF32(w, r * rb, x);
case Q8_0 -> KernelLibrary.Kernels.dotQ8_0(w, r * rb, xq, cols);
case Q4_0 -> KernelLibrary.Kernels.dotQ4_0(w, r * rb, xq, cols);
}
}
long perRow = (mx.getCurrentThreadAllocatedBytes() - before) / rows;
if (pass <= 3 || pass % 10 == 0) line.append(" [").append(pass).append("] ").append(perRow);
}
System.out.println(line);
}
}
}
}Output from the reference machine (pass 1 differs by machine and JVM version). The numbers in brackets are pass numbers, followed by bytes allocated per row:
F32 bytes allocated per row, passes 1 to 30: [1] 55003 [2] 48 [3] 48 [10] 48 [20] 48 [30] 48
Q8_0 bytes allocated per row, passes 1 to 30: [1] 515369 [2] 0 [3] 0 [10] 0 [20] 0 [30] 0
Q4_0 bytes allocated per row, passes 1 to 30: [1] 473690 [2] 0 [3] 0 [10] 0 [20] 0 [30] 0In pass 1, every kernel allocates a lot: about 55 KB per row for F32 and about 500 KB per row for the quantized kernels. The code is not yet compiled by C2, so every fromMemorySegment, convertShape, mul, and add creates a vector object and temporary arrays in the fallback path. The quantized kernels allocate more because they run more vector operations per row (8 conversions per block, 128 blocks per row).
From pass 2 on, the Q8_0 and Q4_0 kernels allocate 0 bytes per row. C2 compiled them with the vector operations as intrinsics (replaced by single CPU instructions), and escape analysis removed every vector object. Escape analysis is a JIT optimization that notices an object never leaves the method and keeps its fields in registers instead of allocating it. Pass 2 starts after 1,024 calls in pass 1, which is enough for the JIT to compile a method that hot.
The F32 kernel keeps allocating 48 bytes per row. That matches one float[8] (a 16-byte array header plus 32 bytes of data): the acc.toArray() call that adds the 8 lanes in a fixed order (Example 6). The array escapes into the summing loop, so escape analysis does not remove it. Over a 4,096-row matrix that is 196,608 bytes per matrix-vector product, which the JVM allocates quickly but which the garbage collector must clear later. reduceLanes(ADD) would not allocate, at the cost of a hardware-dependent order of additions. intoArray into a buffer that each thread allocates once would keep both the fixed order and zero allocation.
Benchmarks discard their first passes for this reason: pass 1 runs different code from pass 2. After warm-up, a kernel that still allocates per call either uses an API that returns a new object or has a vector the compiler could not keep in registers.
Interview points
- A quantized dot product sums integer products inside each block and multiplies once by both FP16 scales. For Q4_0 and Q8_0 weights, ggml quantizes the activations to Q8_0 once per matrix-vector product, so both sides of the block sum are integers.
- When the integer sum is exact and the float operations run in the same order, a vector kernel gives the same bits as the scalar kernel. It can then be tested for equality instead of with a tolerance.
- Float kernels need a tolerance. The error is measured against the sum of absolute products rather than against the result, because the result can be close to zero even when the terms are large.
reduceLanes(ADD)on floats has an unspecified order of additions. Adding the lanes in a fixed loop gives the same bits on every machine.- Use a logical shift (
LSHR) to extract high nibbles from byte lanes. An arithmetic shift copies the sign bit into bytes whose top bit is set and produces values outside the valid range. - FMA rounds once instead of twice. Several accumulators break the dependency chain between FMAs, so the CPU can overlap them.
- Parallelize a matrix-vector product by rows. When each row is computed by one thread, the results are identical for any thread count.
- A kernel is memory-bound when its bytes per second reach the bandwidth limit. Quantized kernels written with the Vector API are often instruction-bound, because widening and per-block reductions cost more instructions per byte.
- Segments shared by pool threads need a shared arena. A confined arena throws
WrongThreadExceptionwhen another thread uses it. - In a JFR profile, Vector API fallback methods (
ldOp,bOpTemplate,toArray) near the top after warm-up mean the kernel is not being compiled to SIMD instructions. - A compiled vector kernel allocates nothing per call. Measure allocation after warm-up, because the first calls run the allocating fallback path.
- On one thread, the Java kernels took about the same time per weight for F32, Q8_0, and Q4_0. Quantization saved bytes, not time, because the number of instructions per weight was the limit.
Summary
ggml_vec_dot_q8_0_q8_0_genericcomputessumf += sumi * (dx * dy)per block.ggml_vec_dot_q4_0_q8_0_genericcomputessumf += sumi * dx * dywith two integer sums, low nibbles against activations 0 to 15 and high nibbles against activations 16 to 31.- Q4_0 and Q8_0 weights both use Q8_0 as
vec_dot_typein the CPU backend, which means the activation vector is converted to Q8_0 before the dot product. FloatVector.SPECIES_256has 8 float lanes.SPECIES_PREFERREDis 512 bits on this AVX-512 machine.loopBound(n)gives the end of the full-vector part of a loop, andindexInRange(i, n)gives a mask for the tail.fmacomputesa * b + cwith one rounding, whilemulfollowed byaddrounds twice.convertShape(B2I, IntVector.SPECIES_256, part)widens 8 bytes to 8 sign-extended ints.castShape(FloatVector.SPECIES_256, part)widens 8 bytes to 8 floats.- The vector Q8_0 and Q4_0 kernels with integer block sums matched the scalar references bit for bit. The F32 kernels and the kernel with F32 activations matched within about 1e-7 of the sum of absolute products.
ByteVector.fromMemorySegmentloads from any byte offset of a mapped or off-heap segment, with bounds and lifetime checks.- ggml splits a decode-step matrix-vector product into 64-row chunks handed out by an atomic counter. The Java version gives identical outputs for 1, 2, 4, and 8 threads.
- On the reference machine, one thread processed 10.8 GB/s of F32, 3.2 GB/s of Q8_0, and 1.3 GB/s of Q4_0, which is about 0.34 to 0.44 ns per weight for all three. F32 reached 39.4 GB/s with 4 threads, so memory was not the limit.
- The vector Q8_0 and Q4_0 kernels were about 2.2 and 2.3 times faster than the scalar kernels on one thread.
- Project 19,
KernelLibrary, passed 10 of 10 tests: bit-identical quantized kernels, F32 within tolerance, thread-count independence, and rejection of bad row lengths. jfr view hot-methodslists methods by top-frame samples.jfr print --events jdk.ExecutionSample --stack-depth Nattributes a hot method to its callers.- Before C2 compiles them, the kernels allocated about 55 KB (F32) to 515 KB (Q8_0) per row. After warm-up, Q8_0 and Q4_0 allocate 0 bytes, and F32 allocates 48 bytes (the
toArraylane buffer).