Skip to content

HTTPS clone URL

Subversion checkout URL

You can clone with HTTPS or Subversion.

Download ZIP

Comparing changes

Choose two branches to see what's changed or to start a new pull request. If you need to, you can also compare across forks.

Open a pull request

Create a new pull request by comparing changes across two branches. If you need to, you can also compare across forks.
base fork: actor-framework/actor-framework
...
head fork: actor-framework/actor-framework
Checking mergeability… Don't worry, you can still create the pull request.
  • 5 commits
  • 9 files changed
  • 0 commit comments
  • 2 contributors
View
2  cppa.files
@@ -303,3 +303,5 @@ unit_testing/test_sync_send.cpp
unit_testing/test_tuple.cpp
unit_testing/test_uniform_type.cpp
unit_testing/test_yield_interface.cpp
+examples/opencl/simple_matrix.cpp
+examples/opencl/proper_matrix.cpp
View
32 cppa/opencl.hpp
@@ -56,7 +56,7 @@ template<typename MapArgs, typename MapResult>
struct get_cl_spawn_helper;
template<typename R, typename... Ts>
-struct get_cl_spawn_helper<std::function<option<cow_tuple<Ts...>> (const any_tuple&)>,
+struct get_cl_spawn_helper<std::function<option<cow_tuple<Ts...>> (any_tuple)>,
std::function<any_tuple (R&)>> {
typedef cl_spawn_helper<R (const Ts&...)> type;
};
@@ -91,6 +91,36 @@ actor_ptr spawn_cl(const char* source,
f0{map_args}, f1{map_result});
}
+template<typename Signature, typename... Ts>
+actor_ptr spawn_cl(const opencl::program& prog,
+ const char* fun_name,
+ std::vector<size_t> dimensions,
+ std::vector<size_t> offset = {},
+ std::vector<size_t> local_dims = {}) {
+ using std::move;
+ cl_spawn_helper<Signature> f;
+ return f(prog, fun_name, move(dimensions), move(offset), move(local_dims));
+}
+
+template<typename MapArgs, typename MapResult>
+actor_ptr spawn_cl(const opencl::program& prog,
+ const char* fun_name,
+ MapArgs map_args,
+ MapResult map_result,
+ std::vector<size_t> dimensions,
+ std::vector<size_t> offset = {},
+ std::vector<size_t> local_dims = {}) {
+ using std::move;
+ typedef typename util::get_callable_trait<MapArgs>::type t0;
+ typedef typename util::get_callable_trait<MapResult>::type t1;
+ typedef typename t0::fun_type f0;
+ typedef typename t1::fun_type f1;
+ typename get_cl_spawn_helper<f0,f1>::type f;
+ return f(prog, fun_name,
+ move(dimensions), move(offset), move(local_dims),
+ f0{map_args}, f1{map_result});
+}
+
} // namespace cppa
#endif // CPPA_OPENCL_HPP
View
4 cppa/opencl/actor_facade.hpp
@@ -69,7 +69,7 @@ class actor_facade<Ret(Args...)> : public actor {
public:
typedef cow_tuple<typename util::rm_ref<Args>::type...> args_tuple;
- typedef std::function<option<args_tuple>(const any_tuple&)> arg_mapping;
+ typedef std::function<option<args_tuple>(any_tuple)> arg_mapping;
typedef std::function<any_tuple(Ret&)> result_mapping;
static actor_facade* create(command_dispatcher* dispatcher,
@@ -152,7 +152,7 @@ class actor_facade<Ret(Args...)> : public actor {
void enqueue_impl(const actor_ptr& sender, any_tuple msg, message_id id, util::int_list<Is...>) {
auto opt = m_map_args(msg);
if (opt) {
- response_handle handle{this, sender, id};
+ response_handle handle{this, sender, id.response_id()};
size_t number_of_values{1};
std::for_each(m_global_dimensions.begin(),
m_global_dimensions.end(),
View
6 cppa/opencl/command.hpp
@@ -148,13 +148,13 @@ class command_impl : public command {
/* get results from gpu */
cl_int err{0};
cl_event read_event;
- T results(m_number_of_values);
+ T result(m_number_of_values);
err = clEnqueueReadBuffer(m_queue.get(),
m_arguments[0].get(),
CL_TRUE,
0,
sizeof(typename T::value_type) * m_number_of_values,
- results.data(),
+ result.data(),
0,
NULL,
&read_event);
@@ -164,7 +164,7 @@ class command_impl : public command {
+ get_opencl_error(err)
+ "'.");
}
- auto mapped_result = m_map_result(results);
+ auto mapped_result = m_map_result(result);
reply_tuple_to(m_handle, mapped_result);
//reply_to(m_handle, results);
}
View
2  cppa/opencl/command_dispatcher.hpp
@@ -87,7 +87,7 @@ class command_dispatcher {
std::vector<size_t> global_dims,
std::vector<size_t> global_offs,
std::vector<size_t> local_dims,
- std::function<option<cow_tuple<typename util::rm_ref<Args>::type...>>(const any_tuple&)> map_args,
+ std::function<option<cow_tuple<typename util::rm_ref<Args>::type...>>(any_tuple)> map_args,
std::function<any_tuple(Ret&)> map_result)
{
return actor_facade<Ret (Args...)>::create(this,
View
11 examples/CMakeLists.txt
@@ -46,3 +46,14 @@ if (NOT "${CPPA_NO_QT_EXAMPLES}" STREQUAL "yes")
add_dependencies(qt_group_chat all_examples)
endif (QT4_FOUND)
endif()
+
+if (ENABLE_OPENCL)
+ macro(add_opencl name folder)
+ add_executable(${name} ${folder}/${name}.cpp ${ARGN})
+ target_link_libraries(${name} ${CMAKE_DL_LIBS} ${CPPA_LIBRARY} ${PTHREAD_LIBRARIES} ${OPENCL_LIBRARIES})
+ add_dependencies(${name} all_examples)
+ endmacro()
+
+ add_opencl(simple_matrix opencl)
+ add_opencl(proper_matrix opencl)
+endif (ENABLE_OPENCL)
View
201 examples/opencl/proper_matrix.cpp
@@ -0,0 +1,201 @@
+/******************************************************************************\
+ * ___ __ *
+ * /\_ \ __/\ \ *
+ * \//\ \ /\_\ \ \____ ___ _____ _____ __ *
+ * \ \ \ \/\ \ \ '__`\ /'___\/\ '__`\/\ '__`\ /'__`\ *
+ * \_\ \_\ \ \ \ \L\ \/\ \__/\ \ \L\ \ \ \L\ \/\ \L\.\_ *
+ * /\____\\ \_\ \_,__/\ \____\\ \ ,__/\ \ ,__/\ \__/.\_\ *
+ * \/____/ \/_/\/___/ \/____/ \ \ \/ \ \ \/ \/__/\/_/ *
+ * \ \_\ \ \_\ *
+ * \/_/ \/_/ *
+ * *
+ * Copyright (C) 2011-2013 *
+ * Dominik Charousset <dominik.charousset@haw-hamburg.de> *
+ * Raphael Hiesgen <raphael.hiesgen@haw-hamburg.de> *
+ * *
+ * This file is part of libcppa. *
+ * libcppa is free software: you can redistribute it and/or modify it under *
+ * the terms of the GNU Lesser General Public License as published by the *
+ * Free Software Foundation, either version 3 of the License *
+ * or (at your option) any later version. *
+ * *
+ * libcppa is distributed in the hope that it will be useful, *
+ * but WITHOUT ANY WARRANTY; without even the implied warranty of *
+ * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. *
+ * See the GNU Lesser General Public License for more details. *
+ * *
+ * You should have received a copy of the GNU Lesser General Public License *
+ * along with libcppa. If not, see <http://www.gnu.org/licenses/>. *
+\******************************************************************************/
+
+#include <vector>
+#include <iomanip>
+#include <numeric>
+#include <iostream>
+
+#include "cppa/cppa.hpp"
+#include "cppa/opencl.hpp"
+
+using namespace std;
+using namespace cppa;
+
+static constexpr size_t matrix_size = 8;
+static constexpr const char* kernel_name = "matrix_mult";
+
+
+// opencl kernel, multiplies matrix1 and matrix2
+namespace { constexpr const char* kernel_source = R"__(
+ __kernel void matrix_mult(__global float* matrix1,
+ __global float* matrix2,
+ __global float* output) {
+ int size = get_global_size(0); // == get_global_size_(1);
+ int x = get_global_id(0); // get position in matrix
+ int y = get_global_id(1); // this work item should calculate
+ int idx = 0;
+ float result = 0;
+ while (idx < size) { // calculate one element of the matrix
+ float i = matrix1[idx+y*size];
+ float j = matrix2[x+idx*size];
+ float tmp = i*j;
+ result = result + tmp;
+ ++idx;
+ }
+ output[x+y*size] = result;
+ }
+)__"; }
+
+// represensts a square matrix
+template<size_t Size>
+class square_matrix {
+
+ public:
+
+ static constexpr size_t num_elements = Size * Size;
+
+ square_matrix(square_matrix&&) = default;
+ square_matrix(const square_matrix&) = default;
+ square_matrix& operator=(square_matrix&&) = default;
+ square_matrix& operator=(const square_matrix&) = default;
+
+ square_matrix() {
+ m_data.resize(num_elements);
+ }
+
+ square_matrix(vector<float> d) : m_data(std::move(d)) { }
+
+ square_matrix(const std::initializer_list<float>& args) : m_data(args) {
+ m_data.resize(num_elements);
+ }
+
+ inline float& operator()(size_t row, size_t column) {
+ return m_data[row + column * Size];
+ }
+
+ inline const float& operator()(size_t row, size_t column) const {
+ return m_data[row + column * Size];
+ }
+
+ inline void zeroize() {
+ std::fill(m_data.begin(), m_data.end(), 0);
+ }
+
+ inline void iota_fill() {
+ std::iota(m_data.begin(), m_data.end(), 0);
+ }
+
+ typedef typename vector<float>::const_iterator const_iterator;
+
+ const_iterator begin() const { return m_data.begin(); }
+
+ const_iterator end() const { return m_data.end(); }
+
+ vector<float>& data() { return m_data; }
+
+ const vector<float>& data() const { return m_data; }
+
+ private:
+
+ vector<float> m_data;
+
+};
+
+// to annouce the square_matrix to lobccpa
+// these operaters have to be implemented
+template<size_t Size>
+inline bool operator==(const square_matrix<Size>& lhs, const square_matrix<Size>& rhs) {
+ return std::equal(lhs.begin(), lhs.end(), rhs.begin());
+}
+
+template<size_t Size>
+inline bool operator!=(const square_matrix<Size>& lhs, const square_matrix<Size>& rhs) {
+ return !(lhs == rhs);
+}
+
+using matrix_type = square_matrix<matrix_size>;
+
+void multiplier() {
+
+ // create two matrix and fill them with values 0, 1, 2, 3, 4, ...
+ matrix_type m1;
+ matrix_type m2;
+ m1.iota_fill();
+ m2.iota_fill();
+
+ // print "source" matrix
+ cout << "calculating square of matrix:" << endl;
+ for (size_t y = 0; y < matrix_size; ++y) {
+ for (size_t x = 0; x < matrix_size; ++x) {
+ cout << fixed << setprecision(2) << setw(6) << m1(x,y);
+ }
+ cout << endl;
+ }
+ cout << endl;
+
+ // spawn an opencl actor
+ // 1st arg: source code of one or more opencl kernels
+ // 2nd arg: name of the kernel to use
+ auto worker = spawn_cl(kernel_source, kernel_name,
+ // 3rd arg: the opencl actor can only handle vectors,
+ // this function creates vectors from a tuple of matrix_types
+ [] (any_tuple msg) -> option<cow_tuple<vector<float>,vector<float>>> {
+ auto opt = tuple_cast<matrix_type,matrix_type>(msg);
+ if (opt) {
+ return {move(get_ref<0>(*opt).data()),
+ move(get_ref<1>(*opt).data())};
+ }
+ return {};
+ },
+ // 4th arg: builds a matrix_type from a vector
+ // allows the actor to pass a matrix_type back instead of a vector
+ [] (vector<float>& result) -> any_tuple {
+ return make_any_tuple<matrix_type>(move(result));
+ },
+ // 5th arg: global dimension arguments for opencl's enqueue
+ // creates matrix_size * matrix_size global work items
+ {matrix_size, matrix_size}
+ );
+
+ // send both matrices to the actor and
+ // wait for results in form of a matrix_type
+ sync_send(worker, move(m1), move(m2)).then(
+ [](const matrix_type& result) {
+ cout << "result:" << endl;
+ for (size_t y = 0; y < matrix_size; ++y) {
+ for (size_t x = 0; x < matrix_size; ++x) {
+ cout << fixed << setprecision(2) << setw(9) << result(x,y);
+ }
+ cout << endl;
+ }
+ }
+ );
+}
+
+int main() {
+ // matrix_type ist not a simple type,
+ // it must be annouced to libcppa
+ announce<matrix_type>();
+ spawn(multiplier);
+ await_all_others_done();
+ shutdown();
+ return 0;
+}
View
117 examples/opencl/simple_matrix.cpp
@@ -0,0 +1,117 @@
+/******************************************************************************\
+ * ___ __ *
+ * /\_ \ __/\ \ *
+ * \//\ \ /\_\ \ \____ ___ _____ _____ __ *
+ * \ \ \ \/\ \ \ '__`\ /'___\/\ '__`\/\ '__`\ /'__`\ *
+ * \_\ \_\ \ \ \ \L\ \/\ \__/\ \ \L\ \ \ \L\ \/\ \L\.\_ *
+ * /\____\\ \_\ \_,__/\ \____\\ \ ,__/\ \ ,__/\ \__/.\_\ *
+ * \/____/ \/_/\/___/ \/____/ \ \ \/ \ \ \/ \/__/\/_/ *
+ * \ \_\ \ \_\ *
+ * \/_/ \/_/ *
+ * *
+ * Copyright (C) 2011-2013 *
+ * Dominik Charousset <dominik.charousset@haw-hamburg.de> *
+ * Raphael Hiesgen <raphael.hiesgen@haw-hamburg.de> *
+ * *
+ * This file is part of libcppa. *
+ * libcppa is free software: you can redistribute it and/or modify it under *
+ * the terms of the GNU Lesser General Public License as published by the *
+ * Free Software Foundation, either version 3 of the License *
+ * or (at your option) any later version. *
+ * *
+ * libcppa is distributed in the hope that it will be useful, *
+ * but WITHOUT ANY WARRANTY; without even the implied warranty of *
+ * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. *
+ * See the GNU Lesser General Public License for more details. *
+ * *
+ * You should have received a copy of the GNU Lesser General Public License *
+ * along with libcppa. If not, see <http://www.gnu.org/licenses/>. *
+\******************************************************************************/
+
+#include <vector>
+#include <iomanip>
+#include <numeric>
+#include <iostream>
+
+#include "cppa/cppa.hpp"
+#include "cppa/opencl.hpp"
+
+using namespace std;
+using namespace cppa;
+
+static constexpr size_t matrix_size = 8;
+static constexpr const char* kernel_name = "matrix_mult";
+
+// opencl kernel, multiplies matrix1 and matrix2
+namespace { constexpr const char* kernel_source = R"__(
+ __kernel void matrix_mult(__global float* matrix1,
+ __global float* matrix2,
+ __global float* output) {
+ int size = get_global_size(0); // == get_global_size_(1);
+ int x = get_global_id(0);
+ int y = get_global_id(1);
+ int idx = 0;
+ float result = 0;
+ while (idx < size) {
+ float i = matrix1[idx+y*size];
+ float j = matrix2[x+idx*size];
+ float tmp = i*j;
+ result = result + tmp;
+ ++idx;
+ }
+ output[x+y*size] = result;
+ }
+)__"; }
+
+void multiplier() {
+ // the opencl actor only understands vectors
+ // so these vectors represent the matrices
+ vector<float> m1(matrix_size * matrix_size);
+ vector<float> m2(matrix_size * matrix_size);
+
+ // fill each with values from 0 to matix_size * matrix_size - 1
+ iota(m1.begin(), m1.end(), 0);
+ iota(m2.begin(), m2.end(), 0);
+
+ // print "source" matrix
+ cout << "calculating square of matrix:" << endl;
+ for (size_t y = 0; y < matrix_size; ++y) {
+ for (size_t x = 0; x < matrix_size; ++x) {
+ cout << fixed << setprecision(2) << setw(6) << m1[x+y*matrix_size];
+ }
+ cout << endl;
+ }
+ cout << endl;
+
+ // spawn an opencl actor
+ // 1st arg: sourec code of one or more kernels
+ // 2nd arg: name of the kernel to use
+ // 3rd arg: global dimension arguments for opencl's enqueue
+ // creates matrix_size * matrix_size global work items
+ auto worker =
+ spawn_cl<vector<float>(vector<float>&,vector<float>&)>(kernel_source,
+ kernel_name,
+ {matrix_size, matrix_size});
+ // send both matrices to the actor and wait for a result
+ sync_send(worker, move(m1), move(m2)).then(
+ [](const vector<float>& result) {
+ cout << "result:" << endl;
+ for (size_t column = 0; column < matrix_size; ++column) {
+ for (size_t row = 0; row < matrix_size; ++row) {
+ cout << fixed << setprecision(2) << setw(9)
+ << result[row+column*matrix_size];
+ }
+ cout << endl;
+ }
+ }
+
+ );
+}
+
+int main() {
+ announce<vector<float>>();
+ spawn(multiplier);
+ await_all_others_done();
+ shutdown();
+ return 0;
+}
View
2  src/response_handle.cpp
@@ -42,7 +42,7 @@ response_handle::response_handle(const actor_ptr& from,
const actor_ptr& to,
const message_id& id)
: m_from(from), m_to(to), m_id(id) {
- CPPA_REQUIRE(id.is_response());
+ CPPA_REQUIRE(id.is_response() || !id.valid());
}
bool response_handle::valid() const {

No commit comments for this range

Something went wrong with that request. Please try again.