Skip to content

Commit

Permalink
add basic example
Browse files Browse the repository at this point in the history
  • Loading branch information
neilkichler committed Feb 26, 2025
1 parent c831eaf commit 7dbdb84
Show file tree
Hide file tree
Showing 3 changed files with 67 additions and 0 deletions.
1 change: 1 addition & 0 deletions examples/CMakeLists.txt
Original file line number Diff line number Diff line change
Expand Up @@ -17,3 +17,4 @@ target_compile_features(example_interval_procedures INTERFACE cxx_std_17 cuda_st
target_include_directories(example_interval_procedures INTERFACE "${cuinterval_SOURCE_DIR}/examples")
target_link_libraries(example_interval_procedures INTERFACE cuinterval)

add_subdirectory(basic)
4 changes: 4 additions & 0 deletions examples/basic/CMakeLists.txt
Original file line number Diff line number Diff line change
@@ -0,0 +1,4 @@
add_executable(basic basic.cu)
target_compile_features(basic PRIVATE cxx_std_20 cuda_std_20)
set_target_properties(basic PROPERTIES CUDA_ARCHITECTURES native)
target_link_libraries(basic PRIVATE ${PROJECT_NAME})
62 changes: 62 additions & 0 deletions examples/basic/basic.cu
Original file line number Diff line number Diff line change
@@ -0,0 +1,62 @@
#include <cuinterval/cuinterval.h>

#include <cuda_runtime.h>

#define CUDA_CHECK(x) \
do { \
cudaError_t err = x; \
if (err != cudaSuccess) { \
fprintf(stderr, "CUDA error in %s at %s:%d: %s (%s=%d)\n", __FUNCTION__, \
__FILE__, __LINE__, cudaGetErrorString(err), \
cudaGetErrorName(err), err); \
abort(); \
} \
} while (0)

__device__ auto f(auto x, auto y)
{
return pow(x - 1.0, 3) - sqr(x) + 4.0;
}

__global__ void kernel(auto *xs, auto *ys, auto *res, std::integral auto n)
{
int i = threadIdx.x + blockIdx.x * blockDim.x;
if (i < n) {
res[i] = f(xs[i], ys[i]);
}
}

int main()
{
constexpr int n = 256;
using T = cu::interval<double>;
T xs[n], ys[n], res[n];

// generate dummy data
for (int i = 0; i < n; i++) {
double v = i;
xs[i] = { { .lb = 0.0, .ub = v } };
ys[i] = { 0.0, v };
}

T *d_xs, *d_ys, *d_res;
CUDA_CHECK(cudaMalloc(&d_xs, n * sizeof(*xs)));
CUDA_CHECK(cudaMalloc(&d_ys, n * sizeof(*ys)));
CUDA_CHECK(cudaMalloc(&d_res, n * sizeof(*res)));

CUDA_CHECK(cudaMemcpy(d_xs, xs, n * sizeof(*xs), cudaMemcpyHostToDevice));
CUDA_CHECK(cudaMemcpy(d_ys, ys, n * sizeof(*ys), cudaMemcpyHostToDevice));

kernel<<<n, 1>>>(d_xs, d_ys, d_res, n);

CUDA_CHECK(cudaMemcpy(res, d_res, n * sizeof(*res), cudaMemcpyDeviceToHost));

auto r = res[3];
printf("f([0,3], [0,3]) = [%g, %g]\n", r.lb, r.ub);

CUDA_CHECK(cudaFree(d_xs));
CUDA_CHECK(cudaFree(d_ys));
CUDA_CHECK(cudaFree(d_res));

return 0;
}

0 comments on commit 7dbdb84

Please sign in to comment.