cl3/
program.rs

1// Copyright (c) 2020-2025 Via Technology Ltd.
2//
3// Licensed under the Apache License, Version 2.0 (the "License");
4// you may not use this file except in compliance with the License.
5// You may obtain a copy of the License at
6//
7//    http://www.apache.org/licenses/LICENSE-2.0
8//
9// Unless required by applicable law or agreed to in writing, software
10// distributed under the License is distributed on an "AS IS" BASIS,
11// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
12// See the License for the specific language governing permissions and
13// limitations under the License.
14
15//! `OpenCL` Program Object API.
16
17#![allow(unused_unsafe)]
18#![allow(non_camel_case_types)]
19#![allow(clippy::not_unsafe_ptr_arg_deref, clippy::wildcard_in_or_patterns)]
20
21pub use opencl_sys::{
22    CL_BUILD_ERROR, CL_BUILD_IN_PROGRESS, CL_BUILD_NONE, CL_BUILD_SUCCESS, CL_FALSE,
23    CL_INVALID_VALUE, CL_PROGRAM_BINARIES, CL_PROGRAM_BINARY_SIZES, CL_PROGRAM_BINARY_TYPE,
24    CL_PROGRAM_BINARY_TYPE_COMPILED_OBJECT, CL_PROGRAM_BINARY_TYPE_EXECUTABLE,
25    CL_PROGRAM_BINARY_TYPE_LIBRARY, CL_PROGRAM_BINARY_TYPE_NONE,
26    CL_PROGRAM_BUILD_GLOBAL_VARIABLE_TOTAL_SIZE, CL_PROGRAM_BUILD_LOG, CL_PROGRAM_BUILD_OPTIONS,
27    CL_PROGRAM_BUILD_STATUS, CL_PROGRAM_CONTEXT, CL_PROGRAM_DEVICES, CL_PROGRAM_IL,
28    CL_PROGRAM_KERNEL_NAMES, CL_PROGRAM_NUM_DEVICES, CL_PROGRAM_NUM_KERNELS,
29    CL_PROGRAM_REFERENCE_COUNT, CL_PROGRAM_SOURCE, CL_SUCCESS, CL_TRUE, cl_context, cl_device_id,
30    cl_int, cl_platform_id, cl_program, cl_program_build_info, cl_program_info, cl_uchar, cl_uint,
31};
32
33use super::info_type::InfoType;
34use super::{
35    api_info_size, api_info_value, api_info_vector, api2_info_size, api2_info_value,
36    api2_info_vector,
37};
38use libc::{c_char, c_uchar, c_void, intptr_t, size_t};
39use std::ffi::CStr;
40use std::mem;
41use std::ptr;
42
43// Missing from cl_sys
44pub const CL_PROGRAM_SCOPE_GLOBAL_CTORS_PRESENT: cl_program_info = 0x116A;
45pub const CL_PROGRAM_SCOPE_GLOBAL_DTORS_PRESENT: cl_program_info = 0x116B;
46
47/// Create an `OpenCL` program object for a context and load source code into that object.
48/// Calls `clCreateProgramWithSource` to create an `OpenCL` program object.
49///
50/// * `context` - a valid `OpenCL` context.
51/// * `sources` - an array of slices of source code strings.
52///
53/// returns a Result containing the new `OpenCL` program object
54/// or the error code from the `OpenCL` C API function.
55#[allow(clippy::cast_possible_truncation)]
56#[inline]
57pub fn create_program_with_source(
58    context: cl_context,
59    sources: &[&str],
60) -> Result<cl_program, cl_int> {
61    let lengths: Vec<size_t> = sources.iter().map(|src| src.len()).collect();
62    let mut status: cl_int = CL_INVALID_VALUE;
63    let program: cl_program = unsafe {
64        cl_call!(clCreateProgramWithSource(
65            context,
66            sources.len() as cl_uint,
67            sources.as_ptr().cast::<*const c_char>(),
68            lengths.as_ptr(),
69            &raw mut status,
70        ))
71    };
72
73    if CL_SUCCESS == status {
74        Ok(program)
75    } else {
76        Err(status)
77    }
78}
79
80/// Create an `OpenCL` program object for a context and load binary bits into that object.
81/// Calls `clCreateProgramWithBinary` to create an `OpenCL` program object.
82///
83/// * `context` - a valid `OpenCL` context.
84/// * `devices` - a slice of devices that are in context.
85/// * `binaries` - a slice of program binaries slices.
86///
87/// returns a Result containing the new `OpenCL` program object
88/// or the error code from the `OpenCL` C API function.
89///
90/// # Safety
91///
92/// This is unsafe when a device is not a member of context.
93#[allow(clippy::cast_possible_truncation)]
94pub unsafe fn create_program_with_binary(
95    context: cl_context,
96    devices: &[cl_device_id],
97    binaries: &[&[u8]],
98) -> Result<cl_program, cl_int> {
99    let binaries_length = binaries.len();
100    let lengths: Vec<size_t> = binaries.iter().map(|bin| bin.len()).collect();
101    let mut binary_status: Vec<cl_int> = Vec::with_capacity(binaries_length);
102    let mut status: cl_int = CL_INVALID_VALUE;
103    let program: cl_program = cl_call!(clCreateProgramWithBinary(
104        context,
105        devices.len() as cl_uint,
106        devices.as_ptr(),
107        lengths.as_ptr(),
108        binaries.as_ptr().cast::<*const c_uchar>(),
109        binary_status.as_mut_ptr(),
110        &raw mut status,
111    ));
112    if CL_SUCCESS == status {
113        Ok(program)
114    } else {
115        Err(status)
116    }
117}
118
119/// Create an `OpenCL` program object for a context and  loads the information
120/// related to the built-in kernels into that object.
121///
122/// Calls `clCreateProgramWithBuiltInKernels` to create an `OpenCL` program object.
123///
124/// * `context` - a valid `OpenCL` context.
125/// * `devices` - a slice of devices that are in context.
126/// * `kernel_names` - a semi-colon separated list of built-in kernel names.
127///
128/// returns a Result containing the new `OpenCL` program object
129/// or the error code from the `OpenCL` C API function.
130///
131/// # Safety
132///
133/// This is unsafe when a device is not a member of context.
134#[cfg(any(feature = "CL_VERSION_1_2", feature = "dynamic"))]
135#[allow(clippy::cast_possible_truncation)]
136#[inline]
137pub unsafe fn create_program_with_builtin_kernels(
138    context: cl_context,
139    devices: &[cl_device_id],
140    kernel_names: &CStr,
141) -> Result<cl_program, cl_int> {
142    let mut status: cl_int = CL_INVALID_VALUE;
143    let program: cl_program = cl_call!(clCreateProgramWithBuiltInKernels(
144        context,
145        devices.len() as cl_uint,
146        devices.as_ptr(),
147        kernel_names.as_ptr(),
148        &raw mut status,
149    ));
150    if CL_SUCCESS == status {
151        Ok(program)
152    } else {
153        Err(status)
154    }
155}
156
157/// Create an `OpenCL` program object for a context and load code in an intermediate
158/// language into that object.
159/// Calls `clCreateProgramWithIL` to create an `OpenCL` program object.
160/// `CL_VERSION_2_1`
161///
162/// * `context` - a valid `OpenCL` context.
163/// * `il` - a slice of program intermediate language code.
164///
165/// returns a Result containing the new `OpenCL` program object
166/// or the error code from the `OpenCL` C API function.
167#[cfg(any(feature = "CL_VERSION_2_1", feature = "dynamic"))]
168#[inline]
169pub fn create_program_with_il(context: cl_context, il: &[u8]) -> Result<cl_program, cl_int> {
170    let mut status: cl_int = CL_INVALID_VALUE;
171    let program: cl_program = unsafe {
172        cl_call!(clCreateProgramWithIL(
173            context,
174            il.as_ptr().cast::<c_void>(),
175            il.len() as size_t,
176            &raw mut status,
177        ))
178    };
179    if CL_SUCCESS == status {
180        Ok(program)
181    } else {
182        Err(status)
183    }
184}
185
186/// Retain an `OpenCL` program.
187/// Calls `clRetainProgram` to increment the program reference count.
188///
189/// * `program` - the `OpenCL` program.
190///
191/// returns an empty Result or the error code from the `OpenCL` C API function.
192///
193/// # Safety
194///
195/// This function is unsafe because it changes the `OpenCL` object reference count.
196#[inline]
197pub unsafe fn retain_program(program: cl_program) -> Result<(), cl_int> {
198    let status: cl_int = cl_call!(clRetainProgram(program));
199    if CL_SUCCESS == status {
200        Ok(())
201    } else {
202        Err(status)
203    }
204}
205
206/// Release an `OpenCL` program.
207/// Calls `clReleaseProgram` to decrement the program reference count.
208///
209/// * `program` - the `OpenCL` program.
210///
211/// returns an empty Result or the error code from the `OpenCL` C API function.
212///
213/// # Safety
214///
215/// This function is unsafe because it changes the `OpenCL` object reference count.
216#[inline]
217pub unsafe fn release_program(program: cl_program) -> Result<(), cl_int> {
218    let status: cl_int = cl_call!(clReleaseProgram(program));
219    if CL_SUCCESS == status {
220        Ok(())
221    } else {
222        Err(status)
223    }
224}
225
226/// Build (compile & link) a program executable.
227/// Calls `clBuildProgram` to build an `OpenCL` program object.
228///
229/// * `program` - a valid `OpenCL` program.
230/// * `devices` - a slice of devices that are in context.
231/// * `options` - the build options in a null-terminated string.
232/// * `pfn_notify` - an optional function pointer to a notification routine.
233/// * `user_data` - passed as an argument when `pfn_notify` is called, or `ptr::null_mut()`.
234///
235/// returns a Result containing the new `OpenCL` program object
236/// or the error code from the `OpenCL` C API function.
237#[allow(clippy::cast_possible_truncation)]
238#[inline]
239pub fn build_program(
240    program: cl_program,
241    devices: &[cl_device_id],
242    options: &CStr,
243    pfn_notify: Option<unsafe extern "C" fn(cl_program, *mut c_void)>,
244    user_data: *mut c_void,
245) -> Result<(), cl_int> {
246    let status: cl_int = unsafe {
247        let devices_ptr = if devices.is_empty() {
248            ptr::null()
249        } else {
250            devices.as_ptr()
251        };
252        cl_call!(clBuildProgram(
253            program,
254            devices.len() as cl_uint,
255            devices_ptr,
256            options.as_ptr(),
257            pfn_notify,
258            user_data,
259        ))
260    };
261    if CL_SUCCESS == status {
262        Ok(())
263    } else {
264        Err(status)
265    }
266}
267
268/// Compile a program’s source for the devices the `OpenCL` context associated
269/// with the program.
270/// Calls clCompileProgram to compile an `OpenCL` program object.
271///
272/// * `program` - a valid `OpenCL` program.
273/// * `devices` - a slice of devices that are in context.
274/// * `options` - the compilation options in a null-terminated string.
275/// * `input_headers` - a slice of programs that describe headers in the `input_headers`.
276/// * `header_include_names` - an array that has a one to one correspondence with
277///   `input_headers`.
278/// * `pfn_notify` - an optional function pointer to a notification routine.
279/// * `user_data` - passed as an argument when `pfn_notify` is called, or `ptr::null_mut()`.
280///
281/// returns a Result containing the new `OpenCL` program object
282/// or the error code from the `OpenCL` C API function.
283///
284/// # Panics
285///
286/// Panics if `input_headers.len()` != `header_include_names.len()`.
287#[cfg(any(feature = "CL_VERSION_1_2", feature = "dynamic"))]
288#[allow(clippy::cast_possible_truncation)]
289#[inline]
290pub fn compile_program(
291    program: cl_program,
292    devices: &[cl_device_id],
293    options: &CStr,
294    input_headers: &[cl_program],
295    header_include_names: &[&CStr],
296    pfn_notify: Option<unsafe extern "C" fn(program: cl_program, user_data: *mut c_void)>,
297    user_data: *mut c_void,
298) -> Result<(), cl_int> {
299    assert!(input_headers.len() == header_include_names.len());
300    let status: cl_int = unsafe {
301        let devices_ptr = if devices.is_empty() {
302            ptr::null()
303        } else {
304            devices.as_ptr()
305        };
306        let input_headers_ptr = if input_headers.is_empty() {
307            ptr::null()
308        } else {
309            input_headers.as_ptr()
310        };
311        let header_include_names_ptr = if header_include_names.is_empty() {
312            ptr::null()
313        } else {
314            header_include_names.as_ptr()
315        };
316        cl_call!(clCompileProgram(
317            program,
318            devices.len() as cl_uint,
319            devices_ptr,
320            options.as_ptr(),
321            input_headers.len() as cl_uint,
322            input_headers_ptr,
323            header_include_names_ptr.cast::<*const c_char>(),
324            pfn_notify,
325            user_data,
326        ))
327    };
328    if CL_SUCCESS == status {
329        Ok(())
330    } else {
331        Err(status)
332    }
333}
334
335/// Link a set of compiled program objects and libraries for the devices in the
336/// `OpenCL` context associated with the program.
337/// Calls clLinkProgram to link an `OpenCL` program object.
338///
339/// * `context` - a valid `OpenCL` context.
340/// * `devices` - a slice of devices that are in context.
341/// * `options` - the link options in a null-terminated string.
342/// * `input_programs` - a slice of programs that are to be linked to create the program executable.
343/// * `pfn_notify` - an optional function pointer to a notification routine.
344/// * `user_data` - passed as an argument when `pfn_notify` is called, or `ptr::null_mut()`.
345///
346/// returns a Result containing the new `OpenCL` program object
347/// or the error code from the `OpenCL` C API function.
348///
349/// # Panics
350///
351/// Panics if `input_programs.is_empty()`.
352///
353/// # Safety
354///
355/// This is unsafe when a device is not a member of context.
356#[cfg(any(feature = "CL_VERSION_1_2", feature = "dynamic"))]
357#[allow(clippy::cast_possible_truncation)]
358#[inline]
359pub unsafe fn link_program(
360    context: cl_context,
361    devices: &[cl_device_id],
362    options: &CStr,
363    input_programs: &[cl_program],
364    pfn_notify: Option<unsafe extern "C" fn(program: cl_program, user_data: *mut c_void)>,
365    user_data: *mut c_void,
366) -> Result<cl_program, cl_int> {
367    assert!(!input_programs.is_empty());
368    let devices_ptr = if devices.is_empty() {
369        ptr::null()
370    } else {
371        devices.as_ptr()
372    };
373    let mut status: cl_int = CL_INVALID_VALUE;
374    let programme: cl_program = cl_call!(clLinkProgram(
375        context,
376        devices.len() as cl_uint,
377        devices_ptr,
378        options.as_ptr(),
379        input_programs.len() as cl_uint,
380        input_programs.as_ptr(),
381        pfn_notify,
382        user_data,
383        &raw mut status,
384    ));
385    if CL_SUCCESS == status {
386        Ok(programme)
387    } else {
388        Err(status)
389    }
390}
391
392/// Set the value of a specialization constant.
393/// Calls `clSetProgramSpecializationConstant`.
394/// `CL_VERSION_2_2`
395///
396/// * `program` - the program.
397/// * `spec_id` - the specialization constant whose value will be set.
398/// * `spec_size` - size in bytes of the data pointed to by `spec_value`.
399/// * `spec_value` - pointer to the memory location that contains the value
400///   of the specialization constant.
401///
402/// returns an empty Result or the error code from the `OpenCL` C API function.
403///
404/// # Safety
405///
406/// This function is unsafe because `spec_size` and `spec_value` must be valid.
407#[cfg(any(feature = "CL_VERSION_2_2", feature = "dynamic"))]
408#[inline]
409pub unsafe fn set_program_specialization_constant(
410    program: cl_program,
411    spec_id: cl_uint,
412    spec_size: size_t,
413    spec_value: *const c_void,
414) -> Result<(), cl_int> {
415    let status: cl_int = cl_call!(clSetProgramSpecializationConstant(
416        program, spec_id, spec_size, spec_value
417    ));
418    if CL_SUCCESS == status {
419        Ok(())
420    } else {
421        Err(status)
422    }
423}
424
425/// Release the resources allocated by the `OpenCL` compiler for platform.
426/// Calls clUnloadPlatformCompiler.
427///
428/// * `platform` - the platform.
429///
430/// returns an empty Result or the error code from the `OpenCL` C API function.
431///
432/// # Safety
433///
434/// This function is unsafe because the platform compiler is not valid after this call.
435#[cfg(any(feature = "CL_VERSION_1_2", feature = "dynamic"))]
436#[inline]
437pub unsafe fn unload_platform_compiler(platform: cl_platform_id) -> Result<(), cl_int> {
438    let status: cl_int = cl_call!(clUnloadPlatformCompiler(platform));
439    if CL_SUCCESS == status {
440        Ok(())
441    } else {
442        Err(status)
443    }
444}
445
446/// Get data about an `OpenCL` program.
447/// Calls clGetProgramInfo to get the desired data about the program.
448pub fn get_program_data(
449    program: cl_program,
450    param_name: cl_program_info,
451) -> Result<Vec<u8>, cl_int> {
452    api_info_size!(get_size, clGetProgramInfo);
453    let size = get_size(program, param_name)?;
454    api_info_vector!(get_vector, u8, clGetProgramInfo);
455    get_vector(program, param_name, size)
456}
457
458/// Get specific information about an `OpenCL` program.
459/// Calls clGetProgramInfo to get the desired information about the program.
460///
461/// * `program` - the `OpenCL` program.
462/// * `param_name` - the type of program information being queried, see:
463///   [Program Object Queries](https://www.khronos.org/registry/OpenCL/specs/3.0-unified/html/OpenCL_API.html#program-info-table).
464///
465/// returns a Result containing the desired information in an `InfoType` enum
466/// or the error code from the `OpenCL` C API function.
467pub fn get_program_info(
468    program: cl_program,
469    param_name: cl_program_info,
470) -> Result<InfoType, cl_int> {
471    api_info_size!(get_size, clGetProgramInfo);
472
473    match param_name {
474        CL_PROGRAM_REFERENCE_COUNT
475        | CL_PROGRAM_NUM_DEVICES
476        | CL_PROGRAM_SCOPE_GLOBAL_CTORS_PRESENT // CL_VERSION_2_2 only
477        | CL_PROGRAM_SCOPE_GLOBAL_DTORS_PRESENT // CL_VERSION_2_2 only
478        => {
479            api_info_value!(get_value, cl_uint, clGetProgramInfo);
480            Ok(InfoType::Uint(get_value(program, param_name)?))
481        }
482
483        CL_PROGRAM_CONTEXT => {
484            api_info_value!(get_value, intptr_t, clGetProgramInfo);
485            Ok(InfoType::Ptr(get_value(program, param_name)?))
486        }
487
488        CL_PROGRAM_DEVICES => {
489            api_info_vector!(get_vec, intptr_t, clGetProgramInfo);
490            let size = get_size(program, param_name)?;
491            Ok(InfoType::VecIntPtr(get_vec(program, param_name, size)?))
492        }
493
494        CL_PROGRAM_BINARY_SIZES => {
495            api_info_vector!(get_vec, size_t, clGetProgramInfo);
496            let size = get_size(program, param_name)?;
497            Ok(InfoType::VecSize(get_vec(program, param_name, size)?))
498        }
499
500        CL_PROGRAM_BINARIES => {
501            // Gets the binaries for all the devices in the context
502
503            // get the binary sizes, as the case above
504            api_info_vector!(get_size_vec, size_t, clGetProgramInfo);
505            let size = get_size(program, CL_PROGRAM_BINARY_SIZES as cl_program_info)?;
506            let binary_sizes = get_size_vec(program, CL_PROGRAM_BINARY_SIZES as cl_program_info, size)?;
507
508            // A vector of vectors to hold the binaries of each device
509            let binaries = binary_sizes.into_iter().map(|size| {
510                vec![0u8; size]
511            }).collect::<Vec<Vec<u8>>>();
512
513            // Create a vector of pointers to the vectors in binaries
514            let mut binary_ptrs = binaries.iter().map(|vec| {
515                vec.as_ptr()
516            }).collect::<Vec<_>>();
517
518            let status = unsafe {
519                cl_call!(clGetProgramInfo(
520                    program,
521                    param_name,
522                    binary_ptrs.len() * mem::size_of::<*mut c_void>(),
523                    binary_ptrs.as_mut_ptr().cast(),
524                    ptr::null_mut(),
525                ))
526            };
527            if CL_SUCCESS == status {
528                Ok(InfoType::VecVecUchar(binaries))
529            } else {
530                Err(status)
531            }
532        }
533
534        CL_PROGRAM_NUM_KERNELS => {
535            api_info_value!(get_value, size_t, clGetProgramInfo);
536            Ok(InfoType::Size(get_value(program, param_name)?))
537        }
538
539        CL_PROGRAM_SOURCE
540        | CL_PROGRAM_KERNEL_NAMES
541        | CL_PROGRAM_IL
542        | _ => {
543            Ok(InfoType::VecUchar(get_program_data(program, param_name)?))
544        }
545    }
546}
547
548/// Get data about an `OpenCL` program build.
549/// Calls clGetProgramBuildInfo to get the desired data about the program build.
550pub fn get_program_build_data(
551    program: cl_program,
552    device: cl_device_id,
553    param_name: cl_program_info,
554) -> Result<Vec<u8>, cl_int> {
555    api2_info_size!(get_size, cl_device_id, clGetProgramBuildInfo);
556    let size = get_size(program, device, param_name)?;
557    api2_info_vector!(get_vector, cl_device_id, u8, clGetProgramBuildInfo);
558    get_vector(program, device, param_name, size)
559}
560
561/// Get specific information about an `OpenCL` program build.
562/// Calls clGetProgramBuildInfo to get the desired information about the program build.
563///
564/// * `program` - the `OpenCL` program.
565/// * `device` - -the device for which build information is being queried.
566/// * `param_name` - the type of program build information being queried, see:
567///   [Program Build Queries](https://www.khronos.org/registry/OpenCL/specs/3.0-unified/html/OpenCL_API.html#program-build-info-table).
568///
569/// returns a Result containing the desired information in an `InfoType` enum
570/// or the error code from the `OpenCL` C API function.
571pub fn get_program_build_info(
572    program: cl_program,
573    device: cl_device_id,
574    param_name: cl_program_build_info,
575) -> Result<InfoType, cl_int> {
576    match param_name {
577        CL_PROGRAM_BUILD_STATUS => {
578            api2_info_value!(
579                get_device_value,
580                cl_device_id,
581                cl_int,
582                clGetProgramBuildInfo
583            );
584            Ok(InfoType::Int(get_device_value(
585                program, device, param_name,
586            )?))
587        }
588
589        CL_PROGRAM_BINARY_TYPE => {
590            api2_info_value!(
591                get_device_value,
592                cl_device_id,
593                cl_uint,
594                clGetProgramBuildInfo
595            );
596            Ok(InfoType::Uint(get_device_value(
597                program, device, param_name,
598            )?))
599        }
600
601        // CL_VERSION_2_0
602        CL_PROGRAM_BUILD_GLOBAL_VARIABLE_TOTAL_SIZE => {
603            api2_info_value!(
604                get_device_value,
605                cl_device_id,
606                size_t,
607                clGetProgramBuildInfo
608            );
609            Ok(InfoType::Size(get_device_value(
610                program, device, param_name,
611            )?))
612        }
613
614        CL_PROGRAM_BUILD_OPTIONS | CL_PROGRAM_BUILD_LOG | _ => Ok(InfoType::VecUchar(
615            get_program_build_data(program, device, param_name)?,
616        )),
617    }
618}
619
620#[cfg(test)]
621mod tests {
622    use super::*;
623    use crate::context::{create_context, release_context};
624    use crate::device::{CL_DEVICE_TYPE_ALL, get_device_ids};
625    #[allow(unused_imports)]
626    use crate::error_codes::error_text;
627    use crate::platform::get_platform_ids;
628    use std::ffi::CString;
629
630    #[test]
631    fn test_program() {
632        let platform_ids = get_platform_ids().unwrap();
633
634        let mut platform_id = platform_ids[0];
635        let mut device_count: usize = 0;
636
637        // Search for a platform with the most devices
638        for p in platform_ids {
639            let ids = get_device_ids(p, CL_DEVICE_TYPE_ALL).unwrap();
640            let count = ids.len();
641            if device_count < count {
642                device_count = count;
643                platform_id = p;
644            }
645        }
646
647        println!("Platform device_count: {}", device_count);
648
649        let device_ids = get_device_ids(platform_id, CL_DEVICE_TYPE_ALL).unwrap();
650        let device_id = device_ids[0];
651
652        let context = create_context(&device_ids, ptr::null(), None, ptr::null_mut());
653        let context = context.unwrap();
654
655        let source = r#"
656            kernel void saxpy_float (global float* z,
657                global float const* x,
658                global float const* y,
659                float a)
660            {
661            size_t i = get_global_id(0);
662            z[i] = a*x[i] + y[i];
663            }
664        "#;
665
666        // Convert source to an array
667        let sources = [source];
668        let program = create_program_with_source(context, &sources).unwrap();
669
670        let value = get_program_info(program, CL_PROGRAM_REFERENCE_COUNT).unwrap();
671        let value = cl_uint::from(value);
672        println!("CL_PROGRAM_REFERENCE_COUNT: {}", value);
673        assert!(0 < value);
674
675        let value = get_program_info(program, CL_PROGRAM_CONTEXT).unwrap();
676        let value = intptr_t::from(value);
677        println!("CL_PROGRAM_CONTEXT: {}", value);
678        assert!(0 < value);
679
680        let value = get_program_info(program, CL_PROGRAM_NUM_DEVICES).unwrap();
681        let value = cl_uint::from(value);
682        println!("CL_PROGRAM_NUM_DEVICES: {}", value);
683        assert!(0 < value);
684
685        let value = get_program_info(program, CL_PROGRAM_DEVICES).unwrap();
686        let value = Vec::<intptr_t>::from(value);
687        println!("CL_PROGRAM_DEVICES: {}", value.len());
688        assert!(0 < value.len());
689
690        let value = get_program_info(program, CL_PROGRAM_SOURCE).unwrap();
691        let value = String::from(value);
692        println!("CL_PROGRAM_SOURCE: {}", value);
693        assert!(0 < value.len());
694
695        let options = CString::default();
696        let empty_device_ids = Vec::new();
697        build_program(program, &empty_device_ids, &options, None, ptr::null_mut()).unwrap();
698
699        let value = get_program_build_info(program, device_id, CL_PROGRAM_BUILD_STATUS).unwrap();
700        let value: cl_int = From::from(value);
701        println!("CL_PROGRAM_BUILD_STATUS: {}", value);
702        assert_eq!(CL_BUILD_SUCCESS, value);
703
704        let value = get_program_build_info(program, device_id, CL_PROGRAM_BUILD_OPTIONS).unwrap();
705        let value = String::from(value);
706        println!("CL_PROGRAM_BUILD_OPTIONS: {}", value);
707
708        let value = get_program_build_info(program, device_id, CL_PROGRAM_BUILD_LOG).unwrap();
709        let value = String::from(value);
710        println!("CL_PROGRAM_BUILD_LOG: {}", value);
711
712        let value = get_program_build_info(program, device_id, CL_PROGRAM_BINARY_TYPE).unwrap();
713        let value = cl_uint::from(value);
714        println!("CL_PROGRAM_BINARY_TYPE: {:?}", value);
715        assert_eq!(CL_PROGRAM_BINARY_TYPE_EXECUTABLE as cl_uint, value);
716
717        #[cfg(any(feature = "CL_VERSION_2_0", feature = "dynamic"))]
718        match get_program_build_info(
719            program,
720            device_id,
721            CL_PROGRAM_BUILD_GLOBAL_VARIABLE_TOTAL_SIZE,
722        ) {
723            Ok(value) => {
724                let value = size_t::from(value);
725                println!("CL_PROGRAM_BUILD_GLOBAL_VARIABLE_TOTAL_SIZE: {:?}", value)
726            }
727            Err(e) => println!(
728                "OpenCL error, CL_PROGRAM_BUILD_GLOBAL_VARIABLE_TOTAL_SIZE: {}",
729                error_text(e)
730            ),
731        }
732
733        let value = get_program_info(program, CL_PROGRAM_BINARY_SIZES).unwrap();
734        let value = Vec::<size_t>::from(value);
735        println!("CL_PROGRAM_BINARY_SIZES: {}", value.len());
736        println!("CL_PROGRAM_BINARY_SIZES: {:?}", value);
737        assert!(0 < value.len());
738
739        let value = get_program_info(program, CL_PROGRAM_BINARIES).unwrap();
740        // println!("CL_PROGRAM_BINARIES: {:?}", value);
741        let value = Vec::<Vec<u8>>::from(value);
742        println!("CL_PROGRAM_BINARIES count: {}", value.len());
743        println!("CL_PROGRAM_BINARIES length[0]: {}", value[0].len());
744        assert!(0 < value.len());
745
746        let value = get_program_info(program, CL_PROGRAM_NUM_KERNELS).unwrap();
747        let value = size_t::from(value);
748        println!("CL_PROGRAM_NUM_KERNELS: {}", value);
749        assert!(0 < value);
750
751        let value = get_program_info(program, CL_PROGRAM_KERNEL_NAMES).unwrap();
752        let value = String::from(value);
753        println!("CL_PROGRAM_KERNEL_NAMES: {}", value);
754        assert!(0 < value.len());
755
756        #[cfg(any(feature = "CL_VERSION_2_1", feature = "dynamic"))]
757        match get_program_info(program, CL_PROGRAM_IL) {
758            Ok(value) => {
759                let value = String::from(value);
760                println!("CL_PROGRAM_IL: {}", value)
761            }
762            Err(e) => println!("OpenCL error, CL_PROGRAM_IL: {}", error_text(e)),
763        };
764
765        #[cfg(any(feature = "CL_VERSION_2_2", feature = "dynamic"))]
766        match get_program_info(program, CL_PROGRAM_SCOPE_GLOBAL_CTORS_PRESENT) {
767            Ok(value) => {
768                let value = cl_uint::from(value);
769                println!("CL_PROGRAM_SCOPE_GLOBAL_CTORS_PRESENT: {}", value)
770            }
771            Err(e) => println!(
772                "OpenCL error, CL_PROGRAM_SCOPE_GLOBAL_CTORS_PRESENT: {}",
773                error_text(e)
774            ),
775        };
776
777        #[cfg(any(feature = "CL_VERSION_2_2", feature = "dynamic"))]
778        match get_program_info(program, CL_PROGRAM_SCOPE_GLOBAL_CTORS_PRESENT) {
779            Ok(value) => {
780                let value = cl_uint::from(value);
781                println!("CL_PROGRAM_SCOPE_GLOBAL_DTORS_PRESENT: {}", value)
782            }
783            Err(e) => println!(
784                "OpenCL error, CL_PROGRAM_SCOPE_GLOBAL_DTORS_PRESENT: {}",
785                error_text(e)
786            ),
787        };
788
789        #[cfg(any(feature = "CL_VERSION_1_2", feature = "dynamic"))]
790        if let Err(e) = unsafe { unload_platform_compiler(platform_id) } {
791            println!("OpenCL error, clUnloadPlatformCompiler: {}", error_text(e));
792        }
793
794        unsafe {
795            release_program(program).unwrap();
796            release_context(context).unwrap();
797        }
798    }
799
800    #[test]
801    fn test_compile_and_link_program() {
802        let platform_ids = get_platform_ids().unwrap();
803
804        let mut platform_id = platform_ids[0];
805        let mut device_count: usize = 0;
806
807        // Search for a platform with the most devices
808        for p in platform_ids {
809            let ids = get_device_ids(p, CL_DEVICE_TYPE_ALL).unwrap();
810            let count = ids.len();
811            if device_count < count {
812                device_count = count;
813                platform_id = p;
814            }
815        }
816
817        println!("Platform device_count: {}", device_count);
818
819        let device_ids = get_device_ids(platform_id, CL_DEVICE_TYPE_ALL).unwrap();
820
821        let context = create_context(&device_ids, ptr::null(), None, ptr::null_mut());
822        let context = context.unwrap();
823
824        let source = r#"
825            kernel void saxpy_float (global float* z,
826                global float const* x,
827                global float const* y,
828                float a)
829            {
830            size_t i = get_global_id(0);
831            z[i] = a*x[i] + y[i];
832            }
833        "#;
834
835        // Convert source to an array
836        let sources = [source];
837        let program = create_program_with_source(context, &sources).unwrap();
838
839        use std::ffi::CString;
840        let no_options = CString::new("").unwrap();
841        compile_program(
842            program,
843            &device_ids,
844            &no_options,
845            &[],
846            &[],
847            None,
848            ptr::null_mut(),
849        )
850        .unwrap();
851
852        let programs = [program];
853        unsafe {
854            link_program(
855                context,
856                &device_ids,
857                &no_options,
858                &programs,
859                None,
860                ptr::null_mut(),
861            )
862            .unwrap()
863        };
864
865        unsafe {
866            release_program(program).unwrap();
867            release_context(context).unwrap();
868        }
869    }
870}
cl3/program.rs

cl3/
program.rs
