From 14930567dbc3ea155c3eb63d9023752c9ddbabba Mon Sep 17 00:00:00 2001 From: neoblizz Date: Tue, 20 Oct 2020 12:09:27 -0700 Subject: [PATCH] Clean-up --- .../applications/sssp/sssp_implementation.hxx | 6 ++- gunrock/framework/enactor.hxx | 4 +- gunrock/memory.hxx | 37 ++++++++++++++----- 3 files changed, 33 insertions(+), 14 deletions(-) diff --git a/gunrock/applications/sssp/sssp_implementation.hxx b/gunrock/applications/sssp/sssp_implementation.hxx index 48ce4e98..731b7bf6 100644 --- a/gunrock/applications/sssp/sssp_implementation.hxx +++ b/gunrock/applications/sssp/sssp_implementation.hxx @@ -150,8 +150,10 @@ struct sssp_enactor_t : enactor_t { }; // Execute advance operator on the provided lambda - operators::advance::execute( - G, enactor_type::get_enactor(), shortest_path); + operators::advance::execute( + G, enactor_type::get_enactor(), shortest_path, context); // Execute filter operator on the provided lambda operators::filter::execute( diff --git a/gunrock/framework/enactor.hxx b/gunrock/framework/enactor.hxx index 4fceb9fc..6799e9df 100644 --- a/gunrock/framework/enactor.hxx +++ b/gunrock/framework/enactor.hxx @@ -47,7 +47,7 @@ struct enactor_t { // or we can move this within the actual context. algorithm_problem_t* problem; thrust::host_vector frontiers; - thrust::device_vector scanned_row_offsets; + thrust::device_vector scanned_work_domain; frontier_type* active_frontier; frontier_type* inactive_frontier; int buffer_selector; @@ -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(); diff --git a/gunrock/memory.hxx b/gunrock/memory.hxx index 384b2646..6029f8d9 100644 --- a/gunrock/memory.hxx +++ b/gunrock/memory.hxx @@ -4,7 +4,6 @@ #include #include - #include namespace gunrock { @@ -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. @@ -34,8 +33,6 @@ enum memory_space_t { device, host, unknown }; template 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); @@ -54,22 +51,42 @@ inline type_t* allocate(std::size_t size, memory_space_t space) { */ template 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 inline type_t* raw_pointer_cast(thrust::device_ptr 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 __host__ __device__ inline type_t* raw_pointer_cast(type_t* pointer) { return thrust::raw_pointer_cast(pointer);