Cutlass CUDA Templates for Linear Algebra Subroutines and Solvers

complex.h
Go to the documentation of this file.
1 /***************************************************************************************************
2 * Copyright (c) 2017-2018, NVIDIA CORPORATION. All rights reserved.
3 *
4 * Redistribution and use in source and binary forms, with or without modification, are permitted
5 * provided that the following conditions are met:
6 * * Redistributions of source code must retain the above copyright notice, this list of
7 * conditions and the following disclaimer.
8 * * Redistributions in binary form must reproduce the above copyright notice, this list of
9 * conditions and the following disclaimer in the documentation and/or other materials
10 * provided with the distribution.
11 * * Neither the name of the NVIDIA CORPORATION nor the names of its contributors may be used
12 * to endorse or promote products derived from this software without specific prior written
13 * permission.
14 *
15 * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS" AND ANY EXPRESS OR
16 * IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND
17 * FITNESS FOR A PARTICULAR PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL NVIDIA CORPORATION BE LIABLE
18 * FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING,
19 * BUT NOT LIMITED TO, PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR PROFITS;
20 * OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN CONTRACT,
21 * STRICT LIABILITY, OR TOR (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE
22 * OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
23 *
24 **************************************************************************************************/
25 #pragma once
26
27 #include <cuComplex.h>
28 #include "cutlass/cutlass.h"
29 #include <iosfwd>
30
31 namespace cutlass {
32 namespace platform {
33
35
36 //
37 // Accessors for CUDA complex types
38 //
39
41 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
42 // host-only type
43 CUTLASS_HOST_DEVICE
44 float const &real(cuFloatComplex const &z) { return z.x; }
45
47 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
48 // host-only type
49 CUTLASS_HOST_DEVICE
50 float &real(cuFloatComplex &z) { return z.x; }
51
53 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
54 // host-only type
55 CUTLASS_HOST_DEVICE
56 double const &real(cuDoubleComplex const &z) { return z.x; }
57
59 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
60 // host-only type
61 CUTLASS_HOST_DEVICE
62 double &real(cuDoubleComplex &z) { return z.x; }
63
65 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
66 // host-only type
67 CUTLASS_HOST_DEVICE
68 float const &imag(cuFloatComplex const &z) { return z.y; }
69
71 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
72 // host-only type
73 CUTLASS_HOST_DEVICE
74 float &imag(cuFloatComplex &z) { return z.y; }
75
77 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
78 // host-only type
79 CUTLASS_HOST_DEVICE
80 double const &imag(cuDoubleComplex const &z) { return z.y; }
81
83 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
84 // host-only type
85 CUTLASS_HOST_DEVICE
86 double &imag(cuDoubleComplex &z) { return z.y; }
87
89
92 template <typename T>
93 class complex {
94 public:
96 typedef T value_type;
97
98 private:
99 //
100 // Data members
101 //
102
104 T _real;
105
107 T _imag;
108
109 public:
110 //
111 // Methods
112 //
113
115 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
116 // host-only type
117 CUTLASS_HOST_DEVICE
118 complex(T r = T(0), T i = T(0)) : _real(r), _imag(i) {}
119
121 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
122 // host-only type
123 CUTLASS_HOST_DEVICE
124 complex(cuFloatComplex const &z) : _real(platform::real(z)), _imag(platform::imag(z)) {}
125
127 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
128 // host-only type
129 CUTLASS_HOST_DEVICE
130 complex(cuDoubleComplex const &z) : _real(platform::real(z)), _imag(platform::imag(z)) {}
131
133 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
134 // host-only type
135 CUTLASS_HOST_DEVICE
136 T const &real() const { return _real; }
137
139 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
140 // host-only type
141 CUTLASS_HOST_DEVICE
142 T &real() { return _real; }
143
145 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
146 // host-only type
147 CUTLASS_HOST_DEVICE
148 T const &imag() const { return _imag; }
149
151 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
152 // host-only type
153 CUTLASS_HOST_DEVICE
154 T &imag() { return _imag; }
155
157 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
158 // host-only type
159 CUTLASS_HOST_DEVICE
160 operator cuFloatComplex() const { return make_cuFloatComplex(real(), imag()); }
161
163 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
164 // host-only type
165 CUTLASS_HOST_DEVICE
166 operator cuDoubleComplex() const { return make_cuDoubleComplex(real(), imag()); }
167 };
168
169 //
170 // Accessors for complex template
171 //
172
174 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
175 // host-only type
176 template <typename T>
177 CUTLASS_HOST_DEVICE T const &real(complex<T> const &z) {
178 return z.real();
179 }
180
182 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
183 // host-only type
184 template <typename T>
185 CUTLASS_HOST_DEVICE T &real(complex<T> &z) {
186 return z.real();
187 }
188
190 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
191 // host-only type
192 template <typename T>
193 CUTLASS_HOST_DEVICE T const &imag(complex<T> const &z) {
194 return z.imag();
195 }
196
198 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
199 // host-only type
200 template <typename T>
201 CUTLASS_HOST_DEVICE T &imag(complex<T> &z) {
202 return z.imag();
203 }
204
205 //
206 // Output operators
207 //
208
209 template <typename T>
210 std::ostream &operator<<(std::ostream &out, complex<T> const &z) {
211 T _r = real(z);
212 T _i = imag(z);
213 return out << _r << "+i" << _i;
214 }
215
216 //
217 // Non-member operators defined for complex types
218 //
219
221 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
222 // host-only type
223 template <typename T>
224 CUTLASS_HOST_DEVICE bool operator==(complex<T> const &lhs, complex<T> const &rhs) {
225 return real(lhs) == (rhs) && imag(lhs) == imag(rhs);
226 }
227
229 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
230 // host-only type
231 template <typename T>
232 CUTLASS_HOST_DEVICE bool operator!=(complex<T> const &lhs, complex<T> const &rhs) {
233 return !(lhs == rhs);
234 }
235
237 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
238 // host-only type
239 template <typename T>
240 CUTLASS_HOST_DEVICE complex<T> operator+(complex<T> const &lhs, complex<T> const &rhs) {
241 return complex<T>(real(lhs) + real(rhs), imag(lhs) + imag(rhs));
242 }
243
245 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
246 // host-only type
247 template <typename T>
248 CUTLASS_HOST_DEVICE complex<T> operator-(complex<T> const &lhs, complex<T> const &rhs) {
249 return complex<T>(real(lhs) - real(rhs), imag(lhs) - imag(rhs));
250 }
251
253 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
254 // host-only type
255 template <typename T>
256 CUTLASS_HOST_DEVICE complex<T> operator*(complex<T> const &lhs, complex<T> const &rhs) {
257 return complex<T>(real(lhs) * real(rhs) - imag(lhs) * imag(rhs),
258 real(lhs) * imag(rhs) + imag(lhs) * real(rhs));
259 }
260
262 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
263 // host-only type
264 template <typename T>
265 CUTLASS_HOST_DEVICE complex<T> operator*(complex<T> const &lhs, T const &s) {
266 return complex<T>(real(lhs) * s, imag(lhs) * s);
267 }
268
270 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
271 // host-only type
272 template <typename T>
273 CUTLASS_HOST_DEVICE complex<T> operator*(T const &s, complex<T> const &rhs) {
274 return complex<T>(s * real(rhs), s * imag(rhs));
275 }
276
278 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
279 // host-only type
280 template <typename T>
281 CUTLASS_HOST_DEVICE complex<T> operator/(complex<T> const &lhs, complex<T> const &rhs) {
282 T d = (real(rhs) * (rhs) + imag(rhs) * imag(rhs));
283
284 return complex<T>((real(lhs) * (rhs) + imag(lhs) * imag(rhs)) / d,
285 (imag(lhs) * (rhs)-real(lhs) * imag(rhs)) / d);
286 }
287
289 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
290 // host-only type
291 template <typename T>
292 CUTLASS_HOST_DEVICE complex<T> operator/(complex<T> const &lhs, T const &s) {
293 return complex<T>(real(lhs) / s, imag(lhs) / s);
294 }
295
297 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
298 // host-only type
299 template <typename T>
300 CUTLASS_HOST_DEVICE complex<T> operator/(T const &s, complex<T> const &rhs) {
301 T d = (real(rhs) * (rhs) + imag(rhs) * imag(rhs));
302
303 return complex<T>((s * (rhs)) / d, -(s * imag(rhs)) / d);
304 }
305
307 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
308 // host-only type
309 template <typename T>
310 CUTLASS_HOST_DEVICE complex<T> &operator+=(complex<T> &lhs, complex<T> const &rhs) {
311 lhs = (lhs + rhs);
312 return lhs;
313 }
314
316 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
317 // host-only type
318 template <typename T>
319 CUTLASS_HOST_DEVICE complex<T> &operator-=(complex<T> &lhs, complex<T> const &rhs) {
320 lhs = (lhs - rhs);
321 return lhs;
322 }
323
325 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
326 // host-only type
327 template <typename T>
328 CUTLASS_HOST_DEVICE complex<T> &operator*=(complex<T> &lhs, complex<T> const &rhs) {
329 lhs = (lhs * rhs);
330 return lhs;
331 }
332
334 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
335 // host-only type
336 template <typename T>
337 CUTLASS_HOST_DEVICE complex<T> &operator*=(complex<T> &lhs, T s) {
338 lhs = (lhs * s);
339 return lhs;
340 }
341
343 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
344 // host-only type
345 template <typename T>
346 CUTLASS_HOST_DEVICE complex<T> &operator/=(complex<T> &lhs, complex<T> const &rhs) {
347 lhs = (lhs / rhs);
348 return lhs;
349 }
350
351 //
352 // Non-member functions defined for complex numbers
353 //
354
356 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
357 // host-only type
358 template <typename T>
359 CUTLASS_HOST_DEVICE T abs(complex<T> const &z) {
360 return sqrt(norm(z));
361 }
362
364 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
365 // host-only type
366 template <typename T>
367 CUTLASS_HOST_DEVICE T arg(complex<T> const &z) {
368 return atan2(imag(z), real(z));
369 }
370
372 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
373 // host-only type
374 template <typename T>
375 CUTLASS_HOST_DEVICE T norm(complex<T> const &z) {
376 return real(z) * real(z) + imag(z) * imag(z);
377 }
378
380 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
381 // host-only type
382 template <typename T>
383 CUTLASS_HOST_DEVICE complex<T> conj(complex<T> const &z) {
384 return complex<T>(real(z), -imag(z));
385 }
386
388 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
389 // host-only type
390 template <typename T>
391 CUTLASS_HOST_DEVICE complex<T> proj(complex<T> const &z) {
392 T d = real(z) * real(z) + imag(z) * imag(z) + T(1);
393 return complex<T>((T(2) * real(z)) / d, (T(2) * imag(z)) / d);
394 }
395
397 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
398 // host-only type
399 template <typename T>
400 CUTLASS_HOST_DEVICE complex<T> polar(T const &r, T const &theta = T()) {
401 return complex<T>(r * cos(theta), r * sin(theta));
402 }
403
405 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
406 // host-only type
407 template <typename T>
408 CUTLASS_HOST_DEVICE complex<T> exp(complex<T> const &z) {
409 return complex<T>(real(z) * cos(imag(z)), real(z) * sin(imag(z)));
410 }
411
413 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
414 // host-only type
415 template <typename T>
416 CUTLASS_HOST_DEVICE complex<T> log(complex<T> const &z) {
417 return complex<T>(log(abs(z)), arg(z));
418 }
419
421 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
422 // host-only type
423 template <typename T>
424 CUTLASS_HOST_DEVICE complex<T> log10(complex<T> const &z) {
425 return log(z) / T(log(T(10)));
426 }
427
429 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
430 // host-only type
431 template <typename T>
432 CUTLASS_HOST_DEVICE complex<T> sqrt(complex<T> const &z) {
433 return sqrt(T(2)) / T(2) *
434 complex<T>(sqrt(sqrt(norm(z)) + real(z)),
435 (imag(z) < 0 ? T(-1) : T(1)) * sqrt(sqrt(norm(z)) - real(z)));
436 }
437
439 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
440 // host-only type
441 template <typename T>
442 CUTLASS_HOST_DEVICE complex<T> cos(complex<T> const &z) {
443 return (exp(z) + exp(-z)) / T(2);
444 }
445
447 #pragma hd_warning_disable // Suppresses warnings when attempting to instantiate complex<T> with a
448 // host-only type
449 template <typename T>
450 CUTLASS_HOST_DEVICE complex<T> sin(complex<T> const &z) {
451 return (exp(-z) - exp(z)) * complex<T>(T(0), T(1) / T(2));
452 }
453
455
456 } // namespace platform
457 } // namespace cutlass
cutlass::platform::proj
CUTLASS_HOST_DEVICE complex< T > proj(complex< T > const &z)
Projects the complex number z onto the Riemann sphere.
Definition: complex.h:391
cutlass
Definition: convert.h:33
cutlass::platform::complex::imag
CUTLASS_HOST_DEVICE T & imag()
Accesses the imaginary part of the complex number.
Definition: complex.h:154
cutlass::platform::operator==
CUTLASS_HOST_DEVICE bool operator==(complex< T > const &lhs, complex< T > const &rhs)
Equality operator.
Definition: complex.h:224
cutlass::platform::complex::imag
CUTLASS_HOST_DEVICE T const & imag() const
Accesses the imaginary part of the complex number.
Definition: complex.h:148
cutlass::platform::operator*
CUTLASS_HOST_DEVICE complex< T > operator*(complex< T > const &lhs, complex< T > const &rhs)
Multiplication.
Definition: complex.h:256
cutlass::platform::operator-=
CUTLASS_HOST_DEVICE complex< T > & operator-=(complex< T > &lhs, complex< T > const &rhs)
Subtraction.
Definition: complex.h:319
cutlass::platform::operator-
CUTLASS_HOST_DEVICE complex< T > operator-(complex< T > const &lhs, complex< T > const &rhs)
Subtraction.
Definition: complex.h:248
cutlass::platform::complex::real
CUTLASS_HOST_DEVICE T & real()
Accesses the real part of the complex number.
Definition: complex.h:142
cutlass::platform::real
CUTLASS_HOST_DEVICE float const & real(cuFloatComplex const &z)
Returns the real part of the complex number.
Definition: complex.h:44
cutlass::platform::sin
CUTLASS_HOST_DEVICE complex< T > sin(complex< T > const &z)
Computes the sin of complex z.
Definition: complex.h:450
cutlass::platform::complex::complex
CUTLASS_HOST_DEVICE complex(cuFloatComplex const &z)
Conversion from cuFloatComplex.
Definition: complex.h:124
cutlass::platform::cos
CUTLASS_HOST_DEVICE complex< T > cos(complex< T > const &z)
Computes the cosine of complex z.
Definition: complex.h:442
cutlass::platform::operator+
CUTLASS_HOST_DEVICE complex< T > operator+(complex< T > const &lhs, complex< T > const &rhs)
Addition.
Definition: complex.h:240
cutlass::platform::polar
CUTLASS_HOST_DEVICE complex< T > polar(T const &r, T const &theta=T())
Returns a complex number with magnitude r and phase theta.
Definition: complex.h:400
cutlass::platform::complex::real
CUTLASS_HOST_DEVICE T const & real() const
Accesses the real part of the complex number.
Definition: complex.h:136
cutlass::platform::operator/=
CUTLASS_HOST_DEVICE complex< T > & operator/=(complex< T > &lhs, complex< T > const &rhs)
Division.
Definition: complex.h:346
cutlass::platform::sqrt
CUTLASS_HOST_DEVICE complex< T > sqrt(complex< T > const &z)
Computes the square root of complex number z.
Definition: complex.h:432
cutlass::platform::operator+=
CUTLASS_HOST_DEVICE complex< T > & operator+=(complex< T > &lhs, complex< T > const &rhs)
Addition.
Definition: complex.h:310
CUTLASS_HOST_DEVICE
#define CUTLASS_HOST_DEVICE
Definition: cutlass.h:46
cutlass::platform::imag
CUTLASS_HOST_DEVICE float const & imag(cuFloatComplex const &z)
Returns the imaginary part of the complex number.
Definition: complex.h:68
cutlass::platform::exp
CUTLASS_HOST_DEVICE complex< T > exp(complex< T > const &z)
Computes the complex exponential of z.
Definition: complex.h:408
cutlass::platform::log10
CUTLASS_HOST_DEVICE complex< T > log10(complex< T > const &z)
Computes the complex exponential of z.
Definition: complex.h:424
cutlass::platform::norm
CUTLASS_HOST_DEVICE T norm(complex< T > const &z)
Returns the squared magnitude.
Definition: complex.h:375
cutlass::platform::operator!=
CUTLASS_HOST_DEVICE bool operator!=(complex< T > const &lhs, complex< T > const &rhs)
Inequality operator.
Definition: complex.h:232
cutlass::platform::abs
CUTLASS_HOST_DEVICE T abs(complex< T > const &z)
Returns the magnitude of the complex number.
Definition: complex.h:359
cutlass::platform::operator*=
CUTLASS_HOST_DEVICE complex< T > & operator*=(complex< T > &lhs, complex< T > const &rhs)
Multiplication.
Definition: complex.h:328
cutlass::platform::complex::complex
CUTLASS_HOST_DEVICE complex(cuDoubleComplex const &z)
Conversion from cuDoubleComplex.
Definition: complex.h:130
cutlass::platform::arg
CUTLASS_HOST_DEVICE T arg(complex< T > const &z)
Returns the magnitude of the complex number.
Definition: complex.h:367
cutlass::platform::complex::complex
CUTLASS_HOST_DEVICE complex(T r=T(0), T i=T(0))
Constructor.
Definition: complex.h:118
cutlass::platform::complex
Definition: complex.h:93
cutlass::platform::log
CUTLASS_HOST_DEVICE complex< T > log(complex< T > const &z)
Computes the complex exponential of z.
Definition: complex.h:416
cutlass::platform::complex::value_type
T value_type
Type alias for scalar type.
Definition: complex.h:96
cutlass.h
Basic include for CUTLASS macros.
cutlass::platform::operator/
CUTLASS_HOST_DEVICE complex< T > operator/(complex< T > const &lhs, complex< T > const &rhs)
Division.
Definition: complex.h:281
cutlass::platform::conj
CUTLASS_HOST_DEVICE complex< T > conj(complex< T > const &z)
Returns the complex conjugate.
Definition: complex.h:383