Skip to content

Commit

Permalink
Clean-up
Browse files Browse the repository at this point in the history
  • Loading branch information
neoblizz committed Oct 20, 2020
1 parent 0bd0118 commit 1493056
Show file tree
Hide file tree
Showing 3 changed files with 33 additions and 14 deletions.
6 changes: 4 additions & 2 deletions gunrock/applications/sssp/sssp_implementation.hxx
Original file line number Diff line number Diff line change
Expand Up @@ -150,8 +150,10 @@ struct sssp_enactor_t : enactor_t<algorithm_problem_t> {
};

// Execute advance operator on the provided lambda
operators::advance::execute<operators::advance_type_t::vertex_to_vertex>(
G, enactor_type::get_enactor(), shortest_path);
operators::advance::execute<operators::advance_type_t::vertex_to_vertex,
operators::advance_direction_t::forward,
operators::load_balance_t::merge_path>(
G, enactor_type::get_enactor(), shortest_path, context);

// Execute filter operator on the provided lambda
operators::filter::execute<operators::filter_type_t::predicated>(
Expand Down
4 changes: 2 additions & 2 deletions gunrock/framework/enactor.hxx
Original file line number Diff line number Diff line change
Expand Up @@ -47,7 +47,7 @@ struct enactor_t {
// or we can move this within the actual context.
algorithm_problem_t* problem;
thrust::host_vector<frontier_type> frontiers;
thrust::device_vector<vertex_t> scanned_row_offsets;
thrust::device_vector<vertex_t> scanned_work_domain;
frontier_type* active_frontier;
frontier_type* inactive_frontier;
int buffer_selector;
Expand All @@ -67,7 +67,7 @@ struct enactor_t {
inactive_frontier(&frontiers[1]),
buffer_selector(0),
iteration(0),
scanned_row_offsets(
scanned_work_domain(
problem->get_host_graph_pointer()->get_number_of_vertices()) {
// Set temporary buffer to be at least the number of edges
auto g = problem->get_host_graph_pointer();
Expand Down
37 changes: 27 additions & 10 deletions gunrock/memory.hxx
Original file line number Diff line number Diff line change
Expand Up @@ -4,7 +4,6 @@
#include <memory>

#include <thrust/device_ptr.h>

#include <gunrock/error.hxx>

namespace gunrock {
Expand All @@ -21,7 +20,7 @@ namespace memory {
* for this.
*
*/
enum memory_space_t { device, host, unknown };
enum memory_space_t { device, host };

/**
* @brief allocate a pointer with size on a specfied memory space.
Expand All @@ -34,8 +33,6 @@ enum memory_space_t { device, host, unknown };
template <typename type_t>
inline type_t* allocate(std::size_t size, memory_space_t space) {
void* pointer = nullptr;
if (space == unknown)
throw error::exception_t(cudaErrorUnknown);
if (size) {
error::error_t status = (device == space) ? cudaMalloc(&pointer, size)
: cudaMallocHost(&pointer, size);
Expand All @@ -54,22 +51,42 @@ inline type_t* allocate(std::size_t size, memory_space_t space) {
*/
template <typename type_t>
inline void free(type_t* pointer, memory_space_t space) {
if (space == unknown)
throw error::exception_t(cudaErrorUnknown);
if (pointer) {
error::error_t status =
(device == space)
? cudaFree((void*)pointer)
: cudaFreeHost((void*)pointer); // XXX: reinterpret_cast?
error::error_t status = (device == space) ? cudaFree((void*)pointer)
: cudaFreeHost((void*)pointer);
error::throw_if_exception(status);
}
}

/**
* @brief Wrapper around thrust::raw_pointer_cast() to accept .data() or raw
* pointer and return a raw pointer. Useful when we would like to return a raw
* pointer of either a thrust device vector or a host vector. Because thrust
* device vector's raw pointer is accessed by `.data().get()`, whereas thrust
* host vector's raw pointer is simply `data()`. So, when calling these
* functions on `.data()`, it can cast either a host or device vector.
*
* @tparam type_t
* @param pointer
* @return type_t*
*/
template <typename type_t>
inline type_t* raw_pointer_cast(thrust::device_ptr<type_t> pointer) {
return thrust::raw_pointer_cast(pointer);
}

/**
* @brief Wrapper around thrust::raw_pointer_cast() to accept .data() or raw
* pointer and return a raw pointer. Useful when we would like to return a raw
* pointer of either a thrust device vector or a host vector. Because thrust
* device vector's raw pointer is accessed by `.data().get()`, whereas thrust
* host vector's raw pointer is simply `data()`. So, when calling these
* functions on `.data()`, it can cast either a host or device vector.
*
* @tparam type_t
* @param pointer
* @return type_t*
*/
template <typename type_t>
__host__ __device__ inline type_t* raw_pointer_cast(type_t* pointer) {
return thrust::raw_pointer_cast(pointer);
Expand Down

0 comments on commit 1493056

Please sign in to comment.