cl3/
program.rs

1// Copyright (c) 2020-2024 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_context, cl_device_id, cl_int, cl_platform_id, cl_program, cl_program_build_info,
23    cl_program_info, cl_uchar, cl_uint, CL_BUILD_ERROR, CL_BUILD_IN_PROGRESS, CL_BUILD_NONE,
24    CL_BUILD_SUCCESS, CL_FALSE, CL_INVALID_VALUE, CL_PROGRAM_BINARIES, CL_PROGRAM_BINARY_SIZES,
25    CL_PROGRAM_BINARY_TYPE, CL_PROGRAM_BINARY_TYPE_COMPILED_OBJECT,
26    CL_PROGRAM_BINARY_TYPE_EXECUTABLE, CL_PROGRAM_BINARY_TYPE_LIBRARY, CL_PROGRAM_BINARY_TYPE_NONE,
27    CL_PROGRAM_BUILD_GLOBAL_VARIABLE_TOTAL_SIZE, CL_PROGRAM_BUILD_LOG, CL_PROGRAM_BUILD_OPTIONS,
28    CL_PROGRAM_BUILD_STATUS, CL_PROGRAM_CONTEXT, CL_PROGRAM_DEVICES, CL_PROGRAM_IL,
29    CL_PROGRAM_KERNEL_NAMES, CL_PROGRAM_NUM_DEVICES, CL_PROGRAM_NUM_KERNELS,
30    CL_PROGRAM_REFERENCE_COUNT, CL_PROGRAM_SOURCE, CL_SUCCESS, CL_TRUE,
31};
32
33use super::info_type::InfoType;
34use super::{
35    api2_info_size, api2_info_value, api2_info_vector, api_info_size, api_info_value,
36    api_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            &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        &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        &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            &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        cl_call!(clBuildProgram(
248            program,
249            devices.len() as cl_uint,
250            devices.as_ptr(),
251            options.as_ptr(),
252            pfn_notify,
253            user_data,
254        ))
255    };
256    if CL_SUCCESS == status {
257        Ok(())
258    } else {
259        Err(status)
260    }
261}
262
263/// Compile a program’s source for the devices the `OpenCL` context associated
264/// with the program.
265/// Calls clCompileProgram to compile an `OpenCL` program object.
266///
267/// * `program` - a valid `OpenCL` program.
268/// * `devices` - a slice of devices that are in context.
269/// * `options` - the compilation options in a null-terminated string.
270/// * `input_headers` - a slice of programs that describe headers in the `input_headers`.
271/// * `header_include_names` - an array that has a one to one correspondence with
272///   `input_headers`.
273/// * `pfn_notify` - an optional function pointer to a notification routine.
274/// * `user_data` - passed as an argument when `pfn_notify` is called, or `ptr::null_mut()`.
275///
276/// returns a Result containing the new `OpenCL` program object
277/// or the error code from the `OpenCL` C API function.
278///
279/// # Panics
280///
281/// Panics if `input_headers.len()` != `header_include_names.len()`.
282#[cfg(any(feature = "CL_VERSION_1_2", feature = "dynamic"))]
283#[allow(clippy::cast_possible_truncation)]
284#[inline]
285pub fn compile_program(
286    program: cl_program,
287    devices: &[cl_device_id],
288    options: &CStr,
289    input_headers: &[cl_program],
290    header_include_names: &[&CStr],
291    pfn_notify: Option<unsafe extern "C" fn(program: cl_program, user_data: *mut c_void)>,
292    user_data: *mut c_void,
293) -> Result<(), cl_int> {
294    assert!(input_headers.len() == header_include_names.len());
295    let status: cl_int = unsafe {
296        let input_headers_ptr = if input_headers.is_empty() {
297            ptr::null()
298        } else {
299            input_headers.as_ptr()
300        };
301        let header_include_names_ptr = if header_include_names.is_empty() {
302            ptr::null()
303        } else {
304            header_include_names.as_ptr()
305        };
306        cl_call!(clCompileProgram(
307            program,
308            devices.len() as cl_uint,
309            devices.as_ptr(),
310            options.as_ptr(),
311            input_headers.len() as cl_uint,
312            input_headers_ptr,
313            header_include_names_ptr.cast::<*const c_char>(),
314            pfn_notify,
315            user_data,
316        ))
317    };
318    if CL_SUCCESS == status {
319        Ok(())
320    } else {
321        Err(status)
322    }
323}
324
325/// Link a set of compiled program objects and libraries for the devices in the
326/// `OpenCL` context associated with the program.
327/// Calls clLinkProgram to link an `OpenCL` program object.
328///
329/// * `context` - a valid `OpenCL` context.
330/// * `devices` - a slice of devices that are in context.
331/// * `options` - the link options in a null-terminated string.
332/// * `input_programs` - a slice of programs that are to be linked to create the program executable.
333/// * `pfn_notify` - an optional function pointer to a notification routine.
334/// * `user_data` - passed as an argument when `pfn_notify` is called, or `ptr::null_mut()`.
335///
336/// returns a Result containing the new `OpenCL` program object
337/// or the error code from the `OpenCL` C API function.
338///
339/// # Panics
340///
341/// Panics if `input_programs.is_empty()`.
342///
343/// # Safety
344///
345/// This is unsafe when a device is not a member of context.
346#[cfg(any(feature = "CL_VERSION_1_2", feature = "dynamic"))]
347#[allow(clippy::cast_possible_truncation)]
348#[inline]
349pub unsafe fn link_program(
350    context: cl_context,
351    devices: &[cl_device_id],
352    options: &CStr,
353    input_programs: &[cl_program],
354    pfn_notify: Option<unsafe extern "C" fn(program: cl_program, user_data: *mut c_void)>,
355    user_data: *mut c_void,
356) -> Result<cl_program, cl_int> {
357    assert!(!input_programs.is_empty());
358    let mut status: cl_int = CL_INVALID_VALUE;
359    let programme: cl_program = cl_call!(clLinkProgram(
360        context,
361        devices.len() as cl_uint,
362        devices.as_ptr(),
363        options.as_ptr(),
364        input_programs.len() as cl_uint,
365        input_programs.as_ptr(),
366        pfn_notify,
367        user_data,
368        &mut status,
369    ));
370    if CL_SUCCESS == status {
371        Ok(programme)
372    } else {
373        Err(status)
374    }
375}
376
377/// Set the value of a specialization constant.
378/// Calls `clSetProgramSpecializationConstant`.
379/// `CL_VERSION_2_2`
380///
381/// * `program` - the program.
382/// * `spec_id` - the specialization constant whose value will be set.
383/// * `spec_size` - size in bytes of the data pointed to by `spec_value`.
384/// * `spec_value` - pointer to the memory location that contains the value
385///   of the specialization constant.
386///
387/// returns an empty Result or the error code from the `OpenCL` C API function.
388///
389/// # Safety
390///
391/// This function is unsafe because `spec_size` and `spec_value` must be valid.
392#[cfg(any(feature = "CL_VERSION_2_2", feature = "dynamic"))]
393#[inline]
394pub unsafe fn set_program_specialization_constant(
395    program: cl_program,
396    spec_id: cl_uint,
397    spec_size: size_t,
398    spec_value: *const c_void,
399) -> Result<(), cl_int> {
400    let status: cl_int = cl_call!(clSetProgramSpecializationConstant(
401        program, spec_id, spec_size, spec_value
402    ));
403    if CL_SUCCESS == status {
404        Ok(())
405    } else {
406        Err(status)
407    }
408}
409
410/// Release the resources allocated by the `OpenCL` compiler for platform.
411/// Calls clUnloadPlatformCompiler.
412///
413/// * `platform` - the platform.
414///
415/// returns an empty Result or the error code from the `OpenCL` C API function.
416///
417/// # Safety
418///
419/// This function is unsafe because the platform compiler is not valid after this call.
420#[cfg(any(feature = "CL_VERSION_1_2", feature = "dynamic"))]
421#[inline]
422pub unsafe fn unload_platform_compiler(platform: cl_platform_id) -> Result<(), cl_int> {
423    let status: cl_int = cl_call!(clUnloadPlatformCompiler(platform));
424    if CL_SUCCESS == status {
425        Ok(())
426    } else {
427        Err(status)
428    }
429}
430
431/// Get data about an `OpenCL` program.
432/// Calls clGetProgramInfo to get the desired data about the program.
433pub fn get_program_data(
434    program: cl_program,
435    param_name: cl_program_info,
436) -> Result<Vec<u8>, cl_int> {
437    api_info_size!(get_size, clGetProgramInfo);
438    let size = get_size(program, param_name)?;
439    api_info_vector!(get_vector, u8, clGetProgramInfo);
440    get_vector(program, param_name, size)
441}
442
443/// Get specific information about an `OpenCL` program.
444/// Calls clGetProgramInfo to get the desired information about the program.
445///
446/// * `program` - the `OpenCL` program.
447/// * `param_name` - the type of program information being queried, see:
448///   [Program Object Queries](https://www.khronos.org/registry/OpenCL/specs/3.0-unified/html/OpenCL_API.html#program-info-table).
449///
450/// returns a Result containing the desired information in an `InfoType` enum
451/// or the error code from the `OpenCL` C API function.
452pub fn get_program_info(
453    program: cl_program,
454    param_name: cl_program_info,
455) -> Result<InfoType, cl_int> {
456    api_info_size!(get_size, clGetProgramInfo);
457
458    match param_name {
459        CL_PROGRAM_REFERENCE_COUNT
460        | CL_PROGRAM_NUM_DEVICES
461        | CL_PROGRAM_SCOPE_GLOBAL_CTORS_PRESENT // CL_VERSION_2_2 only
462        | CL_PROGRAM_SCOPE_GLOBAL_DTORS_PRESENT // CL_VERSION_2_2 only
463        => {
464            api_info_value!(get_value, cl_uint, clGetProgramInfo);
465            Ok(InfoType::Uint(get_value(program, param_name)?))
466        }
467
468        CL_PROGRAM_CONTEXT => {
469            api_info_value!(get_value, intptr_t, clGetProgramInfo);
470            Ok(InfoType::Ptr(get_value(program, param_name)?))
471        }
472
473        CL_PROGRAM_DEVICES => {
474            api_info_vector!(get_vec, intptr_t, clGetProgramInfo);
475            let size = get_size(program, param_name)?;
476            Ok(InfoType::VecIntPtr(get_vec(program, param_name, size)?))
477        }
478
479        CL_PROGRAM_BINARY_SIZES => {
480            api_info_vector!(get_vec, size_t, clGetProgramInfo);
481            let size = get_size(program, param_name)?;
482            Ok(InfoType::VecSize(get_vec(program, param_name, size)?))
483        }
484
485        CL_PROGRAM_BINARIES => {
486            // Gets the binaries for all the devices in the context
487
488            // get the binary sizes, as the case above
489            api_info_vector!(get_size_vec, size_t, clGetProgramInfo);
490            let size = get_size(program, CL_PROGRAM_BINARY_SIZES as cl_program_info)?;
491            let binary_sizes = get_size_vec(program, CL_PROGRAM_BINARY_SIZES as cl_program_info, size)?;
492
493            // A vector of vectors to hold the binaries of each device
494            let binaries = binary_sizes.into_iter().map(|size| {
495                vec![0u8; size]
496            }).collect::<Vec<Vec<u8>>>();
497
498            // Create a vector of pointers to the vectors in binaries
499            let mut binary_ptrs = binaries.iter().map(|vec| {
500                vec.as_ptr()
501            }).collect::<Vec<_>>();
502
503            let status = unsafe {
504                cl_call!(clGetProgramInfo(
505                    program,
506                    param_name,
507                    binary_ptrs.len() * mem::size_of::<*mut c_void>(),
508                    binary_ptrs.as_mut_ptr().cast(),
509                    ptr::null_mut(),
510                ))
511            };
512            if CL_SUCCESS == status {
513                Ok(InfoType::VecVecUchar(binaries))
514            } else {
515                Err(status)
516            }
517        }
518
519        CL_PROGRAM_NUM_KERNELS => {
520            api_info_value!(get_value, size_t, clGetProgramInfo);
521            Ok(InfoType::Size(get_value(program, param_name)?))
522        }
523
524        CL_PROGRAM_SOURCE
525        | CL_PROGRAM_KERNEL_NAMES
526        | CL_PROGRAM_IL
527        | _ => {
528            Ok(InfoType::VecUchar(get_program_data(program, param_name)?))
529        }
530    }
531}
532
533/// Get data about an `OpenCL` program build.
534/// Calls clGetProgramBuildInfo to get the desired data about the program build.
535pub fn get_program_build_data(
536    program: cl_program,
537    device: cl_device_id,
538    param_name: cl_program_info,
539) -> Result<Vec<u8>, cl_int> {
540    api2_info_size!(get_size, cl_device_id, clGetProgramBuildInfo);
541    let size = get_size(program, device, param_name)?;
542    api2_info_vector!(get_vector, cl_device_id, u8, clGetProgramBuildInfo);
543    get_vector(program, device, param_name, size)
544}
545
546/// Get specific information about an `OpenCL` program build.
547/// Calls clGetProgramBuildInfo to get the desired information about the program build.
548///
549/// * `program` - the `OpenCL` program.
550/// * `device` - -the device for which build information is being queried.
551/// * `param_name` - the type of program build information being queried, see:
552///   [Program Build Queries](https://www.khronos.org/registry/OpenCL/specs/3.0-unified/html/OpenCL_API.html#program-build-info-table).
553///
554/// returns a Result containing the desired information in an `InfoType` enum
555/// or the error code from the `OpenCL` C API function.
556pub fn get_program_build_info(
557    program: cl_program,
558    device: cl_device_id,
559    param_name: cl_program_build_info,
560) -> Result<InfoType, cl_int> {
561    match param_name {
562        CL_PROGRAM_BUILD_STATUS => {
563            api2_info_value!(
564                get_device_value,
565                cl_device_id,
566                cl_int,
567                clGetProgramBuildInfo
568            );
569            Ok(InfoType::Int(get_device_value(
570                program, device, param_name,
571            )?))
572        }
573
574        CL_PROGRAM_BINARY_TYPE => {
575            api2_info_value!(
576                get_device_value,
577                cl_device_id,
578                cl_uint,
579                clGetProgramBuildInfo
580            );
581            Ok(InfoType::Uint(get_device_value(
582                program, device, param_name,
583            )?))
584        }
585
586        // CL_VERSION_2_0
587        CL_PROGRAM_BUILD_GLOBAL_VARIABLE_TOTAL_SIZE => {
588            api2_info_value!(
589                get_device_value,
590                cl_device_id,
591                size_t,
592                clGetProgramBuildInfo
593            );
594            Ok(InfoType::Size(get_device_value(
595                program, device, param_name,
596            )?))
597        }
598
599        CL_PROGRAM_BUILD_OPTIONS | CL_PROGRAM_BUILD_LOG | _ => Ok(InfoType::VecUchar(
600            get_program_build_data(program, device, param_name)?,
601        )),
602    }
603}
604
605#[cfg(test)]
606mod tests {
607    use super::*;
608    use crate::context::{create_context, release_context};
609    use crate::device::{get_device_ids, CL_DEVICE_TYPE_ALL};
610    #[allow(unused_imports)]
611    use crate::error_codes::error_text;
612    use crate::platform::get_platform_ids;
613    use std::ffi::CString;
614
615    #[test]
616    fn test_program() {
617        let platform_ids = get_platform_ids().unwrap();
618
619        let mut platform_id = platform_ids[0];
620        let mut device_count: usize = 0;
621
622        // Search for a platform with the most devices
623        for p in platform_ids {
624            let ids = get_device_ids(p, CL_DEVICE_TYPE_ALL).unwrap();
625            let count = ids.len();
626            if device_count < count {
627                device_count = count;
628                platform_id = p;
629            }
630        }
631
632        println!("Platform device_count: {}", device_count);
633
634        let device_ids = get_device_ids(platform_id, CL_DEVICE_TYPE_ALL).unwrap();
635        let device_id = device_ids[0];
636
637        let context = create_context(&device_ids, ptr::null(), None, ptr::null_mut());
638        let context = context.unwrap();
639
640        let source = r#"
641            kernel void saxpy_float (global float* z,
642                global float const* x,
643                global float const* y,
644                float a)
645            {
646            size_t i = get_global_id(0);
647            z[i] = a*x[i] + y[i];
648            }
649        "#;
650
651        // Convert source to an array
652        let sources = [source];
653        let program = create_program_with_source(context, &sources).unwrap();
654
655        let value = get_program_info(program, CL_PROGRAM_REFERENCE_COUNT).unwrap();
656        let value = cl_uint::from(value);
657        println!("CL_PROGRAM_REFERENCE_COUNT: {}", value);
658        assert!(0 < value);
659
660        let value = get_program_info(program, CL_PROGRAM_CONTEXT).unwrap();
661        let value = intptr_t::from(value);
662        println!("CL_PROGRAM_CONTEXT: {}", value);
663        assert!(0 < value);
664
665        let value = get_program_info(program, CL_PROGRAM_NUM_DEVICES).unwrap();
666        let value = cl_uint::from(value);
667        println!("CL_PROGRAM_NUM_DEVICES: {}", value);
668        assert!(0 < value);
669
670        let value = get_program_info(program, CL_PROGRAM_DEVICES).unwrap();
671        let value = Vec::<intptr_t>::from(value);
672        println!("CL_PROGRAM_DEVICES: {}", value.len());
673        assert!(0 < value.len());
674
675        let value = get_program_info(program, CL_PROGRAM_SOURCE).unwrap();
676        let value = String::from(value);
677        println!("CL_PROGRAM_SOURCE: {}", value);
678        assert!(0 < value.len());
679
680        let options = CString::default();
681        build_program(program, &device_ids, &options, None, ptr::null_mut()).unwrap();
682
683        let value = get_program_build_info(program, device_id, CL_PROGRAM_BUILD_STATUS).unwrap();
684        let value: cl_int = From::from(value);
685        println!("CL_PROGRAM_BUILD_STATUS: {}", value);
686        assert_eq!(CL_BUILD_SUCCESS, value);
687
688        let value = get_program_build_info(program, device_id, CL_PROGRAM_BUILD_OPTIONS).unwrap();
689        let value = String::from(value);
690        println!("CL_PROGRAM_BUILD_OPTIONS: {}", value);
691
692        let value = get_program_build_info(program, device_id, CL_PROGRAM_BUILD_LOG).unwrap();
693        let value = String::from(value);
694        println!("CL_PROGRAM_BUILD_LOG: {}", value);
695
696        let value = get_program_build_info(program, device_id, CL_PROGRAM_BINARY_TYPE).unwrap();
697        let value = cl_uint::from(value);
698        println!("CL_PROGRAM_BINARY_TYPE: {:?}", value);
699        assert_eq!(CL_PROGRAM_BINARY_TYPE_EXECUTABLE as cl_uint, value);
700
701        #[cfg(any(feature = "CL_VERSION_2_0", feature = "dynamic"))]
702        match get_program_build_info(
703            program,
704            device_id,
705            CL_PROGRAM_BUILD_GLOBAL_VARIABLE_TOTAL_SIZE,
706        ) {
707            Ok(value) => {
708                let value = size_t::from(value);
709                println!("CL_PROGRAM_BUILD_GLOBAL_VARIABLE_TOTAL_SIZE: {:?}", value)
710            }
711            Err(e) => println!(
712                "OpenCL error, CL_PROGRAM_BUILD_GLOBAL_VARIABLE_TOTAL_SIZE: {}",
713                error_text(e)
714            ),
715        }
716
717        let value = get_program_info(program, CL_PROGRAM_BINARY_SIZES).unwrap();
718        let value = Vec::<size_t>::from(value);
719        println!("CL_PROGRAM_BINARY_SIZES: {}", value.len());
720        println!("CL_PROGRAM_BINARY_SIZES: {:?}", value);
721        assert!(0 < value.len());
722
723        let value = get_program_info(program, CL_PROGRAM_BINARIES).unwrap();
724        // println!("CL_PROGRAM_BINARIES: {:?}", value);
725        let value = Vec::<Vec<u8>>::from(value);
726        println!("CL_PROGRAM_BINARIES count: {}", value.len());
727        println!("CL_PROGRAM_BINARIES length[0]: {}", value[0].len());
728        assert!(0 < value.len());
729
730        let value = get_program_info(program, CL_PROGRAM_NUM_KERNELS).unwrap();
731        let value = size_t::from(value);
732        println!("CL_PROGRAM_NUM_KERNELS: {}", value);
733        assert!(0 < value);
734
735        let value = get_program_info(program, CL_PROGRAM_KERNEL_NAMES).unwrap();
736        let value = String::from(value);
737        println!("CL_PROGRAM_KERNEL_NAMES: {}", value);
738        assert!(0 < value.len());
739
740        #[cfg(any(feature = "CL_VERSION_2_1", feature = "dynamic"))]
741        match get_program_info(program, CL_PROGRAM_IL) {
742            Ok(value) => {
743                let value = String::from(value);
744                println!("CL_PROGRAM_IL: {}", value)
745            }
746            Err(e) => println!("OpenCL error, CL_PROGRAM_IL: {}", error_text(e)),
747        };
748
749        #[cfg(any(feature = "CL_VERSION_2_2", feature = "dynamic"))]
750        match get_program_info(program, CL_PROGRAM_SCOPE_GLOBAL_CTORS_PRESENT) {
751            Ok(value) => {
752                let value = cl_uint::from(value);
753                println!("CL_PROGRAM_SCOPE_GLOBAL_CTORS_PRESENT: {}", value)
754            }
755            Err(e) => println!(
756                "OpenCL error, CL_PROGRAM_SCOPE_GLOBAL_CTORS_PRESENT: {}",
757                error_text(e)
758            ),
759        };
760
761        #[cfg(any(feature = "CL_VERSION_2_2", feature = "dynamic"))]
762        match get_program_info(program, CL_PROGRAM_SCOPE_GLOBAL_CTORS_PRESENT) {
763            Ok(value) => {
764                let value = cl_uint::from(value);
765                println!("CL_PROGRAM_SCOPE_GLOBAL_DTORS_PRESENT: {}", value)
766            }
767            Err(e) => println!(
768                "OpenCL error, CL_PROGRAM_SCOPE_GLOBAL_DTORS_PRESENT: {}",
769                error_text(e)
770            ),
771        };
772
773        #[cfg(any(feature = "CL_VERSION_1_2", feature = "dynamic"))]
774        if let Err(e) = unsafe { unload_platform_compiler(platform_id) } {
775            println!("OpenCL error, clUnloadPlatformCompiler: {}", error_text(e));
776        }
777
778        unsafe {
779            release_program(program).unwrap();
780            release_context(context).unwrap();
781        }
782    }
783
784    #[test]
785    fn test_compile_and_link_program() {
786        let platform_ids = get_platform_ids().unwrap();
787
788        let mut platform_id = platform_ids[0];
789        let mut device_count: usize = 0;
790
791        // Search for a platform with the most devices
792        for p in platform_ids {
793            let ids = get_device_ids(p, CL_DEVICE_TYPE_ALL).unwrap();
794            let count = ids.len();
795            if device_count < count {
796                device_count = count;
797                platform_id = p;
798            }
799        }
800
801        println!("Platform device_count: {}", device_count);
802
803        let device_ids = get_device_ids(platform_id, CL_DEVICE_TYPE_ALL).unwrap();
804
805        let context = create_context(&device_ids, ptr::null(), None, ptr::null_mut());
806        let context = context.unwrap();
807
808        let source = r#"
809            kernel void saxpy_float (global float* z,
810                global float const* x,
811                global float const* y,
812                float a)
813            {
814            size_t i = get_global_id(0);
815            z[i] = a*x[i] + y[i];
816            }
817        "#;
818
819        // Convert source to an array
820        let sources = [source];
821        let program = create_program_with_source(context, &sources).unwrap();
822
823        use std::ffi::CString;
824        let no_options = CString::new("").unwrap();
825        compile_program(
826            program,
827            &device_ids,
828            &no_options,
829            &[],
830            &[],
831            None,
832            ptr::null_mut(),
833        )
834        .unwrap();
835
836        let programs = [program];
837        unsafe {
838            link_program(
839                context,
840                &device_ids,
841                &no_options,
842                &programs,
843                None,
844                ptr::null_mut(),
845            )
846            .unwrap()
847        };
848
849        unsafe {
850            release_program(program).unwrap();
851            release_context(context).unwrap();
852        }
853    }
854}
cl3/program.rs

cl3/
program.rs
