Adding float buffer, Better Erlang API

c_src/pcuda_buffer.h

@@ -8,7 +8,8 @@
 
 enum PCudaBufferTypes {
     BUF_TYPE_INTEGER,
-    BUF_TYPE_STRING
+    BUF_TYPE_STRING,
+    BUF_TYPE_FLOAT
 };
 
 class PCudaBuffer {
@@ -49,6 +50,27 @@ class PCudaIntBuffer : public PCudaBuffer {
     std::vector<long> *data;
 };
 
+class PCudaFloatBuffer : public PCudaBuffer {
+public:
+    PCudaFloatBuffer();
+    virtual ~PCudaFloatBuffer();
+    virtual unsigned int size();
+    virtual PCudaBufferTypes type() { return BUF_TYPE_FLOAT; };
+    virtual bool sort();
+    virtual bool contains(ErlNifEnv *env, ERL_NIF_TERM rawTarget);
+    virtual ERL_NIF_TERM toErlTerms(ErlNifEnv *env);
+    virtual void write(ErlNifEnv *env, ERL_NIF_TERM data);
+    virtual void delete_at(unsigned long position);
+    virtual bool insert_at(unsigned long position, ErlNifEnv *env, ERL_NIF_TERM value);
+    virtual void clear();
+    virtual bool copy(PCudaBuffer *src);
+    virtual ERL_NIF_TERM intersect(ErlNifEnv *env, PCudaBuffer *other);
+    virtual ERL_NIF_TERM minmax(ErlNifEnv *env);
+
+protected:
+    std::vector<double> *data;
+};
+
 class PCudaStringBuffer : public PCudaBuffer {
 public:
     PCudaStringBuffer();

c_src/pcuda_float_buffer.cpp

@@ -0,0 +1,111 @@
+#include <stdio.h>
+#include "pcuda_buffer.h"
+#include "pcuda_ops.h"
+
+PCudaFloatBuffer::PCudaFloatBuffer() {
+    this->data = new std::vector<double>();
+}
+
+PCudaFloatBuffer::~PCudaFloatBuffer() {
+    delete this->data;
+}
+
+unsigned int PCudaFloatBuffer::size() {
+    return this->data->size();
+}
+
+void PCudaFloatBuffer::write(ErlNifEnv *env, ERL_NIF_TERM data) {
+    ERL_NIF_TERM head;
+    double value;
+
+    while (enif_get_list_cell(env, data, &head, &data)) {
+        if (enif_get_double(env, head, &value)) {
+            this->data->push_back(value);
+        }
+    }
+}
+
+void PCudaFloatBuffer::delete_at(unsigned long position) {
+    std::vector<double>::iterator iter = this->data->begin();
+    for (unsigned long i = 0; i < position; i++) {
+        iter++;
+    }
+    this->data->erase(iter);
+}
+
+bool PCudaFloatBuffer::insert_at(unsigned long position, ErlNifEnv *env, ERL_NIF_TERM rawValue) {
+    double value;
+    if (enif_get_double(env, rawValue, &value)) {
+        std::vector<double>::iterator iter = this->data->begin();
+        for (unsigned long i = 0; i < position; i++) {
+            iter++;
+        }
+        this->data->insert(iter, 1, value);
+        return true;
+    }
+    return false;
+}
+
+bool PCudaFloatBuffer::sort() {
+    return pcuda_float_sort(this->data);
+}
+
+bool PCudaFloatBuffer::contains(ErlNifEnv *env, ERL_NIF_TERM rawTarget) {
+    double target;
+    if (enif_get_double(env, rawTarget, &target)) {
+        return pcuda_float_binary_search(this->data, target);
+    }
+    else {
+        return false;
+    }
+}
+
+ERL_NIF_TERM PCudaFloatBuffer::toErlTerms(ErlNifEnv *env) {
+    std::vector<double>::iterator iter;
+    ERL_NIF_TERM retval = enif_make_list(env, 0);
+    if (this->data->size() > 0) {
+        for (iter = this->data->end(); iter != this->data->begin();) {
+            --iter;
+            retval = enif_make_list_cell(env, enif_make_double(env, *iter), retval);
+        }
+    }
+    return retval;
+}
+
+void PCudaFloatBuffer::clear() {
+    this->data->clear();
+}
+
+bool PCudaFloatBuffer::copy(PCudaBuffer *src) {
+    if (src->type() == BUF_TYPE_FLOAT) {
+        PCudaFloatBuffer *source = (PCudaFloatBuffer *) src;
+        std::vector<double>::iterator iter;
+        for (iter = source->data->begin(); iter != source->data->end(); ++iter) {
+            this->data->push_back(*iter);
+        }
+        return true;
+    }
+    return false;
+}
+
+ERL_NIF_TERM PCudaFloatBuffer::intersect(ErlNifEnv *env, PCudaBuffer *otherBuffer) {
+    ERL_NIF_TERM retval = enif_make_list(env, 0);
+    std::vector<double> intersection;
+    if (otherBuffer->type() == BUF_TYPE_FLOAT) {
+        PCudaFloatBuffer *other = (PCudaFloatBuffer *) otherBuffer;
+        pcuda_float_intersection(this->data, other->data, &intersection);
+        if (intersection.size() > 0) {
+            for (std::vector<double>::iterator iter = intersection.end(); iter != intersection.begin();) {
+                --iter;
+                retval = enif_make_list_cell(env, enif_make_double(env, *iter), retval);
+            }
+        }
+    }
+    return retval;
+}
+
+ERL_NIF_TERM PCudaFloatBuffer::minmax(ErlNifEnv *env) {
+    double minmax[2];
+    pcuda_float_minmax(this->data, &minmax[0]);
+    return enif_make_tuple2(env, enif_make_long(env, minmax[0]), enif_make_long(env, minmax[1]));
+}

c_src/pcuda_ops.cu

@@ -72,6 +72,13 @@ bool pcuda_integer_sort(std::vector<long> *data) {
     return true;
 }
 
+bool pcuda_float_sort(std::vector<double> *data) {
+    thrust::device_vector<double> device = *data;
+    thrust::sort(device.begin(), device.end());
+    thrust::copy(device.begin(), device.end(), data->begin());
+    return true;
+}
+
 bool pcuda_string_sort(std::vector<std::string> *data) {
     printf("In pcuda_string_sort\n");
     thrust::device_vector<PCudaString> device;
@@ -104,15 +111,33 @@ bool pcuda_integer_binary_search(std::vector<long> *data, long target) {
     return thrust::binary_search(device.begin(), device.end(), target, thrust::less<long>());
 }
 
+bool pcuda_float_binary_search(std::vector<double> *data, double target) {
+    thrust::device_vector<double> device = *data;
+    return thrust::binary_search(device.begin(), device.end(), target, thrust::less<double>());
+}
+
 void pcuda_integer_intersection(std::vector<long> *first, std::vector<long> *second,
                                 std::vector<long> *intersection) {
     thrust::set_intersection(first->begin(), first->end(),
                              second->begin(), second->end(), std::back_inserter(*intersection));
 }
 
+void pcuda_float_intersection(std::vector<double> *first, std::vector<double> *second,
+                                std::vector<double> *intersection) {
+    thrust::set_intersection(first->begin(), first->end(),
+                             second->begin(), second->end(), std::back_inserter(*intersection));
+}
+
 void pcuda_integer_minmax(std::vector<long> *data, long *minmax) {
     thrust::pair<std::vector<long>::iterator,
                  std::vector<long>::iterator> result = thrust::minmax_element(data->begin(), data->end());
     minmax[0] = *result.first;
     minmax[1] = *result.second;
 }
+
+void pcuda_float_minmax(std::vector<double> *data, double *minmax) {
+    thrust::pair<std::vector<double>::iterator,
+                 std::vector<double>::iterator> result = thrust::minmax_element(data->begin(), data->end());
+    minmax[0] = *result.first;
+    minmax[1] = *result.second;
+}

c_src/pcuda_ops.h

@@ -9,5 +9,11 @@ bool pcuda_integer_binary_search(std::vector<long> *data, long target);
 void pcuda_integer_intersection(std::vector<long> *first, std::vector<long> *second, std::vector<long> *intersection);
 void pcuda_integer_minmax(std::vector<long> *data, long *minmax);
 
+bool pcuda_float_sort(std::vector<double> *data);
+bool pcuda_float_binary_search(std::vector<double> *data, double target);
+void pcuda_float_intersection(std::vector<double> *first, std::vector<double> *second, std::vector<double> *intersection);
+void pcuda_float_minmax(std::vector<double> *data, double *minmax);
+
+// Work in progress
 bool pcuda_string_sort(std::vector<std::string> *data);
 #endif

c_src/pteracuda_nifs.cpp

@@ -35,6 +35,8 @@ extern "C" {
 
     ERL_NIF_TERM pteracuda_nifs_new_int_buffer(ErlNifEnv *env, int argc, const ERL_NIF_TERM argv[]);
     ERL_NIF_TERM pteracuda_nifs_new_string_buffer(ErlNifEnv *env, int argc, const ERL_NIF_TERM argv[]);
+    ERL_NIF_TERM pteracuda_nifs_new_float_buffer(ErlNifEnv *env, int argc, const ERL_NIF_TERM argv[]);
+
     ERL_NIF_TERM pteracuda_nifs_destroy_buffer(ErlNifEnv *env, int argc, const ERL_NIF_TERM argv[]);
     ERL_NIF_TERM pteracuda_nifs_buffer_size(ErlNifEnv *env, int argc, const ERL_NIF_TERM argv[]);
 
@@ -55,6 +57,7 @@ extern "C" {
         {"destroy_context", 1, pteracuda_nifs_destroy_context},
         {"new_int_buffer", 0, pteracuda_nifs_new_int_buffer},
         {"new_string_buffer", 0, pteracuda_nifs_new_string_buffer},
+        {"new_float_buffer", 0, pteracuda_nifs_new_float_buffer},
         {"destroy_buffer", 1, pteracuda_nifs_destroy_buffer},
         {"buffer_size", 1, pteracuda_nifs_buffer_size},
         {"write_buffer", 2, pteracuda_nifs_write_buffer},
@@ -170,6 +173,18 @@ ERL_NIF_TERM pteracuda_nifs_new_string_buffer(ErlNifEnv *env, int argc, const ER
     return enif_make_tuple2(env, ATOM_OK, res);
 }
 
+ERL_NIF_TERM pteracuda_nifs_new_float_buffer(ErlNifEnv *env, int argc, const ERL_NIF_TERM argv[]) {
+    PCudaBufferRef *ref = (PCudaBufferRef *) enif_alloc_resource(pteracuda_buffer_resource, sizeof(PCudaBufferRef));
+    if (!ref) {
+        return OOM_ERROR;
+    }
+    ref->buffer = new PCudaFloatBuffer();
+    ref->destroyed = false;
+    ERL_NIF_TERM res = enif_make_resource(env, ref);
+    enif_release_resource(ref);
+    return enif_make_tuple2(env, ATOM_OK, res);
+}
+
 ERL_NIF_TERM pteracuda_nifs_destroy_buffer(ErlNifEnv *env, int argc, const ERL_NIF_TERM argv[]) {
     PCudaBufferRef *ref;
     if (argc != 1 || !enif_get_resource(env, argv[0], pteracuda_buffer_resource, (void **) &ref)) {

src/pteracuda_buffer.erl

@@ -0,0 +1,62 @@
+-module(pteracuda_buffer).
+
+-include("pteracuda_internals.hrl").
+
+-export([new/1,
+         destroy/1,
+         size/1,
+         write/2,
+         read/1,
+         duplicate/1,
+         clear/1,
+         sort/2,
+         contains/3,
+         intersection/3,
+         minmax/2]).
+
+new(integer) ->
+    {ok, Buf} = pteracuda_nifs:new_int_buffer(),
+    {ok, #pc_buffer{type=integer, ref=Buf}};
+new(float) ->
+    {ok, Buf} = pteracuda_nifs:new_float_buffer(),
+    {ok, #pc_buffer{type=float, ref=Buf}};
+new(string) ->
+    {ok, Buf} = pteracuda_nifs:new_string_buffer(),
+    {ok, #pc_buffer{type=string, ref=Buf}}.
+
+destroy(#pc_buffer{ref=Ref}) ->
+    pteracuda_nifs:destroy_buffer(Ref),
+    ok.
+
+size(#pc_buffer{ref=Ref}) ->
+    pteracuda_nifs:buffer_size(Ref).
+
+write(#pc_buffer{ref=Ref, type=Type}, Data) when Type =:= integer orelse
+                                                 Type =:= string orelse
+                                                 Type =:= float ->
+    pteracuda_nifs:write_buffer(Ref, Data).
+
+read(#pc_buffer{ref=Ref}) ->
+    pteracuda_nifs:read_buffer(Ref).
+
+duplicate(#pc_buffer{ref=Ref, type=Type}) when Type =:= integer orelse
+                                               Type =:= string orelse
+                                               Type =:= float ->
+    {ok, OtherBuf} = new(Type),
+    pteracuda_nifs:copy_buffer(Ref, OtherBuf#pc_buffer.ref),
+    {ok, OtherBuf}.
+
+clear(#pc_buffer{ref=Ref}) ->
+    pteracuda_nifs:clear_buffer(Ref).
+
+sort(#pc_context{ref=Ctx}, #pc_buffer{ref=Buf}) ->
+    pteracuda_nifs:sort_buffer(Ctx, Buf).
+
+contains(#pc_context{ref=Ctx}, #pc_buffer{ref=Buf}, Value) ->
+    pteracuda_nifs:buffer_contains(Ctx, Buf, Value).
+
+intersection(#pc_context{ref=Ctx}, #pc_buffer{ref=Buf1}, #pc_buffer{ref=Buf2}) ->
+    pteracuda_nifs:buffer_intersection(Ctx, Buf1, Buf2).
+
+minmax(#pc_context{ref=Ctx}, #pc_buffer{ref=Buf}) ->
+    pteracuda_nifs:buffer_minmax(Ctx, Buf).

src/pteracuda_context.erl

@@ -0,0 +1,18 @@
+-module(pteracuda_context).
+
+-include("pteracuda_internals.hrl").
+
+-export([new/0,
+         new/1,
+         destroy/1]).
+
+new() ->
+    {ok, Ctx} = pteracuda_nifs:new_context(),
+    {ok, #pc_context{ref=Ctx}}.
+
+new(Device) when is_integer(Device) ->
+    {ok, Ctx} = pteracuda_nifs:new_context(Device),
+    {ok, #pc_context{ref=Ctx}}.
+
+destroy(#pc_context{ref=Ctx}) ->
+    pteracuda_nifs:destroy_context(Ctx).

src/pteracuda_demo.erl

@@ -0,0 +1,24 @@
+-module(pteracuda_demo).
+
+-compile([export_all,
+          native]).
+
+start(N) ->
+    {T1, T2, T3} = erlang:now(),
+    random:seed(T1, T2, T3),
+    io:format("Generating test data: ~p~n", [N]),
+    D = [random:uniform(N) || _ <- lists:seq(1, N)],
+    io:format("Measuring performance "),
+    {Time1, _} = timer:tc(lists, sort, [D]),
+    io:format("."),
+    {ok, C} = pteracuda_context:new(),
+    {ok, B} = pteracuda_buffer:new(integer),
+    pteracuda_buffer:write(B, D),
+    {Time2, _} = timer:tc(pteracuda_demo, pteracuda_sort, [C, B, D]),
+    io:format(".~n"),
+    io:format("Erlang: ~pms, CUDA: ~pms~n", [Time1 / 1000, Time2 / 1000]).
+
+pteracuda_sort(C, B, D) ->
+    pteracuda_buffer:write(B, D),
+    pteracuda_buffer:sort(C, B),
+    pteracuda_buffer:read(B).

src/pteracuda_internals.hrl

@@ -0,0 +1,4 @@
+-record(pc_buffer, {type,
+                    ref}).
+
+-record(pc_context, {ref}).