Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

GPU Batch 4 #5

Merged
merged 39 commits into from
Jul 30, 2024
Merged
Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
39 commits
Select commit Hold shift + click to select a range
ce331c9
Add CUDA compatible alias to thrust::pair or std::pair
mborland Jul 26, 2024
f323141
Fix arcsine dist range and support with thrust
mborland Jul 26, 2024
eb7ca3b
Add CUDA testing of range and support
mborland Jul 26, 2024
1bf0c94
Fix bernoulli dist use of std::pair for range and support
mborland Jul 26, 2024
815950d
Add CUDA testing of range and support on bernoulli dist
mborland Jul 26, 2024
31c0db0
Fix more uses of std::pair in binomial and cauchy distributions
mborland Jul 26, 2024
e48aec5
Fix usages of std::numeric_limits and add testing
mborland Jul 26, 2024
f5e9a31
Add GPU support to the exponential dist
mborland Jul 26, 2024
0cf32de
Add exponential dist CUDA testing
mborland Jul 26, 2024
f7a82bb
Adjust test tools definition of throw in GPU environment
mborland Jul 26, 2024
a054903
Add SYCL testing of exponential dist
mborland Jul 26, 2024
21649b9
Fix use of std::numeric_limits
mborland Jul 26, 2024
f381d62
Add GPU support to landau dist
mborland Jul 29, 2024
7855e66
Add landau dist CUDA tests
mborland Jul 29, 2024
899399b
Remove recursion from quantile dispatch function
mborland Jul 29, 2024
8365145
Add SYCL test for Landau distribution
mborland Jul 29, 2024
5e6326b
Add GPU support to mapairy dist
mborland Jul 29, 2024
f3bb34c
Add CUDA testing to mapairy dist
mborland Jul 29, 2024
45338a2
Add sycl testing and adjust a few tolerances
mborland Jul 29, 2024
2f7e833
Add GPU support to holtsmark dist
mborland Jul 29, 2024
9aa19fa
Fix static local variable
mborland Jul 29, 2024
70f3332
Add CUDA testing to holtsmark dist
mborland Jul 29, 2024
3428e53
Add SYCL testing for holtsmark dist
mborland Jul 29, 2024
0ec1d5f
Add GPU support to saspoint5 dist
mborland Jul 29, 2024
32af9cd
Add CUDA testing to saspoint5 dist
mborland Jul 29, 2024
322a2f4
Eliminate recursion in quantile
mborland Jul 29, 2024
ae553a1
Add saspoint5 CUDA testing
mborland Jul 29, 2024
a520aa4
Replace workaround macro with new numeric_limits
mborland Jul 29, 2024
ecee617
Add GPU support to extreme value dist
mborland Jul 29, 2024
ab9ccdf
Fix GPU support for extreme value dist quantile
mborland Jul 29, 2024
7b80fa3
Add CUDA testing of extreme value dist
mborland Jul 29, 2024
1a9ed1b
Add extreme value dist SYCL testing
mborland Jul 29, 2024
27eb3aa
Add GPU support to laplace dist
mborland Jul 29, 2024
2986468
Add laplace dist CUDA testing
mborland Jul 29, 2024
b3535b2
Add laplace dist SYCL testing
mborland Jul 29, 2024
f289225
Add GPU support to logistic dist
mborland Jul 29, 2024
1348a14
Add logistic dist CUDA testing
mborland Jul 29, 2024
c6edfc1
Add logistic dist SYCL testing
mborland Jul 29, 2024
f1305a3
Fix macro naming
mborland Jul 29, 2024
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
10 changes: 6 additions & 4 deletions include/boost/math/distributions/arcsine.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -2,6 +2,7 @@

// Copyright John Maddock 2014.
// Copyright Paul A. Bristow 2014.
// Copyright Matt Borland 2024.

// Use, modification and distribution are subject to the
// Boost Software License, Version 1.0.
Expand Down Expand Up @@ -31,6 +32,7 @@

#include <cmath>
#include <boost/math/tools/config.hpp>
#include <boost/math/tools/tuple.hpp>
#include <boost/math/distributions/fwd.hpp>
#include <boost/math/distributions/complement.hpp> // complements.
#include <boost/math/distributions/detail/common_error_handling.hpp> // error checks.
Expand Down Expand Up @@ -204,17 +206,17 @@ namespace boost
#endif

template <class RealType, class Policy>
BOOST_MATH_GPU_ENABLED inline const std::pair<RealType, RealType> range(const arcsine_distribution<RealType, Policy>& dist)
BOOST_MATH_GPU_ENABLED inline const boost::math::pair<RealType, RealType> range(const arcsine_distribution<RealType, Policy>& dist)
{ // Range of permissible values for random variable x.
using boost::math::tools::max_value;
return std::pair<RealType, RealType>(static_cast<RealType>(dist.x_min()), static_cast<RealType>(dist.x_max()));
return boost::math::pair<RealType, RealType>(static_cast<RealType>(dist.x_min()), static_cast<RealType>(dist.x_max()));
}

template <class RealType, class Policy>
BOOST_MATH_GPU_ENABLED inline const std::pair<RealType, RealType> support(const arcsine_distribution<RealType, Policy>& dist)
BOOST_MATH_GPU_ENABLED inline const boost::math::pair<RealType, RealType> support(const arcsine_distribution<RealType, Policy>& dist)
{ // Range of supported values for random variable x.
// This is range where cdf rises from 0 to 1, and outside it, the pdf is zero.
return std::pair<RealType, RealType>(static_cast<RealType>(dist.x_min()), static_cast<RealType>(dist.x_max()));
return boost::math::pair<RealType, RealType>(static_cast<RealType>(dist.x_min()), static_cast<RealType>(dist.x_max()));
}

template <class RealType, class Policy>
Expand Down
9 changes: 5 additions & 4 deletions include/boost/math/distributions/bernoulli.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -30,6 +30,7 @@

#include <boost/math/distributions/fwd.hpp>
#include <boost/math/tools/config.hpp>
#include <boost/math/tools/tuple.hpp>
#include <boost/math/distributions/complement.hpp> // complements
#include <boost/math/distributions/detail/common_error_handling.hpp> // error checks
#include <boost/math/special_functions/fpclassify.hpp> // isnan.
Expand Down Expand Up @@ -133,17 +134,17 @@ namespace boost
#endif

template <class RealType, class Policy>
BOOST_MATH_GPU_ENABLED inline const std::pair<RealType, RealType> range(const bernoulli_distribution<RealType, Policy>& /* dist */)
BOOST_MATH_GPU_ENABLED inline const boost::math::pair<RealType, RealType> range(const bernoulli_distribution<RealType, Policy>& /* dist */)
{ // Range of permissible values for random variable k = {0, 1}.
using boost::math::tools::max_value;
return std::pair<RealType, RealType>(static_cast<RealType>(0), static_cast<RealType>(1));
return boost::math::pair<RealType, RealType>(static_cast<RealType>(0), static_cast<RealType>(1));
}

template <class RealType, class Policy>
BOOST_MATH_GPU_ENABLED inline const std::pair<RealType, RealType> support(const bernoulli_distribution<RealType, Policy>& /* dist */)
BOOST_MATH_GPU_ENABLED inline const boost::math::pair<RealType, RealType> support(const bernoulli_distribution<RealType, Policy>& /* dist */)
{ // Range of supported values for random variable k = {0, 1}.
// This is range where cdf rises from 0 to 1, and outside it, the pdf is zero.
return std::pair<RealType, RealType>(static_cast<RealType>(0), static_cast<RealType>(1));
return boost::math::pair<RealType, RealType>(static_cast<RealType>(0), static_cast<RealType>(1));
}

template <class RealType, class Policy>
Expand Down
9 changes: 5 additions & 4 deletions include/boost/math/distributions/binomial.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -80,6 +80,7 @@
#define BOOST_MATH_SPECIAL_BINOMIAL_HPP

#include <boost/math/tools/config.hpp>
#include <boost/math/tools/tuple.hpp>
#include <boost/math/distributions/fwd.hpp>
#include <boost/math/special_functions/beta.hpp> // for incomplete beta.
#include <boost/math/distributions/complement.hpp> // complements
Expand Down Expand Up @@ -420,17 +421,17 @@ namespace boost
#endif

template <class RealType, class Policy>
BOOST_MATH_CUDA_ENABLED const std::pair<RealType, RealType> range(const binomial_distribution<RealType, Policy>& dist)
BOOST_MATH_CUDA_ENABLED const boost::math::pair<RealType, RealType> range(const binomial_distribution<RealType, Policy>& dist)
{ // Range of permissible values for random variable k.
using boost::math::tools::max_value;
return std::pair<RealType, RealType>(static_cast<RealType>(0), dist.trials());
return boost::math::pair<RealType, RealType>(static_cast<RealType>(0), dist.trials());
}

template <class RealType, class Policy>
BOOST_MATH_CUDA_ENABLED const std::pair<RealType, RealType> support(const binomial_distribution<RealType, Policy>& dist)
BOOST_MATH_CUDA_ENABLED const boost::math::pair<RealType, RealType> support(const binomial_distribution<RealType, Policy>& dist)
{ // Range of supported values for random variable k.
// This is range where cdf rises from 0 to 1, and outside it, the pdf is zero.
return std::pair<RealType, RealType>(static_cast<RealType>(0), dist.trials());
return boost::math::pair<RealType, RealType>(static_cast<RealType>(0), dist.trials());
}

template <class RealType, class Policy>
Expand Down
39 changes: 19 additions & 20 deletions include/boost/math/distributions/cauchy.hpp
Original file line number Diff line number Diff line change
@@ -1,5 +1,6 @@
// Copyright John Maddock 2006, 2007.
// Copyright Paul A. Bristow 2007.
// Copyright Matt Borland 2024.

// Use, modification and distribution are subject to the
// Boost Software License, Version 1.0. (See accompanying file
Expand All @@ -14,6 +15,8 @@
#endif

#include <boost/math/tools/config.hpp>
#include <boost/math/tools/tuple.hpp>
#include <boost/math/tools/numeric_limits.hpp>
#include <boost/math/distributions/fwd.hpp>
#include <boost/math/constants/constants.hpp>
#include <boost/math/distributions/complement.hpp>
Expand Down Expand Up @@ -77,11 +80,11 @@ BOOST_MATH_GPU_ENABLED RealType cdf_imp(const cauchy_distribution<RealType, Poli
return static_cast<RealType>((complement) ? 1 : 0);
}
#else
if(std::numeric_limits<RealType>::has_infinity && x == std::numeric_limits<RealType>::infinity())
if(boost::math::numeric_limits<RealType>::has_infinity && x == boost::math::numeric_limits<RealType>::infinity())
{ // cdf +infinity is unity.
return static_cast<RealType>((complement) ? 0 : 1);
}
if(std::numeric_limits<RealType>::has_infinity && x == -std::numeric_limits<RealType>::infinity())
if(boost::math::numeric_limits<RealType>::has_infinity && x == -boost::math::numeric_limits<RealType>::infinity())
{ // cdf -infinity is zero.
return static_cast<RealType>((complement) ? 1 : 0);
}
Expand Down Expand Up @@ -196,35 +199,31 @@ cauchy_distribution(RealType,RealType)->cauchy_distribution<typename boost::math
#endif

template <class RealType, class Policy>
BOOST_MATH_GPU_ENABLED inline const std::pair<RealType, RealType> range(const cauchy_distribution<RealType, Policy>&)
BOOST_MATH_GPU_ENABLED inline const boost::math::pair<RealType, RealType> range(const cauchy_distribution<RealType, Policy>&)
{ // Range of permissible values for random variable x.
#ifndef BOOST_MATH_HAS_GPU_SUPPORT
BOOST_MATH_IF_CONSTEXPR (std::numeric_limits<RealType>::has_infinity)
BOOST_MATH_IF_CONSTEXPR (boost::math::numeric_limits<RealType>::has_infinity)
{
return std::pair<RealType, RealType>(-std::numeric_limits<RealType>::infinity(), std::numeric_limits<RealType>::infinity()); // - to + infinity.
return boost::math::pair<RealType, RealType>(-boost::math::numeric_limits<RealType>::infinity(), boost::math::numeric_limits<RealType>::infinity()); // - to + infinity.
}
else
#endif
{ // Can only use max_value.
using boost::math::tools::max_value;
return std::pair<RealType, RealType>(-max_value<RealType>(), max_value<RealType>()); // - to + max.
return boost::math::pair<RealType, RealType>(-max_value<RealType>(), max_value<RealType>()); // - to + max.
}
}

template <class RealType, class Policy>
BOOST_MATH_GPU_ENABLED inline const std::pair<RealType, RealType> support(const cauchy_distribution<RealType, Policy>& )
BOOST_MATH_GPU_ENABLED inline const boost::math::pair<RealType, RealType> support(const cauchy_distribution<RealType, Policy>& )
{ // Range of supported values for random variable x.
// This is range where cdf rises from 0 to 1, and outside it, the pdf is zero.
#ifndef BOOST_MATH_HAS_GPU_SUPPORT
BOOST_MATH_IF_CONSTEXPR (std::numeric_limits<RealType>::has_infinity)
BOOST_MATH_IF_CONSTEXPR (boost::math::numeric_limits<RealType>::has_infinity)
{
return std::pair<RealType, RealType>(-std::numeric_limits<RealType>::infinity(), std::numeric_limits<RealType>::infinity()); // - to + infinity.
return boost::math::pair<RealType, RealType>(-boost::math::numeric_limits<RealType>::infinity(), boost::math::numeric_limits<RealType>::infinity()); // - to + infinity.
}
else
#endif
{ // Can only use max_value.
using boost::math::tools::max_value;
return std::pair<RealType, RealType>(-tools::max_value<RealType>(), max_value<RealType>()); // - to + max.
return boost::math::pair<RealType, RealType>(-tools::max_value<RealType>(), max_value<RealType>()); // - to + max.
}
}

Expand All @@ -250,7 +249,7 @@ BOOST_MATH_GPU_ENABLED inline RealType pdf(const cauchy_distribution<RealType, P
return 0; // pdf + and - infinity is zero.
}
// These produce MSVC 4127 warnings, so the above used instead.
//if(std::numeric_limits<RealType>::has_infinity && abs(x) == std::numeric_limits<RealType>::infinity())
//if(boost::math::numeric_limits<RealType>::has_infinity && abs(x) == boost::math::numeric_limits<RealType>::infinity())
//{ // pdf + and - infinity is zero.
// return 0;
//}
Expand Down Expand Up @@ -299,7 +298,7 @@ BOOST_MATH_GPU_ENABLED inline RealType mean(const cauchy_distribution<RealType,
"boost::math::mean(cauchy<%1%>&)",
"The Cauchy distribution does not have a mean: "
"the only possible return value is %1%.",
std::numeric_limits<RealType>::quiet_NaN(), Policy());
boost::math::numeric_limits<RealType>::quiet_NaN(), Policy());
}

template <class RealType, class Policy>
Expand All @@ -313,7 +312,7 @@ BOOST_MATH_GPU_ENABLED inline RealType variance(const cauchy_distribution<RealTy
"boost::math::variance(cauchy<%1%>&)",
"The Cauchy distribution does not have a variance: "
"the only possible return value is %1%.",
std::numeric_limits<RealType>::quiet_NaN(), Policy());
boost::math::numeric_limits<RealType>::quiet_NaN(), Policy());
}

template <class RealType, class Policy>
Expand All @@ -339,7 +338,7 @@ BOOST_MATH_GPU_ENABLED inline RealType skewness(const cauchy_distribution<RealTy
"boost::math::skewness(cauchy<%1%>&)",
"The Cauchy distribution does not have a skewness: "
"the only possible return value is %1%.",
std::numeric_limits<RealType>::quiet_NaN(), Policy()); // infinity?
boost::math::numeric_limits<RealType>::quiet_NaN(), Policy()); // infinity?
}

template <class RealType, class Policy>
Expand All @@ -353,7 +352,7 @@ BOOST_MATH_GPU_ENABLED inline RealType kurtosis(const cauchy_distribution<RealTy
"boost::math::kurtosis(cauchy<%1%>&)",
"The Cauchy distribution does not have a kurtosis: "
"the only possible return value is %1%.",
std::numeric_limits<RealType>::quiet_NaN(), Policy());
boost::math::numeric_limits<RealType>::quiet_NaN(), Policy());
}

template <class RealType, class Policy>
Expand All @@ -367,7 +366,7 @@ BOOST_MATH_GPU_ENABLED inline RealType kurtosis_excess(const cauchy_distribution
"boost::math::kurtosis_excess(cauchy<%1%>&)",
"The Cauchy distribution does not have a kurtosis: "
"the only possible return value is %1%.",
std::numeric_limits<RealType>::quiet_NaN(), Policy());
boost::math::numeric_limits<RealType>::quiet_NaN(), Policy());
}

template <class RealType, class Policy>
Expand Down
Loading