Skip to content

Commit 1f52533

Browse files
tangjj11tangjj11
authored
[SYCLomatic oneapi-src#1351] Disable double cases on some gpu. (oneapi-src#528)
Signed-off-by: Tang, Jiajun [email protected]
1 parent f00e6e0 commit 1f52533

File tree

6 files changed

+175
-101
lines changed

6 files changed

+175
-101
lines changed

‎features/feature_case/math/math-emu-bf16-conv-double.cu

Lines changed: 86 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,86 @@
1+
// ====---------- math-emu-bf16-conv-double.cu---------- *- CUDA -* -------===//
2+
//
3+
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
4+
// See https://llvm.org/LICENSE.txt for license information.
5+
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
6+
//
7+
//
8+
// ===---------------------------------------------------------------------===//
9+
10+
#include <iomanip>
11+
#include <iostream>
12+
#include <vector>
13+
14+
#include "cuda_bf16.h"
15+
16+
using namespace std;
17+
18+
typedef pair<__nv_bfloat16, int> bf16i_pair;
19+
20+
int passed = 0;
21+
int failed = 0;
22+
23+
void check(bool IsPassed) {
24+
if (IsPassed) {
25+
cout << " ---- passed" << endl;
26+
passed++;
27+
} else {
28+
cout << " ---- failed" << endl;
29+
failed++;
30+
}
31+
}
32+
33+
void checkResult(const string &FuncName, const vector<float> &Inputs,
34+
const float &Expect, const float &Result,
35+
const int precision) {
36+
cout << FuncName << "(" << Inputs[0] << "";
37+
for (size_t i = 1; i < Inputs.size(); ++i) {
38+
cout << ", " << Inputs[i];
39+
}
40+
cout << ") = " << fixed << setprecision(precision < 0 ? 0 : precision)
41+
<< Result << " (expect " << Expect - pow(10, -precision) << " ~ "
42+
<< Expect + pow(10, -precision) << ")";
43+
cout.unsetf(ios::fixed);
44+
check(abs(Result - Expect) < pow(10, -precision));
45+
}
46+
47+
void checkResult(const string &FuncName, const vector<float> &Inputs,
48+
const __nv_bfloat16 &Expect, const float &Result,
49+
const int precision) {
50+
float FExpect = __bfloat162float(Expect);
51+
checkResult(FuncName, Inputs, FExpect, Result, precision);
52+
}
53+
54+
// Bfloat16 Precision Conversion and Data Movement
55+
56+
__global__ void double2bfloat16(float *const Result, double Input1) {
57+
*Result = __double2bfloat16(Input1);
58+
}
59+
60+
void testDouble2bfloat16Cases(
61+
const vector<pair<double, bf16i_pair>> &TestCases) {
62+
float *Result;
63+
cudaMallocManaged(&Result, sizeof(*Result));
64+
for (const auto &TestCase : TestCases) {
65+
double2bfloat16<<<1, 1>>>(Result, TestCase.first);
66+
cudaDeviceSynchronize();
67+
checkResult("__double2bfloat16", {(float)TestCase.first},
68+
TestCase.second.first, *Result, TestCase.second.second);
69+
}
70+
}
71+
72+
int main() {
73+
testDouble2bfloat16Cases({
74+
{-0.3, {-0.30078125, 16}},
75+
{0.3, {0.30078125, 16}},
76+
{30, {30, 14}},
77+
{0.432643, {0.43359375, 16}},
78+
{1, {1, 15}},
79+
{10.7, {10.6875, 15}},
80+
});
81+
cout << "passed " << passed << "/" << passed + failed << " cases!" << endl;
82+
if (failed) {
83+
cout << "failed!" << endl;
84+
}
85+
return failed;
86+
}

‎features/feature_case/math/math-emu-double.cu

Lines changed: 0 additions & 50 deletions
Original file line numberDiff line numberDiff line change
@@ -11,11 +11,8 @@
1111
#include <iostream>
1212
#include <vector>
1313

14-
#include "cuda_bf16.h"
15-
1614
using namespace std;
1715

18-
typedef pair<__nv_bfloat16, int> bf16i_pair;
1916
typedef vector<double> d_vector;
2017
typedef tuple<double, double, double> d_tuple3;
2118
typedef tuple<double, double, double, double> d_tuple4;
@@ -34,27 +31,6 @@ void check(bool IsPassed) {
3431
}
3532
}
3633

37-
void checkResult(const string &FuncName, const vector<float> &Inputs,
38-
const float &Expect, const float &Result,
39-
const int precision) {
40-
cout << FuncName << "(" << Inputs[0] << "";
41-
for (size_t i = 1; i < Inputs.size(); ++i) {
42-
cout << ", " << Inputs[i];
43-
}
44-
cout << ") = " << fixed << setprecision(precision) << Result << " (expect "
45-
<< Expect - pow(10, -precision) << " ~ " << Expect + pow(10, -precision)
46-
<< ")";
47-
cout.unsetf(ios::fixed);
48-
check(abs(Result - Expect) < pow(10, -precision));
49-
}
50-
51-
void checkResult(const string &FuncName, const vector<float> &Inputs,
52-
const __nv_bfloat16 &Expect, const float &Result,
53-
const int precision) {
54-
float FExpect = __bfloat162float(Expect);
55-
checkResult(FuncName, Inputs, FExpect, Result, precision);
56-
}
57-
5834
template <typename T = double>
5935
void checkResult(const string &FuncName, const vector<T> &Inputs,
6036
const double &Expect, const double &DeviceResult,
@@ -74,24 +50,6 @@ __global__ void setVecValue(double *Input1, const double Input2) {
7450
*Input1 = Input2;
7551
}
7652

77-
// Bfloat16 Precision Conversion and Data Movement
78-
79-
__global__ void double2bfloat16(float *const Result, double Input1) {
80-
*Result = __double2bfloat16(Input1);
81-
}
82-
83-
void testDouble2bfloat16Cases(
84-
const vector<pair<double, bf16i_pair>> &TestCases) {
85-
float *Result;
86-
cudaMallocManaged(&Result, sizeof(*Result));
87-
for (const auto &TestCase : TestCases) {
88-
double2bfloat16<<<1, 1>>>(Result, TestCase.first);
89-
cudaDeviceSynchronize();
90-
checkResult("__double2bfloat16", {(float)TestCase.first},
91-
TestCase.second.first, *Result, TestCase.second.second);
92-
}
93-
}
94-
9553
// Double Precision Mathematical Functions
9654

9755
__global__ void _norm(double *const DeviceResult, int Input1,
@@ -495,14 +453,6 @@ void testDsub_rzCases(
495453
}
496454

497455
int main() {
498-
testDouble2bfloat16Cases({
499-
{-0.3, {-0.30078125, 16}},
500-
{0.3, {0.30078125, 16}},
501-
{30, {30, 14}},
502-
{0.432643, {0.43359375, 16}},
503-
{1, {1, 15}},
504-
{10.7, {10.6875, 15}},
505-
});
506456
testNormCases({
507457
{{-0.3, -0.34, -0.98}, {1.079814798935447, 15}},
508458
{{0.3, 0.34, 0.98}, {1.079814798935447, 15}},

‎features/feature_case/math/math-ext-bf16-conv-double.cu

Lines changed: 86 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,86 @@
1+
// ====---------- math-ext-bf16-conv-double.cu---------- *- CUDA -* -------===//
2+
//
3+
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
4+
// See https://llvm.org/LICENSE.txt for license information.
5+
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
6+
//
7+
//
8+
// ===---------------------------------------------------------------------===//
9+
10+
#include <iomanip>
11+
#include <iostream>
12+
#include <vector>
13+
14+
#include "cuda_bf16.h"
15+
16+
using namespace std;
17+
18+
typedef pair<__nv_bfloat16, int> bf16i_pair;
19+
20+
int passed = 0;
21+
int failed = 0;
22+
23+
void check(bool IsPassed) {
24+
if (IsPassed) {
25+
cout << " ---- passed" << endl;
26+
passed++;
27+
} else {
28+
cout << " ---- failed" << endl;
29+
failed++;
30+
}
31+
}
32+
33+
void checkResult(const string &FuncName, const vector<float> &Inputs,
34+
const float &Expect, const float &Result,
35+
const int precision) {
36+
cout << FuncName << "(" << Inputs[0] << "";
37+
for (size_t i = 1; i < Inputs.size(); ++i) {
38+
cout << ", " << Inputs[i];
39+
}
40+
cout << ") = " << fixed << setprecision(precision < 0 ? 0 : precision)
41+
<< Result << " (expect " << Expect - pow(10, -precision) << " ~ "
42+
<< Expect + pow(10, -precision) << ")";
43+
cout.unsetf(ios::fixed);
44+
check(abs(Result - Expect) < pow(10, -precision));
45+
}
46+
47+
void checkResult(const string &FuncName, const vector<float> &Inputs,
48+
const __nv_bfloat16 &Expect, const float &Result,
49+
const int precision) {
50+
float FExpect = __bfloat162float(Expect);
51+
checkResult(FuncName, Inputs, FExpect, Result, precision);
52+
}
53+
54+
// Bfloat16 Precision Conversion and Data Movement
55+
56+
__global__ void double2bfloat16(float *const Result, double Input1) {
57+
*Result = __double2bfloat16(Input1);
58+
}
59+
60+
void testDouble2bfloat16Cases(
61+
const vector<pair<double, bf16i_pair>> &TestCases) {
62+
float *Result;
63+
cudaMallocManaged(&Result, sizeof(*Result));
64+
for (const auto &TestCase : TestCases) {
65+
double2bfloat16<<<1, 1>>>(Result, TestCase.first);
66+
cudaDeviceSynchronize();
67+
checkResult("__double2bfloat16", {(float)TestCase.first},
68+
TestCase.second.first, *Result, TestCase.second.second);
69+
}
70+
}
71+
72+
int main() {
73+
testDouble2bfloat16Cases({
74+
{-0.3, {-0.30078125, 16}},
75+
{0.3, {0.30078125, 16}},
76+
{30, {30, 14}},
77+
{0.432643, {0.43359375, 16}},
78+
{1, {1, 15}},
79+
{10.7, {10.6875, 15}},
80+
});
81+
cout << "passed " << passed << "/" << passed + failed << " cases!" << endl;
82+
if (failed) {
83+
cout << "failed!" << endl;
84+
}
85+
return failed;
86+
}

‎features/feature_case/math/math-ext-double.cu

Lines changed: 0 additions & 50 deletions
Original file line numberDiff line numberDiff line change
@@ -11,11 +11,8 @@
1111
#include <iostream>
1212
#include <vector>
1313

14-
#include "cuda_bf16.h"
15-
1614
using namespace std;
1715

18-
typedef pair<__nv_bfloat16, int> bf16i_pair;
1916
typedef vector<double> d_vector;
2017
typedef pair<double, int> di_pair;
2118

@@ -32,27 +29,6 @@ void check(bool IsPassed) {
3229
}
3330
}
3431

35-
void checkResult(const string &FuncName, const vector<float> &Inputs,
36-
const float &Expect, const float &Result,
37-
const int precision) {
38-
cout << FuncName << "(" << Inputs[0] << "";
39-
for (size_t i = 1; i < Inputs.size(); ++i) {
40-
cout << ", " << Inputs[i];
41-
}
42-
cout << ") = " << fixed << setprecision(precision < 0 ? 0 : precision)
43-
<< Result << " (expect " << Expect - pow(10, -precision) << " ~ "
44-
<< Expect + pow(10, -precision) << ")";
45-
cout.unsetf(ios::fixed);
46-
check(abs(Result - Expect) < pow(10, -precision));
47-
}
48-
49-
void checkResult(const string &FuncName, const vector<float> &Inputs,
50-
const __nv_bfloat16 &Expect, const float &Result,
51-
const int precision) {
52-
float FExpect = __bfloat162float(Expect);
53-
checkResult(FuncName, Inputs, FExpect, Result, precision);
54-
}
55-
5632
template <typename T = double>
5733
void checkResult(const string &FuncName, const vector<T> &Inputs,
5834
const double &Expect, const double &DeviceResult,
@@ -68,24 +44,6 @@ void checkResult(const string &FuncName, const vector<T> &Inputs,
6844
check(abs(DeviceResult - Expect) < pow(10, -precision));
6945
}
7046

71-
// Bfloat16 Precision Conversion and Data Movement
72-
73-
__global__ void double2bfloat16(float *const Result, double Input1) {
74-
*Result = __double2bfloat16(Input1);
75-
}
76-
77-
void testDouble2bfloat16Cases(
78-
const vector<pair<double, bf16i_pair>> &TestCases) {
79-
float *Result;
80-
cudaMallocManaged(&Result, sizeof(*Result));
81-
for (const auto &TestCase : TestCases) {
82-
double2bfloat16<<<1, 1>>>(Result, TestCase.first);
83-
cudaDeviceSynchronize();
84-
checkResult("__double2bfloat16", {(float)TestCase.first},
85-
TestCase.second.first, *Result, TestCase.second.second);
86-
}
87-
}
88-
8947
// Double Precision Mathematical Functions
9048

9149
__global__ void cylBesselI0(double *const Result, double Input1) {
@@ -549,14 +507,6 @@ void testDsub_rzCases(
549507
}
550508

551509
int main() {
552-
testDouble2bfloat16Cases({
553-
{-0.3, {-0.30078125, 16}},
554-
{0.3, {0.30078125, 16}},
555-
{30, {30, 14}},
556-
{0.432643, {0.43359375, 16}},
557-
{1, {1, 15}},
558-
{10.7, {10.6875, 15}},
559-
});
560510
testCylBesselI0Cases({
561511
{0.3, {1.022626879351597, 15}},
562512
{0.5, {1.063483370741324, 15}},

‎features/features.xml

Lines changed: 2 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -107,6 +107,7 @@
107107
<test testName="math-drcp" configFile="config/TEMPLATE_math_drcp.xml" splitGroup="double" />
108108
<test testName="math-emu" configFile="config/TEMPLATE_math.xml" />
109109
<test testName="math-emu-bf16" configFile="config/TEMPLATE_math_after_11.xml" />
110+
<test testName="math-emu-bf16-conv-double" configFile="config/TEMPLATE_math_after_11_skip_double.xml" splitGroup="double"/>
110111
<test testName="math-emu-bf162-after12" configFile="config/TEMPLATE_math_after_12.xml" />
111112
<test testName="math-emu-bf162" configFile="config/TEMPLATE_math_after_11.xml" />
112113
<test testName="math-emu-double" configFile="config/TEMPLATE_math_skip_double.xml" splitGroup="double"/>
@@ -124,6 +125,7 @@
124125
<test testName="math-experimental-bf16" configFile="config/TEMPLATE_math_after_11.xml" />
125126
<test testName="math-experimental-bf162" configFile="config/TEMPLATE_math_after_11.xml" />
126127
<test testName="math-ext-bf16-conv" configFile="config/TEMPLATE_math_after_11.xml" />
128+
<test testName="math-ext-bf16-conv-double" configFile="config/TEMPLATE_math_after_11_skip_double.xml" splitGroup="double"/>
127129
<test testName="math-ext-double" configFile="config/TEMPLATE_math_skip_double.xml" splitGroup="double"/>
128130
<test testName="math-ext-float" configFile="config/TEMPLATE_math_skip_double.xml" splitGroup="double"/>
129131
<test testName="math-ext-half" configFile="config/TEMPLATE_math_after_9.xml" />

‎features/test_feature.py

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -35,7 +35,7 @@
3535
'math-ext-bf16-conv', 'math-ext-double', 'math-ext-float', 'math-ext-half', 'math-ext-half-after11', 'math-ext-half-conv', 'math-ext-half2', 'math-ext-half2-after11', 'math-ext-simd', 'cudnn-activation',
3636
'cudnn-fill', 'cudnn-lrn', 'cudnn-memory', 'cudnn-pooling', 'cudnn-reorder', 'cudnn-scale', 'cudnn-softmax',
3737
'cudnn-sum', 'math-funnelshift', 'thrust-sort_by_key', 'thrust-find', 'thrust-inner_product', 'thrust-reduce_by_key',
38-
'math-bf16-conv', 'math-half-conv',
38+
'math-bf16-conv', 'math-emu-bf16-conv-double', 'math-ext-bf16-conv-double', 'math-half-conv',
3939
'math-bfloat16', 'libcu_atomic', 'test_shared_memory', 'cudnn-reduction', 'cudnn-binary', 'cudnn-bnp1', 'cudnn-bnp2', 'cudnn-bnp3',
4040
'cudnn-normp1', 'cudnn-normp2', 'cudnn-normp3', 'cudnn-convp1', 'cudnn-convp2', 'cudnn-convp3', 'cudnn-convp4', 'cudnn-convp5', 'cudnn-convp6',
4141
'cudnn_mutilple_files', "cusparse_1", "cusparse_2", "cusparse_3", "cusparse_4", "cusparse_5", "cusparse_6", "cusparse_7",

0 commit comments

Comments
 (0)