Compiler Error saying 'identifier "__builtin_ia32_serialize" is undefined' while running CUDA

-1

I am using GEC Library (GitHub - HareInWeed/gec: elliptic curve cryptography with GPU acceleration) to perform point operations on secp26k1. This project has the provision of using CUDA.

#include <gec/utils/macros.hpp>
#include <gec/bigint.hpp>
#include <gec/curve.hpp>

// nvcc i_hope.cu

using namespace gec::bigint::literal;

using Bigint256 = gec::bigint::ArrayBE<uint64_t, 4>;

GEC_DEF_GLOBAL(MOD, Bigint256,     // cardinality of finite field
    0xfffffffffffffffffffffffffffffffffffffffffffffffffffffffefffffc2f_int);
constexpr Bigint256::LimbT MOD_P = // -MOD^-1 mod 2^64
    0xd838091dd2253531ull;
GEC_DEF_GLOBAL(RR, Bigint256,      // 2^512 mod MOD
    0x01000007a2000e90a1_int);
GEC_DEF_GLOBAL(ONE_R, Bigint256,   // 2^256 mod MOD
    0x1000003d1_int);


using Field = GEC_BASE_FIELD(Bigint256, MOD, MOD_P, RR, ONE_R);

 const Field B(0x700001ab7_int);
__constant__ const Field d_B(0x700001ab7_int);

using Secp256k1   = GEC_CURVE_B(gec::curve::JacobianCurve, Field, B);

 const Secp256k1 Gen(
    Field(0x9981e643e9089f48979f48c033fd129c231e295329bc66dbd7362e5a487e2097_int),
    Field(0xcf3f851fd4a582d670b6b59aac19c1368dfc5d5d1f1dc64db15ea6d2d3dbabe2_int),
    Field(0x1000003d1_int)
);

__constant__ const Secp256k1 d_Gen(
    Field(0x9981e643e9089f48979f48c033fd129c231e295329bc66dbd7362e5a487e2097_int),
    Field(0xcf3f851fd4a582d670b6b59aac19c1368dfc5d5d1f1dc64db15ea6d2d3dbabe2_int),
    Field(0x1000003d1_int)
);


__global__ void cuda_kernel() 
{
    Secp256k1 p1;
    Secp256k1::mul(p1, 1, d_Gen);

    p1.print();
}


int main()
{


 cuda_kernel<<<1,1>>>();
 cudaDeviceSynchronize();
 return 0;
}

I am getting the below error

/usr/lib/gcc/x86_64-linux-gnu/11/include/serializeintrin.h(41): error: identifier "__builtin_ia32_serialize" is undefined
1 error detected in the compilation

CUDA Version

nvcc: NVIDIA (R) Cuda compiler driver
Copyright (c) 2005-2021 NVIDIA Corporation
Built on Thu_Nov_18_09:45:30_PST_2021
Cuda compilation tools, release 11.5, V11.5.119
Build cuda_11.5.r11.5/compiler.30672275_0

I am using Ubuntu 22.04

The Contents of serializeintrin.h located at /usr/lib/gcc/x86_64-linux-gnu/11/include/ is below.

#ifndef _X86GPRINTRIN_H_INCLUDED
# error "Never use <serializeintrin.h> directly; include <x86gprintrin.h> instead."
#endif

#ifndef _SERIALIZE_H_INCLUDED
#define _SERIALIZE_H_INCLUDED

#ifndef __SERIALIZE__
#pragma GCC push_options
#pragma GCC target("serialize")
#define __DISABLE_SERIALIZE__
#endif /* __SERIALIZE__ */

extern __inline void
__attribute__((__gnu_inline__, __always_inline__, __artificial__))
_serialize (void)
{
  __builtin_ia32_serialize ();
}

#ifdef __DISABLE_SERIALIZE__
#undef __DISABLE_SERIALIZE__
#pragma GCC pop_options
#endif /* __DISABLE_SERIALIZE__ */

#endif /* _SERIALIZE_H_INCLUDED.  */

gcc (Ubuntu 11.3.0-1ubuntu1~22.04.1) 11.3.0

Thanks for the response. I am the one who has asked that question on Stack Overflow also… GCC is up to date.

Can you please reshare the solution, the stackoverflow article is removed.

Original SO post:

I am using GEC Library (GitHub - HareInWeed/gec: elliptic curve cryptography with GPU acceleration) to perform point operations on secp26k1.

#include <gec/utils/macros.hpp>
#include <gec/bigint.hpp>
#include <gec/curve.hpp>

// nvcc -O3 i_hope.cu

using namespace gec::bigint::literal;

using Bigint256 = gec::bigint::ArrayBE<uint64_t, 4>;

GEC_DEF_GLOBAL(MOD, Bigint256,     // cardinality of finite field
    0xfffffffffffffffffffffffffffffffffffffffffffffffffffffffefffffc2f_int);
constexpr Bigint256::LimbT MOD_P = // -MOD^-1 mod 2^64
    0xd838091dd2253531ull;
GEC_DEF_GLOBAL(RR, Bigint256,      // 2^512 mod MOD
    0x01000007a2000e90a1_int);
GEC_DEF_GLOBAL(ONE_R, Bigint256,   // 2^256 mod MOD
    0x1000003d1_int);


using Field = GEC_BASE_FIELD(Bigint256, MOD, MOD_P, RR, ONE_R);

 const Field B(0x700001ab7_int);
__constant__ const Field d_B(0x700001ab7_int);

using Secp256k1   = GEC_CURVE_B(gec::curve::JacobianCurve, Field, B);

 const Secp256k1 Gen(
    Field(0x9981e643e9089f48979f48c033fd129c231e295329bc66dbd7362e5a487e2097_int),
    Field(0xcf3f851fd4a582d670b6b59aac19c1368dfc5d5d1f1dc64db15ea6d2d3dbabe2_int),
    Field(0x1000003d1_int)
);

__constant__ const Secp256k1 d_Gen(
    Field(0x9981e643e9089f48979f48c033fd129c231e295329bc66dbd7362e5a487e2097_int),
    Field(0xcf3f851fd4a582d670b6b59aac19c1368dfc5d5d1f1dc64db15ea6d2d3dbabe2_int),
    Field(0x1000003d1_int)
);


__global__ void cuda_kernel() 
{
    Secp256k1 p1;
    Secp256k1::mul(p1, 1, d_Gen);

    p1.print();
}


int main()
{


 cuda_kernel<<<1,1>>>();
 cudaDeviceSynchronize();
 return 0;
}

I am getting the below error

/usr/lib/gcc/x86_64-linux-gnu/11/include/serializeintrin.h(41): error: identifier "__builtin_ia32_serialize" is undefined
1 error detected in the compilation

CUDA Version

nvcc: NVIDIA (R) Cuda compiler driver
Copyright (c) 2005-2021 NVIDIA Corporation
Built on Thu_Nov_18_09:45:30_PST_2021
Cuda compilation tools, release 11.5, V11.5.119
Build cuda_11.5.r11.5/compiler.30672275_0

I am using Ubuntu 22.04

I tried understanding these steps, but I didn’t.

The Contents of serializeintrin.h located at /usr/lib/gcc/x86_64-linux-gnu/11/include/ is below.

#ifndef _X86GPRINTRIN_H_INCLUDED
# error "Never use <serializeintrin.h> directly; include <x86gprintrin.h> instead."
#endif

#ifndef _SERIALIZE_H_INCLUDED
#define _SERIALIZE_H_INCLUDED

#ifndef __SERIALIZE__
#pragma GCC push_options
#pragma GCC target("serialize")
#define __DISABLE_SERIALIZE__
#endif /* __SERIALIZE__ */

extern __inline void
__attribute__((__gnu_inline__, __always_inline__, __artificial__))
_serialize (void)
{
  __builtin_ia32_serialize ();
}

#ifdef __DISABLE_SERIALIZE__
#undef __DISABLE_SERIALIZE__
#pragma GCC pop_options
#endif /* __DISABLE_SERIALIZE__ */

#endif /* _SERIALIZE_H_INCLUDED.  */

gcc (Ubuntu 11.3.0-1ubuntu1~22.04.1) 11.3.0

asked Jun 19 at 8:09

This is a bug in gcc and/or a clash between whatever that code is you are compiling at gcc. It has nothing to do with CUDA

– talonmies

Jun 19 at 10:14

  • @talonmies Your’e right. CUDA is working perfectly for other examples. This error has been reported before as shown in the link. I don’t think it’s the problem with the module. I have updated the file contents of serializeintrin.h along with gcc version. Is there any workaround to this?

– Knm

Jun 19 at 10:34

  • Install a version of gcc with the patch, I guess. I know nothing about the specifics of the bug, you might need to look at the patch history for gcc and then find whether your OS distro has included the patch or not

– talonmies

Jun 19 at 10:43