mirror of
https://github.com/boostorg/compute.git
synced 2026-01-21 04:42:41 +00:00
5d77bbebee
refs kylelutz/compute#9 device, context, and queue are initialized statically in `context_setup.hpp`. With this change all tests are able to complete when an NVIDIA GPU is in exclusive compute mode. Side effect of the change: Time for all tests to complete reduced from 15.71 to 13.03 sec Tesla C2075.
98 lines
3.2 KiB
C++
98 lines
3.2 KiB
C++
//---------------------------------------------------------------------------//
|
|
// Copyright (c) 2013 Kyle Lutz <kyle.r.lutz@gmail.com>
|
|
//
|
|
// Distributed under the Boost Software License, Version 1.0
|
|
// See accompanying file LICENSE_1_0.txt or copy at
|
|
// http://www.boost.org/LICENSE_1_0.txt
|
|
//
|
|
// See http://kylelutz.github.com/compute for more information.
|
|
//---------------------------------------------------------------------------//
|
|
|
|
#define BOOST_TEST_MODULE TestKernel
|
|
#include <boost/test/unit_test.hpp>
|
|
|
|
#include <boost/compute/kernel.hpp>
|
|
#include <boost/compute/system.hpp>
|
|
|
|
#include "context_setup.hpp"
|
|
|
|
namespace bc = boost::compute;
|
|
|
|
BOOST_AUTO_TEST_CASE(name)
|
|
{
|
|
bc::kernel foo = bc::kernel::create_with_source("__kernel void foo(int x) { }",
|
|
"foo",
|
|
context);
|
|
BOOST_CHECK_EQUAL(foo.name(), "foo");
|
|
|
|
bc::kernel bar = bc::kernel::create_with_source("__kernel void bar(float x) { }",
|
|
"bar",
|
|
context);
|
|
BOOST_CHECK_EQUAL(bar.name(), "bar");
|
|
}
|
|
|
|
BOOST_AUTO_TEST_CASE(num_args)
|
|
{
|
|
bc::kernel foo = bc::kernel::create_with_source("__kernel void foo(int x) { }",
|
|
"foo",
|
|
context);
|
|
BOOST_CHECK_EQUAL(foo.num_args(), size_t(1));
|
|
|
|
bc::kernel bar = bc::kernel::create_with_source("__kernel void bar(float x, float y) { }",
|
|
"bar",
|
|
context);
|
|
BOOST_CHECK_EQUAL(bar.num_args(), size_t(2));
|
|
|
|
bc::kernel baz = bc::kernel::create_with_source("__kernel void baz(char x, char y, char z) { }",
|
|
"baz",
|
|
context);
|
|
BOOST_CHECK_EQUAL(baz.num_args(), size_t(3));
|
|
}
|
|
|
|
BOOST_AUTO_TEST_CASE(get_work_group_info)
|
|
{
|
|
const char source[] =
|
|
"__kernel void sum(__global const float *input,\n"
|
|
" __global float *output)\n"
|
|
"{\n"
|
|
" __local float scratch[16];\n"
|
|
" const uint gid = get_global_id(0);\n"
|
|
" const uint lid = get_local_id(0);\n"
|
|
" if(lid < 16)\n"
|
|
" scratch[lid] = input[gid];\n"
|
|
"}\n";
|
|
|
|
boost::compute::program program =
|
|
boost::compute::program::create_with_source(source, context);
|
|
|
|
program.build();
|
|
|
|
boost::compute::kernel kernel = program.create_kernel("sum");
|
|
|
|
using boost::compute::ulong_;
|
|
|
|
// get local memory size
|
|
kernel.get_work_group_info<ulong_>(device, CL_KERNEL_LOCAL_MEM_SIZE);
|
|
|
|
// check work group size
|
|
size_t work_group_size =
|
|
kernel.get_work_group_info<size_t>(device, CL_KERNEL_WORK_GROUP_SIZE);
|
|
BOOST_CHECK(work_group_size >= 1);
|
|
}
|
|
|
|
#ifndef BOOST_NO_VARIADIC_TEMPLATES
|
|
BOOST_AUTO_TEST_CASE(kernel_set_args)
|
|
{
|
|
bc::kernel k =
|
|
bc::kernel::create_with_source(
|
|
"__kernel void test(int x, float y, char z) { }",
|
|
"test",
|
|
context
|
|
);
|
|
|
|
k.set_args(4, 2.4f, 'a');
|
|
}
|
|
#endif
|
|
|
|
BOOST_AUTO_TEST_SUITE_END()
|