Skip to content
This repository has been archived by the owner on Dec 22, 2022. It is now read-only.

Commit

Permalink
Merge pull request #18 from bkj/bkj/sssp_cpu
Browse files Browse the repository at this point in the history
Add SSSP CPU implementation
  • Loading branch information
neoblizz authored Jan 14, 2021
2 parents bff2fff + defcbb0 commit 6457a35
Show file tree
Hide file tree
Showing 4 changed files with 126 additions and 19 deletions.
16 changes: 8 additions & 8 deletions CMakeLists.txt
Original file line number Diff line number Diff line change
Expand Up @@ -139,11 +139,11 @@ target_compile_options(essentials INTERFACE
####################################################
############ BUILD EXAMPLE APPLICATIONS ############
####################################################
# option(ESSENTIALS_BUILD_EXAMPLES
# "If on, builds the example graph applications."
# OFF)

# # Subdirectories for examples, testing and documentation
# if(ESSENTIALS_BUILD_EXAMPLES)
# add_subdirectory(examples)
# endif(ESSENTIALS_BUILD_EXAMPLES)
option(ESSENTIALS_BUILD_EXAMPLES
"If on, builds the example graph applications."
ON)

# Subdirectories for examples, testing and documentation
if(ESSENTIALS_BUILD_EXAMPLES)
add_subdirectory(examples)
endif(ESSENTIALS_BUILD_EXAMPLES)
4 changes: 2 additions & 2 deletions examples/CMakeLists.txt
Original file line number Diff line number Diff line change
@@ -1,5 +1,5 @@
# begin /* Add premitives' subdirectories */
# begin /* Add examples' subdirectories */
add_subdirectory(sssp)
# add_subdirectory(bfs)
# add_subdirectory(color)
# end /* Add premitives' subdirectories */
# end /* Add examples' subdirectories */
39 changes: 30 additions & 9 deletions examples/sssp/sssp.cu
Original file line number Diff line number Diff line change
@@ -1,6 +1,6 @@
#include <cstdlib> // EXIT_SUCCESS

#include <gunrock/applications/sssp.hxx>
#include "sssp_cpu.hxx" // Reference implementation

using namespace gunrock;
using namespace memory;
Expand All @@ -24,7 +24,9 @@ void test_sssp(int num_arguments, char** argument_array) {
std::string filename = argument_array[1];

io::matrix_market_t<vertex_t, edge_t, weight_t> mm;
format::csr_t<memory::memory_space_t::device, vertex_t, edge_t, weight_t> csr;

using csr_t = format::csr_t<memory::memory_space_t::device, vertex_t, edge_t, weight_t>;
csr_t csr;
csr.from_coo(mm.load(filename));

// --
Expand All @@ -49,19 +51,38 @@ void test_sssp(int num_arguments, char** argument_array) {
thrust::device_vector<vertex_t> predecessors(n_vertices);

// --
// Run problem
// GPU Run

float elapsed = gunrock::sssp::run(G, single_source, distances.data().get(),
float gpu_elapsed = gunrock::sssp::run(G, single_source, distances.data().get(),
predecessors.data().get());

// --
// Log

std::cout << "Distances (output) = ";
// CPU Run

thrust::host_vector<weight_t> h_distances(n_vertices);
thrust::host_vector<vertex_t> h_predecessors(n_vertices);

float cpu_elapsed = sssp_cpu::run<csr_t, vertex_t, edge_t, weight_t>(
csr, single_source, h_distances.data(), h_predecessors.data());

int n_errors = sssp_cpu::compute_error(distances, h_distances);

// --
// Log + Validate

std::cout << "GPU Distances (output) = ";
thrust::copy(distances.begin(), distances.end(),
std::ostream_iterator<weight_t>(std::cout, " "));
std::cout << std::endl;
std::cout << "SSSP Elapsed Time: " << elapsed << " (ms)" << std::endl;

std::cout << "CPU Distances (output) = ";
thrust::copy(h_distances.begin(), h_distances.end(),
std::ostream_iterator<weight_t>(std::cout, " "));
std::cout << std::endl;

std::cout << "GPU Elapsed Time : " << gpu_elapsed << " (ms)" << std::endl;
std::cout << "CPU Elapsed Time : " << cpu_elapsed << " (ms)" << std::endl;
std::cout << "Number of errors : " << n_errors << std::endl;
}

int main(int argc, char** argv) {
Expand Down
86 changes: 86 additions & 0 deletions examples/sssp/sssp_cpu.hxx
Original file line number Diff line number Diff line change
@@ -0,0 +1,86 @@
#ifndef __SSSP_CPU_H__
#define __SSSP_CPU_H__

#include <chrono>
#include <vector>
#include <queue>

namespace sssp_cpu {

using namespace std;
using namespace std::chrono;

template <typename vertex_t, typename weight_t>
class prioritize {
public:
bool operator()(pair<vertex_t, weight_t> &p1, pair<vertex_t, weight_t> &p2) {
return p1.second > p2.second;
}
};

template <typename csr_t, typename vertex_t, typename edge_t, typename weight_t>
float run(
csr_t& csr,
vertex_t& single_source,
weight_t* distances,
vertex_t* predecessors
) {

thrust::host_vector<edge_t> row_offsets(csr.row_offsets); // Copy data to CPU
thrust::host_vector<vertex_t> column_indices(csr.column_indices);
thrust::host_vector<weight_t> nonzero_values(csr.nonzero_values);

for(vertex_t i = 0; i < csr.number_of_rows; i++)
distances[i] = std::numeric_limits<weight_t>::max();

auto t_start = high_resolution_clock::now();

distances[single_source] = 0;

priority_queue<pair<vertex_t,weight_t>, std::vector<pair<vertex_t,weight_t>>, prioritize<vertex_t, weight_t>> pq;
pq.push(make_pair(single_source, 0.0));

while(!pq.empty()) {
pair<vertex_t, weight_t> curr = pq.top();
pq.pop();

vertex_t curr_node = curr.first;
weight_t curr_dist = curr.second;

vertex_t start = row_offsets[curr_node];
vertex_t end = row_offsets[curr_node + 1];

for(vertex_t offset = start; offset < end; offset++) {
vertex_t neib = column_indices[offset];
weight_t new_dist = curr_dist + nonzero_values[offset];
if(new_dist < distances[neib]) {
distances[neib] = new_dist;
pq.push(make_pair(neib, new_dist));
}
}
}

auto t_stop = high_resolution_clock::now();
auto elapsed = duration_cast<microseconds>(t_stop - t_start).count();
return (float)elapsed / 1000;
}

template <typename val_t>
int compute_error(
thrust::device_vector<val_t> _gpu_result,
thrust::host_vector<val_t> cpu_result
) {
thrust::host_vector<val_t> gpu_result(_gpu_result);

int n_errors = 0;
for(int i = 0 ; i < cpu_result.size(); i++) {
if(gpu_result[i] != cpu_result[i]) {
n_errors++;
}
}
return n_errors;
}

}

#endif

0 comments on commit 6457a35

Please sign in to comment.