char-fork/llama.cpp

git clone https://git.t4t.associates/char-fork/llama.cpp

Nick Farrellsycl: attribute device allocations by site (GGML_SYCL_MEMTRACE) (#27631)cd8cdf397

char/aliceai-t5-moe
brancheschar/aliceai-t5-moeall refs →
1.6 KiB60 linesraw
1//
2// MIT license
3// Copyright (C) 2025 Intel Corporation
4// SPDX-License-Identifier: MIT
5//
6
7//
8// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
9// See https://llvm.org/LICENSE.txt for license information.
10// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
11//
12
13#include "common.hpp"
14
15sycl::half * ggml_sycl_fattn_kv_buffers::kv_buffer::ensure_half(size_t n_elems) {
16    const size_t need_bytes = n_elems * sizeof(sycl::half);
17
18    if (capacity >= need_bytes) {
19        return ptr;
20    }
21
22    if (ptr) {
23        SYCL_CHECK(CHECK_TRY_ERROR(qptr->wait()));
24        ggml_sycl_memtrace_del(ptr);
25        SYCL_CHECK(CHECK_TRY_ERROR(sycl::free(ptr, *qptr)));
26        ptr = nullptr;
27        capacity = 0;
28    }
29
30    size_t cap = 0;
31    while (cap < need_bytes) {
32        cap += CHUNK_SIZE;
33    }
34
35    void * dev_ptr;
36    SYCL_CHECK(
37        CHECK_TRY_ERROR(dev_ptr = sycl::malloc_device(
38                        cap, *qptr)));
39
40    if (!dev_ptr) {
41        GGML_LOG_ERROR("%s: can't allocate %lu Bytes of memory on device\n", __func__, cap);
42        ggml_sycl_memtrace_fail(GGML_SYCL_MEM_FATTN_KV, cap);
43        GGML_ABORT("fattn buffer alloc failed");
44    }
45
46    ptr = static_cast<sycl::half *>(dev_ptr);
47    capacity = cap;
48    ggml_sycl_memtrace_add(GGML_SYCL_MEM_FATTN_KV, ptr, cap);
49    return ptr;
50}
51
52ggml_sycl_fattn_kv_buffers::kv_buffer::~kv_buffer() {
53#ifdef DEBUG_SYCL_POOL
54    GGML_LOG_INFO("ggml_sycl_fattn_kv_buffer[%d]: %.2f MiB\n", device, capacity / 1024.0 / 1024.0);
55#endif
56    if (ptr) {
57        ggml_sycl_memtrace_del(ptr);
58        SYCL_CHECK(CHECK_TRY_ERROR(sycl::free(ptr, *qptr)));
59    }
60}