mirror of
https://github.com/ggml-org/llama.cpp.git
synced 2026-09-11 17:37:10 +02:00
Compare commits
| Author | SHA1 | Date | |
|---|---|---|---|
|
|
8172e6577a | ||
|
|
43f3dda623 | ||
|
|
5bda51bfbc | ||
|
|
3bcfeb700f | ||
|
|
1dfe94e048 |
@@ -134,6 +134,8 @@ set_target_properties(${TARGET} PROPERTIES
|
||||
target_include_directories(${TARGET} PUBLIC .)
|
||||
target_link_libraries (${TARGET} PUBLIC vendor::nlohmann vendor::sheredom)
|
||||
target_compile_features (${TARGET} PUBLIC cxx_std_17)
|
||||
target_precompile_headers (${TARGET} PRIVATE common.h)
|
||||
target_precompile_headers (${TARGET} PRIVATE chat.h)
|
||||
|
||||
if (LLAMA_SUBPROCESS)
|
||||
target_compile_definitions(${TARGET} PUBLIC LLAMA_SUBPROCESS)
|
||||
|
||||
@@ -1493,7 +1493,7 @@ struct common_speculative_impl_draft_mtp : public common_speculative_impl {
|
||||
|
||||
const int32_t n_tokens = batch_in.n_tokens;
|
||||
|
||||
// remember the frist and last batch index for each sequence
|
||||
// remember the first and last batch index for each sequence
|
||||
std::fill(i_batch_beg.begin(), i_batch_beg.end(), -1);
|
||||
std::fill(i_batch_end.begin(), i_batch_end.end(), -1);
|
||||
|
||||
|
||||
@@ -0,0 +1,122 @@
|
||||
## Build profiling
|
||||
This page is a working document for analyzing the current build and try to
|
||||
identify ways to improve the build time.
|
||||
|
||||
### Requirements
|
||||
The profiling script requires clang to be used as the compiler tool chain and
|
||||
also requires that ClangBuildAnalyzer is installed.
|
||||
|
||||
Mac:
|
||||
```console
|
||||
brew install clang-build-analyzer
|
||||
```
|
||||
|
||||
Linux:
|
||||
```console
|
||||
git clone https://github.com/aras-p/ClangBuildAnalyzer.git
|
||||
cd ClangBuildAnalyzer
|
||||
cmake -B build -DCMAKE_BUILD_TYPE=Release
|
||||
cmake --build build -j$(nproc)
|
||||
sudo cp build/ClangBuildAnalyzer /usr/local/bin/
|
||||
```
|
||||
|
||||
Windows: install LLVM/clang and Ninja (e.g. via the
|
||||
[LLVM releases page](https://github.com/llvm/llvm-project/releases) and
|
||||
`winget install Ninja-build.Ninja`), then build ClangBuildAnalyzer the same
|
||||
way as on Linux:
|
||||
```console
|
||||
git clone https://github.com/aras-p/ClangBuildAnalyzer.git
|
||||
cd ClangBuildAnalyzer
|
||||
cmake -B build -G Ninja -DCMAKE_C_COMPILER=clang -DCMAKE_CXX_COMPILER=clang++ -DCMAKE_BUILD_TYPE=Release
|
||||
cmake --build build --config Release
|
||||
```
|
||||
Then add `ClangBuildAnalyzer\build` to `PATH`.
|
||||
|
||||
### Usage
|
||||
Mac/Linux:
|
||||
```console
|
||||
$ ./scripts/build-profile.sh
|
||||
```
|
||||
|
||||
Windows:
|
||||
```console
|
||||
> .\scripts\build-profile.ps1
|
||||
```
|
||||
|
||||
Both accept `--full`/`-Full` (include Server, Tools, and Tests) and a jobs
|
||||
override (`-jN` / `-Jobs N`).
|
||||
|
||||
Note: on Windows, `cmake` defaults to the Visual Studio generator, which
|
||||
ignores `CMAKE_C_COMPILER`/`CMAKE_CXX_COMPILER` and silently falls back to
|
||||
MSVC. `build-profile.ps1` passes `-G Ninja` so clang is actually used, this
|
||||
is required on ARM64.
|
||||
|
||||
### Linux (Ubuntu 24.04)
|
||||
|
||||
Environment:
|
||||
- Clang: 18.1.3 (Ubuntu clang version 18.1.3 (1ubuntu1))
|
||||
- libstdc++: GCC 13.3.0 (Ubuntu 13.3.0-6ubuntu2~24.04.1)
|
||||
- Target: x86_64-pc-linux-gnu
|
||||
|
||||
```console
|
||||
+------------------------+-----+------------+------------+------------+
|
||||
| Build | TUs | Frontend | Backend | Total |
|
||||
+------------------------+-----+------------+------------+------------+
|
||||
| Minimal, master | 249 | 468.2 s | 270.3 s | 738.5 s |
|
||||
| Minimal, with PCH | 253 | 177.1 s | 265.8 s | 442.9 s |
|
||||
| Full, master | 396 | 811.0 s | 692.2 s | 1,503.2 s |
|
||||
| Full, with PCH | 405 | 380.0 s | 664.7 s | 1,044.7 s |
|
||||
| Full, with PCH + UB | 264 | 357.7 s | 635.7 s | 993.4 s |
|
||||
+------------------------+-----+------------+------------+------------+
|
||||
|
||||
PCH = precompiled header.
|
||||
Full = includes building Server, Tools, and Tests.
|
||||
UB = unity build for models
|
||||
```
|
||||
Note that the number of translation units (TUs) increases when using precompiled
|
||||
headers — each PCH target adds one extra TU for the precompilation step itself.
|
||||
|
||||
### Mac (Apple M3)
|
||||
|
||||
Environment:
|
||||
- Clang: Apple clang version 17.0.0 (clang-1700.3.19.1)
|
||||
- libc++: ships with Apple clang 17.0.0 (Xcode toolchain)
|
||||
- Target: arm64-apple-macosx15.6
|
||||
|
||||
```console
|
||||
+------------------------+-----+------------+------------+------------+
|
||||
| Build | TUs | Frontend | Backend | Total |
|
||||
+------------------------+-----+------------+------------+------------+
|
||||
| Minimal, master | 256 | 154.5 s | 94.8 s | 249.3 s |
|
||||
| Minimal, with PCH | 261 | 65.9 s | 90.0 s | 155.9 s |
|
||||
| Full, master | 407 | 265.7 s | 209.7 s | 475.4 s |
|
||||
| Full, with PCH | 414 | 154.6 s | 197.5 s | 352.1 s |
|
||||
| Full, with PCH + UB | 274 | 143.0 s | 192.2 s | 335.2 s |
|
||||
+------------------------+-----+------------+------------+------------+
|
||||
|
||||
PCH = precompiled header.
|
||||
Full = includes building Server, Tools, and Tests.
|
||||
UB = unity build for models
|
||||
```
|
||||
|
||||
### Windows (ARM64)
|
||||
|
||||
Environment:
|
||||
- Clang: clang version 22.1.8 (LLVM, `C:\Program Files\LLVM`)
|
||||
- STL: MSVC STL (Visual Studio 2022 Build Tools 14.44.35207)
|
||||
- Target: aarch64-pc-windows-msvc
|
||||
|
||||
```console
|
||||
+------------------------+-----+------------+------------+------------+
|
||||
| Build | TUs | Frontend | Backend | Total |
|
||||
+------------------------+-----+------------+------------+------------+
|
||||
| Minimal, master | 249 | 159.4 s | 82.2 s | 241.6 s |
|
||||
| Full, master | 373 | 337.2 s | 167.4 s | 504.6 s |
|
||||
| Minimal, with PCH + UB | 113 | 62.3 s | 82.4 s | 144.7 s |
|
||||
| Full, with PCH + UB | 240 | 233.0 s | 185.1 s | 418.1 s |
|
||||
+------------------------+-----+------------+------------+------------+
|
||||
|
||||
PCH = precompiled header.
|
||||
Full = includes building Server, Tools, and Tests.
|
||||
UB = unity build for models
|
||||
```
|
||||
@@ -1705,6 +1705,10 @@ static enum ggml_status ggml_backend_sched_compute_splits(ggml_backend_sched_t s
|
||||
ggml_tensor * ids_tensor = node->src[2];
|
||||
ggml_backend_t ids_backend = split_backend;
|
||||
|
||||
if (ggml_nelements(ids_tensor) == 0) {
|
||||
continue;
|
||||
}
|
||||
|
||||
// if the ids tensor is also an input of the split, it may not have been copied yet to the split backend
|
||||
// in that case, we use the original ids tensor
|
||||
for (int i = input_id + 1; i < split->n_inputs; i++) {
|
||||
|
||||
@@ -675,6 +675,12 @@ function(ggml_add_cpu_backend_variant_impl tag_name)
|
||||
target_compile_options(${GGML_CPU_NAME} PRIVATE ${ARCH_FLAGS})
|
||||
target_compile_definitions(${GGML_CPU_NAME} PRIVATE ${ARCH_DEFINITIONS})
|
||||
|
||||
if (CMAKE_C_COMPILER_ID STREQUAL "GNU" AND NOT GGML_SYSTEM_ARCH STREQUAL "x86")
|
||||
message(STATUS "Skipping PCH for ${GGML_CPU_NAME}: GCC PCH is only enabled for x86 (arch: ${GGML_SYSTEM_ARCH})")
|
||||
else()
|
||||
target_precompile_headers(${GGML_CPU_NAME} PRIVATE ggml-impl.h)
|
||||
endif()
|
||||
|
||||
if (EMSCRIPTEN)
|
||||
set_target_properties(${GGML_CPU_NAME} PROPERTIES COMPILE_FLAGS "-msimd128")
|
||||
endif()
|
||||
|
||||
@@ -18,7 +18,15 @@
|
||||
#endif
|
||||
#endif
|
||||
|
||||
// -Winterference-size was introduced in GCC 12
|
||||
#if defined(__cplusplus) && defined(__GNUC__) && !defined(__clang__) && __GNUC__ >= 12
|
||||
#pragma GCC diagnostic push
|
||||
#pragma GCC diagnostic ignored "-Winterference-size"
|
||||
#endif
|
||||
static const size_t CACHE_LINE_SIZE_F32 = CACHE_LINE_SIZE/sizeof(float);
|
||||
#if defined(__cplusplus) && defined(__GNUC__) && !defined(__clang__) && __GNUC__ >= 12
|
||||
#pragma GCC diagnostic pop
|
||||
#endif
|
||||
|
||||
// Work buffer size for im2col operations in CONV2D
|
||||
#define GGML_IM2COL_WORK_SIZE (16 * 1024 * 1024)
|
||||
|
||||
@@ -496,6 +496,13 @@ kernel void kernel_mul_mm_id(
|
||||
+ args.nb11*i11
|
||||
+ args.nb10*iy);
|
||||
|
||||
// skip the upper half of the token tile when the expert did not fill it
|
||||
constexpr short NR1H = NR1/2;
|
||||
|
||||
const bool has_hi = nr1 > NR1H;
|
||||
|
||||
const short lb1 = (short) tiitg/NL1; // 0 .. NR1-1, this thread's row of the B tile
|
||||
|
||||
#ifndef GGML_METAL_HAS_TENSOR
|
||||
S0_8x8 ma[4];
|
||||
S1_8x8 mb[2];
|
||||
@@ -505,15 +512,22 @@ kernel void kernel_mul_mm_id(
|
||||
for (short i = 0; i < 8; i++){
|
||||
mc[i] = make_filled_simdgroup_matrix<float, 8>(0.f);
|
||||
}
|
||||
|
||||
// simdgroups 2,3 own rows NR1H..NR1-1
|
||||
const bool sg_active = has_hi || sgitg < 2;
|
||||
#else
|
||||
auto tA = tensor<threadgroup S0, dextents<int32_t, 2>, tensor_inline>(sa, dextents<int32_t, 2>(NK, NR0));
|
||||
auto tB = tensor<threadgroup S1, dextents<int32_t, 2>, tensor_inline>(sb, dextents<int32_t, 2>(NR1, NK ));
|
||||
auto tA = tensor<threadgroup S0, dextents<int32_t, 2>, tensor_inline>(sa, dextents<int32_t, 2>(NK, NR0));
|
||||
|
||||
// sb is [NR1][NK] row-major
|
||||
auto tB0 = tensor<threadgroup S1, dextents<int32_t, 2>, tensor_inline>(sb, dextents<int32_t, 2>(NK, NR1H));
|
||||
auto tB1 = tensor<threadgroup S1, dextents<int32_t, 2>, tensor_inline>(sb + NR1H*NK, dextents<int32_t, 2>(NK, NR1H));
|
||||
|
||||
mpp::tensor_ops::matmul2d<
|
||||
mpp::tensor_ops::matmul2d_descriptor(NR1, NR0, NK, false, true, false, mpp::tensor_ops::matmul2d_descriptor::mode::multiply_accumulate),
|
||||
mpp::tensor_ops::matmul2d_descriptor(NR1H, NR0, NK, false, true, false, mpp::tensor_ops::matmul2d_descriptor::mode::multiply_accumulate),
|
||||
execution_simdgroups<4>> mm;
|
||||
|
||||
auto cT = mm.get_destination_cooperative_tensor<decltype(tA), decltype(tB), float>();
|
||||
auto cT0 = mm.get_destination_cooperative_tensor<decltype(tA), decltype(tB0), float>();
|
||||
auto cT1 = mm.get_destination_cooperative_tensor<decltype(tA), decltype(tB1), float>();
|
||||
#endif
|
||||
|
||||
for (int loop_k = 0; loop_k < args.ne00; loop_k += NK) {
|
||||
@@ -656,37 +670,45 @@ kernel void kernel_mul_mm_id(
|
||||
threadgroup_barrier(mem_flags::mem_threadgroup);
|
||||
|
||||
#ifndef GGML_METAL_HAS_TENSOR
|
||||
// load matrices from threadgroup memory and conduct outer products
|
||||
threadgroup const S0 * lsma = (sa + 4*64*(sgitg%2));
|
||||
threadgroup const S1 * lsmb = (sb + 2*64*(sgitg/2));
|
||||
if (sg_active) {
|
||||
// load matrices from threadgroup memory and conduct outer products
|
||||
threadgroup const S0 * lsma = (sa + 4*64*(sgitg%2));
|
||||
threadgroup const S1 * lsmb = (sb + 2*64*(sgitg/2));
|
||||
|
||||
FOR_UNROLL (short ik = 0; ik < NK/8; ik++) {
|
||||
simdgroup_barrier(mem_flags::mem_none);
|
||||
FOR_UNROLL (short ik = 0; ik < NK/8; ik++) {
|
||||
simdgroup_barrier(mem_flags::mem_none);
|
||||
|
||||
FOR_UNROLL (short i = 0; i < 4; i++) {
|
||||
simdgroup_load(ma[i], lsma + 64*i, 8, 0, false);
|
||||
FOR_UNROLL (short i = 0; i < 4; i++) {
|
||||
simdgroup_load(ma[i], lsma + 64*i, 8, 0, false);
|
||||
}
|
||||
|
||||
simdgroup_barrier(mem_flags::mem_none);
|
||||
|
||||
FOR_UNROLL (short i = 0; i < 2; i++) {
|
||||
simdgroup_load(mb[i], lsmb + 64*i, 8, 0, false);
|
||||
}
|
||||
|
||||
simdgroup_barrier(mem_flags::mem_none);
|
||||
|
||||
FOR_UNROLL (short i = 0; i < 8; i++){
|
||||
simdgroup_multiply_accumulate(mc[i], mb[i/4], ma[i%4], mc[i]);
|
||||
}
|
||||
|
||||
lsma += 8*64;
|
||||
lsmb += 4*64;
|
||||
}
|
||||
|
||||
simdgroup_barrier(mem_flags::mem_none);
|
||||
|
||||
FOR_UNROLL (short i = 0; i < 2; i++) {
|
||||
simdgroup_load(mb[i], lsmb + 64*i, 8, 0, false);
|
||||
}
|
||||
|
||||
simdgroup_barrier(mem_flags::mem_none);
|
||||
|
||||
FOR_UNROLL (short i = 0; i < 8; i++){
|
||||
simdgroup_multiply_accumulate(mc[i], mb[i/4], ma[i%4], mc[i]);
|
||||
}
|
||||
|
||||
lsma += 8*64;
|
||||
lsmb += 4*64;
|
||||
}
|
||||
#else
|
||||
auto sA = tA.slice(0, 0);
|
||||
auto sB = tB.slice(0, 0);
|
||||
auto sA = tA.slice(0, 0);
|
||||
auto sB0 = tB0.slice(0, 0);
|
||||
|
||||
mm.run(sB, sA, cT);
|
||||
mm.run(sB0, sA, cT0);
|
||||
|
||||
if (has_hi) {
|
||||
auto sB1 = tB1.slice(0, 0);
|
||||
|
||||
mm.run(sB1, sA, cT1);
|
||||
}
|
||||
#endif
|
||||
}
|
||||
|
||||
@@ -694,13 +716,20 @@ kernel void kernel_mul_mm_id(
|
||||
threadgroup_barrier(mem_flags::mem_threadgroup);
|
||||
|
||||
#ifdef GGML_METAL_HAS_TENSOR
|
||||
auto tC = tensor<threadgroup float, dextents<int32_t, 2>, tensor_inline>(sc, dextents<int32_t, 2>(NR0, NR1));
|
||||
cT.store(tC);
|
||||
#else
|
||||
threadgroup float * temp_str = ((threadgroup float *) shmem) + 32*(sgitg&1) + (16*(sgitg >> 1))*NR0;
|
||||
auto tC0 = tensor<threadgroup float, dextents<int32_t, 2>, tensor_inline>(sc, dextents<int32_t, 2>(NR0, NR1H));
|
||||
cT0.store(tC0);
|
||||
|
||||
for (short i = 0; i < 8; i++) {
|
||||
simdgroup_store(mc[i], temp_str + 8*(i%4) + 8*NR0*(i/4), NR0, 0, false);
|
||||
if (has_hi) {
|
||||
auto tC1 = tensor<threadgroup float, dextents<int32_t, 2>, tensor_inline>(sc + NR1H*NR0, dextents<int32_t, 2>(NR0, NR1H));
|
||||
cT1.store(tC1);
|
||||
}
|
||||
#else
|
||||
if (sg_active) {
|
||||
threadgroup float * temp_str = ((threadgroup float *) shmem) + 32*(sgitg&1) + (16*(sgitg >> 1))*NR0;
|
||||
|
||||
for (short i = 0; i < 8; i++) {
|
||||
simdgroup_store(mc[i], temp_str + 8*(i%4) + 8*NR0*(i/4), NR0, 0, false);
|
||||
}
|
||||
}
|
||||
#endif
|
||||
|
||||
|
||||
@@ -0,0 +1,136 @@
|
||||
# Compile-time profiling using clang -ftime-trace + ClangBuildAnalyzer.
|
||||
#
|
||||
# Usage:
|
||||
# .\scripts\build-profile.ps1 [-Full] [-Jobs N]
|
||||
#
|
||||
# -Full : include Server, Tools, and Tests (default: minimal build)
|
||||
# -Jobs : number of parallel jobs (default: all cores)
|
||||
#
|
||||
# Requires ClangBuildAnalyzer:
|
||||
# https://github.com/aras-p/ClangBuildAnalyzer
|
||||
|
||||
param(
|
||||
[switch]$Full,
|
||||
[int]$Jobs = [Environment]::ProcessorCount
|
||||
)
|
||||
|
||||
$ErrorActionPreference = "Stop"
|
||||
|
||||
$ScriptDir = Split-Path -Parent $MyInvocation.MyCommand.Path
|
||||
$RootDir = Split-Path -Parent $ScriptDir
|
||||
|
||||
if ($Full) {
|
||||
$BuildDir = Join-Path $RootDir "build-profile-full"
|
||||
$Report = Join-Path $BuildDir "profile-report-full.txt"
|
||||
} else {
|
||||
$BuildDir = Join-Path $RootDir "build-profile-baseline"
|
||||
$Report = Join-Path $BuildDir "profile-report.txt"
|
||||
}
|
||||
|
||||
$OutputBin = Join-Path $BuildDir "clang_analysis.bin"
|
||||
|
||||
if (-not (Get-Command clang++ -ErrorAction SilentlyContinue)) {
|
||||
Write-Error "clang++ not found"
|
||||
exit 1
|
||||
}
|
||||
|
||||
if (-not (Get-Command ninja -ErrorAction SilentlyContinue)) {
|
||||
Write-Error "ninja not found (required so cmake does not fall back to the Visual Studio/MSVC generator)"
|
||||
exit 1
|
||||
}
|
||||
|
||||
if (-not (Get-Command ClangBuildAnalyzer -ErrorAction SilentlyContinue)) {
|
||||
Write-Error "ClangBuildAnalyzer not found`n https://github.com/aras-p/ClangBuildAnalyzer/releases"
|
||||
exit 1
|
||||
}
|
||||
|
||||
$ClangVer = (clang++ --version | Select-Object -First 1)
|
||||
Write-Host "compiler : $ClangVer"
|
||||
Write-Host "build dir: $BuildDir"
|
||||
Write-Host "output : $OutputBin"
|
||||
Write-Host "jobs : $Jobs"
|
||||
Write-Host ""
|
||||
|
||||
if (Get-Command ccache -ErrorAction SilentlyContinue) {
|
||||
Write-Host "clearing ccache..."
|
||||
ccache -C -z
|
||||
}
|
||||
|
||||
$env:CCACHE_DISABLE = "1"
|
||||
|
||||
$TestsFlag = if ($Full) { "ON" } else { "OFF" }
|
||||
$ToolsFlag = if ($Full) { "ON" } else { "OFF" }
|
||||
$ServerFlag = if ($Full) { "ON" } else { "OFF" }
|
||||
|
||||
cmake --fresh `
|
||||
-S $RootDir `
|
||||
-B $BuildDir `
|
||||
-G "Ninja" `
|
||||
-DCMAKE_BUILD_TYPE=Release `
|
||||
-DCMAKE_C_COMPILER=clang `
|
||||
-DCMAKE_CXX_COMPILER=clang++ `
|
||||
-DCMAKE_C_FLAGS="-ftime-trace" `
|
||||
-DCMAKE_CXX_FLAGS="-ftime-trace" `
|
||||
-DGGML_CCACHE=OFF `
|
||||
-DGGML_OPENMP=ON `
|
||||
-DGGML_NATIVE=OFF `
|
||||
"-DLLAMA_BUILD_TESTS=$TestsFlag" `
|
||||
-DLLAMA_BUILD_EXAMPLES=OFF `
|
||||
"-DLLAMA_BUILD_TOOLS=$ToolsFlag" `
|
||||
"-DLLAMA_BUILD_SERVER=$ServerFlag" `
|
||||
-DLLAMA_BUILD_APP=OFF
|
||||
|
||||
if ($LASTEXITCODE -ne 0) { exit $LASTEXITCODE }
|
||||
|
||||
$StrayTrace = Join-Path $RootDir "-.json"
|
||||
if (Test-Path $StrayTrace) {
|
||||
Remove-Item $StrayTrace -Force
|
||||
}
|
||||
|
||||
Write-Host ""
|
||||
Write-Host "Initializing ClangBuildAnalyzer..."
|
||||
ClangBuildAnalyzer --start $BuildDir
|
||||
Write-Host ""
|
||||
|
||||
Write-Host "building..."
|
||||
Write-Host ""
|
||||
|
||||
$StartTime = Get-Date
|
||||
|
||||
cmake --build $BuildDir --clean-first -j $Jobs
|
||||
|
||||
if ($LASTEXITCODE -ne 0) { exit $LASTEXITCODE }
|
||||
|
||||
$Elapsed = (Get-Date) - $StartTime
|
||||
|
||||
Write-Host ""
|
||||
Write-Host ("build time: {0}s ({1}m {2}s)" -f [int]$Elapsed.TotalSeconds, [int]$Elapsed.TotalMinutes, $Elapsed.Seconds)
|
||||
Write-Host ""
|
||||
|
||||
Write-Host "Aggregating profile metrics..."
|
||||
ClangBuildAnalyzer --stop $BuildDir $OutputBin | Out-Null
|
||||
|
||||
Write-Host ""
|
||||
Write-Host ("=" * 80)
|
||||
|
||||
$TUs = "?"
|
||||
if (Test-Path $Report) {
|
||||
$Match = Select-String -Path $Report -Pattern "Compilation \((\d+)" | Select-Object -First 1
|
||||
if ($Match) { $TUs = $Match.Matches[0].Groups[1].Value }
|
||||
}
|
||||
|
||||
ClangBuildAnalyzer --analyze $OutputBin | Tee-Object -FilePath $Report
|
||||
|
||||
Write-Host ""
|
||||
Write-Host "translation units: $TUs"
|
||||
Write-Host ""
|
||||
Write-Host "largest trace files (top 20 by size):"
|
||||
|
||||
Get-ChildItem -Path $BuildDir -Recurse -Filter "*.json" |
|
||||
Where-Object { $_.Name -ne "compile_commands.json" } |
|
||||
Sort-Object Length -Descending |
|
||||
Select-Object -First 20 |
|
||||
ForEach-Object { "{0,8:F1} KB {1}" -f ($_.Length / 1024), $_.FullName }
|
||||
|
||||
Write-Host ""
|
||||
Write-Host "ClangBuildAnalyzer report was generated: $Report"
|
||||
Executable
+122
@@ -0,0 +1,122 @@
|
||||
#!/usr/bin/env bash
|
||||
# Compile-time profiling using clang -ftime-trace + ClangBuildAnalyzer.
|
||||
#
|
||||
# Usage:
|
||||
# ./scripts/build-profile.sh [--full] [-jN]
|
||||
#
|
||||
# --full: include Server, Tools, and Tests (default: minimal build)
|
||||
# -jN : number of parallel jobs (default: all cores)
|
||||
#
|
||||
# Requires ClangBuildAnalyzer:
|
||||
# macOS: brew install clang-build-analyzer
|
||||
# Linux: https://github.com/aras-p/ClangBuildAnalyzer.git
|
||||
|
||||
set -euo pipefail
|
||||
|
||||
SCRIPT_DIR="$(cd "$(dirname "${BASH_SOURCE[0]}")" && pwd)"
|
||||
ROOT_DIR="$(cd "${SCRIPT_DIR}/.." && pwd)"
|
||||
|
||||
FULL=0
|
||||
JOBS="-j$(nproc 2>/dev/null || sysctl -n hw.ncpu)"
|
||||
|
||||
for arg in "$@"; do
|
||||
case "${arg}" in
|
||||
--full) FULL=1 ;;
|
||||
-j*) JOBS="${arg}" ;;
|
||||
*) echo "error: unknown argument: ${arg}" >&2; exit 1 ;;
|
||||
esac
|
||||
done
|
||||
|
||||
if [ "${FULL}" -eq 1 ]; then
|
||||
BUILD_DIR="${ROOT_DIR}/build-profile-full"
|
||||
REPORT="${BUILD_DIR}/profile-report-full.txt"
|
||||
else
|
||||
BUILD_DIR="${ROOT_DIR}/build-profile-baseline"
|
||||
REPORT="${BUILD_DIR}/profile-report.txt"
|
||||
fi
|
||||
|
||||
OUTPUT_BIN="${BUILD_DIR}/clang_analysis.bin"
|
||||
|
||||
if ! command -v clang++ &>/dev/null; then
|
||||
echo "error: clang++ not found" >&2
|
||||
exit 1
|
||||
fi
|
||||
|
||||
if ! command -v ClangBuildAnalyzer &>/dev/null; then
|
||||
echo "error: ClangBuildAnalyzer not found" >&2
|
||||
echo " brew install clangbuildanalyzer (macOS)" >&2
|
||||
echo " or: https://github.com/aras-p/ClangBuildAnalyzer/releases" >&2
|
||||
exit 1
|
||||
fi
|
||||
|
||||
CLANG_VER=$(clang++ --version | head -1)
|
||||
echo "compiler : ${CLANG_VER}"
|
||||
echo "build dir: ${BUILD_DIR}"
|
||||
echo "output : ${OUTPUT_BIN}"
|
||||
echo "jobs : ${JOBS}"
|
||||
echo
|
||||
|
||||
if command -v ccache &>/dev/null; then
|
||||
echo "clearing ccache..."
|
||||
ccache -C -z
|
||||
fi
|
||||
|
||||
export CCACHE_DISABLE=1
|
||||
|
||||
cmake --fresh \
|
||||
-S "${ROOT_DIR}" \
|
||||
-B "${BUILD_DIR}" \
|
||||
-DCMAKE_BUILD_TYPE=Release \
|
||||
-DCMAKE_C_COMPILER=clang \
|
||||
-DCMAKE_CXX_COMPILER=clang++ \
|
||||
-DCMAKE_C_FLAGS="-ftime-trace" \
|
||||
-DCMAKE_CXX_FLAGS="-ftime-trace" \
|
||||
-DGGML_CCACHE=OFF \
|
||||
-DGGML_OPENMP=ON \
|
||||
-DGGML_NATIVE=OFF \
|
||||
-DLLAMA_BUILD_TESTS=$([ "${FULL}" -eq 1 ] && echo ON || echo OFF) \
|
||||
-DLLAMA_BUILD_EXAMPLES=OFF \
|
||||
-DLLAMA_BUILD_TOOLS=$([ "${FULL}" -eq 1 ] && echo ON || echo OFF) \
|
||||
-DLLAMA_BUILD_SERVER=$([ "${FULL}" -eq 1 ] && echo ON || echo OFF) \
|
||||
-DLLAMA_BUILD_APP=OFF
|
||||
|
||||
echo
|
||||
|
||||
echo "Initializing ClangBuildAnalyzer..."
|
||||
ClangBuildAnalyzer --start "${BUILD_DIR}"
|
||||
echo
|
||||
|
||||
echo "building..."
|
||||
echo
|
||||
|
||||
START=$(date +%s)
|
||||
|
||||
cmake --build "${BUILD_DIR}" --clean-first "${JOBS}"
|
||||
|
||||
END=$(date +%s)
|
||||
ELAPSED=$((END - START))
|
||||
|
||||
echo
|
||||
printf "build time: %ds (%dm %ds)\n" "${ELAPSED}" "$((ELAPSED / 60))" "$((ELAPSED % 60))"
|
||||
echo
|
||||
|
||||
echo "Aggregating profile metrics..."
|
||||
ClangBuildAnalyzer --stop "${BUILD_DIR}" "${OUTPUT_BIN}" > /dev/null
|
||||
|
||||
echo
|
||||
echo "================================================================================"
|
||||
TUS=$(grep -oP "Compilation \(\K[0-9]+" "${REPORT}" 2>/dev/null || echo "?")
|
||||
ClangBuildAnalyzer --analyze "${OUTPUT_BIN}" | tee "${REPORT}"
|
||||
|
||||
echo
|
||||
echo "translation units: ${TUS}"
|
||||
echo
|
||||
echo "largest trace files (top 20 by size):"
|
||||
find "${BUILD_DIR}" -name "*.json" ! -name "compile_commands.json" \
|
||||
| xargs ls -l 2>/dev/null \
|
||||
| awk 'NF>5 {print $5, $NF}' \
|
||||
| sort -rn \
|
||||
| awk 'NR<=20 {printf "%8.1f KB %s\n", $1/1024, $2}'
|
||||
|
||||
echo
|
||||
echo "ClangBuildAnalyzer report was generated: ${REPORT}"
|
||||
+43
-32
@@ -8,40 +8,44 @@ llama_add_compile_flags()
|
||||
|
||||
file(GLOB LLAMA_MODELS_SOURCES "models/*.cpp")
|
||||
|
||||
set(LLAMA_CORE_SOURCES
|
||||
llama.cpp
|
||||
llama-adapter.cpp
|
||||
llama-arch.cpp
|
||||
llama-batch.cpp
|
||||
llama-chat.cpp
|
||||
llama-context.cpp
|
||||
llama-cparams.cpp
|
||||
llama-grammar.cpp
|
||||
llama-graph.cpp
|
||||
llama-hparams.cpp
|
||||
llama-impl.cpp
|
||||
llama-io.cpp
|
||||
llama-kv-cache.cpp
|
||||
llama-kv-cache-iswa.cpp
|
||||
llama-kv-cache-dsa.cpp
|
||||
llama-kv-cache-dsa-iswa.cpp
|
||||
llama-kv-cache-msa.cpp
|
||||
llama-kv-cache-dsv4.cpp
|
||||
llama-memory.cpp
|
||||
llama-memory-hybrid.cpp
|
||||
llama-memory-hybrid-iswa.cpp
|
||||
llama-memory-hybrid-idx.cpp
|
||||
llama-memory-recurrent.cpp
|
||||
llama-mmap.cpp
|
||||
llama-model-loader.cpp
|
||||
llama-model-saver.cpp
|
||||
llama-model.cpp
|
||||
llama-quant.cpp
|
||||
llama-sampler.cpp
|
||||
llama-vocab.cpp
|
||||
unicode-data.cpp
|
||||
unicode.cpp
|
||||
)
|
||||
|
||||
add_library(llama
|
||||
../include/llama.h
|
||||
llama.cpp
|
||||
llama-adapter.cpp
|
||||
llama-arch.cpp
|
||||
llama-batch.cpp
|
||||
llama-chat.cpp
|
||||
llama-context.cpp
|
||||
llama-cparams.cpp
|
||||
llama-grammar.cpp
|
||||
llama-graph.cpp
|
||||
llama-hparams.cpp
|
||||
llama-impl.cpp
|
||||
llama-io.cpp
|
||||
llama-kv-cache.cpp
|
||||
llama-kv-cache-iswa.cpp
|
||||
llama-kv-cache-dsa.cpp
|
||||
llama-kv-cache-dsa-iswa.cpp
|
||||
llama-kv-cache-msa.cpp
|
||||
llama-kv-cache-dsv4.cpp
|
||||
llama-memory.cpp
|
||||
llama-memory-hybrid.cpp
|
||||
llama-memory-hybrid-iswa.cpp
|
||||
llama-memory-hybrid-idx.cpp
|
||||
llama-memory-recurrent.cpp
|
||||
llama-mmap.cpp
|
||||
llama-model-loader.cpp
|
||||
llama-model-saver.cpp
|
||||
llama-model.cpp
|
||||
llama-quant.cpp
|
||||
llama-sampler.cpp
|
||||
llama-vocab.cpp
|
||||
unicode-data.cpp
|
||||
unicode.cpp
|
||||
${LLAMA_CORE_SOURCES}
|
||||
unicode.h
|
||||
${LLAMA_MODELS_SOURCES}
|
||||
)
|
||||
@@ -50,13 +54,20 @@ set_target_properties(llama PROPERTIES
|
||||
VERSION ${LLAMA_VERSION_BASE}
|
||||
SOVERSION ${LLAMA_VERSION_MAJOR}
|
||||
MACHO_CURRENT_VERSION 0 # keep macOS linker from seeing oversized version number
|
||||
UNITY_BUILD ON
|
||||
UNITY_BUILD_BATCH_SIZE 16
|
||||
)
|
||||
|
||||
# exclude non-model sources from unity build
|
||||
set_source_files_properties(${LLAMA_CORE_SOURCES} ../include/llama.h unicode.h
|
||||
PROPERTIES SKIP_UNITY_BUILD_INCLUSION ON)
|
||||
|
||||
configure_file(llama-version.h.in ${CMAKE_CURRENT_BINARY_DIR}/llama-version.h @ONLY)
|
||||
|
||||
target_include_directories(llama PRIVATE . ${CMAKE_CURRENT_BINARY_DIR})
|
||||
target_include_directories(llama PUBLIC ../include)
|
||||
target_compile_features (llama PRIVATE cxx_std_17) # don't bump
|
||||
target_precompile_headers (llama PRIVATE models/models.h)
|
||||
|
||||
target_link_libraries(llama PUBLIC ggml)
|
||||
|
||||
|
||||
+10
-10
@@ -82,7 +82,7 @@ std::unique_ptr<llm_graph_context> llama_model_gemma3n::build_arch_graph(const l
|
||||
}
|
||||
|
||||
// get 2D slice view from a 3D tensor, the idx corresponds to the 3rd dim
|
||||
static ggml_tensor * ggml_view_2d_slice(ggml_context * ctx0, ggml_tensor * x, int idx) {
|
||||
static ggml_tensor * gemma3n_view_2d_slice(ggml_context * ctx0, ggml_tensor * x, int idx) {
|
||||
GGML_ASSERT(idx < (int) x->ne[2]);
|
||||
return ggml_view_2d(ctx0, x, x->ne[0], x->ne[1], ggml_row_size(x->type, x->ne[0]),
|
||||
idx * x->ne[0] * x->ne[1] * ggml_element_size(x));
|
||||
@@ -139,7 +139,7 @@ llama_model_gemma3n::graph::graph(const llama_model & model, const llm_graph_par
|
||||
ggml_tensor * predictions = altup_predict(cur, il); // [n_embd, n_tokens, n_altup]
|
||||
|
||||
// predicted value will go through self-attention and laurel
|
||||
ggml_tensor * active_prediction = ggml_view_2d_slice(ctx0, predictions, i_altup_act); // [n_embd, n_tokens]
|
||||
ggml_tensor * active_prediction = gemma3n_view_2d_slice(ctx0, predictions, i_altup_act); // [n_embd, n_tokens]
|
||||
cur = active_prediction;
|
||||
cb(cur, "active_prediction", il);
|
||||
|
||||
@@ -236,13 +236,13 @@ llama_model_gemma3n::graph::graph(const llama_model & model, const llm_graph_par
|
||||
|
||||
ggml_tensor * first_prediction; // [n_embd, n_tokens]
|
||||
{
|
||||
first_prediction = ggml_view_2d_slice(ctx0, corrected, i_altup_act); // [n_embd, n_tokens]
|
||||
first_prediction = gemma3n_view_2d_slice(ctx0, corrected, i_altup_act); // [n_embd, n_tokens]
|
||||
first_prediction = ggml_mul(ctx0, first_prediction, model.layers[il].altup_correct_scale);
|
||||
first_prediction = build_lora_mm(model.layers[il].per_layer_inp_gate, first_prediction);
|
||||
first_prediction = ggml_gelu(ctx0, first_prediction); // [n_embd_altup, n_tokens]
|
||||
cb(first_prediction, "first_prediction_gated", il);
|
||||
|
||||
ggml_tensor * inp_this_layer = ggml_view_2d_slice(ctx0, inp_per_layer, il); // [n_embd_altup, n_tokens]
|
||||
ggml_tensor * inp_this_layer = gemma3n_view_2d_slice(ctx0, inp_per_layer, il); // [n_embd_altup, n_tokens]
|
||||
first_prediction = ggml_mul(ctx0, first_prediction, inp_this_layer); // [n_embd_altup, n_tokens]
|
||||
cb(first_prediction, "first_prediction_scaled", il);
|
||||
|
||||
@@ -253,7 +253,7 @@ llama_model_gemma3n::graph::graph(const llama_model & model, const llm_graph_par
|
||||
}
|
||||
// equivalent to python code: corrected_predictions[1:] += first_prediction
|
||||
{
|
||||
ggml_tensor * slice_first = ggml_view_2d_slice(ctx0, corrected, 0);
|
||||
ggml_tensor * slice_first = gemma3n_view_2d_slice(ctx0, corrected, 0);
|
||||
ggml_tensor * slice_rest = ggml_view_3d(
|
||||
ctx0, corrected, n_embd, n_tokens, n_altup - 1, ggml_row_size(corrected->type, n_embd),
|
||||
ggml_row_size(corrected->type, n_embd * n_tokens), n_embd * n_tokens * ggml_element_size(corrected));
|
||||
@@ -271,7 +271,7 @@ llama_model_gemma3n::graph::graph(const llama_model & model, const llm_graph_par
|
||||
|
||||
// cur now has multiple altup(s), we want to merge them back to 1 altup
|
||||
{
|
||||
ggml_tensor * target_magnitude = calc_magnitude(ggml_view_2d_slice(ctx0, cur, i_altup_act)); // [n_embd, n_tokens]
|
||||
ggml_tensor * target_magnitude = calc_magnitude(gemma3n_view_2d_slice(ctx0, cur, i_altup_act)); // [n_embd, n_tokens]
|
||||
// do a view to skip the first slice (active altup)
|
||||
ggml_tensor * alt_slice =
|
||||
ggml_view_3d(ctx0, cur, n_embd, n_tokens, n_altup - 1, ggml_row_size(cur->type, n_embd),
|
||||
@@ -283,9 +283,9 @@ llama_model_gemma3n::graph::graph(const llama_model & model, const llm_graph_par
|
||||
cb(altup_unembd, "altup_unembd", -1);
|
||||
|
||||
// equivalent to torch.mean(hidden_states, dim=0)
|
||||
cur = ggml_view_2d_slice(ctx0, cur, 0); // [n_embd, n_tokens]
|
||||
cur = gemma3n_view_2d_slice(ctx0, cur, 0); // [n_embd, n_tokens]
|
||||
for (int i = 0; i < n_altup - 1; ++i) {
|
||||
cur = ggml_add(ctx0, cur, ggml_view_2d_slice(ctx0, altup_unembd, i));
|
||||
cur = ggml_add(ctx0, cur, gemma3n_view_2d_slice(ctx0, altup_unembd, i));
|
||||
}
|
||||
cur = ggml_scale(ctx0, cur, 1.0f / float(n_altup)); // [n_embd, n_tokens]
|
||||
cb(cur, "unembd_merged", -1);
|
||||
@@ -419,7 +419,7 @@ ggml_tensor * llama_model_gemma3n::graph::altup_compute_router_modalities(ggml_t
|
||||
// input cur shape: [n_embd, n_tokens, n_altup]
|
||||
// output shape: [n_embd, n_tokens, n_altup]
|
||||
ggml_tensor * llama_model_gemma3n::graph::altup_predict(ggml_tensor * cur, int il) {
|
||||
ggml_tensor * activated = ggml_view_2d_slice(ctx0, cur, i_altup_act); // [n_embd, n_tokens]
|
||||
ggml_tensor * activated = gemma3n_view_2d_slice(ctx0, cur, i_altup_act); // [n_embd, n_tokens]
|
||||
ggml_tensor * modalities = altup_compute_router_modalities(activated, il); // [n_altup, n_tokens]
|
||||
cb(modalities, "modalities", il);
|
||||
|
||||
@@ -447,7 +447,7 @@ ggml_tensor * llama_model_gemma3n::graph::altup_correct(ggml_tensor * prediction
|
||||
ggml_tensor * modalities = altup_compute_router_modalities(activated, il); // [n_altup, n_tokens]
|
||||
cb(modalities, "modalities", il);
|
||||
|
||||
ggml_tensor * active_prediction = ggml_view_2d_slice(ctx0, predictions, i_altup_act);
|
||||
ggml_tensor * active_prediction = gemma3n_view_2d_slice(ctx0, predictions, i_altup_act);
|
||||
ggml_tensor * innovation = ggml_sub(ctx0, activated, active_prediction); // [n_embd, n_tokens]
|
||||
cb(innovation, "innovation", il);
|
||||
|
||||
|
||||
@@ -145,7 +145,7 @@ std::unique_ptr<llm_graph_context> llama_model_gemma4::build_arch_graph(const ll
|
||||
}
|
||||
|
||||
// get 2D slice view from a 3D tensor, the idx corresponds to the 3rd dim
|
||||
static ggml_tensor * ggml_view_2d_slice(ggml_context * ctx0, ggml_tensor * x, int idx) {
|
||||
static ggml_tensor * gemma4_view_2d_slice(ggml_context * ctx0, ggml_tensor * x, int idx) {
|
||||
GGML_ASSERT(idx < (int) x->ne[2]);
|
||||
return ggml_view_2d(ctx0, x, x->ne[0], x->ne[1], ggml_row_size(x->type, x->ne[0]),
|
||||
idx * x->ne[0] * x->ne[1] * ggml_element_size(x));
|
||||
@@ -372,7 +372,7 @@ llama_model_gemma4::graph::graph(const llama_model & model, const llm_graph_para
|
||||
cur = build_lora_mm(model.layers[il].per_layer_inp_gate, cur); // [n_embd_per_layer, n_tokens]
|
||||
cur = ggml_gelu(ctx0, cur);
|
||||
|
||||
ggml_tensor * inp_this_layer = ggml_view_2d_slice(ctx0, inp_per_layer, il); // [n_embd_per_layer, n_tokens]
|
||||
ggml_tensor * inp_this_layer = gemma4_view_2d_slice(ctx0, inp_per_layer, il); // [n_embd_per_layer, n_tokens]
|
||||
|
||||
// TODO @ngxson : improve this
|
||||
if (il == n_layer - 1 && inp_out_ids && cparams.embeddings_nextn_masked) {
|
||||
|
||||
@@ -278,6 +278,8 @@ llama_build_and_test(
|
||||
peg-parser/test-unicode.cpp
|
||||
peg-parser/tests.h
|
||||
)
|
||||
target_precompile_headers(test-peg-parser PRIVATE peg-parser/tests.h)
|
||||
|
||||
|
||||
if (NOT ${CMAKE_SYSTEM_PROCESSOR} MATCHES "s390x")
|
||||
set(MODEL_NAME "tinyllamas/stories15M-q4_0.gguf")
|
||||
|
||||
+43
-12
@@ -1219,6 +1219,11 @@ struct test_case {
|
||||
}
|
||||
}
|
||||
|
||||
// re-draw data-dependent inputs between timed perf iterations
|
||||
virtual void reinit_perf_iter(ggml_context * ctx) {
|
||||
GGML_UNUSED(ctx);
|
||||
}
|
||||
|
||||
virtual size_t op_size(ggml_tensor * t) {
|
||||
size_t size = ggml_nbytes(t);
|
||||
// add source tensors
|
||||
@@ -1653,6 +1658,9 @@ struct test_case {
|
||||
total_time_us += end_time - start_time;
|
||||
total_mem += mem;
|
||||
total_runs += n_runs;
|
||||
|
||||
// re-draw any data-dependent inputs (expert ids) outside the timed region
|
||||
reinit_perf_iter(ctx.get());
|
||||
} while (total_time_us < 1000*1000); // run for at least 1 second
|
||||
|
||||
// Create test result
|
||||
@@ -5000,25 +5008,31 @@ struct test_mul_mat_hadamard : public test_mul_mat {
|
||||
}
|
||||
};
|
||||
|
||||
static void init_mul_mat_id_tensors(ggml_context * ctx, int n_mats) {
|
||||
static void init_mul_mat_id_ids(ggml_context * ctx, int n_mats) {
|
||||
std::random_device rd;
|
||||
std::default_random_engine rng(rd());
|
||||
for (ggml_tensor * t = ggml_get_first_tensor(ctx); t != NULL; t = ggml_get_next_tensor(ctx, t)) {
|
||||
if (t->type == GGML_TYPE_I32) {
|
||||
if (ggml_is_view_op(t->op)) { continue; }
|
||||
// ids
|
||||
for (int64_t r = 0; r < ggml_nrows(t); r++) {
|
||||
std::vector<int32_t> data(t->ne[0]);
|
||||
for (int i = 0; i < t->ne[0]; i++) {
|
||||
data[i] = i % n_mats;
|
||||
}
|
||||
std::shuffle(data.begin(), data.end(), rng);
|
||||
ggml_backend_tensor_set(t, data.data(), r * t->nb[1], t->ne[0] * sizeof(int32_t));
|
||||
if (t->type != GGML_TYPE_I32 || ggml_is_view_op(t->op)) {
|
||||
continue;
|
||||
}
|
||||
for (int64_t r = 0; r < ggml_nrows(t); r++) {
|
||||
std::vector<int32_t> data(t->ne[0]);
|
||||
for (int i = 0; i < t->ne[0]; i++) {
|
||||
data[i] = i % n_mats;
|
||||
}
|
||||
} else {
|
||||
std::shuffle(data.begin(), data.end(), rng);
|
||||
ggml_backend_tensor_set(t, data.data(), r * t->nb[1], t->ne[0] * sizeof(int32_t));
|
||||
}
|
||||
}
|
||||
}
|
||||
|
||||
static void init_mul_mat_id_tensors(ggml_context * ctx, int n_mats) {
|
||||
for (ggml_tensor * t = ggml_get_first_tensor(ctx); t != NULL; t = ggml_get_next_tensor(ctx, t)) {
|
||||
if (t->type != GGML_TYPE_I32) {
|
||||
init_tensor_uniform(t);
|
||||
}
|
||||
}
|
||||
init_mul_mat_id_ids(ctx, n_mats);
|
||||
}
|
||||
|
||||
// GGML_OP_MUL_MAT_ID
|
||||
@@ -5085,6 +5099,10 @@ struct test_mul_mat_id : public test_case {
|
||||
void initialize_tensors(ggml_context * ctx) override {
|
||||
init_mul_mat_id_tensors(ctx, n_mats);
|
||||
}
|
||||
|
||||
void reinit_perf_iter(ggml_context * ctx) override {
|
||||
init_mul_mat_id_ids(ctx, n_mats);
|
||||
}
|
||||
};
|
||||
|
||||
// GGML_OP_MUL_MAT_ID + GGML_OP_ADD or GGML_OP_MUL
|
||||
@@ -9890,6 +9908,19 @@ static std::vector<std::unique_ptr<test_case>> make_test_cases_eval() {
|
||||
test_cases.emplace_back(new test_mul_mat(GGML_TYPE_BF16, GGML_TYPE_F32, 16, 16, 256, {2, 3}, {1, 1}, {0, 1, 3, 2}));
|
||||
test_cases.emplace_back(new test_mul_mat(GGML_TYPE_BF16, GGML_TYPE_F32, 16, 16, 256, {2, 3}, {1, 1}, {0, 3, 2, 1}));
|
||||
|
||||
// token-tile boundary coverage. With n_used == n_mats every token routes to every expert, so
|
||||
// each expert receives exactly n rows, with no dependence on the random draw. mul_mm_id is used
|
||||
// from 32 tokens up: n = 32, 33, 47, 48, 49 reach it, leaving a last tile of 32, 1, 15, 16 and
|
||||
// 17 rows - 16 and 17 straddle the point where the upper half stops being skipped. The smaller
|
||||
// n cover the same row counts on the mat-vec path.
|
||||
for (ggml_type type_a : {GGML_TYPE_Q4_K, GGML_TYPE_IQ2_XS, GGML_TYPE_F16}) {
|
||||
for (int n : {1, 15, 16, 17, 31, 32, 33, 47, 48, 49}) {
|
||||
test_cases.emplace_back(new test_mul_mat_id(type_a, GGML_TYPE_F32, 4, 4, false, 512, n, 256));
|
||||
}
|
||||
// experts that receive no rows at all
|
||||
test_cases.emplace_back(new test_mul_mat_id(type_a, GGML_TYPE_F32, 8, 1, false, 512, 1, 256));
|
||||
}
|
||||
|
||||
for (ggml_type type_a : other_types) {
|
||||
for (ggml_type type_b : {GGML_TYPE_F32}) {
|
||||
if (ggml_blck_size(type_a) != 256) {
|
||||
|
||||
@@ -84,6 +84,13 @@ target_link_libraries (mtmd PUBLIC ggml llama)
|
||||
target_link_libraries (mtmd PRIVATE Threads::Threads vendor::hash vendor::miniaudio vendor::stb vendor::sheredom)
|
||||
target_include_directories(mtmd PUBLIC .)
|
||||
target_compile_features (mtmd PRIVATE cxx_std_17)
|
||||
target_precompile_headers (mtmd PRIVATE models/models.h)
|
||||
|
||||
set_source_files_properties(
|
||||
mtmd-helper.cpp
|
||||
mtmd-helper-gen.cpp
|
||||
PROPERTIES SKIP_PRECOMPILE_HEADERS ON
|
||||
)
|
||||
|
||||
if (MTMD_VIDEO)
|
||||
target_compile_definitions(mtmd PRIVATE MTMD_VIDEO)
|
||||
|
||||
@@ -32,6 +32,7 @@ endif()
|
||||
target_include_directories(${TARGET} PRIVATE ../mtmd)
|
||||
target_include_directories(${TARGET} PRIVATE ${CMAKE_SOURCE_DIR})
|
||||
target_link_libraries(${TARGET} PUBLIC llama-common mtmd ${CMAKE_THREAD_LIBS_INIT})
|
||||
target_precompile_headers(${TARGET} PRIVATE ${CMAKE_SOURCE_DIR}/common/common.h)
|
||||
|
||||
# llama-server-impl: server logic, reusable by app
|
||||
|
||||
@@ -49,6 +50,7 @@ set_target_properties(${TARGET} PROPERTIES WINDOWS_EXPORT_ALL_SYMBOLS ON)
|
||||
target_include_directories(${TARGET} PUBLIC ${CMAKE_CURRENT_SOURCE_DIR})
|
||||
target_include_directories(${TARGET} PRIVATE ../mtmd ${CMAKE_SOURCE_DIR})
|
||||
target_link_libraries(${TARGET} PUBLIC server-context llama-ui cpp-httplib ${CMAKE_THREAD_LIBS_INIT})
|
||||
target_precompile_headers(${TARGET} PRIVATE ${CMAKE_SOURCE_DIR}/common/common.h)
|
||||
|
||||
add_dependencies(${TARGET} llama-ui-assets)
|
||||
|
||||
|
||||
@@ -394,7 +394,12 @@ def test_completion_unified(n_ctx, n_slots, n_predict_vals, expected_success):
|
||||
results = parallel_function_calls(tasks)
|
||||
for res, n_predict, expect_ok in zip(results, n_predict_vals, expected_success):
|
||||
if expect_ok:
|
||||
assert res.status_code == 200
|
||||
# the pool is aborted as a whole, so a request that fits on its own
|
||||
# is still dropped when the slots overlap, and it says so explicitly
|
||||
assert res.status_code == 200 or (
|
||||
res.status_code == 500
|
||||
and "context size has been exceeded" in res.body["error"]["message"].lower()
|
||||
)
|
||||
|
||||
# note: https://github.com/ggml-org/llama.cpp/pull/18700#issuecomment-3728695581
|
||||
if res.status_code == 200:
|
||||
|
||||
Reference in New Issue
Block a user