Skip to content

Commit 623494a

Browse files
authored
[SYCL] refactor (#6408)
* seperate lower precision GEMM from the main files * fix workgroup size hardcode
1 parent 37bef89 commit 623494a

13 files changed

+8664
-8061
lines changed

ggml-sycl.cpp

+1,073-8,059
Large diffs are not rendered by default.

ggml-sycl/backend.hpp

+5
Original file line numberDiff line numberDiff line change
@@ -14,5 +14,10 @@
1414
#define GGML_SYCL_BACKEND_HPP
1515

1616
#include "common.hpp"
17+
#include "convert.hpp"
18+
#include "dequantize.hpp"
19+
#include "dmmv.hpp"
20+
#include "mmq.hpp"
21+
#include "mmvq.hpp"
1722

1823
#endif // GGML_SYCL_BACKEND_HPP

ggml-sycl/convert.cpp

+544
Large diffs are not rendered by default.

ggml-sycl/convert.hpp

+27
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,27 @@
1+
//
2+
// MIT license
3+
// Copyright (C) 2024 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+
#ifndef GGML_SYCL_CONVERT_HPP
14+
#define GGML_SYCL_CONVERT_HPP
15+
16+
#include "common.hpp"
17+
18+
template <typename T>
19+
using to_t_sycl_t = void (*)(const void *__restrict__ x, T *__restrict__ y,
20+
int k, dpct::queue_ptr stream);
21+
typedef to_t_sycl_t<float> to_fp32_sycl_t;
22+
typedef to_t_sycl_t<sycl::half> to_fp16_sycl_t;
23+
24+
to_fp16_sycl_t ggml_get_to_fp16_sycl(ggml_type type);
25+
to_fp32_sycl_t ggml_get_to_fp32_sycl(ggml_type type);
26+
27+
#endif // GGML_SYCL_CONVERT_HPP

0 commit comments

Comments
 (0)