[igt-dev] [RFC 0/5] Add support for fp16 formats

[igt-dev] [RFC 0/5] Add support for fp16 formats

[Kevin Strasser]

This series adds support for testing new fp16 formats. As these formats
haven't yet landed in the kernel I am sharing this as an RFC.

In particular coverage is added to the following tests which
already enumerate through all formats advertised by each plane:
  kms_available_modes_crc
  kms_rotation_crc
  kms_plane_scaling

Kevin Strasser (5):
  lib: Add halffloat implementation
  include: Add fp16 format defines
  tests/kms_available_modes_crc: Add cases for 64bit formats
  lib/igt_fb: Add support for fp16 formats through conversion
  tests/kms_plane_scaling: Skip testing unsupported fp16 features

 COPYING                         |  22 ++++
 include/drm-uapi/drm_fourcc.h   |  10 ++
 lib/Makefile.sources            |   2 +
 lib/igt_fb.c                    | 188 ++++++++++++++++++++++++++++++++--
 lib/igt_fb.h                    |   1 +
 lib/igt_halffloat.c             | 218 ++++++++++++++++++++++++++++++++++++++++
 lib/igt_halffloat.h             |  27 +++++
 lib/igt_x86.c                   |   9 ++
 lib/igt_x86.h                   |   1 +
 lib/meson.build                 |   1 +
 tests/kms_available_modes_crc.c |  19 +++-
 tests/kms_plane_scaling.c       |  14 +++
 12 files changed, 502 insertions(+), 10 deletions(-)
 create mode 100644 lib/igt_halffloat.c
 create mode 100644 lib/igt_halffloat.h

-- 
2.7.4

_______________________________________________
igt-dev mailing list
igt-dev@lists.freedesktop.org
https://lists.freedesktop.org/mailman/listinfo/igt-dev

[igt-dev] [RFC 1/5] lib: Add halffloat implementation

[Kevin Strasser]

Probe for and make an API available for tests to use f16c intrinsics to
generate needed fp16 pixel data.

Also import a pure c fp32 <-> fp16 conversion implementation from Mesa
18.3.4, which will act as a fallback when f16c is unavailable.

Signed-off-by: Kevin Strasser <kevin.strasser@intel.com>
---
 COPYING              |  22 ++++++
 lib/Makefile.sources |   2 +
 lib/igt_halffloat.c  | 218 +++++++++++++++++++++++++++++++++++++++++++++++++++
 lib/igt_halffloat.h  |  27 +++++++
 lib/igt_x86.c        |   9 +++
 lib/igt_x86.h        |   1 +
 lib/meson.build      |   1 +
 7 files changed, 280 insertions(+)
 create mode 100644 lib/igt_halffloat.c
 create mode 100644 lib/igt_halffloat.h

diff --git a/COPYING b/COPYING
index 29e0238..0fe6584 100644
--- a/COPYING
+++ b/COPYING
@@ -126,3 +126,25 @@ SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES
 WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN
 ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF
 OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.
+
+Copyright (C) 1999-2007  Brian Paul   All Rights Reserved.
+Copyright 2015 Philip Taylor <philip@zaynar.co.uk>
+Copyright 2018 Advanced Micro Devices, Inc.
+
+Permission is hereby granted, free of charge, to any person obtaining a
+copy of this software and associated documentation files (the "Software"),
+to deal in the Software without restriction, including without limitation
+the rights to use, copy, modify, merge, publish, distribute, sublicense,
+and/or sell copies of the Software, and to permit persons to whom the
+Software is furnished to do so, subject to the following conditions:
+
+The above copyright notice and this permission notice shall be included
+in all copies or substantial portions of the Software.
+
+THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS
+OR IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL
+THE AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR
+OTHER LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE,
+ARISING FROM, OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR
+OTHER DEALINGS IN THE SOFTWARE.
diff --git a/lib/Makefile.sources b/lib/Makefile.sources
index cf27209..89f6ce5 100644
--- a/lib/Makefile.sources
+++ b/lib/Makefile.sources
@@ -30,6 +30,8 @@ lib_source_list =	 	\
 	igt_gt.h		\
 	igt_gvt.c		\
 	igt_gvt.h		\
+	igt_halffloat.c		\
+	igt_halffloat.h		\
 	igt_matrix.c		\
 	igt_matrix.h		\
 	igt_primes.c		\
diff --git a/lib/igt_halffloat.c b/lib/igt_halffloat.c
new file mode 100644
index 0000000..e2ab00a
--- /dev/null
+++ b/lib/igt_halffloat.c
@@ -0,0 +1,218 @@
+/*
+ * Copyright (C) 1999-2007  Brian Paul   All Rights Reserved.
+ * Copyright 2015 Philip Taylor <philip@zaynar.co.uk>
+ * Copyright 2018 Advanced Micro Devices, Inc.
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a
+ * copy of this software and associated documentation files (the "Software"),
+ * to deal in the Software without restriction, including without limitation
+ * the rights to use, copy, modify, merge, publish, distribute, sublicense,
+ * and/or sell copies of the Software, and to permit persons to whom the
+ * Software is furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included
+ * in all copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS
+ * OR IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL
+ * THE AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR
+ * OTHER LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE,
+ * ARISING FROM, OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR
+ * OTHER DEALINGS IN THE SOFTWARE.
+ */
+
+#include <assert.h>
+#include <math.h>
+
+#include "igt_halffloat.h"
+#include "igt_x86.h"
+
+typedef union { float f; int32_t i; uint32_t u; } fi_type;
+
+/**
+ * Convert a 4-byte float to a 2-byte half float.
+ *
+ * Not all float32 values can be represented exactly as a float16 value. We
+ * round such intermediate float32 values to the nearest float16. When the
+ * float32 lies exactly between to float16 values, we round to the one with
+ * an even mantissa.
+ *
+ * This rounding behavior has several benefits:
+ *   - It has no sign bias.
+ *
+ *   - It reproduces the behavior of real hardware: opcode F32TO16 in Intel's
+ *     GPU ISA.
+ *
+ *   - By reproducing the behavior of the GPU (at least on Intel hardware),
+ *     compile-time evaluation of constant packHalf2x16 GLSL expressions will
+ *     result in the same value as if the expression were executed on the GPU.
+ */
+static uint16_t float_to_half(float val)
+{
+	const fi_type fi = {val};
+	const int flt_m = fi.i & 0x7fffff;
+	const int flt_e = (fi.i >> 23) & 0xff;
+	const int flt_s = (fi.i >> 31) & 0x1;
+	int s, e, m = 0;
+	uint16_t result;
+
+	/* sign bit */
+	s = flt_s;
+
+	/* handle special cases */
+	if ((flt_e == 0) && (flt_m == 0)) {
+		/* zero */
+		/* m = 0; - already set */
+		e = 0;
+	} else if ((flt_e == 0) && (flt_m != 0)) {
+		/* denorm -- denorm float maps to 0 half */
+		/* m = 0; - already set */
+		e = 0;
+	} else if ((flt_e == 0xff) && (flt_m == 0)) {
+		/* infinity */
+		/* m = 0; - already set */
+		e = 31;
+	} else if ((flt_e == 0xff) && (flt_m != 0)) {
+		/* NaN */
+		m = 1;
+		e = 31;
+	} else {
+		/* regular number */
+		const int new_exp = flt_e - 127;
+		if (new_exp < -14) {
+			/* The float32 lies in the range (0.0, min_normal16) and
+			 * is rounded to a nearby float16 value. The result will
+			 * be either zero, subnormal, or normal.
+			 */
+			e = 0;
+			m = lrintf((1 << 24) * fabsf(fi.f));
+		} else if (new_exp > 15) {
+			/* map this value to infinity */
+			/* m = 0; - already set */
+			e = 31;
+		} else {
+			/* The float32 lies in the range
+			 *   [min_normal16, max_normal16 + max_step16)
+			 * and is rounded to a nearby float16 value. The result
+			 * will be either normal or infinite.
+			 */
+			e = new_exp + 15;
+			m = lrintf(flt_m / (float)(1 << 13));
+		}
+	}
+
+	assert(0 <= m && m <= 1024);
+	if (m == 1024) {
+		/* The float32 was rounded upwards into the range of the next
+		 * exponent, so bump the exponent. This correctly handles the
+		 * case where f32 should be rounded up to float16 infinity.
+		 */
+		++e;
+		m = 0;
+	}
+
+	result = (s << 15) | (e << 10) | m;
+	return result;
+}
+
+/**
+ * Convert a 2-byte half float to a 4-byte float.
+ * Based on code from:
+ * http://www.opengl.org/discussion_boards/ubb/Forum3/HTML/008786.html
+ */
+static float half_to_float(uint16_t val)
+{
+	/* XXX could also use a 64K-entry lookup table */
+	const int m = val & 0x3ff;
+	const int e = (val >> 10) & 0x1f;
+	const int s = (val >> 15) & 0x1;
+	int flt_m, flt_e, flt_s;
+	fi_type fi;
+
+	/* sign bit */
+	flt_s = s;
+
+	/* handle special cases */
+	if ((e == 0) && (m == 0)) {
+		/* zero */
+		flt_m = 0;
+		flt_e = 0;
+	} else if ((e == 0) && (m != 0)) {
+		/* denorm -- denorm half will fit in non-denorm single */
+		const float half_denorm = 1.0f / 16384.0f; /* 2^-14 */
+		float mantissa = ((float) (m)) / 1024.0f;
+		float sign = s ? -1.0f : 1.0f;
+		return sign * mantissa * half_denorm;
+	} else if ((e == 31) && (m == 0)) {
+		/* infinity */
+		flt_e = 0xff;
+		flt_m = 0;
+	} else if ((e == 31) && (m != 0)) {
+		/* NaN */
+		flt_e = 0xff;
+		flt_m = 1;
+	} else {
+		/* regular */
+		flt_e = e + 112;
+		flt_m = m << 13;
+	}
+
+	fi.i = (flt_s << 31) | (flt_e << 23) | flt_m;
+	return fi.f;
+}
+
+#if defined(__x86_64__) && !defined(__clang__)
+#pragma GCC push_options
+#pragma GCC target("f16c")
+
+#include <immintrin.h>
+
+static uint16_t float_to_half_f16c(float f)
+{
+	return _cvtss_sh(f, 0);
+}
+
+static float half_to_float_f16c(uint16_t h)
+{
+	return _cvtsh_ss(h);
+}
+
+#pragma GCC pop_options
+
+static uint16_t (*resolve_float_to_half(void))(float f)
+{
+	if (igt_x86_features() & F16C)
+		return float_to_half_f16c;
+
+	return float_to_half;
+}
+
+uint16_t igt_float_to_half(float f)
+	__attribute__((ifunc("resolve_float_to_half")));
+
+static float (*resolve_half_to_float(void))(uint16_t h)
+{
+	if (igt_x86_features() & F16C)
+		return half_to_float_f16c;
+
+	return half_to_float;
+}
+
+float igt_half_to_float(uint16_t h)
+	__attribute__((ifunc("resolve_half_to_float")));
+
+#else
+
+igt_half igt_float_to_half(float f)
+{
+	return float_to_half(f);
+}
+
+float igt_half_to_float(uint16_t h)
+{
+	return half_to_float(h);
+}
+
+#endif
+
diff --git a/lib/igt_halffloat.h b/lib/igt_halffloat.h
new file mode 100644
index 0000000..2ae189e
--- /dev/null
+++ b/lib/igt_halffloat.h
@@ -0,0 +1,27 @@
+/*
+ * Copyright (C) 1999-2007  Brian Paul   All Rights Reserved.
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a
+ * copy of this software and associated documentation files (the "Software"),
+ * to deal in the Software without restriction, including without limitation
+ * the rights to use, copy, modify, merge, publish, distribute, sublicense,
+ * and/or sell copies of the Software, and to permit persons to whom the
+ * Software is furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included
+ * in all copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS
+ * OR IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL
+ * THE AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR
+ * OTHER LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE,
+ * ARISING FROM, OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR
+ * OTHER DEALINGS IN THE SOFTWARE.
+ */
+
+#include <stdint.h>
+
+uint16_t igt_float_to_half(float f);
+float igt_half_to_float(uint16_t h);
+
diff --git a/lib/igt_x86.c b/lib/igt_x86.c
index 88e514d..9d28e6b 100644
--- a/lib/igt_x86.c
+++ b/lib/igt_x86.c
@@ -88,6 +88,10 @@
 #define bit_AVX		(1 << 28)
 #endif
 
+#ifndef bit_F16C
+#define bit_F16C	(1 << 29)
+#endif
+
 #ifndef bit_AVX2
 #define bit_AVX2	(1<<5)
 #endif
@@ -138,6 +142,9 @@ unsigned igt_x86_features(void)
 
 		if (edx & bit_SSE2)
 			features |= SSE2;
+
+		if (edx & bit_F16C)
+			features |= F16C;
 	}
 
 	if (max >= 7) {
@@ -174,6 +181,8 @@ char *igt_x86_features_to_string(unsigned features, char *line)
 		line += sprintf(line, ", avx");
 	if (features & AVX2)
 		line += sprintf(line, ", avx2");
+	if (features & F16C)
+		line += sprintf(line, ", f16c");
 
 	(void)line;
 
diff --git a/lib/igt_x86.h b/lib/igt_x86.h
index d4f8c34..c7b84de 100644
--- a/lib/igt_x86.h
+++ b/lib/igt_x86.h
@@ -39,6 +39,7 @@
 #define SSE4_2	0x40
 #define AVX	0x80
 #define AVX2	0x100
+#define F16C	0x200
 
 #if defined(__x86_64__) || defined(__i386__)
 unsigned igt_x86_features(void);
diff --git a/lib/meson.build b/lib/meson.build
index 0eb5585..edf0a66 100644
--- a/lib/meson.build
+++ b/lib/meson.build
@@ -11,6 +11,7 @@ lib_sources = [
 	'igt_aux.c',
 	'igt_gt.c',
 	'igt_gvt.c',
+	'igt_halffloat.c',
 	'igt_matrix.c',
 	'igt_primes.c',
 	'igt_rand.c',
-- 
2.7.4

_______________________________________________
igt-dev mailing list
igt-dev@lists.freedesktop.org
https://lists.freedesktop.org/mailman/listinfo/igt-dev

Re: [igt-dev] [RFC 1/5] lib: Add halffloat implementation

[Maarten Lankhorst]

Op 03-03-2019 om 02:39 schreef Kevin Strasser:
> Probe for and make an API available for tests to use f16c intrinsics to
> generate needed fp16 pixel data.
>
> Also import a pure c fp32 <-> fp16 conversion implementation from Mesa
> 18.3.4, which will act as a fallback when f16c is unavailable.
>
> Signed-off-by: Kevin Strasser <kevin.strasser@intel.com>
> ---
>  COPYING              |  22 ++++++
>  lib/Makefile.sources |   2 +
>  lib/igt_halffloat.c  | 218 +++++++++++++++++++++++++++++++++++++++++++++++++++
>  lib/igt_halffloat.h  |  27 +++++++
>  lib/igt_x86.c        |   9 +++
>  lib/igt_x86.h        |   1 +
>  lib/meson.build      |   1 +
>  7 files changed, 280 insertions(+)
>  create mode 100644 lib/igt_halffloat.c
>  create mode 100644 lib/igt_halffloat.h
>
> diff --git a/COPYING b/COPYING
> index 29e0238..0fe6584 100644
> --- a/COPYING
> +++ b/COPYING
> @@ -126,3 +126,25 @@ SPECIAL, DIRECT, INDIRECT, OR CONSEQUENTIAL DAMAGES OR ANY DAMAGES
>  WHATSOEVER RESULTING FROM LOSS OF USE, DATA OR PROFITS, WHETHER IN AN
>  ACTION OF CONTRACT, NEGLIGENCE OR OTHER TORTIOUS ACTION, ARISING OUT OF
>  OR IN CONNECTION WITH THE USE OR PERFORMANCE OF THIS SOFTWARE.
> +
> +Copyright (C) 1999-2007  Brian Paul   All Rights Reserved.
> +Copyright 2015 Philip Taylor <philip@zaynar.co.uk>
> +Copyright 2018 Advanced Micro Devices, Inc.
> +
> +Permission is hereby granted, free of charge, to any person obtaining a
> +copy of this software and associated documentation files (the "Software"),
> +to deal in the Software without restriction, including without limitation
> +the rights to use, copy, modify, merge, publish, distribute, sublicense,
> +and/or sell copies of the Software, and to permit persons to whom the
> +Software is furnished to do so, subject to the following conditions:
> +
> +The above copyright notice and this permission notice shall be included
> +in all copies or substantial portions of the Software.
> +
> +THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS
> +OR IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
> +FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL
> +THE AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR
> +OTHER LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE,
> +ARISING FROM, OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR
> +OTHER DEALINGS IN THE SOFTWARE.
> diff --git a/lib/Makefile.sources b/lib/Makefile.sources
> index cf27209..89f6ce5 100644
> --- a/lib/Makefile.sources
> +++ b/lib/Makefile.sources
> @@ -30,6 +30,8 @@ lib_source_list =	 	\
>  	igt_gt.h		\
>  	igt_gvt.c		\
>  	igt_gvt.h		\
> +	igt_halffloat.c		\
> +	igt_halffloat.h		\
>  	igt_matrix.c		\
>  	igt_matrix.h		\
>  	igt_primes.c		\
> diff --git a/lib/igt_halffloat.c b/lib/igt_halffloat.c
> new file mode 100644
> index 0000000..e2ab00a
> --- /dev/null
> +++ b/lib/igt_halffloat.c
> @@ -0,0 +1,218 @@
> +/*
> + * Copyright (C) 1999-2007  Brian Paul   All Rights Reserved.
> + * Copyright 2015 Philip Taylor <philip@zaynar.co.uk>
> + * Copyright 2018 Advanced Micro Devices, Inc.
> + *
> + * Permission is hereby granted, free of charge, to any person obtaining a
> + * copy of this software and associated documentation files (the "Software"),
> + * to deal in the Software without restriction, including without limitation
> + * the rights to use, copy, modify, merge, publish, distribute, sublicense,
> + * and/or sell copies of the Software, and to permit persons to whom the
> + * Software is furnished to do so, subject to the following conditions:
> + *
> + * The above copyright notice and this permission notice shall be included
> + * in all copies or substantial portions of the Software.
> + *
> + * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS
> + * OR IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
> + * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL
> + * THE AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR
> + * OTHER LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE,
> + * ARISING FROM, OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR
> + * OTHER DEALINGS IN THE SOFTWARE.
> + */
> +
> +#include <assert.h>
> +#include <math.h>
> +
> +#include "igt_halffloat.h"
> +#include "igt_x86.h"
> +
> +typedef union { float f; int32_t i; uint32_t u; } fi_type;
> +
> +/**
> + * Convert a 4-byte float to a 2-byte half float.
> + *
> + * Not all float32 values can be represented exactly as a float16 value. We
> + * round such intermediate float32 values to the nearest float16. When the
> + * float32 lies exactly between to float16 values, we round to the one with
> + * an even mantissa.
> + *
> + * This rounding behavior has several benefits:
> + *   - It has no sign bias.
> + *
> + *   - It reproduces the behavior of real hardware: opcode F32TO16 in Intel's
> + *     GPU ISA.
> + *
> + *   - By reproducing the behavior of the GPU (at least on Intel hardware),
> + *     compile-time evaluation of constant packHalf2x16 GLSL expressions will
> + *     result in the same value as if the expression were executed on the GPU.
> + */
> +static uint16_t float_to_half(float val)
> +{
> +	const fi_type fi = {val};
> +	const int flt_m = fi.i & 0x7fffff;
> +	const int flt_e = (fi.i >> 23) & 0xff;
> +	const int flt_s = (fi.i >> 31) & 0x1;
> +	int s, e, m = 0;
> +	uint16_t result;
> +
> +	/* sign bit */
> +	s = flt_s;
> +
> +	/* handle special cases */
> +	if ((flt_e == 0) && (flt_m == 0)) {
> +		/* zero */
> +		/* m = 0; - already set */
> +		e = 0;
> +	} else if ((flt_e == 0) && (flt_m != 0)) {
> +		/* denorm -- denorm float maps to 0 half */
> +		/* m = 0; - already set */
> +		e = 0;
> +	} else if ((flt_e == 0xff) && (flt_m == 0)) {
> +		/* infinity */
> +		/* m = 0; - already set */
> +		e = 31;
> +	} else if ((flt_e == 0xff) && (flt_m != 0)) {
> +		/* NaN */
> +		m = 1;
> +		e = 31;
> +	} else {
> +		/* regular number */
> +		const int new_exp = flt_e - 127;
> +		if (new_exp < -14) {
> +			/* The float32 lies in the range (0.0, min_normal16) and
> +			 * is rounded to a nearby float16 value. The result will
> +			 * be either zero, subnormal, or normal.
> +			 */
> +			e = 0;
> +			m = lrintf((1 << 24) * fabsf(fi.f));
> +		} else if (new_exp > 15) {
> +			/* map this value to infinity */
> +			/* m = 0; - already set */
> +			e = 31;
> +		} else {
> +			/* The float32 lies in the range
> +			 *   [min_normal16, max_normal16 + max_step16)
> +			 * and is rounded to a nearby float16 value. The result
> +			 * will be either normal or infinite.
> +			 */
> +			e = new_exp + 15;
> +			m = lrintf(flt_m / (float)(1 << 13));
> +		}
> +	}
> +
> +	assert(0 <= m && m <= 1024);
> +	if (m == 1024) {
> +		/* The float32 was rounded upwards into the range of the next
> +		 * exponent, so bump the exponent. This correctly handles the
> +		 * case where f32 should be rounded up to float16 infinity.
> +		 */
> +		++e;
> +		m = 0;
> +	}
> +
> +	result = (s << 15) | (e << 10) | m;
> +	return result;
> +}
> +
> +/**
> + * Convert a 2-byte half float to a 4-byte float.
> + * Based on code from:
> + * http://www.opengl.org/discussion_boards/ubb/Forum3/HTML/008786.html
> + */
> +static float half_to_float(uint16_t val)
> +{
> +	/* XXX could also use a 64K-entry lookup table */
> +	const int m = val & 0x3ff;
> +	const int e = (val >> 10) & 0x1f;
> +	const int s = (val >> 15) & 0x1;
> +	int flt_m, flt_e, flt_s;
> +	fi_type fi;
> +
> +	/* sign bit */
> +	flt_s = s;
> +
> +	/* handle special cases */
> +	if ((e == 0) && (m == 0)) {
> +		/* zero */
> +		flt_m = 0;
> +		flt_e = 0;
> +	} else if ((e == 0) && (m != 0)) {
> +		/* denorm -- denorm half will fit in non-denorm single */
> +		const float half_denorm = 1.0f / 16384.0f; /* 2^-14 */
> +		float mantissa = ((float) (m)) / 1024.0f;
> +		float sign = s ? -1.0f : 1.0f;
> +		return sign * mantissa * half_denorm;
> +	} else if ((e == 31) && (m == 0)) {
> +		/* infinity */
> +		flt_e = 0xff;
> +		flt_m = 0;
> +	} else if ((e == 31) && (m != 0)) {
> +		/* NaN */
> +		flt_e = 0xff;
> +		flt_m = 1;
> +	} else {
> +		/* regular */
> +		flt_e = e + 112;
> +		flt_m = m << 13;
> +	}
> +
> +	fi.i = (flt_s << 31) | (flt_e << 23) | flt_m;
> +	return fi.f;
> +}
> +
> +#if defined(__x86_64__) && !defined(__clang__)
> +#pragma GCC push_options
> +#pragma GCC target("f16c")
> +
> +#include <immintrin.h>
> +
> +static uint16_t float_to_half_f16c(float f)
> +{
> +	return _cvtss_sh(f, 0);
> +}
> +
> +static float half_to_float_f16c(uint16_t h)
> +{
> +	return _cvtsh_ss(h);
> +}
> +
> +#pragma GCC pop_options
> +
> +static uint16_t (*resolve_float_to_half(void))(float f)
> +{
> +	if (igt_x86_features() & F16C)
> +		return float_to_half_f16c;
> +
> +	return float_to_half;
> +}
> +
> +uint16_t igt_float_to_half(float f)
> +	__attribute__((ifunc("resolve_float_to_half")));
> +
> +static float (*resolve_half_to_float(void))(uint16_t h)
> +{
> +	if (igt_x86_features() & F16C)
> +		return half_to_float_f16c;
> +
> +	return half_to_float;
> +}
> +
> +float igt_half_to_float(uint16_t h)
> +	__attribute__((ifunc("resolve_half_to_float")));
> +
> +#else
> +
> +igt_half igt_float_to_half(float f)
> +{
> +	return float_to_half(f);
> +}
> +
> +float igt_half_to_float(uint16_t h)
> +{
> +	return half_to_float(h);
> +}
> +
> +#endif
> +
> diff --git a/lib/igt_halffloat.h b/lib/igt_halffloat.h
> new file mode 100644
> index 0000000..2ae189e
> --- /dev/null
> +++ b/lib/igt_halffloat.h
> @@ -0,0 +1,27 @@
> +/*
> + * Copyright (C) 1999-2007  Brian Paul   All Rights Reserved.
> + *
> + * Permission is hereby granted, free of charge, to any person obtaining a
> + * copy of this software and associated documentation files (the "Software"),
> + * to deal in the Software without restriction, including without limitation
> + * the rights to use, copy, modify, merge, publish, distribute, sublicense,
> + * and/or sell copies of the Software, and to permit persons to whom the
> + * Software is furnished to do so, subject to the following conditions:
> + *
> + * The above copyright notice and this permission notice shall be included
> + * in all copies or substantial portions of the Software.
> + *
> + * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS
> + * OR IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
> + * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL
> + * THE AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR
> + * OTHER LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE,
> + * ARISING FROM, OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR
> + * OTHER DEALINGS IN THE SOFTWARE.
> + */
> +
> +#include <stdint.h>
> +
> +uint16_t igt_float_to_half(float f);
> +float igt_half_to_float(uint16_t h);
> +
> diff --git a/lib/igt_x86.c b/lib/igt_x86.c
> index 88e514d..9d28e6b 100644
> --- a/lib/igt_x86.c
> +++ b/lib/igt_x86.c
> @@ -88,6 +88,10 @@
>  #define bit_AVX		(1 << 28)
>  #endif
>  
> +#ifndef bit_F16C
> +#define bit_F16C	(1 << 29)
> +#endif
> +
>  #ifndef bit_AVX2
>  #define bit_AVX2	(1<<5)
>  #endif
> @@ -138,6 +142,9 @@ unsigned igt_x86_features(void)
>  
>  		if (edx & bit_SSE2)
>  			features |= SSE2;
> +
> +		if (edx & bit_F16C)
> +			features |= F16C;
>  	}
>  
>  	if (max >= 7) {
> @@ -174,6 +181,8 @@ char *igt_x86_features_to_string(unsigned features, char *line)
>  		line += sprintf(line, ", avx");
>  	if (features & AVX2)
>  		line += sprintf(line, ", avx2");
> +	if (features & F16C)
> +		line += sprintf(line, ", f16c");
>  
>  	(void)line;
>  
> diff --git a/lib/igt_x86.h b/lib/igt_x86.h
> index d4f8c34..c7b84de 100644
> --- a/lib/igt_x86.h
> +++ b/lib/igt_x86.h
> @@ -39,6 +39,7 @@
>  #define SSE4_2	0x40
>  #define AVX	0x80
>  #define AVX2	0x100
> +#define F16C	0x200
>  
>  #if defined(__x86_64__) || defined(__i386__)
>  unsigned igt_x86_features(void);
> diff --git a/lib/meson.build b/lib/meson.build
> index 0eb5585..edf0a66 100644
> --- a/lib/meson.build
> +++ b/lib/meson.build
> @@ -11,6 +11,7 @@ lib_sources = [
>  	'igt_aux.c',
>  	'igt_gt.c',
>  	'igt_gvt.c',
> +	'igt_halffloat.c',
>  	'igt_matrix.c',
>  	'igt_primes.c',
>  	'igt_rand.c',

4 function calls for every pixel is a bit much. Is there any cleaner way we could accomplish the same?

void igt_float_to_half(const float *, uint16_t *, int num)

void igt_half_to_float(const uint16_t *, float *, int num)

If no swizzle is required, it could do an entire line in 1 go, else per 4 or 3.

Might be worth it to still do a line in 1 go, even for xrgb, with format always set to rba128f, hardware will ignore the X component anyway. :)

And then if required reswizzle.

~Maarten

_______________________________________________
igt-dev mailing list
igt-dev@lists.freedesktop.org
https://lists.freedesktop.org/mailman/listinfo/igt-dev

Re: [igt-dev] [RFC 1/5] lib: Add halffloat implementation

[Strasser, Kevin]

Maarten Lankhorst wrote:
> 4 function calls for every pixel is a bit much. Is there any cleaner way we
> could accomplish the same?
> 
> void igt_float_to_half(const float *, uint16_t *, int num)
> 
> void igt_half_to_float(const uint16_t *, float *, int num)
> 
> If no swizzle is required, it could do an entire line in 1 go, else per 4 or 3.
> 
> Might be worth it to still do a line in 1 go, even for xrgb, with format
> always set to rba128f, hardware will ignore the X component anyway. :)
> 
> And then if required reswizzle.

Interesting ideas, I agree we should be able to shave off some overhead with the
function calls. I'm not so sure the payoff would be worth the extra conversions
in the xrgb case, but I'll do some experimentation and fix accordingly.

Thanks,
Kevin
_______________________________________________
igt-dev mailing list
igt-dev@lists.freedesktop.org
https://lists.freedesktop.org/mailman/listinfo/igt-dev

Re: [igt-dev] [RFC 1/5] lib: Add halffloat implementation

[Maarten Lankhorst]

Op 04-03-2019 om 18:03 schreef Strasser, Kevin:
> Maarten Lankhorst wrote:
>> 4 function calls for every pixel is a bit much. Is there any cleaner way we
>> could accomplish the same?
>>
>> void igt_float_to_half(const float *, uint16_t *, int num)
>>
>> void igt_half_to_float(const uint16_t *, float *, int num)
>>
>> If no swizzle is required, it could do an entire line in 1 go, else per 4 or 3.
>>
>> Might be worth it to still do a line in 1 go, even for xrgb, with format
>> always set to rba128f, hardware will ignore the X component anyway. :)
>>
>> And then if required reswizzle.
> Interesting ideas, I agree we should be able to shave off some overhead with the
> function calls. I'm not so sure the payoff would be worth the extra conversions
> in the xrgb case, but I'll do some experimentation and fix accordingly.
>
> Thanks,
> Kevin

Just treat xrgb as argb, alpha channel should be ignored by hw. :)

_______________________________________________
igt-dev mailing list
igt-dev@lists.freedesktop.org
https://lists.freedesktop.org/mailman/listinfo/igt-dev

[igt-dev] [RFC 2/5] include: Add fp16 format defines

[Kevin Strasser]

Adds fp16 formats needed for testing.

NOTE: This should not be merged until defines have actually landed in the
drm-next uapi headers.

Signed-off-by: Kevin Strasser <kevin.strasser@intel.com>
---
 include/drm-uapi/drm_fourcc.h | 10 ++++++++++
 1 file changed, 10 insertions(+)

diff --git a/include/drm-uapi/drm_fourcc.h b/include/drm-uapi/drm_fourcc.h
index bab2029..6302070 100644
--- a/include/drm-uapi/drm_fourcc.h
+++ b/include/drm-uapi/drm_fourcc.h
@@ -144,6 +144,16 @@ extern "C" {
 #define DRM_FORMAT_RGBA1010102	fourcc_code('R', 'A', '3', '0') /* [31:0] R:G:B:A 10:10:10:2 little endian */
 #define DRM_FORMAT_BGRA1010102	fourcc_code('B', 'A', '3', '0') /* [31:0] B:G:R:A 10:10:10:2 little endian */
 
+/* Floating point 64 bpp RGB
+ * IEEE 754-2008 binary16 half-precision float
+ * [15:0] sign:exponent:mantissa 1:5:10
+ */
+#define DRM_FORMAT_XRGB16161616F fourcc_code('X', 'R', '4', 'H') /* [63:0] x:R:G:B 16:16:16:16 little endian */
+#define DRM_FORMAT_XBGR16161616F fourcc_code('X', 'B', '4', 'H') /* [63:0] x:B:G:R 16:16:16:16 little endian */
+
+#define DRM_FORMAT_ARGB16161616F fourcc_code('A', 'R', '4', 'H') /* [63:0] A:R:G:B 16:16:16:16 little endian */
+#define DRM_FORMAT_ABGR16161616F fourcc_code('A', 'B', '4', 'H') /* [63:0] A:B:G:R 16:16:16:16 little endian */
+
 /* packed YCbCr */
 #define DRM_FORMAT_YUYV		fourcc_code('Y', 'U', 'Y', 'V') /* [31:0] Cr0:Y1:Cb0:Y0 8:8:8:8 little endian */
 #define DRM_FORMAT_YVYU		fourcc_code('Y', 'V', 'Y', 'U') /* [31:0] Cb0:Y1:Cr0:Y0 8:8:8:8 little endian */
-- 
2.7.4

_______________________________________________
igt-dev mailing list
igt-dev@lists.freedesktop.org
https://lists.freedesktop.org/mailman/listinfo/igt-dev

[igt-dev] [RFC 3/5] tests/kms_available_modes_crc: Add cases for 64bit formats

[Kevin Strasser]

This test iterates through all supported formats, add cases for new 64 bit
formats so it continues passing.

Signed-off-by: Kevin Strasser <kevin.strasser@intel.com>
---
 tests/kms_available_modes_crc.c | 19 ++++++++++++++++++-
 1 file changed, 18 insertions(+), 1 deletion(-)

diff --git a/tests/kms_available_modes_crc.c b/tests/kms_available_modes_crc.c
index 7ff385f..fd20f28 100644
--- a/tests/kms_available_modes_crc.c
+++ b/tests/kms_available_modes_crc.c
@@ -124,10 +124,11 @@ static const struct {
 	enum		{ BYTES_PP_1=1,
 				BYTES_PP_2=2,
 				BYTES_PP_4=4,
+				BYTES_PP_8=8,
 				NV12,
 				P010,
 				SKIP4 } bpp;
-	uint32_t	value;
+	uint64_t	value;
 } fillers[] = {
 	{ DRM_FORMAT_C8, 0, BYTES_PP_1, 0xff},
 	{ DRM_FORMAT_RGB565, 0, BYTES_PP_2, 0xffff},
@@ -146,6 +147,11 @@ static const struct {
 	{ DRM_FORMAT_XRGB2101010, 0, BYTES_PP_4, 0xffffffff},
 	{ DRM_FORMAT_XBGR2101010, 0, BYTES_PP_4, 0xffffffff},
 
+	{ DRM_FORMAT_XRGB16161616F, 0, BYTES_PP_8, 0x3c003c003c003c00},
+	{ DRM_FORMAT_XBGR16161616F, 0, BYTES_PP_8, 0x3c003c003c003c00},
+	{ DRM_FORMAT_ARGB16161616F, 0, BYTES_PP_8, 0x3c003c003c003c00},
+	{ DRM_FORMAT_ABGR16161616F, 0, BYTES_PP_8, 0x3c003c003c003c00},
+
 	{ DRM_FORMAT_YUYV, 0, BYTES_PP_4, 0x80eb80eb},
 	{ DRM_FORMAT_YVYU, 0, BYTES_PP_4, 0x80eb80eb},
 	{ DRM_FORMAT_VYUY, 0, BYTES_PP_4, 0xeb80eb80},
@@ -177,6 +183,7 @@ static bool fill_in_fb(data_t *data, igt_output_t *output, igt_plane_t *plane,
 	signed i, c, writesize;
 	unsigned short* ptemp_16_buf;
 	unsigned int* ptemp_32_buf;
+	unsigned long int* ptemp_64_buf;
 
 	for( i = 0; fillers[i].fourcc != 0; i++ ) {
 		if( fillers[i].fourcc == format )
@@ -184,6 +191,12 @@ static bool fill_in_fb(data_t *data, igt_output_t *output, igt_plane_t *plane,
 	}
 
 	switch (fillers[i].bpp) {
+	case BYTES_PP_8:
+		ptemp_64_buf = (unsigned long int*)data->buf;
+		for (c = 0; c < data->size/8; c++)
+			ptemp_64_buf[c] = fillers[i].value;
+		writesize = data->size;
+		break;
 	case BYTES_PP_4:
 		ptemp_32_buf = (unsigned int*)data->buf;
 		for (c = 0; c < data->size/4; c++)
@@ -291,6 +304,10 @@ static bool setup_fb(data_t *data, igt_output_t *output, igt_plane_t *plane,
 	case BYTES_PP_4:
 		bpp = 32;
 		break;
+
+	case BYTES_PP_8:
+		bpp = 64;
+		break;
 	}
 
 	igt_get_fb_tile_size(data->gfx_fd, tiling, bpp,
-- 
2.7.4

_______________________________________________
igt-dev mailing list
igt-dev@lists.freedesktop.org
https://lists.freedesktop.org/mailman/listinfo/igt-dev

Re: [igt-dev] [RFC 3/5] tests/kms_available_modes_crc: Add cases for 64bit formats

[Maarten Lankhorst]

Op 03-03-2019 om 02:39 schreef Kevin Strasser:
> This test iterates through all supported formats, add cases for new 64 bit
> formats so it continues passing.
>
> Signed-off-by: Kevin Strasser <kevin.strasser@intel.com>
> ---
>  tests/kms_available_modes_crc.c | 19 ++++++++++++++++++-
>  1 file changed, 18 insertions(+), 1 deletion(-)
>
> diff --git a/tests/kms_available_modes_crc.c b/tests/kms_available_modes_crc.c
> index 7ff385f..fd20f28 100644
> --- a/tests/kms_available_modes_crc.c
> +++ b/tests/kms_available_modes_crc.c
> @@ -124,10 +124,11 @@ static const struct {
>  	enum		{ BYTES_PP_1=1,
>  				BYTES_PP_2=2,
>  				BYTES_PP_4=4,
> +				BYTES_PP_8=8,
>  				NV12,
>  				P010,
>  				SKIP4 } bpp;
> -	uint32_t	value;
> +	uint64_t	value;
>  } fillers[] = {
>  	{ DRM_FORMAT_C8, 0, BYTES_PP_1, 0xff},
>  	{ DRM_FORMAT_RGB565, 0, BYTES_PP_2, 0xffff},
> @@ -146,6 +147,11 @@ static const struct {
>  	{ DRM_FORMAT_XRGB2101010, 0, BYTES_PP_4, 0xffffffff},
>  	{ DRM_FORMAT_XBGR2101010, 0, BYTES_PP_4, 0xffffffff},
>  
> +	{ DRM_FORMAT_XRGB16161616F, 0, BYTES_PP_8, 0x3c003c003c003c00},
> +	{ DRM_FORMAT_XBGR16161616F, 0, BYTES_PP_8, 0x3c003c003c003c00},
> +	{ DRM_FORMAT_ARGB16161616F, 0, BYTES_PP_8, 0x3c003c003c003c00},
> +	{ DRM_FORMAT_ABGR16161616F, 0, BYTES_PP_8, 0x3c003c003c003c00},
> +
>  	{ DRM_FORMAT_YUYV, 0, BYTES_PP_4, 0x80eb80eb},
>  	{ DRM_FORMAT_YVYU, 0, BYTES_PP_4, 0x80eb80eb},
>  	{ DRM_FORMAT_VYUY, 0, BYTES_PP_4, 0xeb80eb80},
> @@ -177,6 +183,7 @@ static bool fill_in_fb(data_t *data, igt_output_t *output, igt_plane_t *plane,
>  	signed i, c, writesize;
>  	unsigned short* ptemp_16_buf;
>  	unsigned int* ptemp_32_buf;
> +	unsigned long int* ptemp_64_buf;
>  
>  	for( i = 0; fillers[i].fourcc != 0; i++ ) {
>  		if( fillers[i].fourcc == format )
> @@ -184,6 +191,12 @@ static bool fill_in_fb(data_t *data, igt_output_t *output, igt_plane_t *plane,
>  	}
>  
>  	switch (fillers[i].bpp) {
> +	case BYTES_PP_8:
> +		ptemp_64_buf = (unsigned long int*)data->buf;
Use uint64_t ? Otherwise looks good.
> +		for (c = 0; c < data->size/8; c++)
> +			ptemp_64_buf[c] = fillers[i].value;
> +		writesize = data->size;
> +		break;
>  	case BYTES_PP_4:
>  		ptemp_32_buf = (unsigned int*)data->buf;
>  		for (c = 0; c < data->size/4; c++)
> @@ -291,6 +304,10 @@ static bool setup_fb(data_t *data, igt_output_t *output, igt_plane_t *plane,
>  	case BYTES_PP_4:
>  		bpp = 32;
>  		break;
> +
> +	case BYTES_PP_8:
> +		bpp = 64;
> +		break;
>  	}
>  
>  	igt_get_fb_tile_size(data->gfx_fd, tiling, bpp,


_______________________________________________
igt-dev mailing list
igt-dev@lists.freedesktop.org
https://lists.freedesktop.org/mailman/listinfo/igt-dev

[igt-dev] [RFC 4/5] lib/igt_fb: Add support for fp16 formats through conversion

[Kevin Strasser]

Follow design of P01x conversion to support tests needing pixel data in fp16
(half float 64 bpp).

Signed-off-by: Kevin Strasser <kevin.strasser@intel.com>
---
 lib/igt_fb.c | 188 ++++++++++++++++++++++++++++++++++++++++++++++++++++++++---
 lib/igt_fb.h |   1 +
 2 files changed, 180 insertions(+), 9 deletions(-)

diff --git a/lib/igt_fb.c b/lib/igt_fb.c
index 9dca2a4..e441201 100644
--- a/lib/igt_fb.c
+++ b/lib/igt_fb.c
@@ -35,6 +35,7 @@
 #include "igt_aux.h"
 #include "igt_color_encoding.h"
 #include "igt_fb.h"
+#include "igt_halffloat.h"
 #include "igt_kms.h"
 #include "igt_matrix.h"
 #include "igt_vc4.h"
@@ -161,6 +162,22 @@ static const struct format_desc_struct {
 	  .num_planes = 1, .plane_bpp = { 32, },
 	  .hsub = 1, .vsub = 1,
 	},
+	{ .name = "XRGB16161616F", .depth = -1, .drm_id = DRM_FORMAT_XRGB16161616F,
+	  .cairo_id = CAIRO_FORMAT_RGB96F,
+	  .num_planes = 1, .plane_bpp = { 64, },
+	},
+	{ .name = "ARGB16161616F", .depth = -1, .drm_id = DRM_FORMAT_ARGB16161616F,
+	  .cairo_id = CAIRO_FORMAT_RGBA128F,
+	  .num_planes = 1, .plane_bpp = { 64, },
+	},
+	{ .name = "XBGR16161616F", .depth = -1, .drm_id = DRM_FORMAT_XBGR16161616F,
+	  .cairo_id = CAIRO_FORMAT_RGB96F,
+	  .num_planes = 1, .plane_bpp = { 64, },
+	},
+	{ .name = "ABGR16161616F", .depth = -1, .drm_id = DRM_FORMAT_ABGR16161616F,
+	  .cairo_id = CAIRO_FORMAT_RGBA128F,
+	  .num_planes = 1, .plane_bpp = { 64, },
+	},
 	{ .name = "NV12", .depth = -1, .drm_id = DRM_FORMAT_NV12,
 	  .cairo_id = CAIRO_FORMAT_RGB24,
 	  .num_planes = 2, .plane_bpp = { 8, 16, },
@@ -661,7 +678,8 @@ static int create_bo_for_fb(struct igt_fb *fb)
 	 * them, so we need to make sure to use a device BO then.
 	 */
 	if (fb->tiling || fb->size || fb->strides[0] ||
-	    (is_i915_device(fd) && igt_format_is_yuv(fb->drm_format)))
+	    (is_i915_device(fd) && igt_format_is_yuv(fb->drm_format)) ||
+	    (is_i915_device(fd) && igt_format_is_fp16(fb->drm_format)))
 		device_bo = true;
 
 	/* Sets offets and stride if necessary. */
@@ -2030,19 +2048,35 @@ static void convert_rgb24_to_yuv(struct fb_convert *cvt)
 	}
 }
 
-static void read_rgbf(struct igt_vec4 *rgb, const float *rgb24)
+static void read_rgbf(struct igt_vec4 *rgb, const float *rgbf)
 {
-	rgb->d[0] = rgb24[0];
-	rgb->d[1] = rgb24[1];
-	rgb->d[2] = rgb24[2];
+	rgb->d[0] = rgbf[0];
+	rgb->d[1] = rgbf[1];
+	rgb->d[2] = rgbf[2];
 	rgb->d[3] = 1.0f;
 }
 
-static void write_rgbf(float *rgb24, const struct igt_vec4 *rgb)
+static void read_rgbaf(struct igt_vec4 *rgba, const float *rgbaf)
+{
+	rgba->d[0] = rgbaf[0];
+	rgba->d[1] = rgbaf[1];
+	rgba->d[2] = rgbaf[2];
+	rgba->d[3] = rgbaf[3];
+}
+
+static void write_rgbf(float *rgbf, const struct igt_vec4 *rgb)
 {
-	rgb24[0] = rgb->d[0];
-	rgb24[1] = rgb->d[1];
-	rgb24[2] = rgb->d[2];
+	rgbf[0] = rgb->d[0];
+	rgbf[1] = rgb->d[1];
+	rgbf[2] = rgb->d[2];
+}
+
+static void write_rgbaf(float *rgbaf, const struct igt_vec4 *rgba)
+{
+	rgbaf[0] = rgba->d[0];
+	rgbaf[1] = rgba->d[1];
+	rgbaf[2] = rgba->d[2];
+	rgbaf[3] = rgba->d[2];
 }
 
 static void convert_yuv16_to_float(struct fb_convert *cvt)
@@ -2202,6 +2236,106 @@ static void convert_float_to_yuv16(struct fb_convert *cvt)
 	}
 }
 
+/* { R, G, B, X } */
+static const unsigned char swizzle_rgbx[] = { 0, 1, 2, 3 };
+static const unsigned char swizzle_bgrx[] = { 2, 1, 0, 3 };
+
+static const unsigned char *rgbx_swizzle(uint32_t format)
+{
+	switch (format) {
+	default:
+	case DRM_FORMAT_XRGB16161616F:
+	case DRM_FORMAT_ARGB16161616F:
+		return swizzle_bgrx;
+	case DRM_FORMAT_XBGR16161616F:
+	case DRM_FORMAT_ABGR16161616F:
+		return swizzle_rgbx;
+	}
+}
+
+static void convert_fp16_to_float(struct fb_convert *cvt, bool alpha)
+{
+	int i, j;
+	uint16_t *fp16;
+	float *ptr = cvt->dst.ptr;
+	unsigned int float_stride = cvt->dst.fb->strides[0] / sizeof(*ptr);
+	unsigned int fp16_stride = cvt->src.fb->strides[0] / sizeof(*fp16);
+	const unsigned char *swz = rgbx_swizzle(cvt->src.fb->drm_format);
+	uint8_t fpp = alpha ? 4 : 3;
+
+	uint16_t *buf = convert_src_get(cvt);
+	fp16 = buf + cvt->src.fb->offsets[0] / sizeof(*buf);
+
+	for (i = 0; i < cvt->dst.fb->height; i++) {
+		const uint16_t *fp16_tmp = fp16;
+		float *rgb_tmp = ptr;
+
+		for (j = 0; j < cvt->dst.fb->width; j++) {
+			struct igt_vec4 rgb;
+
+			rgb.d[0] = igt_half_to_float(fp16_tmp[swz[0]]);
+			rgb.d[1] = igt_half_to_float(fp16_tmp[swz[1]]);
+			rgb.d[2] = igt_half_to_float(fp16_tmp[swz[2]]);
+
+			if (alpha) {
+				rgb.d[3] = igt_half_to_float(fp16_tmp[swz[3]]);
+				write_rgbaf(rgb_tmp, &rgb);
+			} else {
+				write_rgbf(rgb_tmp, &rgb);
+			}
+
+			rgb_tmp += fpp;
+			fp16_tmp += 4;
+		}
+
+		ptr += float_stride;
+		fp16 += fp16_stride;
+	}
+
+	convert_src_put(cvt, buf);
+}
+
+static void convert_float_to_fp16(struct fb_convert *cvt, bool alpha)
+{
+	int i, j;
+	uint16_t *fp16 = cvt->dst.ptr + cvt->dst.fb->offsets[0];
+	const float *ptr = cvt->src.ptr;
+	unsigned float_stride = cvt->src.fb->strides[0] / sizeof(*ptr);
+	unsigned fp16_stride = cvt->dst.fb->strides[0] / sizeof(*fp16);
+	const unsigned char *swz = rgbx_swizzle(cvt->src.fb->drm_format);
+	uint8_t fpp = alpha ? 4 : 3;
+
+	for (i = 0; i < cvt->dst.fb->height; i++) {
+		const float *rgb_tmp = ptr;
+		uint16_t *fp16_tmp = fp16;
+
+		for (j = 0; j < cvt->dst.fb->width; j++) {
+			struct igt_vec4 rgb;
+
+			if (alpha)
+				read_rgbaf(&rgb, rgb_tmp);
+			else
+				read_rgbf(&rgb, rgb_tmp);
+
+			rgb_tmp += fpp;
+
+			fp16_tmp[swz[0]] = igt_float_to_half(rgb.d[0]);
+			fp16_tmp[swz[1]] = igt_float_to_half(rgb.d[1]);
+			fp16_tmp[swz[2]] = igt_float_to_half(rgb.d[2]);
+
+			if (alpha)
+				fp16_tmp[swz[3]] = igt_float_to_half(rgb.d[3]);
+			else
+				fp16_tmp[swz[3]] = 0;
+
+			fp16_tmp += 4;
+		}
+
+		ptr += float_stride;
+		fp16 += fp16_stride;
+	}
+}
+
 static void convert_pixman(struct fb_convert *cvt)
 {
 	pixman_format_code_t src_pixman = drm_format_to_pixman(cvt->src.fb->drm_format);
@@ -2290,6 +2424,14 @@ static void fb_convert(struct fb_convert *cvt)
 		case DRM_FORMAT_P016:
 			convert_yuv16_to_float(cvt);
 			return;
+		case DRM_FORMAT_XRGB16161616F:
+		case DRM_FORMAT_XBGR16161616F:
+			convert_float_to_fp16(cvt, false);
+			return;
+		case DRM_FORMAT_ARGB16161616F:
+		case DRM_FORMAT_ABGR16161616F:
+			convert_fp16_to_float(cvt, true);
+			return;
 		}
 	} else if (cvt->src.fb->drm_format == IGT_FORMAT_FLOAT) {
 		switch (cvt->dst.fb->drm_format) {
@@ -2298,6 +2440,14 @@ static void fb_convert(struct fb_convert *cvt)
 		case DRM_FORMAT_P016:
 			convert_float_to_yuv16(cvt);
 			return;
+		case DRM_FORMAT_XRGB16161616F:
+		case DRM_FORMAT_XBGR16161616F:
+			convert_float_to_fp16(cvt, false);
+			return;
+		case DRM_FORMAT_ARGB16161616F:
+		case DRM_FORMAT_ABGR16161616F:
+			convert_float_to_fp16(cvt, true);
+			return;
 		}
 	}
 
@@ -2453,6 +2603,7 @@ cairo_surface_t *igt_get_cairo_surface(int fd, struct igt_fb *fb)
 
 	if (fb->cairo_surface == NULL) {
 		if (igt_format_is_yuv(fb->drm_format) ||
+		    igt_format_is_fp16(fb->drm_format) ||
 		    ((f->cairo_id == CAIRO_FORMAT_INVALID) &&
 		     (f->pixman_id != PIXMAN_invalid)))
 			create_cairo_surface__convert(fd, fb);
@@ -2762,6 +2913,25 @@ bool igt_format_is_yuv(uint32_t drm_format)
 }
 
 /**
+ * igt_format_is_fp16
+ * @drm_format: drm fourcc
+ *
+ * Check if the format is fp16.
+ */
+bool igt_format_is_fp16(uint32_t drm_format)
+{
+	switch (drm_format) {
+	case DRM_FORMAT_XRGB16161616F:
+	case DRM_FORMAT_ARGB16161616F:
+	case DRM_FORMAT_XBGR16161616F:
+	case DRM_FORMAT_ABGR16161616F:
+		return true;
+	default:
+		return false;
+	}
+}
+
+/**
  * igt_format_plane_bpp:
  * @drm_format: drm fourcc
  * @plane: format plane index
diff --git a/lib/igt_fb.h b/lib/igt_fb.h
index e1d885e..07b6814 100644
--- a/lib/igt_fb.h
+++ b/lib/igt_fb.h
@@ -182,6 +182,7 @@ uint32_t igt_drm_format_to_bpp(uint32_t drm_format);
 const char *igt_format_str(uint32_t drm_format);
 bool igt_fb_supported_format(uint32_t drm_format);
 bool igt_format_is_yuv(uint32_t drm_format);
+bool igt_format_is_fp16(uint32_t drm_format);
 int igt_format_plane_bpp(uint32_t drm_format, int plane);
 void igt_format_array_fill(uint32_t **formats_array, unsigned int *count,
 			   bool allow_yuv);
-- 
2.7.4

_______________________________________________
igt-dev mailing list
igt-dev@lists.freedesktop.org
https://lists.freedesktop.org/mailman/listinfo/igt-dev

[igt-dev] [RFC 5/5] tests/kms_plane_scaling: Skip testing unsupported fp16 features

[Kevin Strasser]

Disallow Yf tiling and Y-tiled 90/270 rotation for fp16 on Intel hardware.

Signed-off-by: Kevin Strasser <kevin.strasser@intel.com>
---
 tests/kms_plane_scaling.c | 14 ++++++++++++++
 1 file changed, 14 insertions(+)

diff --git a/tests/kms_plane_scaling.c b/tests/kms_plane_scaling.c
index 1015c80..fed6b3d 100644
--- a/tests/kms_plane_scaling.c
+++ b/tests/kms_plane_scaling.c
@@ -193,6 +193,14 @@ static void test_scaler_with_rotation_pipe(data_t *d, enum pipe pipe,
 			igt_rotation_t rot = rotations[i];
 			for (int j = 0; j < plane->drm_plane->count_formats; j++) {
 				unsigned format = plane->drm_plane->formats[j];
+
+				// Y-tiled 90/270 rotation isn't supported for fp16 on Intel
+				if (is_i915_device(d->drm_fd) &&
+				    igt_format_is_fp16(format) &&
+				    (rot == IGT_ROTATION_90 ||
+				     rot == IGT_ROTATION_270))
+					continue;
+
 				if (igt_fb_supported_format(format) &&
 				    can_rotate(d, format))
 					check_scaling_pipe_plane_rot(d, plane, format,
@@ -227,6 +235,12 @@ static void test_scaler_with_pixel_format_pipe(data_t *d, enum pipe pipe, igt_ou
 			for (int j = 0; j < plane->drm_plane->count_formats; j++) {
 				uint32_t format = plane->drm_plane->formats[j];
 
+				// Yf Tiling not supported for fp16 on Intel
+				if (is_i915_device(d->drm_fd) &&
+				    igt_format_is_fp16(format) &&
+				    tiling == LOCAL_I915_FORMAT_MOD_Yf_TILED)
+					continue;
+
 				if (igt_fb_supported_format(format))
 					check_scaling_pipe_plane_rot(d, plane,
 								     format, tiling,
-- 
2.7.4

_______________________________________________
igt-dev mailing list
igt-dev@lists.freedesktop.org
https://lists.freedesktop.org/mailman/listinfo/igt-dev

Re: [igt-dev] [RFC 5/5] tests/kms_plane_scaling: Skip testing unsupported fp16 features

[Maarten Lankhorst]

Op 03-03-2019 om 02:39 schreef Kevin Strasser:
> Disallow Yf tiling and Y-tiled 90/270 rotation for fp16 on Intel hardware.
>
> Signed-off-by: Kevin Strasser <kevin.strasser@intel.com>
> ---
>  tests/kms_plane_scaling.c | 14 ++++++++++++++
>  1 file changed, 14 insertions(+)
>
> diff --git a/tests/kms_plane_scaling.c b/tests/kms_plane_scaling.c
> index 1015c80..fed6b3d 100644
> --- a/tests/kms_plane_scaling.c
> +++ b/tests/kms_plane_scaling.c
> @@ -193,6 +193,14 @@ static void test_scaler_with_rotation_pipe(data_t *d, enum pipe pipe,
>  			igt_rotation_t rot = rotations[i];
>  			for (int j = 0; j < plane->drm_plane->count_formats; j++) {
>  				unsigned format = plane->drm_plane->formats[j];
> +
> +				// Y-tiled 90/270 rotation isn't supported for fp16 on Intel
> +				if (is_i915_device(d->drm_fd) &&
> +				    igt_format_is_fp16(format) &&
> +				    (rot == IGT_ROTATION_90 ||
> +				     rot == IGT_ROTATION_270))
> +					continue;
> +
>  				if (igt_fb_supported_format(format) &&
>  				    can_rotate(d, format))
^Use this, and perhaps add the rotation as argument?
>  					check_scaling_pipe_plane_rot(d, plane, format,
> @@ -227,6 +235,12 @@ static void test_scaler_with_pixel_format_pipe(data_t *d, enum pipe pipe, igt_ou
>  			for (int j = 0; j < plane->drm_plane->count_formats; j++) {
>  				uint32_t format = plane->drm_plane->formats[j];
>  
> +				// Yf Tiling not supported for fp16 on Intel
> +				if (is_i915_device(d->drm_fd) &&
> +				    igt_format_is_fp16(format) &&
> +				    tiling == LOCAL_I915_FORMAT_MOD_Yf_TILED)
> +					continue;

This should be testable with igt_plane_has_format_mod() instead?



_______________________________________________
igt-dev mailing list
igt-dev@lists.freedesktop.org
https://lists.freedesktop.org/mailman/listinfo/igt-dev

Re: [igt-dev] [RFC 5/5] tests/kms_plane_scaling: Skip testing unsupported fp16 features

[Strasser, Kevin]

Maarten Lankhorst wrote:
> > diff --git a/tests/kms_plane_scaling.c b/tests/kms_plane_scaling.c 
> > index 1015c80..fed6b3d 100644
> > --- a/tests/kms_plane_scaling.c
> > +++ b/tests/kms_plane_scaling.c
> > @@ -193,6 +193,14 @@ static void test_scaler_with_rotation_pipe(data_t *d, enum pipe pipe,
> >  			igt_rotation_t rot = rotations[i];
> >  			for (int j = 0; j < plane->drm_plane->count_formats; j++) {
> >  				unsigned format = plane->drm_plane->formats[j];
> > +
> > +				// Y-tiled 90/270 rotation isn't supported for fp16 on Intel
> > +				if (is_i915_device(d->drm_fd) &&
> > +				    igt_format_is_fp16(format) &&
> > +				    (rot == IGT_ROTATION_90 ||
> > +				     rot == IGT_ROTATION_270))
> > +					continue;
> > +
> >  				if (igt_fb_supported_format(format) &&
> >  				    can_rotate(d, format))
> ^Use this, and perhaps add the rotation as argument?

Makes sense.

> > @@ static void test_scaler_with_pixel_format_pipe(data_t *d, enum pipe pipe, igt_ou
> >  			for (int j = 0; j < plane->drm_plane->count_formats; j++) {
> >  				uint32_t format = plane->drm_plane->formats[j];
> >  
> > +				// Yf Tiling not supported for fp16 on Intel
> > +				if (is_i915_device(d->drm_fd) &&
> > +				    igt_format_is_fp16(format) &&
> > +				    tiling == LOCAL_I915_FORMAT_MOD_Yf_TILED)
> > +					continue;
> 
> This should be testable with igt_plane_has_format_mod() instead?

Sure, better to take the kernel's lead.

Thanks,
Kevin
_______________________________________________
igt-dev mailing list
igt-dev@lists.freedesktop.org
https://lists.freedesktop.org/mailman/listinfo/igt-dev

[igt-dev] ✓ Fi.CI.BAT: success for Add support for fp16 formats

[Patchwork]

== Series Details ==

Series: Add support for fp16 formats
URL   : https://patchwork.freedesktop.org/series/57473/
State : success

== Summary ==

CI Bug Log - changes from CI_DRM_5686 -> IGTPW_2546
====================================================

Summary
-------

  **SUCCESS**

  No regressions found.

  External URL: https://patchwork.freedesktop.org/api/1.0/series/57473/revisions/1/mbox/

Known issues
------------

  Here are the changes found in IGTPW_2546 that come from known issues:

### IGT changes ###

#### Issues hit ####

  * igt@gem_ctx_create@basic-files:
    - fi-gdg-551:         NOTRUN -> SKIP [fdo#109271] +106

  * igt@i915_pm_rpm@basic-pci-d3-state:
    - fi-bsw-kefka:       PASS -> SKIP [fdo#109271]

  * igt@i915_pm_rpm@basic-rte:
    - fi-bsw-kefka:       PASS -> FAIL [fdo#108800]

  * igt@kms_busy@basic-flip-c:
    - fi-byt-j1900:       NOTRUN -> SKIP [fdo#109271] / [fdo#109278]
    - fi-gdg-551:         NOTRUN -> SKIP [fdo#109271] / [fdo#109278]

  * igt@kms_chamelium@hdmi-crc-fast:
    - fi-byt-j1900:       NOTRUN -> SKIP [fdo#109271] +52

  * igt@kms_pipe_crc_basic@suspend-read-crc-pipe-a:
    - fi-blb-e6850:       PASS -> INCOMPLETE [fdo#107718]

  
#### Possible fixes ####

  * igt@gem_mmap_gtt@basic-write:
    - {fi-icl-y}:         INCOMPLETE -> PASS

  * igt@kms_chamelium@common-hpd-after-suspend:
    - fi-kbl-7567u:       WARN [fdo#109380] -> PASS

  * igt@kms_pipe_crc_basic@nonblocking-crc-pipe-c:
    - fi-kbl-7567u:       SKIP [fdo#109271] -> PASS +33

  * igt@kms_pipe_crc_basic@suspend-read-crc-pipe-c:
    - fi-kbl-7560u:       FAIL [fdo#103375] -> PASS

  
  {name}: This element is suppressed. This means it is ignored when computing
          the status of the difference (SUCCESS, WARNING, or FAILURE).

  [fdo#103375]: https://bugs.freedesktop.org/show_bug.cgi?id=103375
  [fdo#107718]: https://bugs.freedesktop.org/show_bug.cgi?id=107718
  [fdo#108569]: https://bugs.freedesktop.org/show_bug.cgi?id=108569
  [fdo#108800]: https://bugs.freedesktop.org/show_bug.cgi?id=108800
  [fdo#109271]: https://bugs.freedesktop.org/show_bug.cgi?id=109271
  [fdo#109278]: https://bugs.freedesktop.org/show_bug.cgi?id=109278
  [fdo#109284]: https://bugs.freedesktop.org/show_bug.cgi?id=109284
  [fdo#109285]: https://bugs.freedesktop.org/show_bug.cgi?id=109285
  [fdo#109294]: https://bugs.freedesktop.org/show_bug.cgi?id=109294
  [fdo#109315]: https://bugs.freedesktop.org/show_bug.cgi?id=109315
  [fdo#109380]: https://bugs.freedesktop.org/show_bug.cgi?id=109380


Participating hosts (41 -> 39)
------------------------------

  Additional (2): fi-byt-j1900 fi-gdg-551 
  Missing    (4): fi-ilk-m540 fi-byt-squawks fi-bsw-cyan fi-hsw-4200u 


Build changes
-------------

    * IGT: IGT_4868 -> IGTPW_2546

  CI_DRM_5686: 054474a91167b9aa71a85b97e6c70250f05c114c @ git://anongit.freedesktop.org/gfx-ci/linux
  IGTPW_2546: https://intel-gfx-ci.01.org/tree/drm-tip/IGTPW_2546/
  IGT_4868: 27027cf078e5e8c4ced3b7d941890659e4adf1cd @ git://anongit.freedesktop.org/xorg/app/intel-gpu-tools

== Logs ==

For more details see: https://intel-gfx-ci.01.org/tree/drm-tip/IGTPW_2546/
_______________________________________________
igt-dev mailing list
igt-dev@lists.freedesktop.org
https://lists.freedesktop.org/mailman/listinfo/igt-dev

[igt-dev] ✓ Fi.CI.IGT: success for Add support for fp16 formats

[Patchwork]

== Series Details ==

Series: Add support for fp16 formats
URL   : https://patchwork.freedesktop.org/series/57473/
State : success

== Summary ==

CI Bug Log - changes from CI_DRM_5686_full -> IGTPW_2546_full
====================================================

Summary
-------

  **SUCCESS**

  No regressions found.

  External URL: https://patchwork.freedesktop.org/api/1.0/series/57473/revisions/1/mbox/

Known issues
------------

  Here are the changes found in IGTPW_2546_full that come from known issues:

### IGT changes ###

#### Issues hit ####

  * igt@gem_ctx_param@invalid-param-get:
    - shard-snb:          NOTRUN -> FAIL [fdo#109559]

  * igt@gem_eio@unwedge-stress:
    - shard-snb:          PASS -> FAIL [fdo#109661]

  * igt@i915_pm_rpm@modeset-pc8-residency-stress:
    - shard-apl:          NOTRUN -> SKIP [fdo#109271] +26

  * igt@kms_atomic_transition@5x-modeset-transitions-nonblocking:
    - shard-glk:          NOTRUN -> SKIP [fdo#109271] / [fdo#109278]

  * igt@kms_busy@extended-modeset-hang-newfb-render-b:
    - shard-hsw:          PASS -> DMESG-WARN [fdo#107956] +1
    - shard-kbl:          PASS -> DMESG-WARN [fdo#107956]
    - shard-snb:          NOTRUN -> DMESG-WARN [fdo#107956]

  * igt@kms_busy@extended-modeset-hang-newfb-with-reset-render-b:
    - shard-snb:          PASS -> DMESG-WARN [fdo#107956]

  * igt@kms_busy@extended-pageflip-hang-newfb-render-e:
    - shard-kbl:          NOTRUN -> SKIP [fdo#109271] / [fdo#109278]

  * igt@kms_busy@extended-pageflip-hang-oldfb-render-e:
    - shard-apl:          NOTRUN -> SKIP [fdo#109271] / [fdo#109278]

  * igt@kms_color@pipe-a-legacy-gamma:
    - shard-glk:          PASS -> FAIL [fdo#104782] / [fdo#108145]

  * igt@kms_color@pipe-c-degamma:
    - shard-apl:          PASS -> FAIL [fdo#104782]

  * igt@kms_cursor_crc@cursor-128x128-onscreen:
    - shard-kbl:          PASS -> FAIL [fdo#103232]

  * igt@kms_cursor_crc@cursor-128x128-random:
    - shard-apl:          PASS -> FAIL [fdo#103232] +5

  * igt@kms_cursor_crc@cursor-128x128-sliding:
    - shard-apl:          NOTRUN -> FAIL [fdo#103232]

  * igt@kms_cursor_crc@cursor-size-change:
    - shard-glk:          PASS -> FAIL [fdo#103232]

  * igt@kms_flip@2x-flip-vs-panning-vs-hang-interruptible:
    - shard-kbl:          NOTRUN -> SKIP [fdo#109271] +9

  * igt@kms_frontbuffer_tracking@fbc-1p-primscrn-cur-indfb-draw-mmap-cpu:
    - shard-kbl:          PASS -> FAIL [fdo#103167]

  * igt@kms_frontbuffer_tracking@fbc-1p-primscrn-spr-indfb-draw-render:
    - shard-apl:          PASS -> FAIL [fdo#103167] +2

  * igt@kms_frontbuffer_tracking@fbc-1p-rte:
    - shard-apl:          PASS -> FAIL [fdo#103167] / [fdo#105682]

  * igt@kms_frontbuffer_tracking@fbc-2p-primscrn-spr-indfb-draw-mmap-cpu:
    - shard-glk:          PASS -> FAIL [fdo#103167] +5

  * igt@kms_plane@pixel-format-pipe-c-planes:
    - shard-apl:          PASS -> FAIL [fdo#103166] +3

  * igt@kms_plane_multiple@atomic-pipe-a-tiling-x:
    - shard-kbl:          PASS -> FAIL [fdo#103166]

  * igt@kms_plane_multiple@atomic-pipe-b-tiling-none:
    - shard-glk:          PASS -> FAIL [fdo#103166] +7

  * igt@kms_setmode@basic:
    - shard-hsw:          PASS -> FAIL [fdo#99912]

  * igt@kms_vblank@pipe-b-ts-continuation-modeset-rpm:
    - shard-snb:          NOTRUN -> SKIP [fdo#109271] / [fdo#109278] +14

  * igt@kms_vblank@pipe-b-ts-continuation-suspend:
    - shard-kbl:          PASS -> FAIL [fdo#104894]
    - shard-apl:          PASS -> FAIL [fdo#104894]

  * igt@perf_pmu@busy-start-vcs1:
    - shard-snb:          NOTRUN -> SKIP [fdo#109271] +165

  * igt@perf_pmu@multi-client-vcs1:
    - shard-glk:          NOTRUN -> SKIP [fdo#109271] +7

  
#### Possible fixes ####

  * igt@gem_tiled_swapping@non-threaded:
    - shard-glk:          SKIP [fdo#109271] -> PASS
    - shard-kbl:          SKIP [fdo#109271] -> PASS
    - shard-apl:          SKIP [fdo#109271] -> PASS

  * igt@i915_suspend@forcewake:
    - shard-kbl:          INCOMPLETE [fdo#103665] -> PASS +1

  * igt@kms_ccs@pipe-a-crc-sprite-planes-basic:
    - shard-glk:          FAIL [fdo#108145] -> PASS

  * igt@kms_ccs@pipe-b-crc-sprite-planes-basic:
    - shard-apl:          FAIL [fdo#106510] / [fdo#108145] -> PASS

  * igt@kms_cursor_crc@cursor-128x128-suspend:
    - shard-apl:          FAIL [fdo#103191] / [fdo#103232] -> PASS +1

  * igt@kms_cursor_crc@cursor-128x42-onscreen:
    - shard-kbl:          FAIL [fdo#103232] -> PASS +1

  * igt@kms_cursor_crc@cursor-64x21-random:
    - shard-apl:          FAIL [fdo#103232] -> PASS +4

  * igt@kms_cursor_crc@cursor-alpha-opaque:
    - shard-glk:          FAIL [fdo#109350] -> PASS

  * igt@kms_cursor_legacy@2x-long-cursor-vs-flip-atomic:
    - shard-hsw:          FAIL [fdo#105767] -> PASS

  * igt@kms_cursor_legacy@cursor-vs-flip-legacy:
    - shard-hsw:          FAIL [fdo#103355] -> PASS

  * igt@kms_frontbuffer_tracking@fbc-1p-primscrn-spr-indfb-move:
    - shard-glk:          FAIL [fdo#103167] -> PASS +2

  * igt@kms_frontbuffer_tracking@fbc-1p-primscrn-spr-indfb-onoff:
    - shard-apl:          FAIL [fdo#103167] -> PASS +1

  * igt@kms_frontbuffer_tracking@fbc-1p-rte:
    - shard-glk:          FAIL [fdo#103167] / [fdo#105682] -> PASS

  * igt@kms_plane_multiple@atomic-pipe-a-tiling-y:
    - shard-glk:          FAIL [fdo#103166] -> PASS +2

  * igt@kms_plane_multiple@atomic-pipe-a-tiling-yf:
    - shard-kbl:          FAIL [fdo#103166] -> PASS +1

  * igt@kms_plane_multiple@atomic-pipe-c-tiling-yf:
    - shard-apl:          FAIL [fdo#103166] -> PASS +3

  * igt@kms_rotation_crc@multiplane-rotation-cropping-top:
    - shard-kbl:          FAIL [fdo#109016] -> PASS +1

  * igt@kms_vblank@pipe-a-ts-continuation-modeset-hang:
    - shard-kbl:          FAIL [fdo#104894] -> PASS +1

  * igt@kms_vblank@pipe-a-ts-continuation-modeset-rpm:
    - shard-apl:          FAIL [fdo#104894] -> PASS +2

  
  {name}: This element is suppressed. This means it is ignored when computing
          the status of the difference (SUCCESS, WARNING, or FAILURE).

  [fdo#103166]: https://bugs.freedesktop.org/show_bug.cgi?id=103166
  [fdo#103167]: https://bugs.freedesktop.org/show_bug.cgi?id=103167
  [fdo#103191]: https://bugs.freedesktop.org/show_bug.cgi?id=103191
  [fdo#103232]: https://bugs.freedesktop.org/show_bug.cgi?id=103232
  [fdo#103355]: https://bugs.freedesktop.org/show_bug.cgi?id=103355
  [fdo#103665]: https://bugs.freedesktop.org/show_bug.cgi?id=103665
  [fdo#104782]: https://bugs.freedesktop.org/show_bug.cgi?id=104782
  [fdo#104894]: https://bugs.freedesktop.org/show_bug.cgi?id=104894
  [fdo#105682]: https://bugs.freedesktop.org/show_bug.cgi?id=105682
  [fdo#105767]: https://bugs.freedesktop.org/show_bug.cgi?id=105767
  [fdo#106510]: https://bugs.freedesktop.org/show_bug.cgi?id=106510
  [fdo#107956]: https://bugs.freedesktop.org/show_bug.cgi?id=107956
  [fdo#108145]: https://bugs.freedesktop.org/show_bug.cgi?id=108145
  [fdo#109016]: https://bugs.freedesktop.org/show_bug.cgi?id=109016
  [fdo#109271]: https://bugs.freedesktop.org/show_bug.cgi?id=109271
  [fdo#109278]: https://bugs.freedesktop.org/show_bug.cgi?id=109278
  [fdo#109350]: https://bugs.freedesktop.org/show_bug.cgi?id=109350
  [fdo#109559]: https://bugs.freedesktop.org/show_bug.cgi?id=109559
  [fdo#109661]: https://bugs.freedesktop.org/show_bug.cgi?id=109661
  [fdo#99912]: https://bugs.freedesktop.org/show_bug.cgi?id=99912


Participating hosts (7 -> 5)
------------------------------

  Missing    (2): shard-skl shard-iclb 


Build changes
-------------

    * IGT: IGT_4868 -> IGTPW_2546
    * Piglit: piglit_4509 -> None

  CI_DRM_5686: 054474a91167b9aa71a85b97e6c70250f05c114c @ git://anongit.freedesktop.org/gfx-ci/linux
  IGTPW_2546: https://intel-gfx-ci.01.org/tree/drm-tip/IGTPW_2546/
  IGT_4868: 27027cf078e5e8c4ced3b7d941890659e4adf1cd @ git://anongit.freedesktop.org/xorg/app/intel-gpu-tools
  piglit_4509: fdc5a4ca11124ab8413c7988896eec4c97336694 @ git://anongit.freedesktop.org/piglit

== Logs ==

For more details see: https://intel-gfx-ci.01.org/tree/drm-tip/IGTPW_2546/
_______________________________________________
igt-dev mailing list
igt-dev@lists.freedesktop.org
https://lists.freedesktop.org/mailman/listinfo/igt-dev