<html> <head> <meta http-equiv="content-type" content="text/html; charset=iso-8859-15"> </head> <body text="#000000" bgcolor="#FFFFFF"> <tt>It seems that packed structs are completely brok</tt><tt>en in Beignet 1.1.0</tt><tt>, Ivy Bridge code gen. The following kernel</tt><tt> </tt> <blockquote><tt>typedef struct __attribute__((packed)){</tt><tt> </tt><tt> float v;</tt><tt> </tt><tt>} point;</tt><tt> </tt><tt> </tt><tt>__kernel void test(__global point* points){</tt><tt> </tt><tt> __global point* p = points + get_global_id(0);</tt><tt> </tt><tt> p->v = 0.1f * p->v;</tt><tt> </tt><tt>}</tt><tt> </tt></blockquote> <tt>makes the compiler crash with a failed assertion:</tt><tt> </tt> <blockquote><tt>ASSERTION FAILED: elemSize == GEN_BYTE_SCATTER_WORD || elemSize == GEN_BYTE_SCATTER_BYTE</tt><tt> </tt><tt> at file /home/pisto/sorgenti/Beignet-1.1.0-Source/backend/src/backend/gen_insn_selection.cpp, function void gbe::LoadInstructionPattern::emitUnalignedByteGather(gbe::Selection::Opaque&, const gbe::ir::LoadInstruction&, uint32_t, gbe::GenRegister, gbe::ir::BTI) const, line 3561</tt><tt> </tt></blockquote> <tt>Replacing the type of point::v with float2 generates another assertion:</tt><tt> </tt> <blockquote><tt>ASSERTION FAILED: count == 4 || count == 2</tt><tt> </tt><tt> at file /home/pisto/sorgenti/Beignet-1.1.0-Source/backend/src/./backend/gen_register.hpp, function static gbe::GenRegister gbe::GenRegister::splitReg(gbe::GenRegister, uint32_t, uint32_t), line 284</tt><tt> </tt></blockquote> </body> </html>